Declaring what pointer points to Advisory: Cannot tell what pointer point

Is there any way declaring what a pointer points to?

I’d like something like;

shared float * device foo;

but it seems to accomplish nothing. I still get warnings I don’t need;
Advisory: Cannot tell what pointer points to, assuming global memory space

I guess this can be a real problem if I need declaring a pointer to shared data.

– Kuisma

I had the same problem and only found a way to shut things up (in the nvcc documentation)
I got the warning when I had a float *pointer; and let it point to one of my global input parameters, dependend on which was my blockIdx.x.

I cannot think of a situation where it would be beneficial to have such a pointer point to shared memory. Whenever I wanted to sometimes use a different name for shared memory I solved it with a #define cool_shared_mem_name sdata

Therez no need for device as you have already said shared

device merely says that it is on the GPU side. Since anyway, your pointer is within your global kernel, whether or not you use device, NVCC knows where it actually resides.

Now, where it points to and where it resides are two different questions.

The CUDA manual tells that therez no way to tell where a pointer points to. The compiler will figure out by itself. Search for “Pointers” in CUDA manual.

I have seen the compiler generating correct code depending on where the pointer points to. It emits an advisary warning but u need not worry about it.

Different things, as you point out yourself.

Well, obviously NVCC does not, hence the warning (“advisory”).

I find no such claims in the manual. Please supply a pointer (pun intended ;-) to the manual backing this up.

Programmers Guide 1.1 page 20;

Pointers in code that is executed on the device are supported as long as the compiler

is able to resolve whether they point to either the shared memory space or the

global memory space, otherwise they are restricted to only point to memory

allocated or declared in the global memory space.

This brings me back to my original question; how do I tell the compiler how to resolve the pointer reference?

A compiler not possible to write code not generating warnings, is a flawed compiler.

If I get a page of warnings to ignore, what purpose are the warnings? On the contrary, they may hide some important message I really wanted to see. All code shall compile without any warnings at all - and no, disabling the warnings is not correct answer. :)

NVIDIA, is this a compiler misfeature?

– Kuisma

I have asked the same question a week ago. You can NOT tell the compiler which type of memory a pointer points to. I’ll post something in the wishlist.

You have quoted the right passage from the manual. That is what I was referring to.

The manual says “as long as the compiler is able to resolve…”. So, the compiler will figure it out for itself depending on what your code does with the pointer.

So, commenting that printf() inside that compiler will make it good, I suppose – humor intended.

The warnings are advisary. There is a possibility that the compiler might have resolved a pointer incorrectly – in a way that you did NOT intend to use. It is good that the compiler emits that. So, you can go to those places and verify that the compiler’s assumptions are really OK. It is a good thing, in fact.

A better thing would be to provide for “scoping” those pointers. It would probably be very nice.

A message can never hide any other message as there is always a scroll bar – ha ha ha… Fun intended.

Hopefully, NVIDIA would plan for scoping pointers and emit a compiler error when the user uses a pointer in a way in-appropriate with its scoping. (sorry for the complex english sentence)

[codebox]extern shared float foo;[/codebox]

Size of the array is defined as the third argument of the launch configuration <<<blocks, threads, arraysize in bytes (not elements)>>>.

[codebox]extern shared float foo;[/codebox]

Size of the array is defined as the third argument of the launch configuration <<<blocks, threads, arraysize in bytes (not elements)>>>.

I think C simply does not have the concept of different address spaces.

I think C simply does not have the concept of different address spaces.