尽管索引在范围内，但 __global__ 写入无效

Question

我有一个设备功能抛出

Invalid __global__ write of size 4 bytes

，我花了几个小时调试但无济于事。

所讨论的函数如下所示（删除了不相关的位）：

__device__ inline int64_t divider() {
    const uint32_t look_up_table[] = { /* 33 const values*/ };

    struct divider_node {
        int64_t diff;
        int32_t left;
        int32_t sign;
    }

    divider_node* div_nodes = new divider_node[33];

    for (size_t i = 0; i < 32; ++i) {
        div_nodes[i].left = (int32_t)look_up_table[i]; // <--- offending line
       
        // rest of loop
    }

    // rest of function

}

我一生都无法弄清楚为什么这是无效的写入？

new

应该为 33 个分隔符节点分配足够的内存，并且循环仅覆盖索引 0 - 31，因此它应该在范围内。

我想知道是否有我不知道的 Cuda 特定考虑因素（或者只是我忽略了一个潜在的非常愚蠢的错误）

如有任何帮助，我们将不胜感激

Answer 1

为了节省寄存器文件空间，可以使用常量存储器。当所有线程在循环中遵循相同的索引模式时，它几乎与寄存器性能相同。

使用过多的寄存器空间会减少每个 SM 飞行中的最大可能块数，并可能减少占用率。

Answer 2

尝试在循环中将

size_t

更改为

int

。我遇到了同样的错误。

size_t

是

long

的别名，可能太大了。