使用 CUDA-C 生成索引答案

【问题标题】：Generate Index using CUDA-C使用 CUDA-C 生成索引
【发布时间】：2013-06-27 04:37:20
【问题描述】：

我正在尝试在下面生成一组索引：

我有一个由 20 个块组成的 cuda 块（blockIdx：从 0 -19），每个块细分为 4 个块（子块 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);
        }
    }

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

关于如何生成它的逻辑将很有帮助。代码表示赞赏。请帮忙。

提前致谢。

【问题讨论】：

如果您已经知道内核将读取的每个可能的线程索引值的 4 个子索引的值，为什么需要算法来计算它们？您不能将它们放在查找表中并直接读取它们吗？ .看起来每个线程一个 char4 就足够了
这需要这样做，因为我上面只有一个 cuda 块的表示。我正在处理许多这样的 cuda 块，其中块 idx 发生变化并且查找表将是巨大的。因此，查找表方法不是一个好主意，因此想使用算法方法。
但是无论您运行多少块或线程，您显示的代码只能计算这 48 个索引。
感谢您的回复，但不幸的是必须这样做（这是一个限制）。如果可能的话，任何帮助将不胜感激。
耶 .... 找到了出路。

标签： c cuda gpgpu nvidia

【解决方案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);        
}

【讨论】：