使用CUDA在本地内存中的数组上定义变量大小

是否有可能在设备函数中创建列表,数组,列表/数组的大小,以及调用中的参数…或者在调用时初始化的全局变量?

我想像这些列表中的一个工作:

unsigned int size1; __device__ void function(int size2) { int list1[size1]; int list2[size2]; } 

是否有可能做一些聪明的事情来做这样的工作?

有一种方法可以分配动态数量的共享内存 – 使用第三个启动内核参数:

 __global__ void kernel (int * arr) { extern __shared__ int buf []; // size is not stated // copy data to shared mem: buf[threadIdx.x] = arr[blockIdx.x * blockDim.x + threadIdx.x]; // . . . } // . . . // launch kernel, set size of shared mem in bytes (k elements in buf): kernel<<>> (arr); 

许多arrays都有一个hack:

 __device__ void function(int * a, int * b, int k) // k elements in first list { extern __shared__ int list1 []; extern __shared__ int list2 []; // list2 points to the same point as list1 does list1 [threadIdx.x] = a[blockIdx.x * blockDim.x + threadIdx.x]; list2 [k + threadIdx.x] = b[blockIdx.x * blockDim.x + threadIdx.x]; // . . . } 

您必须考虑:分配给所有块的内存。

如果您知道可以期望的大小值,请考虑使用C ++模板。 与boost预处理器一起,您可以轻松生成多个实例/入口点。

您可以做的另一件事是动态分配共享内存并手动分配指针。 显然,如果您需要超过共享内存的线程专用内存,这可能不起作用

如果您想看示例,我可以为您提供链接

当然有可能!

请查看项目的源代码: http : //code.google.com/p/cuda-grayscale/

从main()调用此函数,并根据gpu_image的宽度和高度对gpu_image执行灰度转换: cuda_grayscale(gpu_image,width,height,grid,block);

如果你挖一点,你会在kernel_gpu.cu中找到实现:

 __global__ void grayscale(float4* imagem, int width, int height) { const int i = blockIdx.x * (blockDim.x * blockDim.y) + blockDim.x * threadIdx.y + threadIdx.x; if (i < width * height) { float v = 0.3 * imagem[i].x + 0.6 * imagem[i].y + 0.1 * imagem[i].z; imagem[i] = make_float4(v, v, v, 0); } }