Copy a static array to host in managedCUDA

Question

I've declared a static array in kernel.cu file

__device__ int myStaticArray[5];

I can modify this array from host using

myKernel.SetConstantVariable("myStaticArray", new int[]{1,2,3,4,5});

After a few processing, I want to copy this array to host, how can I do so?

EDIT1: I noticed that the array will be reset every time when I run a new kernel. I cannot use that array to keep intermediate values to be used in the next kernel. Is it possible to keep those values in static ways?

EDIT2: The problem in EDIT1 occur because I load multiple kernels using LoadKernelPTX. The correct way is to load module once, then construct multiple kernels from that module. (as suggested in https://github.com/kunzmi/managedCuda/wiki/CudaKernel) This way, I can have shared static array/variable across multiple kernels.

Answer 1

Either you declare the array as __device__ __constant__ int myStaticArray[5]; and treat it as a constant array (i.e. you don't write to it). Then you can initialize the values from host using myKernel.SetConstantVariable(...).

Or you you keep it as an __device__ int myStaticArray[5];, then you can declare a CudaDeviceVariable using the constructor CudaDeviceVariable(CUmodule module, string name) that gathers the pointer to that static variable. You can then do any copy-to-host or copy-to-device as usual.

Note that names in Cuda get mangled if the variable is not declared as extern "C", so you might have to look up the full mangled name in the PTX-file.