CUDA 共享内存在调试时不排他性阻塞

Question

基本上，我很难准确理解这里出了什么问题。

共享内存在调试时似乎没有以块独占方式运行。正常运行代码时，不会打印任何内容。但是，如果我尝试对其进行调试，则会在块之间共享共享内存并到达 print 语句。

这是一个示例，显然这不是非常有用的代码，但它在我的系统上重现了该问题。难道我做错了什么？这是调试器的错误还是预期行为？

__global__ 
void test()
{
    __shared__ int result[1];
    if (blockIdx.x == 0 && blockIdx.y == 0 && blockIdx.z == 0)
        result[0] = 4444;
    else
    {
        if (result[0] == 4444)
            printf("This should never print if shared memory is unique\n");
    }
}

然后启动它：

test<<<dim3(8,8,1), dim3(8,8,1)>>>();

也完全有可能我完全误解了共享内存。

感谢您的帮助。

其他信息： 我使用的是 GTX 460。为项目设置了 Compute_20 和 sm_20。我正在使用 nsight 3.0 preview 在 Visual Studio 2010 中编写代码。

Answer 1

两者之间有一个微妙但重要的区别

共享内存在块之间共享，打印语句为 到达

和

共享内存被连续的块重复使用并且打印语句是 到达

你假设前者，但后者是真实发生的。

除了第一个块之外，您的代码正在从未初始化的内存中读取。这本身就是未定义的行为。 C++（和 CUDA）不保证静态声明的内存在进入或超出范围时设置为任何值。您不能指望result 的值不会为 4444，尤其是当它可能存储在与可能已将其设置为值 4444 的前一个块相同的共享暂存空间中时。

代码的整个前提和这个问题都是有缺陷的，你不应该从你看到的结果中得出任何结论，即未定义的行为是未定义的。