为内核设置参数非常慢（OpenCL）

Question

在我的 OpenCL Dijkstra 算法实现中，迄今为止最慢的部分是将一维简化图矩阵写入内核参数，即全局内存。

我的图是一个二维数组；对于 OpenCL，它会被简化为一维数组，如下所示：

for (int q = 0; q < numberOfVertices; q++)
{
    for (int t = 0; t < numberOfVertices; t++)
    {
        reducedGraph[q * numberOfVertices + t] = graph[q][t];
    }
}

放入缓冲区：

cl::Buffer graphBuffer = cl::Buffer(context, CL_MEM_READ_WRITE, numberOfVertices * numberOfVertices * sizeof(int));

设置参数需要很长时间。对于我的 5,760,000 个顶点的测试，将数据写入参数需要 3 秒以上，而算法本身需要不到 1 毫秒： p>

kernel_dijkstra.setArg(5, graphBuffer);

内核使用图作为全局参数：

void kernel min_distance(global int* dist, global bool* verticesSet, const int sizeOfChunks, global int* result, const int huge_int, global int* graph, const int numberOfVertices)

有什么办法可以加快速度吗？谢谢！

编辑：我的内核代码：

// Kernel source, calculates minimum distance in segment and relaxes graph.
std::string kernel_code =
       void kernel min_distance(global int* dist, global bool* verticesSet, const int sizeOfChunks, global int* result, const int huge_int, global int* graph, const int numberOfVertices) {
           for (int b = 0; b < numberOfVertices; b++) {
               int gid = get_global_id(0);
               int min = huge_int, min_index = -1;
               for (int v = gid * sizeOfChunks; v < sizeOfChunks * gid + sizeOfChunks; v++) {
                   if (verticesSet[v] == false && dist[v] < min && dist[v] != 0) {
                       min = dist[v];
                       min_index = v;
                    }
               }
               result[gid] = min_index;
               if (gid != 0) continue;
               min = huge_int;
               min_index = -1;
               int current_min;
               for (int a = 0; a < numberOfVertices; a++) {
                   current_min = dist[result[a]];
                   if (current_min < min && current_min != -1 && current_min != 0) { min = current_min; min_index = result[a]; }
               }
               verticesSet[min_index] = true;
     // relax graph with found global min.
               int a = 0;
               int min_dist = dist[min_index];
               int current_dist;
               int compare_dist;
               for (int i = min_index * numberOfVertices; i < min_index * numberOfVertices + numberOfVertices; i++) {
                   current_dist = dist[a];
                   compare_dist = graph[min_index * numberOfVertices + a];
                   if (current_dist > min_dist + compare_dist && !verticesSet[a] && compare_dist != 0) {
                       dist[a] = min_dist + compare_dist;
                   }
                   a++;
               }
           }
       };

我如何入队：

    numberOfComputeUnits = default_device.getInfo<CL_DEVICE_MAX_COMPUTE_UNITS>();
queue.enqueueNDRangeKernel(kernel_dijkstra, 0, cl::NDRange(numberOfVertices), numberOfComputeUnits);

Answer 1

这里的错误是您的内存分配太大：5.76M 顶点需要 133TB 缓冲区，因为缓冲区大小是顶点数的二次方。 C++ 编译器和 OpenCL 都不会将此报告为错误，甚至您的内核看起来也可以正常启动和运行，但实际上它不会计算任何内容，因为内存不足，您会得到随机和未定义的结果。

通常.setArg(...) 的时间不应超过几毫秒。此外，最好在开始时只执行一次初始化部分（包括缓冲区分配、.setArg(...) 等），然后重复运行内核或在缓冲区中交换数据而不重新分配。