对于使用CUDA的嵌套循环

我有一些问题需要嵌套循环，我必须从C / C ++转换为CUDA。 基本上我有4个嵌套循环，它们共享相同的数组并进行位移操作。

#define N 65536 // ---------------------------------------------------------------------------------- int a1,a2,a3,a4, i1,i2,i3,i4; int Bit4CBitmapLookUp[16] = {0, 1, 3, 3, 7, 7, 7, 7, 15, 15, 15, 15, 15, 15, 15, 15}; int _cBitmapLookupTable[N]; int s = 0; // index into the cBitmapLookupTable for (i1 = 0; i1 < 16; i1++) { // first customer a1 = Bit4CBitmapLookUp[i1] << 12; for (i2 = 0; i2 < 16; i2++) { // second customer a2 = Bit4CBitmapLookUp[i2] << 8; for (i3 = 0; i3 < 16; i3++) { // third customer a3 = Bit4CBitmapLookUp[i3] << 4; for (i4 = 0;i4 < 16;i4++) { // fourth customer a4 = Bit4CBitmapLookUp[i4]; // now actually set the sBitmapLookupTable value _cBitmapLookupTable[s] = a1 | a2 | a3 | a4; s++; } // for i4 } // for i3 } // for i2 } // for i1 

这是我应该转换为CUDA的代码。 我尝试了不同的方法，但每次输出错误。 在这里，我发布我的CUDA转换版本（来自内核的部分）

 #define N 16 //---------------------------------------------------------------------------------- // index for the GPU int i1 = blockDim.x * blockIdx.x + threadIdx.x; int i2 = blockDim.y * blockIdx.y + threadIdx.y; int i3 = i1; int i4 = i2; __syncthreads(); for(i1 = i2 = 0; i1 < N, i2 < N; i1++, i2++) { // first customer a1 = Bit4CBitmapLookUp_device[i1] << 12; // second customer a2 = Bit4CBitmapLookUp_device[i2] << 8; for(i3 = i4 = 0; i3 < N, i4 < N; i3++, i4++){ // third customer a3 = Bit4CBitmapLookUp_device[i3] << 4; // fourth customer a4 = Bit4CBitmapLookUp_device[i4]; // now actually set the sBitmapLookupTable value _cBitmapLookupTable[s] = a1 | a2 | a3 | a4; s++; } } 

我是CUDA的新手，我还在学习，但实际上我找不到嵌套循环的解决方案。 先感谢您。

Answer 1

至于左下方已经表明初始化存在问题。 我建议您重写程序如下

 int i1 = blockDim.x * blockIdx.x + threadIdx.x; int i2 = blockDim.y * blockIdx.y + threadIdx.y; int i3; int i4; while(i1 < N && i2 < N){ a1 = ..; a2 = ..; for(i3 = i4 = 0; i3 < N, i4 < N; i3++, i4++){ // third customer a3 = Bit4CBitmapLookUp_device[i3] << 4; // fourth customer a4 = Bit4CBitmapLookUp_device[i4]; // now actually set the sBitmapLookupTable value _cBitmapLookupTable[s] = a1 | a2 | a3 | a4; s ++; } s += blockDim.x*gridDim.x*blockDim.y*gridDim.y; i1 += blockDim.x*gridDim.x; i2 += blockDim.y*gridDim.y; } 

我没有测试过，所以我无法保证索引是正确的。 我会留给你的。

更多解释：在上面的代码中，只有i1和i2上的循环是并行化的。 这假设N ** 2与GPU上的核心数相比足够大。 如果不是这样的话。 所有四个循环都需要并行化以获得有效的程序。 这种方法会有所不同。