c++ - 为什么从 __global__ 函数中引用外部内存会搞砸一切？

Question

我正在用 CUDA 编写一些代码(准确地说是霍夫曼算法，但这与案例完全无关)。我有一个包含两个函数的文件 Paralellel.cu:一个( WriteDictionary )是一个普通函数，第二个( wrtDict )是一个特殊的 CUDA _全局_ 在 CUDA GPU 中运行的函数。以下是这些函数的主体:

//I know body of this function looks kinda not-related 
//   to program main topic, but it's just for tests.
__global__ void wrtDict(Node** nodes, unsigned char* str)
{
    int i = threadIdx.x;

    Node* n = nodes[i];
    char c = n->character;

    str[6 * i] = 1;//c;                         !!!
    str[6 * i + 1] = 2;

    str[6 * i + 2] = 0;
    str[6 * i + 3] = 0;
    str[6 * i + 4] = 0;
    str[6 * i + 5] = 0;
}

我知道这两行似乎毫无意义，因为我不使用 的这个对象 n节点 在这里上课，但让他们待一会儿。还有一条 super secret 的评论，标有“!!!”。这里是 写字典 :

void WriteDictionary(NodeList* nodeList, unsigned char* str)
{
    Node** nodes = nodeList->elements;   
    int N = nodeList->getCount();

    Node** cudaNodes;
    unsigned char* cudaStr;

    cudaMalloc((void**)&cudaStr, 6 * N * sizeof(unsigned char));
    cudaMalloc((void**)&cudaNodes, N * sizeof(Node*));

    cudaMemcpy(cudaStr, str, 6 * N * sizeof(char), cudaMemcpyHostToDevice); 
    cudaMemcpy(cudaNodes, nodes, N * sizeof(Node*), cudaMemcpyHostToDevice);

    dim3 block(1);
    dim3 thread(N);

    std::cout << N << "\n";

    wrtDict<<<block,thread>>>(cudaNodes, cudaStr);

    cudaMemcpy(str, cudaStr, 6 * N * sizeof(unsigned char), cudaMemcpyDeviceToHost);


    cudaFree(cudaNodes);
    cudaFree(cudaStr);
}

可以看到，函数写字典是 CUDA 和程序其余部分之间的代理。我有一堆我类的对象节点 指向的普通内存中的某处节点 * 保存在我的对象中的数组元素 节点列表 .现在了解 就足够了节点 , 它有一个公共(public)字段 字符 特点。一个 字符 * str 现在将填充一些测试数据。它包含 6 * N 为字符分配的内存，其中 N = 元素数组中所有元素的计数。所以我在 CUDA 中为 6 * N 分配了一个内存空间字符 和 N 节点 指针。然后我在那里复制我的节点 指针，它们仍然指向普通内存。我正在运行该功能。在函数内 wrtDict 我将字符提取到 字符 c 变量，这次不尝试将其放入输出数组 str。

因此，当我编写输出数组 str 的内容时(在 WriteDictionary 函数之外)，我得到了完全正确的答案，即:

1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0   1  2  0  0  0  0 
1  2  0  0  0  0

是的，这里我们有 39 个正确的 字符 (以十六进制显示)。但是，当我们稍微更改 内的 super secret 评论时wrtDict 函数，像这样:

__global__ void wrtDict(Node** nodes, unsigned char* str)
{
    int i = threadIdx.x;

    Node* n = nodes[i];
    char c = n->character;

    str[6 * i] = c;//1;                         !!!
    str[6 * i + 1] = 2;

    str[6 * i + 2] = 0;
    str[6 * i + 3] = 0;
    str[6 * i + 4] = 0;
    str[6 * i + 5] = 0;
}

我们会看到奇怪的事情。我现在期待第一个 字符 每六个中的一个字符来自 节点 由数组指向 - 每一个都不同。或者，即使它失败了，我只期待第一个 字符 每六个都搞砸了，但其余的都完好无损:？ 2 0 0 0 0。但是不!当我这样做时，一切都变得一团糟，现在输出数组的内容 字符串 看起来像这样:

70 21 67 b7 70 21  67 b7  0  0  0  0 
 0  0  0  0 18 d7  85  8 b8 d7 85  8 
78 d7 85  8 38 d9  85  8 d8 d7 85  8 
f8 d5 85  8 58 d6  85  8 d8 d5 85  8 
78 d6 85  8 b8 d6  85  8 98 d7 85  8 
98 d6 85  8 38 d6  85  8 d8 d6 85  8 
38 d5 85  8 18 d6  85  8 f8 d6 85  8 
58 d9 85  8 f8 d7  85  8 78 d9 85  8 
98 d9 85  8 d8 d4  85  8 b8 d8 85  8 
38 d8 85  8 38 d7  85  8 78 d8 85  8 
f8 d8 85  8 d8 d8  85  8 18 d5 85  8 
61 20 75 6c 74 72  69 63 65 73 20 6d 
6f 6c 65 73 74 69  65 20 73 69 74 20 
61 6d 65 74 20 69  64 20 73 61 70 69 
65 6e 2e 20 4d 61  75 72 69 73 20 73 
61 70 69 65 6e 20  65 73 74 2c 20 64 
69 67 6e 69 73 73  69 6d 20 61 63 20 
70 6f 72 74 61 20  75 74 2c 20 76 75 
6c 70 75 74 61 74  65 20 61 63 20 61 
6e 74 65 2e 20 46 

我现在问——为什么？是因为我试图从 CUDA GPU 中获取普通内存吗？我收到一个警告，可能正是关于这种情况，说:

Cannot tell what pointer points to, assuming global memory space

我已经用谷歌搜索了这个，只找到了这个，CUDA 它正在达到一个普通的内存，导致无法找到到达的位置，99.99% 的这个警告应该被忽略。所以我忽略了它，认为它会没事的，但事实并非如此——我的情况在那个 0.01% 之内吗？

我怎么解决这个问题？我知道我可以复制 节点 ，而不是指向它们的指针，进入CUDA，但我认为复制它们会花费我更多的时间，而不是我节省并行化内部对它们所做的事情。我还可以从每个 中提取字符节点 ，将它们全部放入一个数组中，然后将其复制到 CUDA，但是 - 与上一条语句中的问题相同。

我完全不知道该怎么办，更糟糕的是，我大学的 CUDA 项目的截止日期是今天，apx。晚上 17 点(我只是没有足够的时间来早点，该死的......)。

PS。如果有帮助:我正在使用非常简单的(没有任何开关)命令进行编译:

nvcc -o huff ArchiveManager.cpp IOManager.cpp Node.cpp NodeList.cpp Program.cpp Paraleller.cu

Answer 1

这是一个可怕的问题，请参阅 talonmies 的评论。

检查每个 CUDA API 调用的错误值。您将在 cudaMemcpy 上收到启动失败消息内核启动后

运行cuda-memcheck帮助调试错误(基本上是段错误)

意识到您正在从 GPU 取消引用(未映射的)指向主机内存的指针，您需要复制节点，而不仅仅是指向节点的指针