tags:

views:

421

answers:

2

I'm trying to set my simulation params in constant memory but without luck (CUDA.NET). cudaMemcpyToSymbol function returns cudaErrorInvalidSymbol. The first parameter in cudaMemcpyToSymbol is string... Is it symbol name? actualy I don't understand how it could be resolved. Any help appreciated.

//init, load .cubin   
float[] arr = new float[1];
    arr[0] = 0.0f;
    int size = Marshal.SizeOf(arr[0]) * arr.Length;
    IntPtr ptr = Marshal.AllocHGlobal(size);
    Marshal.Copy(arr, 0, ptr, arr.Length);
    var error = CUDARuntime.cudaMemcpyToSymbol("param", ptr, 4, 0, cudaMemcpyKind.cudaMemcpyHostToDevice);

my .cu file contain

__constant__ float param;

Working solution

     cuda.LoadModule(Path.Combine(Environment.CurrentDirectory, "name.cubin"));            
 simParams = cuda.GetModuleGlobal("params");
 float[] parameters = new float[N]{...}             
 cuda.CopyHostToDevice<float>(simParams, parameters);
A: 

constant memory has implicit local scope linkage. make sure declaration is in the same file where you use it. it sounds like you have two files. may also have to declare param to array (or maybe not)

aaa
+1  A: 

Unfortunately the __ constant __ must be in the same file scope as the memcpy to the symbol, and in your case your __ constant __ is in a separate .cu file.

The simple way around this is to provide a wrapper function in your .cu file, for example:

__constant__ float param;

// Host function to set the constant
void setParam(float value)
{
  cudaMemcpyToSymbol("param", ptr, 4, 0, cudaMemcpyHostToDevice);
}

// etc.
__global__ void ...
Tom
Thanks idea is clear and works for cpp (--compile). But not in case of nvcc myfile.cu --cubin in post build event of .net app. "(-cubin) Compile all .cu/.ptx/.gpu input files to device-only .cubin files. This step discards the host code for each .cu input file." So when I load this module it doesn't include host functions.
Vlad