10

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.

Ade Miller
  • 13,575
  • 1
  • 42
  • 75

3 Answers3

12

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.

Adi Inbar
  • 12,097
  • 13
  • 56
  • 69
CygnusX1
  • 20,968
  • 5
  • 65
  • 109
4

So... figured it out and thought I'd post it here. The solution is either

Don't use the -G flag on NVCC

or

Compile for arch sm_20 (Fermi) if you are targetting such a device

This is a known limitation of NVCC and not a Thrust bug. See:

http://groups.google.com/group/thrust-users/browse_thread/thread/1914198abf646c6d/8bc00e6030b0030b?#8bc00e6030b0030b

Ade Miller
  • 13,575
  • 1
  • 42
  • 75
1

if you are using mirosoft visual studio: from project->properties->CUDA C/C++->Device->Code Generation; change the compute_10,sm_10 to compute_20,sm_20

Reza
  • 11
  • 1