为内核设置参数的速度非常慢(OpenCL)

Question

在我的OpenCL Dijkstra算法实现中，到目前为止最慢的部分是将1D简化图矩阵写入内核参数，即全局内存。

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

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));

设置参数会花费非常长的时间对于我的5760000个顶点的测试，将数据写入参数需要3秒以上，而算法本身需要不到一毫秒的时间：

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(...)等(，然后重复运行内核或在缓冲区中交换数据而不重新分配也是有益的。