Resolving Thrust/CUDA warnings "Cannot tell what pointer points to..."

Question

I'm trying to build a trivial application using Thrust/CUDA 4.0 and get lots of warnings "warning : Cannot tell what pointer points to, assuming global memory space"

Has anyone else seen this and how do I either disable them or fix my code?

Thanks,

Ade

Here's my code.

Hello.h

class DECLSPECIFIER Hello   
{ 
private:
    thrust::device_vector<unsigned long> m_device_data;

public:
    Hello(const thrust::host_vector<unsigned long>& data);
    unsigned long Sum();
    unsigned long Max();
};

Hello.cu

#include "Hello.h"

Hello::Hello(const thrust::host_vector<unsigned long>& data)
{
    m_device_data = data;
}

unsigned long Hello::Sum()
{
    return thrust::reduce(m_device_data.cbegin(), m_device_data.cend(), 0, thrust::plus<unsigned long>());
}

unsigned long Hello::Max()
{
    return *thrust::max_element(m_device_data.cbegin(), m_device_data.cend(), thrust::less<unsigned long>());
}

The output

1>  Compiling CUDA source file Hello.cu...
1>  
1>  C:\SrcHg\blog\HelloWorld\HelloWorldCuda>"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v4.0\bin\nvcc.exe" -gencode=arch=compute_10,code=\"sm_10,compute_10\" --use-local-env --cl-version 2008 -ccbin "C:\Program Files (x86)\Microsoft Visual Studio 9.0\VC\bin"  -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v4.0\include"  -G0  --keep-dir "Debug" -maxrregcount=32  --machine 32 --compile  -D_NEXUS_DEBUG -g    -Xcompiler "/EHsc /nologo /Od /Zi  /MDd " -o "Debug\Hello.cu.obj" "C:\SrcHg\blog\HelloWorld\HelloWorldCuda\Hello.cu" 
1>  Hello.cu
1>  tmpxft_00001fac_00000000-0_Hello.cudafe1.gpu
1>  tmpxft_00001fac_00000000-5_Hello.cudafe2.gpu
1>  Hello.cu
1>C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v4.0\include\thrust/detail/internal_functional.h(197): warning : Cannot tell what pointer points to, assuming global memory space
1>C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v4.0\include\thrust/detail/internal_functional.h(197): warning : Cannot tell what pointer points to, assuming global memory space
1>C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v4.0\include\thrust/detail/internal_functional.h(197): warning : Cannot tell what pointer points to, assuming global memory space

There's a lot of these.

Answer 1

Fermi uses uniform addressing of shared and global memory space, while pre-Fermi messages don't.

For the pre-Fermi case, when you get an address, you don't know if it should be shared or global. The compiler tries to figure it out, but sometimes it can't. When that happens, the message pops up - "assuming global" is correct in 99.999% of cases, because when you want a pointer to shared memory, you usually explicitly take an address of a shared variable and the compiler can recognise that.

For Fermi cards, shared-or-global can be deduced at runtime (based on the address) and no assumptions have to be made by the compiler.

Suggeston: Ignore those warnings.