c - 使用 CUDA-C 生成索引

Question

我正在尝试生成以下索引集:

我有一个由 20 个 block (blockIdx:从 0 -19)组成的 cuda block ，每个 block 又分为 4 个 block (子 block Idx:0、1、2 和 3)。

我正在尝试生成这样的索引模式:

threadIdx(tid)，SubBlockIdxA(SA)，SubBlockIdxB(SB)，BlockIdxA(BA)，BlockIdxB(BB)

            Required                           Obtained         
   tid  SBA SBB BA  BB    SBA  SBB BA  BB
    0   0   1   0   0       0   1   0   0
    1   1   0   0   1       1   0   0   1
    2   0   1   1   1       0   1   1   1
    3   1   0   1   2       1   0   1   2
    4   0   1   2   2       0   1   2   2
    5   1   0   2   3       1   0   2   3
    6   0   1   3   3       0   1   3   3
    7   1   0   3   4       1   0   3   4
    8   2   3   0   0       2   3   0   0
    9   3   2   0   1       3   2   0   1
    10  2   3   1   1       2   3   1   1
    11  3   2   1   2       3   2   1   2
    12  2   3   2   2       2   3   2   2
    13  3   2   2   3       3   2   2   3
    14  2   3   3   3       2   3   3   3
    15  3   2   3   4       3   2   3   4
    16  0   1   5   5       0   1   5   5
    17  1   0   5   6       1   0   5   6
    18  0   1   6   6       0   1   6   6
    19  1   0   6   7       1   0   6   7
    20  0   1   7   7       0   1   7   7
    21  1   0   7   8       1   0   7   8
    22  0   1   8   8       0   1   8   8
    23  1   0   8   9       1   0   8   9
    24  0   1   10  10      2   3   5   5
    25  1   0   10  11      3   2   5   6
    26  0   1   11  11      2   3   6   6
    27  1   0   11  12      3   2   6   7
    28  0   1   12  12      2   3   7   7
    29  1   0   12  13      3   2   7   8
    30  0   1   13  13      2   3   8   8
    31  1   0   13  14      3   2   8   9
    32  2   3   10  10      0   1   10  10
    33  3   2   10  11      1   0   10  11
    34  2   3   11  11      0   1   11  11
    35  3   2   11  12      1   0   11  12
    36  2   3   12  12      0   1   12  12
    37  3   2   12  13      1   0   12  13
    38  2   3   13  13      0   1   13  13
    39  3   2   13  14      1   0   13  14
    40  0   1   15  15      2   3   10  10
    41  1   0   15  16      3   2   10  11
    42  0   1   16  16      2   3   11  11
    43  1   0   16  17      3   2   11  12
    44  0   1   17  17      2   3   12  12
    45  1   0   17  18      3   2   12  13
    46  0   1   18  18      2   3   13  13
    47  1   0   18  19      3   2   13  14

请参阅下面的代码:

static __device__ void function()
    {
        uint16 uBlockIdxA, uBlockIdxB, uSubBlockIdxA, usubBlockIdxB;
        if threadIdx.x < 48)
        {
            uint16 uY = threadIdx.x / 8;
            uint16 uX = threadIdx.x - (uY * 8);    
            uSubBlockIdxA = ((uY & 0x01) << 1)  + (uX & 0x01); 
            uSubBlockIdxB = ((uY & 0x01) << 1)  + ((uX + 1) & 0x01);  
            uBlockIdxB = (uY >> 1) * 5  + ((1 + uX) >> 1);
            uBlockIdxA = (uY >> 1) * 5  + ((0 + uX) >> 1);    
            func (uBlockIdxA, uBlockIdxB, uSubBlockIdxA, uSubBlockIdxB);
        }
    }

我正在尝试思考实现我正在寻找的目标的逻辑。我不对，但不确定我错过了什么。

关于如何生成它的逻辑会很有帮助。代码受到赞赏。请帮忙。

提前致谢。

Answer 1

生成索引的代码如下:

将线程分成两半，因为前半部分的索引可用于后半部分，偏移量为 10 并进行一些细微调整。

static __device__ void function()
{
    uint16 uBlockIdxA, uBlockIdxB, uSubBlockIdxA, usubBlockIdxB;

    if(threadIdx.x < 48)
    {
        uint16 uY = threadIdx.x / 8;
        uint16 uX = threadIdx.x - (uY * 8);

        if (threadIdx.x < 24)
        {

            uSubBlockIdxA = ((uY & 0x01) << 1)  + (uX & 0x01); 
            uSubBlockIdxB = ((uY & 0x01) << 1)  + ((uX + 1) & 0x01);  
            uBlockIdxB = (uY >> 1) * 5  + ((1 + uX) >> 1);
            uBlockIdxA = (uY >> 1) * 5  + ((0 + uX) >> 1);    

        }
        else if
        {
            uSubBlockIdxA = (((uY - 3) & 0x01) << 1)  + (uX & 0x01); 
            uSubBlockIdxB = (((uY - 3) & 0x01) << 1)  + ((uX + 1) & 0x01);  
            uBlockIdxB = ((uY - 3) >> 1) * 5  + ((1 + uX) >> 1) + 10;
            uBlockIdxA = ((uY - 3) >> 1) * 5  + ((0 + uX) >> 1) + 10;
        }
    }
    func (uBlockIdxA, uBlockIdxB, uSubBlockIdxA, uSubBlockIdxB);        
}