带有 clBuildProgram 错误的 HD 530 上的 MacOs OpenCL (-11)答案

【问题标题】：MacOs OpenCL on HD 530 with clBuildProgram error(-11)带有 clBuildProgram 错误的 HD 530 上的 MacOs OpenCL (-11)
【发布时间】：2017-03-24 07:45:44
【问题描述】：

我有最新的 mac pro(OS:10.12.2)，带有运行 OpenCL 代码的英特尔集成 GPU HD 530(Gen9)。在我的 OpenCL 代码中，我使用 vloadx 和 atomic_add 指令。将我的 OpenCL 内核代码更改为位代码，例如 https://developer.apple.com/library/content/samplecode/OpenCLOfflineCompilation/Introduction/Intro.html#//apple_ref/doc/uid/DTS40011196-Intro-DontLinkElementID_2 .并使用 clCreateProgramWithBinary 创建程序。但是当 clBuildProgram 时，它返回 -11 错误。并且构建日志是“ 错误：未定义对_Z6vload2mPKU3AS1h()' undefined reference to_Z8atom_addPVU3AS3ii()'的引用 " 但是在我的带有 HD 5500(Gen8) 的 mac air 中，代码是可以的。谁能告诉我该怎么办？

【问题讨论】：

您只能在保存它的同一硬件上可靠地使用保存的二进制文件，并且使用保存驱动程序版本。如果在不同的硬件上或使用不同的驱动程序版本，您应该从源代码重新构建。在 OpenCL 2.0 中，SPIR-V 使得发布与硬件和驱动程序无关的字节码成为可能，但在 OpenCL 1.x 中并非真正可行。
我想保护我的 IP，所以我不能从源代码构建。从苹果的例子中，我们可以使用离线编译器生成的位码。但是这种方法不适用于intel Gen9，但适合Gen7.5和Gen8。

标签： macos binary opencl

【解决方案1】：

这里的问题是，您不能在不同的设备中使用不兼容的二进制文件。这意味着如果您为 Intel 编译，例如，您不能使用为 AMD 编译的二进制文件。您需要做的是每次从源代码编译特定设备的代码。

如果您不想在不同的文件中使用 OpenCL 代码，您可以通过对它们进行字符串化将它们放入源文件中。您无需读取文件，而是使用主机代码中的内核字符串作为内核字符串传递。这将允许您保护您的 IP。但是，每次都需要使用clBuildProgram 构建代码。您还可以将构建的程序保存为二进制文件，这样在第一次运行后，您不会因为每次都构建它而降低性能。举个例子，假设你有一个 kernel.cl 文件如下：

__kernel void foo(__global int* in, __global int* out)
{
  int idx = get_global_id(0);
  out[idx] = in[idx] * in[idx];
}

你可能通过读取文件来获得这个内核代码：

char *source_str;
fp = fopen("kernel.cl", "r");
source_str = (char *)malloc(MAX_SOURCE_SIZE);
source_size = fread(source_str, 1, MAX_SOURCE_SIZE, fp);
fclose(fp);

program = clCreateProgramWithSource(context, 1, (const char **)&source_str, (const size_t *)&source_size, &ret);

你可以做的是：

const char* src = "__kernel void foo(__global int* in, __global int* out)\
                   {\
                     int idx = get_global_id(0);\
                     out[idx] = in[idx] * in[idx];\
                   }";

program = clCreateProgramWithSource(context, 1, (const char **)&src, (const size_t *)&src_size, &ret);

当你编译你的 C 代码时，这个字符串会被转换成二进制，这样你就可以保护你的源代码。

【讨论】：