Constant memory problem in CUDA 1.0?

Accelerated Computing CUDA CUDA Programming and Performance
1

I have declared a variable in constant memory with:

__device__ __constant__ float d_A[4]; 

__global__ myKernel()

{

//Do something with d_A

}

I initialize the values of d_A from the host with:

CUDA_SAFE_CALL(cudaMemcpyToSymbol(d_A, h_A, 4*sizeof(float), 0, cudaMemcpyHostToDevice));

But for some reason, the kernel always uses the values of d_A as zero (the initial values). Even in device emulation mode, d_A has always zeros inside the kernel. Something which is even weirder, if I read d_A from the host with cudaMemcpyFromSymbol, the value of d_A are correct (before and also after the execution of the kernel)!!

Is there any bug related to this with CUDA 1.0? I can’t use CUDA 1.1 for some reason in my computer. That’s why I use CUDA 1.0.

2

cudaMemcpyToSymbol() expects name of variable as first parameter, so just do it in this way:

cudaMemcpyToSymbol("d_A", h_A, 4*sizeof(float) );
3

I already tried that but the problem remains. I think I read somewhere that someone was having trouble with constant memory.

4

I am using cudaMemcpyToSymbol() just like you Kubrick except that I am not using device when declaring the constant memory. I am using CUDA 1.1 though.

I have been using this function without the quotations that AndreiB is suggesting, but again that may be a 1.0 vs 1.1 thing.

5

Ok, forget about it. Problem solved.
My makefile is a bit messed up and it didn’t compile what it should.

6

Quotations vs. no quotations is not a 1.0 or 1.1 thing, it’s just a matter of variable scoping, I suppose.
In my programs I don’t want device-side variables to be visible from host, so I prefer to use quotations. It worked just fine for me both in 1.0 and 1.1.

Kubrick
Can you post full source code here?