OpenCL的多GPU开发（顺便提一下OpenGL的多GPU开发）

标签（空格分隔）：加速 opencl

---

转载说明出处：http://blog.csdn.net/hust_sheng/article/details/75912004

需求

在某些加速优化项目中会使用到GPU，且为了追求速度，有时候我们会使用多块GPU。就OpenCL而言，如何充分利用多个GPU的运算能力十分关键。

---

• 多线程（以两个GPU为例）

//传入图像数据errNum = clEnqueueWriteImage(commandQueue_1stGPU, imgIn_mt1[0], CL_TRUE, origin, region, 0, 0, tlImage, 0, NULL, NULL);errNum = clEnqueueNDRangeKernel(commandQueue_1stGPU, step4WarpKernel_1stGPU, 1, NULL, step4GlobalWorkSize, step4LocalWorkSize, 0, NULL, NULL);

注意
1. 避免线程之间的共享写缓冲区操作
2. 共享读缓冲区操作必须要为每个设备申请独立的资源
3. 设备命令队列、可执行内核也应该是独立的

只有这样才可以实现充分的并行，总的来说，这里的多线程不涉及GPU的内部操作，只是子线程会去调用GPU，整体上是完全符合C语言级别的多线程的，测试之后，效果明显。

补充
有时候 GPU->CPU 需要传输很大的数据量，这个时候要使用异步操作，

void tmpfuc(){    ;}

static cl_event event_async;// CL_FALSE 表示 clEnqueueReadImage 是非阻塞的// 事件event_async和clEnqueueReadImage命令绑定，该命令执行结束，event_async状态变为CL_COMPLETEerrNum = clEnqueueReadImage(commandQueue_2ndGPU, imgOut_mt2[1], CL_FALSE, origin, region_imgout,                             0, 0, *(outBuffer + 1), 0, NULL, &event_async);if (errNum != CL_SUCCESS) {    printf("clEnqueueReadImage Error.\n");}// 事件event_async状态变为CL_COMPLETE触发回调函数clSetEventCallback(event_async, CL_COMPLETE, &tmpfuc, NULL);        // 最后一个表示函数入参

需要注意的是，回调函数本身就是一个子线程。在项目开发的过程中，异步操作一直没有成功的感觉，所以这一步暂时也用多线程代替，使用信号量同步。

---

OpemCL本身能否实现多GPU调用？可以看下面的例子，逻辑很简单，但是从代码上来看是不科学的，代码逻辑和单个GPU没有差别，那上述的共享缓冲区问题怎么处理？留有疑问…按照下面的代码实现之后发现没有速度上的提升…很尴尬。

NVIDIA官方demo

---

OpenGL使用多GPU

很多时候，OpenCL实现的加速用OpenGL来做更加方便，比如贴图渲染。那对OpenGL来说，怎么使用多GPU呢？

有点奇怪：

正常来说，如果可以实现，或许我们使用OpenCL提供的接口就可以，但是调研之后发现，OpenCL对GPU的使用是第三方来提供接口的，也就是NVIDIA来提供。

• NVIDIA提供的WGL_NV_gpu_affibity

  该库的使用需要glew（wglew.h文件）：

  #include    "GL/glew.h"  #include    "GL/wglew.h"  

  遗憾的是，目前N卡中只有Quadro支持这个库，GeForce等游戏显卡是不支持的，直接崩掉，我也是醉了…

  我们可以用代码来检测电脑显卡是否支持 WGL_NV_gpu_affibity，见github代码