OpenCL 支持constant memory,这正是您正在寻找的。它的工作方式类似于全局内存,尽管确切的位置可能因实现而异,并且它允许编译器以不同的方式优化事物,因为保证在内核执行期间不会修改内存。
您将创建此缓冲区,将其标记为只读并从主机可访问的数据中复制。然后你只需将它正常传递到内核中:
float[] hugeData = { ... };
void run_kernel() {
// ...
cl_mem cl_hugeData;
hugeDataInOpenCL = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(hugeData), hugeData, &error);
// ...
clSetKernelArg(kernel, 0, sizeof(cl_hugeData), &cl_hugeData);
// ...
}
__kernel void mykernel (__constant float * hugeData, ...) {
// use hugeData however you want
}
大概您希望为每个上下文创建一次缓冲区并重新使用它(就像您使用内核一样)以避免过多的返工。
此外,您的数据足够大,以至于某些设备可能没有足够的可用常量内存,在这种情况下,您可以回退到使用全局内存并将其标记为只读。(这意味着您的内核有两个版本,一个带有任一类型的参数。)在运行时检查clGetDeviceInfo
并CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE
决定。