Troubles with cudaMemcpyToSymbol

Question

I'm trying to copy to constant memory. But I can not because of my misunderstanding of usage of cudaMemcpyToSymbol function. I'm trying to follow this

Here is some code

__device__ __constant__ double var1;
__device__ __constant__ int var2;

int main(){

   //... some code here...

   double var1ToCopy = 10.1;
   int var2ToCopy = 1; 

   void * p1 = &var1ToCopy; 
   void * p2 = &var2ToCopy;

   cudaStatus = cudaMemcpyToSymbol((void*)&var1,p1,sizeof(double),0,cudaMemcpyHostToDevice);
   if (cudaStatus != cudaSuccess){
      return -1;
   }

   cudaStatus = cudaMemcpyToSymbol((void*)&var2,p2,sizeof(int),0,cudaMemcpyHostToDevice);
   if (cudaStatus != cudaSuccess){
      return -1;
   }


   //... and some code here...
}

I know it is a very dumb question, but I have spent several hours googling an answer and did not had any success.

Answer 1

You don't need the ampersand on the symbol name. A symbol is not the same as a pointer or a variable.

Instead of this:

cudaStatus = cudaMemcpyToSymbol((void*)&var1,p1,sizeof(double),0,cudaMemcpyHostToDevice);

Do this:

cudaStatus = cudaMemcpyToSymbol(var1,&var1ToCopy,sizeof(double));

I've also simplified the above call based on the fact that some of the parameters have defaults as indicated in the documentation.

Here's a fully worked example around a modfied version of your code (requires cc2.0+ GPU):

$ cat t626.cu
#include <stdio.h>

__device__ __constant__ double var1;
__device__ __constant__ int var2;

__global__ void kernel(){

  printf("%f\n", var1);
  printf("%d\n", var2);
}


int main(){


   double var1ToCopy = 10.1;
   int var2ToCopy = 1;
   cudaError_t cudaStatus = cudaMemcpyToSymbol(var1,&var1ToCopy,sizeof(double));
   if (cudaStatus != cudaSuccess) {printf("fail1\n"); return 1;}

   cudaStatus = cudaMemcpyToSymbol(var2,&var2ToCopy,sizeof(int));
   if (cudaStatus != cudaSuccess) {printf("fail2\n"); return 1;}
   kernel<<<1,1>>>();
   cudaDeviceSynchronize();
   return 0;

}
$ nvcc -arch=sm_20 -o t626 t626.cu
$ ./t626
10.100000
1
$