1270803596|%Y-%m-%d|agohover
嗯,這就是我最近在做的事。
OpenCL 是一套在異質平台 (heterogeneous platforms) 上編寫平行處理程式的一套 API 標準。如果你平常就對什麼 CUDA 啦或是 GPGPU 之類的東西耳熟能詳,那應該也會聽過這個字眼,如果沒聽過的話就去參考一下 wiki 吧。
OpenCL 和 CUDA 的架構是很類似的,不過在程式編譯的流程上卻走了不同的方向。CUDA 本身是一個新的語言(儘管它非常像 C++),程式設計師使用 CUDA 的語法寫好程式後,要使用 NVIDIA 提供的 nvcc compiler 編譯自己的程式。nvcc 會把 CUDA 程式中給 GPU 執行的部份(也就是宣告為 __globale__ 和 __device__ 的函式)取出來編譯成 GPU 所使用的 binary,然後把程式的其它部份轉換成合法的 C/C++ 後,交給其它的 compiler(VC 或是 GCC)來編譯,最後把這兩塊合起來。
OpenCL 的方式則是依照 OpenGL 的 shader,它定義了一組 C API 讓程式設計師可以編譯 device program。因此程式流程大致上是:
- 初始化 OpenCL context。
- 讀入 device program 的原始碼1並使用 OpenCL 提供的 API 進行編譯。
- 配置資料存取用的 device buffer。
- 設定 kernel function 的參數及 thread 總數並呼叫之。
以上的流程都是使用 C/C++ 呼叫 OpenCL API 來完成的,因此 OpenCL 的 host program 就只是一支單純的 C/C++ 程式。然而上述的前兩項步驟,在 CUDA 中都被 nvcc 自動幫你解決了。因此對懶惰的人來說,CUDA 是踏入 GPGPU 的一條捷徑。不過就我個人來說,是比較喜歡 OpenCL 的,因為:
- OpenCL 是標準。儘管這是一項很新的標準,但至少 GPU 兩大龍頭 NVIDIA 和 AMD 都會支援它。相較之下,CUDA 就只有 NVIDIA 的支援了。
- OpenCL 針對的不止是 GPGPU,而是提供異質平台的平行處理架構。因此除了 GPU 外,物理加速卡或是 cell processor 也是 OpenCL 可以使用的運算資源。
- 我比較喜歡 OpenCL 的 C extension。CUDA 的 C extension 雖然提供向量型別,但向量竟然不能加減乘除,難用爆了。而 OpenCL 的 device program 寫起來和 GLSL 很像,有好用的向量型別,當然 OpenCL 是設計成 general purpose,你不需要寫過 GLSL 也能很快就上手。
與 CUDA 比起來,OpenCL API 為了給使用者更大的彈性,它會顯得比較低階。以一個最簡單的例子來看:
// hello.cl kernel void hello(global int *output) { int i = get_global_id(0); output[i] = i*i; }
// hello.cpp #include <iostream> #include <fstream> #include <CL/cl.h> using namespace std; int main() { // get the first OpenCL platform cl_platform_id platform; clGetPlatformIDs(1, &platform, NULL); // get the first GPU device on the first platform cl_device_id device; clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 1, &device, NULL); // create OpenCL context cl_context_properties properties[] = { CL_CONTEXT_PLATFORM, reinterpret_cast<cl_context_properties>(platform), 0 }; cl_context context; context = clCreateContext(properties, 1, &device, NULL, NULL, NULL); // read device program source ifstream fin("hello.cl", ios::binary); fin.seekg(0, ios::end); size_t len = fin.tellg(); char* src = new char[len]; fin.seekg(0, ios::beg); fin.read(&src[0], len); cl_program program; program = clCreateProgramWithSource(context, 1, const_cast<const char**>(&src), &len, NULL); // build the program clBuildProgram(program, 1, &device, NULL, NULL, NULL); // create a kernel object cl_kernel kernel = clCreateKernel(program, "hello", NULL); // create a buffer object cl_mem buf = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(cl_int)*16, NULL, NULL); // set kernel arguments clSetKernelArg(kernel, 0, sizeof(cl_mem), &buf); // create the command queue object cl_command_queue queue = clCreateCommandQueue(context, device, 0, NULL); // execute the kernel size_t global_size = 16; size_t local_size = 4; cl_int err = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &global_size, &local_size, 0, NULL, NULL); // check the answer cl_int output[16]; clEnqueueReadBuffer(queue, buf, CL_TRUE, 0, sizeof(cl_int)*16, output, 0, NULL, NULL); for(size_t i = 0; i < 16; ++i) cout << output[i] << ' '; cout << endl; // release resources delete [] src; clReleaseCommandQueue(queue); clReleaseMemObject(buf); clReleaseKernel(kernel); clReleaseProgram(program); clReleaseContext(context); return 0; }
未免也太麻煩了,而且這還不包含錯誤處理!因此我嘗試用 C++ 來封裝 OpenCL API,同樣功能的 code 使用 C++ 的版本大概像這樣:
#include <iostream> #include <exception> #include "clpp/clpp.hpp" using namespace std; using namespace clpp; int main() { try{ Context context; Kernel k = context.readProgramSourceFile("test.cl").kernel("hello"); Buffer<cl_int> output = context.createBuffer<cl_int>(16); k.setArgs(output); CommandQueue q = context.queue(); q.exec(k, 16, 4); // check the answer cl_int data[16]; q.read(output, data); for(int i = 0; i < 16; ++i) cout << i << ' '; cout << endl; }catch(exception& err){ cerr << err.what() << endl; return 1; } return 0; }
當然,發生錯誤時會使用 exception 來進行妥善的回報。目前這個 clpp library 正在開發中,如果你有興趣的話也很歡迎來和我討論。
Is source code of clpp available to public?
Hi fr3@k,
Thank you for your interest! Clpp was started from just few days ago so most OpenCL features are still incomplete. I'm going to upload the source after some modification in 2 or 3 days.
Post preview:
Close preview