Hi all. I am having a problem with the setting of constant memory using the cudaCpymemToSymbol function. The guide is not very clear on right way to go about doing this and so I do not know if I am the one making the mistake, or if it is a possible bug.
The code is as follows:
__constant__ float min;
float tmp = 1.0f;
CUDA_SAFE_CALL( cudaMemcpyToSymbol( min , &tmp , sizeof( float ) , 0 ) );
__constant__ float3 pos;
float3 cPos;
... ( cPos is set to some value )
CUDA_SAFE_CALL( cudaMemcpyToSymbol( pos , &cPos , sizeof( float3 ) , 0 ) );
This code is executed on the host. Getting the constant data values in the kernel, resulted in min always having the value 0, but pos had the correct values. I assumed that it might not be possible to allocate a single float value in constant memory, so I constructed a single struct to hold all the constant data I need ( which is just 32 bytes in size ).
typedef struct
{
int voxelGridSize;
float voxelSize;
float originOffset;
float minPos;
float maxPos;
float3 beamPosition;
} ConstantData;
__constant__ ConstantData __constantData;
ConstantData tmp;
... ( values of tmp are set )
CUDA_SAFE_CALL( cudaMemcpyToSymbol( __constantData , &tmp , sizeof( ConstantData ) ) );
This struct shows the exact same behaviour: the float3 pos value is correctly set to the right values, but al single float values in the struct have bogus values assigned to them.
I am using CUDA version 0.8 on a 8800GTX, 8 processor Xeon E5320 machine running Windows XP SP2.
Some help will be sincerely appreciated! Thanks.
__constant__ float min;
float tmp = 1.0f;
CUDA_SAFE_CALL( cudaMemcpyToSymbol( min , &tmp , sizeof( float ) , 0 ) );
From your system spec I see that you are using the VC compiler. It defines min as a macro that you need to avoid by #define NOMINMAX . The CUDA SDK projects do that in the vcproj settings. Make sure this is not an issue in your project.
For the struct - the VC compiler sometimes has trouble figuring the sizeof structs. Try the following: define the struct rather than declaring a type
Peter
Thanks for the tip. I renamed the variable, but I do not think this was a problem.
I was actually using the Intel compiler v9.0 in VS7, which I forgot to mention. There was obviously something going wrong with the struct alignment. I tried some compiler switches, but it did not help. Switching to the VC++ compiler solved the issue, but this reduces the performance of the needed C host code considerably. I could not find much about Intel compiler support in Windows. Will it be supported sometime in the future, or should I stick with the VC compiler?
Also puzzling is the fact that single floats were not being set correctly in constant memory. Is this not supported? I haven’t tested it yet with VC++.
Thanks for the help.
Setting single floats is supported. The Intel compiler might not understand the alignment statements correctly if they are part of a struct declaration (the float3 declaration has one). Try adding a CUDA align(16) to your struct declaration.
Peter
I added the aligns and all seems to be working perfectly right now using CUDA 0.8 and the Intel C++ compiler v9.0 in VS7.0. Thanks for the help!