OpenCL簡介

? ? 開放計算語言（Open Computing Language, OpenCL）是非盈利技術聯盟Khronos Group管理的異構編程框架。該框架充分利用了CPU、DSP、FPGA、GPU的計算能力。OpenCL支持多層次的并行，可以高效的映射到同構或異構的體系結構上。

OpenCL標準

? ? OpenCL API是按照C 的，由C和C++封裝而成，并且有很多第三方語言的綁定。這些語言包括Java、Python以及.NET等。OpenCL是C99語言的子集，并適當地擴展了在眾多異構設備上執行數據并行代碼的能力。 ? ??OpenCL和許多CPU并發編程模型一樣，語法類似于標準的C函數，主要區別在于它擁有額外的一些關鍵字和OpenCL Kernel實現的執行模型。在OpenCL編程中，編程人員應該考慮的是如何細粒度地表示程序中的并行性。一個典型的OpenCL Kernel應該是這樣的： <span style="font-size:10px;">_kernel void vecadd(_global int *C, _global int *A, _global int *C) {int tid = get_global_id(0); //OpenCL intrisic functionC[tid] = A[tid] + B[tid]; }</span>

OpenCL規范

? ? OpenCL規范由四個模塊組成：

平臺模型：定義了一個抽象的硬件模型，描述了宿主機和設備。供編程人員在上面編寫在這些設備上執行的Kernel。

執行模型：定義了如何在宿主機上配置OpenCL環境以及如何在設備上執行Kernel。

內存模型：定義了Kernel中所使用的內存層次，無需考慮實際的底層架構。

編程模型：定義了如何將并發模型映射到物理硬件上。

平臺模型

? ? 如下圖描述，平臺模型由一個宿主機以及一個或多個設備組成，一個設備可以分為多個CUs(Compute Units)，一個CU又可以進一步劃分為多個PEs(Processing Elements)。設備上的計算具體映射到了每個PE之上。
? ? 在平臺模型中編程人員需要關注三種版本號：平臺版本號、設備版本號、OpenCL支持版本號。 ? ? 平臺版本號表明了平臺所支持的運行時功能，包括了OpenCL運行時所有與宿主機交互的APIs，例如上下文、內存對象、設備、命令隊列相關的函數。 ? ? 設備版本號表明了設備的能力，可以使用clGetDeviceInfo函數獲取。設備信息通常包括資源限制和擴展功能等。 ? ? 語言版本號表明了設備最高支持的語言版本。

執行模型

? ? OpenCL的執行模型分為 兩部分，Kernel Program和Host Program，宿主程序定義了Kernel執行的上下文并控制其執行。執行模型中主要定義了kernel是怎樣執行的。 ? ? 這里首先講一下執行模型中非常重要的兩個概念——work-item和work-groups。在kernel由宿主機提交給設備執行的同時，定義了一個索引空間。每個work-item可以看做是kernel在設備上單次執行的實例，該實例在索引空間中擁有唯一的索引值。基于索引值的不同，每個work-item雖然執行同樣的kernel程序，但卻有自己的代碼路徑和數據路徑。work-item被組織為work-groups。在同一work-groups中每個work-item擁有唯一的local ID，如果有多個work-groups，那么在不同的work-groups之間local ID不是唯一的。一個work-item索引必須采用global ID或local ID + work-groups ID才能唯一確定。 ? ?? ? ? 在執行模型中另一個非常重要的概念是NDRange，即N維索引空間，這里的N可以定義為1,2或3。它由一個長度為N的數組組成，每個數組元素表示該維度上工作節點的個數。工作節點的global ID、local ID以及work-groups ID都是N維的。上圖是一個二維索引空間的實例，NDRange(Gx,Gy)，對應的工作組為NDRange(Sx,Sy)。每個維度的工作空間恰好分配為整數個工作組。這里要求Gx必須為Sx的整數倍，Gy必須為Sy的整數倍。

上下文與命令隊列

? ? 在OpenCL中，上下文（context）是一個存在于主機端的抽象容器。負責協調主機——設備之間的交互，管理設備上可用的內存對象，跟蹤每個設備建立的kernel和程序。 ? ? 宿主機負責定義kernel的上下文，上下文包括以下資源：

設備：平臺上可用的設備。

Kernels：運行在設備上的OpenCL C程序。

程序對象：負責執行和實現Kernels的程序資源。

內存對象：Kernel和宿主機可見的內存對象的集合，內存對象中包含kernel實例中用于計算的數據。

? ? 新建上下文的API為clCreateContext()。 cl_context clCreateContext(const cl_context_properties *properties,cl_unit num_devices,void (CL_CALLBACK *pfn_notify)(const char *errinfo,const void *private_info,size_t cb,void *user_data),void *user_data,cl_int *errcode_ret)? ? API函數clCreateCommandQueue()用來建立命令隊列并將其關聯到某個設備。cl_command_queue clCreateCommandQueue(cl_context contextcl_device_id device,cl_command_queue_properties properties,cl_int *errcode_ret)
? ? clCreateCommandQueue()中的屬性參數properties是一個位域，用于開啟程序剖析（profiling）命令（CL_QUEUE_PROFILING_ENABLE）、亂序執行命令（CL_QUEUE_OUT_OF_ORDER_EXEC_ENABLE）。

內存對象

? ? OpenCL應用程序用來處理大型數組或多維矩陣，在kernel開始執行之前需要將這些數據實際映射到設備上。在OpenCL中采用的方法是將這些數據封裝為內存對象。OpenCL定義了兩種內存對象：buffer和image。buffer類似于C語言中的數組，由malloc()分配，在內存中是連續存放的。而image被設計為不透明的對象，以利于進行數據填充和在特定的設備上進行性能優化。 ? ? 創建buffer的內存對象的API為clCreateBuffer()，內存對象作為返回值返回。 cl_mem clCreateBuffer(cl_context context,cl_mem_flags flags,size_t size,void *host_ptr,cl_int *errcode_ret)? ? 將宿主機內存傳入或傳出設備的API分別為clEnqueueWriteBuffer和clEnqueueReadBuffer。可通過設置參數選項blocking_write為CL_TRUE，使寫函數工作在阻塞模式——數據完成到OpenCL buffer的傳輸后才返回。 image對象在后面用到的時候在詳細介紹。

程序對象

? ? OpenCL中的程序（program）特指kernel函數的集合，是kernel被調度安排到設備上運行的單位。 ? ? OpenCL程序運行時通過調用一系列。API進行編譯，編譯系統對具體設備進行優化。OpenCL軟件僅鏈接到公共運行層（稱為ICD），所有平臺特定的SDK通過一個動態鏈接接口委托給某個廠商的運行時。 ? ? 新建kernel的步驟如下：

OpenCL源代碼以字符串形式存儲。如果代碼存放在文件中，則必須讀取到內存中，在內存中以字符串數組的形式存放。

源代碼通過調用clCreateProgramWithSource()轉化為cl_program對象。

調用clBuildProgram()在支持OpenCL的設備上編譯程序對象。

向量加代碼演示

#include<stdio.h> #include<stdlib.h> #include<CL/cl.h>const char *programSource = "_kernel void vecadd(_global int *A, _global int *B, _global int *C) \n" "{ \n" " int idx = get_global_id(0); \n" " C[idx] = A[idx] + B[idx]; \n" "} \n" int main() {int *A = NULL;int *B = NULL;int *C = NULL;const int elements = 2048;size_t datasize = sizeof(int) * elements;A = (int*)malloc(datasize);B = (int*)malloc(datasize);C = (int*)malloc(datasize);int i;for(i=0; i<elements; i++){A[i] = i;B[i] = i;}//Check return valuecl_int status;//Retrieve the number of platforms cl_uint numPlatforms = 0;status = clGetPlatformIDs(0, NULL, &numPlatforms);//Allocate space for each platformcl_plarform_id *platforms = NULL;platforms = (cl_platform_id*)malloc(numPlatforms * sizeof(cl_platform_id));//Fill in the platformsstatus = clGetPlatformIDs(numPlatforms, platforms, NULL);//Retrieve the number of devicescl_unit numDevicdes = 0;status = clGetDeviceIDs(platform[0], CL_DEVICE_TYPE_ALL, 0, NULL, &numDevices);//Allocate space for devicescl_device_id *devices;devices = (cl_device_id*)malloc(numDevices * sizeof(cl_device_id));//Fill in the devicesstatus = clGetDeviceIDs(platform[0], CL_DEVICE_TYPE_ALL, numDevices, deivces, NULL);//Create a context and associate it with devicescl_context context;context = clCreateContext(NULL, numDevices, devices, NULL, NULL, &status);//Create a command queue and associate it with the devicecl_command_queue cmdQueue;cmdQueue = clCreateCommandQueue(context, devices[0], 0, &status);//Create buffer objectcl_mem bufA = clCreateBuffer(context, CL_MEM_READ_ONLY, datasize, NULL, &status);cl_mem bufB = clCreateBuffer(context, CL_MEM_READ_ONLY, datasize, NULL, &status);cl_mem bufC = clCreateBuffer(context, CL_MEM_READ_ONLY, datasize, NULL, &status);//Write to the device bufferstatus = clEnqueueWriteBuffer(cmdQueue, bufA, CL_FALSE, 0, datasize, A, 0, NULL, NULL);status = clEnqueueWriteBuffer(cmdQueue, bufB, CL_FALSE, 0, datasize, B, 0, NULL, NULL);status = clEnqueueWriteBuffer(cmdQueue, bufC, CL_FALSE, 0, datasize, C, 0, NULL, NULL);//Create program from source codecl_program program = clCreateProgramWithSource(context, 1, (const char**)&programSource, NULL, &status);//Compile the programstatus = clBuildProgram(program, "vacadd", &status);//Associate buffers with the kernelstatus = clSetKernelArg(kernel, 0, sizeof(cl_mem), &bufA);status = clSetKernelArg(kernel, 1, sizeof(cl_mem), &bufB);status = clSetKernelArg(kernel, 2, sizeof(cl_mem), &bufC);//Define index spacesize_t globalWorkSize[1];globalWorkSize[0] = elements;//execute the kernelstatus = clEnqueueNDRangeKernel(cmdQueue, kernle, 1, NULL, globalWorkSize, NULL, 0, NULL, NULL);//Read the device output buffer to the host output array clEnqueueReadBuffer(cmdQueue, bufC, CL_TRUE, 0, datasize, C, 0, NULL, NULL);//Verify the outputint result = 1;for(i=0;i<elements;i++){if(C[i] != i+i){result = 0;break;}}if(result){printf("correct!\n");}else{printf("incorrect!\n");}//Release resourcesclReleaseKernel(kernel);clReleaseProgram(program);clReleaseCommandQueue(cmdQueue);clReleaseMemObject(bufA);clReleaseMemObject(bufB);clReleaseMemObject(bufC);clReleaseContext(context);//Free host resourcesfree(A);free(B);free(C);free(platform);free(devices); }

總結

以上是生活随笔為你收集整理的OpenCL编程入门（一）的全部內容，希望文章能夠幫你解決所遇到的問題。

如果覺得生活随笔網站內容還不錯，歡迎將生活随笔推薦給好友。