CUDA的嵌套循环

Question

CUDA的嵌套循环

3

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

#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方面是个新手，仍在学习中，但是对于嵌套循环问题，我真的找不到解决方案。谢谢您提前的帮助。

- spaghettifunk

1

提示：您正在将变量i1...i4初始化为从未被使用的值。 - leftaroundabout

请查看以下链接：https://dev59.com/B1XTa4cB1Zd3GeqPzjgChttp://stackoverflow.com/questions/6479715/nested-loops-to-cudahttps://dev59.com/3mHVa4cB1Zd3GeqPjRIF - user1113134

1个回答

网页内容由stack overflow 提供, 点击上面的

可以查看英文原文，
原文链接

- Azrael3000 · Accepted Answer

正如leftaroundabout已经指出的那样，初始化存在问题。我建议您按照以下方式重写程序

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上的核心数量相比足够大。如果不是这种情况，则所有四个循环都需要并行化才能获得高效的程序。方法会有所不同。