0

I have something like either:

  • __constant__ double PNT[ NUMCOORDS ];
  • __device__ double PNT[ NUMCOORDS ];

depending upon some preprocessor selections. I then use this variable:

cudaMemcpyToSymbol("PNT", point, pntSize)

However, sometimes (and I really CAN'T say when which really confuses me) I get the error message:

duplicate global variable looked up by string name

when checking for CUDA errors. I tried replacing "PNT" with PNT and strangely, this works:

cudaMemcpyToSymbol(PNT, point, pntSize)

Shound I use this solution in practice (instead of using a string "PNT")?

M. Tibbits
  • 8,400
  • 8
  • 44
  • 59
tim
  • 9,896
  • 20
  • 81
  • 137
  • I think it's hard to tell what is going on without some more elaborate example, preferably in a form that someone else could reproduce. Do you parse your file several times, with different macro setting? Maybe nvcc does not update some intermediate files, and it ends up having both __constant__ and __device__ variables? – CygnusX1 Jun 20 '11 at 09:06
  • Thanks already; just the short question: Is the code `cudaMemcpyToSymbol(PNT, point, pntSize)` 100% correct? If so, I'll just stick with it because I don't get any errors when using it! – tim Jun 20 '11 at 09:14
  • [CUDA Library documentation](http://developer.download.nvidia.com/compute/cuda/3_1/toolkit/docs/online/group__CUDART__MEMORY_g58acd070e6f9bb82d7aeee3d0b958460.html) suggests that what you try do to is correct. You might need to pass 2 more arguments to that function. However, I would be careful, if nvcc messes up variables, it may actually mess up in this function as well, by copying the data to the incorrect destination. – CygnusX1 Jun 20 '11 at 09:19
  • Okay thanks, I'll look into that and than just verify the data my program returns. – tim Jun 20 '11 at 09:48
  • Did you define the symbol in a header file and include it in two different files ? – Pavan Yalamanchili Jun 21 '11 at 03:15
  • Nope, just ONE source-code file where everything is defined! – tim Jun 21 '11 at 09:59

1 Answers1

3

The underlying problem has nothing to do with cudaMemcpyToSymbol. The error you are seeing is generated by the CUDA runtime when it searches for the symbol you have supplied because it is multiply defined in the context in which your code is running. CUDA runtime versions have been getting progressively better at detecting duplicate definitions (things like __constant__ declarations, textures, __device__ functions).

The solution is to refactor your code so that symbols are only ever defined once within an application. Because CUDA doesn't have a linker, there will be no compilation time errors if you define a symbol in two files. But when the CUDA runtime loads the resulting binary payloads from the final linked application into a context, duplicate symbol conflicts can occur and runtime errors can result.

talonmies
  • 70,661
  • 34
  • 192
  • 269
  • From the description, I assumed that the variable is defined exactly once and its type (constant or device) is depending only on some compilation flag. – CygnusX1 Jun 21 '11 at 09:33