2

我试图用这段代码对一个数组求和，但我被卡住了。我可能需要一些“CUDA for dummies tutorial”，因为我花了很多时间在这样的基本操作上，但我无法让它工作。

以下是我不明白或不确定的事情的清单：

  1. 我应该使用多少块（dimGrid）？我认为应该是N/dimBlock.x/2（N=输入数组的长度），因为在内核开始时，数据被加载并从全局内存的两个“块”添加到共享内存

  2. 在原始代码中有blockSize. 我将其替换为blockDim.x因为我不知道这些变量有何不同。但是当blockSize=时blockDim.x，gridSize = blockDim.x*2*gridDim.x对我来说没有意义 -gridSize将大于 N。在一维数组的上下文中 *Dim.x 和 *Size 有什么区别？

  3. 主要逻辑 - 在内核中，每个块总和 2*dimBlock（块中的线程）个数。当 N = 262144 和 dimBlock = 128 时，内核返回 1024 个部分和数组。然后我再次运行内核，结果 = 4 部分和。最后，在最后一次运行中，返回单个和，因为数组是由单个块处理的。

  4. 我对二进制数组求和。在第一次运行中，我可以uchar4用于输入数据。在第二次和第三次运行中，我将使用int.

请告诉我我错过了什么

谢谢

__global__ void sum_reduction(uchar4* g_idata, int* g_odata, int N) { 

extern __shared__ int s_data[]; 

unsigned int tid = threadIdx.x;
unsigned int i = blockIdx.x*(blockDim.x*2) + tid;
unsigned int gridSize = blockDim.x*2*gridDim.x;

while (i < N) {
    s_data[tid] += g_idata[i].x + g_idata[i+blockDim.x].x +
            g_idata[i].y + g_idata[i+blockDim.x].y +
            g_idata[i].z + g_idata[i+blockDim.x].z +
            g_idata[i].w + g_idata[i+blockDim.x].w;
    i += gridSize;
}
__syncthreads();

if (tid < 64) {
    s_data[tid] += s_data[tid + 64];
}
__syncthreads(); 

if (tid < 32) { 
    volatile int *s_ptr = s_data; 
    s_ptr[tid] += s_ptr[tid + 32];
    s_ptr[tid] += s_ptr[tid + 16];
    s_ptr[tid] += s_ptr[tid + 8]; 
    s_ptr[tid] += s_ptr[tid + 4];
    s_ptr[tid] += s_ptr[tid + 2]; 
    s_ptr[tid] += s_ptr[tid + 1]; 
} 
if (tid == 0) {
    g_odata[blockIdx.x] = s_data[0];
} 
}


main{
...
dim3 dimBlock(128);
dim3 dimGrid(N/dimBlock.x);
sum_reduction<<<dimGrid, dimBlock>>>(in, out, N);
...
}
4

2 回答 2

4

Calling the kernel like this fixes the problem.

dim3 dimBlock(128);
dim3 dimGrid(N/dimBlock.x);
int smemSize = dimBlock.x * sizeof(int);
sum_reduction<<<dimGrid, dimBlock, smemSize>>>(in, out, N);
于 2012-09-12T02:12:44.077 回答
-3

好吧，我认为你需要重新开始。查看 NVIDiA 提供的关于减少的分步流程指南

于 2012-06-20T07:52:03.213 回答