CUDA与OpenCL在空核心上的性能比较

Question

在对比CUDA和OpenCL上运行相同内核的性能时，我发现了一件奇怪的事情。

当我完全不加任何输入参数和计算代码，仅仅是空内核时，CUDA的性能非常低，相比之下，OpenCL要好很多。

CUDA内核：

__global__ void kernel_empty()
{
}

CUDA主机:

kernel_empty<<<dim3(10000, 10000, 1), dim3(8, 8, 1)>>>();

---

OpenCl内核：

__attribute__((reqd_work_group_size(8, 8, 1)))
__kernel void kernel_empty()
{
}

OpenCL主机：

cl_event perf_event;
size_t global_work_offset[3] = {0, 0, 0};
size_t global_work_size[3] = {10000, 10000, 1};
size_t local_work_size[3] = {8, 8, 1};
clEnqueueNDRangeKernel(queue, kernel, 3, global_work_offset, global_work_size, local_work_size, 0, NULL, &perf_event);

OpenCL提供了6ms

CUDA提供了390ms

• 两个API的内核都正常工作，因为我使用它们来计算我的东西。
• 两边都没有返回错误代码。
• 使用Visual Studio 2010，发布模式
• 使用NVIDIA GPU Computing Toolkit 5.5中的OpenCL 1.1 lib
• 使用NVIDIA GPU Computing Toolkit 5.5中的CUDA lib
• 时间也是正确的，我已经用CPU计时器进行了双重检查。而且当使用巨大的网格时，你可以看到CUDA在没有任何计时器的情况下滞后。
• 对于OpenCL，使用clGetEventProfilingInfo。
• 对于CUDA，使用cudaEventElapsedTime。
• 测试在具有NVIDIA Quadro K4000的同一台PC上运行。

有人能解释一下为什么有这么大的差异吗？

- Rusty Labs

---

你能展示一下你传递给 clEnqueueNDRangeKernel 的全局和本地大小吗？我刚在一些 NVIDIA GPU 上尝试了一下，得到的运行时间是几百毫秒，就像你的 CUDA 实现一样。 - jprice

我已经在初始问题中更新了OpenCL主机代码，谢谢。 - Rusty Labs

顺便问一下，你用的是什么GPU？ - Rusty Labs

Answer 1

在OpenCL和CUDA中，内核的启动方式是不同的，因此实际上您为每种方法启动了不同数量的工作。

在OpenCL中，您指定全局工作大小（要启动的总工作项数）和本地工作大小（工作组大小）。在您的示例中，您以8x8的组为单位启动10000x10000个工作项。

在CUDA中，您指定块大小（类似于工作组大小）和网格大小，即要启动多少块。这意味着您的CUDA示例正在启动10000x10000个块，总共有80000x80000个CUDA线程。

因此，这个CUDA内核启动：

kernel_empty<<<dim3(10000, 10000, 1), dim3(8, 8, 1)>>>();

相当于这个OpenCL内核入队：

size_t global_work_size[3] = {80000, 80000, 1};
size_t local_work_size[3] = {8, 8, 1};
clEnqueueNDRangeKernel(queue, kernel, 3, NULL, global_work_size, local_work_size, 0, NULL, &perf_event);

- jprice

---

谢谢！那是我的巨大错误！ - Rusty Labs

没错，6ms88 = 384ms :) 即使你已经熟悉了API也很难发现。 - DarkZeros