我正在尝试减少CUDA,我确实是一个新手。我目前正在研究NVIDIA的示例代码。
我想我真的不确定如何设置块大小和网格大小,尤其是当我的输入数组大于512 X 512单个块大小时。
512 X 512
这是代码。
template <unsigned int blockSize> __global__ void reduce6(int *g_idata, int *g_odata, unsigned int n) { extern __shared__ int sdata[]; unsigned int tid = threadIdx.x; unsigned int i = blockIdx.x*(blockSize*2) + tid; unsigned int gridSize = blockSize*2*gridDim.x; sdata[tid] = 0; while (i < n) { sdata[tid] += g_idata[i] + g_idata[i+blockSize]; i += gridSize; } __syncthreads(); if (blockSize >= 512) { if (tid < 256) { sdata[tid] += sdata[tid + 256]; } __syncthreads(); } if (blockSize >= 256) { if (tid < 128) { sdata[tid] += sdata[tid + 128]; } __syncthreads(); } if (blockSize >= 128) { if (tid < 64) { sdata[tid] += sdata[tid + 64]; } __syncthreads(); } if (tid < 32) { if (blockSize >= 64) sdata[tid] += sdata[tid + 32]; if (blockSize >= 32) sdata[tid] += sdata[tid + 16]; if (blockSize >= 16) sdata[tid] += sdata[tid + 8]; if (blockSize >= 8) sdata[tid] += sdata[tid + 4]; if (blockSize >= 4) sdata[tid] += sdata[tid + 2]; if (blockSize >= 2) sdata[tid] += sdata[tid + 1]; } if (tid == 0) g_odata[blockIdx.x] = sdata[0]; }
但是,在我看来,g_odata[blockIdx.x]保存了所有块的部分和,并且,如果我想获得最终结果,则需要对数组中的所有项求和g_odata[blockIdx.x]。
g_odata[blockIdx.x]
我想知道:是否有一个内核可以完成整个求和?还是我误会了这里的事情?如果有人可以教育我,我将不胜感激。非常感谢。
为了更好地理解该主题,您可以查看NVIDIA的pdf格式,以图形方式说明您在代码中使用的所有策略。