CODEWITHSUNDEEP
cuda
user1292251
user1292251

Reputation: 1745

cuda error message : invalid device symbol

I'm seeing the message of "invalid device symbol" when I run the cuda binary. There was no error during compilation it. The message is in below.

Cuda error in file 'euler3d.cu' in line 416 : invalid device symbol.

And the related source code is as following.

CUDA_SAFE_CALL( cudaMemcpyToSymbol(ff_variable, h_ff_variable, NVAR*sizeof(float)) );

Is there any fault from source code? Actually this code is from Rodinia v2.1, cfd benchmark program. I'm using cuda version 3.1 and did compile with following options.

nvcc -Xptxas -v -O3 --gpu-architecture=compute_13 --gpu-code=compute_13 euler3d.cu -o euler3d -I$(CUDA_SDK_PATH)/common/inc  -L$(CUDA_SDK_PATH)/lib $(CUTIL_LIB)

The ff_variable related code is here.

#define NDIM 3
#define VAR_MOMENTUM  1
#define VAR_DENSITY_ENERGY (VAR_MOMENTUM+NDIM)
#define NVAR (VAR_DENSITY_ENERGY+1)
__constant__ float ff_variable[NVAR];

Upvotes: 2

Views: 6616

Answers (1)

P Marecki
P Marecki

Reputation: 1138

Code below compiler and runs OK, but reproduces your error on replacement of kk_d with "kk_d" (i.e. it compiles, but reports invalid device symbol in runtime). CUDA reference (v. 4.2) is a little misleading here, as it says first argument should indeed be a const literal. 

#include <cstdio>
#include "XFC_cudaError.cuh"   //my error reporting

__device__ int kk_d;

__global__ void foo() {
  printf("%i ", kk_d);
}

int main() {
  int kk = 10;
  cudaMemcpyToSymbol(kk_d, &kk, 4);
  CUDA_CHK;
  foo<<<1,1>>>();
  CUDA_CHK;
  cudaDeviceSynchronize();
}

You should now be able to modify the code appropriately. Note, that printf will not compile with arch=sm_13; you would need at least sm_20, but this is irrelevant to your issue.

With your edit: Syntax for copying arrays is ... exactly as you have in your code. Specifically adding __constant__ float ff[2]; to global device variables, and float rr[] = {1,2}; to host code allows for a copy via cudaMemcpyToSymbol(ff, rr, 8);, which compiles, and runs OK, even on sm_13.

Maybe the error originates from an earlier code, or from your h_ff_variable?

Upvotes: 2

Related Questions