Allocate constant memory

Question

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);

Answer 1

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)

Answer 2

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 ...

Answer 3

If this question is actual you can use cuModuleGetGlobal and next cudaMemcpy like this:

private bool setValueToSymbol(CUmodule module, string symbol, int value)
{
    CUdeviceptr devPtr = new CUdeviceptr();
    uint lenBytes = 0;
    CUResult result = CUDADriver.cuModuleGetGlobal(ref devPtr, ref lenBytes, module, symbol);
    if (result == CUResult.Success)
    {
        int[] src = new int[] { value };
        cudaError error = CUDARuntime.cudaMemcpy(devPtr, src, lenBytes, cudaMemcpyKind.cudaMemcpyHostToDevice);
        if (error == cudaError.cudaSuccess)
            return true;
        else
            return false;
    }
    else
    {
        return false;
    }
}

where CUmodule module = cuda.LoadModule("MyCode.cubin"); This code works with NVIDIA GPU Computing SDK 3.1 and CUDA.NET 3.0.