本篇介绍几个OpenCL基础API，涉及平台、设备等初始化必备函数其次介绍几个关于缓冲区操作以及工作空间划分的API建议阅读参考书籍，我的推荐是《OpenCL编程指南》和《OpenCL异构计算》，尤其是后者从实践出发，更是适合上手。建议收藏这个API的参考网址，里面也OpenCL API的详细介绍OpenCL 2.1 Reference Pages本章节函数使用举例见系列篇三，这里不再重复；仅介绍API相关参数，以及自己的见解。

一、OpenCL平台API

平台API定义了宿主机程序发现OpenCL设备所用的函数以及这些函数的功能，另外还定义了为OpenCL 应用创建上下文的函数。包括平台、设备、上下文等相关函数。 不过为了方便起见这里将构建程序和创建内核也放在了此章节，即1.4、1.5、1.6。 注意1.7及之后为一些篇外话，不作为和1.1~1.6的连续内容。

1.1、获取平台clGetPlatformIDs()

cl_int clGetPlatformIDs( cl_uint num_entries, //限制返回的平台数 cl_platform_id *platforms, //平台列表保存位置指针 cl_uint *num_platforms //平台个数保存位置指针 );

将num_entries和platforms分别设置为0和NULL可以查询可用的平台个数，保存在num_platforms中。返回的平台数可以用num_entries来限制，获取平台列表时这个参数要大于0并小于等于可用平台数。 平台这个概念应该是不同厂商的实现，比如我一开始安装了AMD APP SDK，运行程序只有1个可用平台；后来有想法就又安装了Intel的SDK，这时运行OpenCL有两个可用平台。区别是一个平台包含GPU和CPU，一个平台只包含CPU；而且打印设备信息，他们包含的CPU名称相同。

1.2、获取设备clGetDeviceIDs()

cl_int clGetDeviceIDs( cl_platform_id platform, // 指定平台 cl_device_type device_type, // 指定设备类型 cl_uint num_entries, cl_device_id *devices, cl_uint *num_devices );

这个命令会得到与platform关联的eOpenCL设备列表。如果参数devices为NULL，则用num_devices得到可用设备数。类似平台函数用num_entries可以限制返回的设备数。 参数device_type用来指定计算设备类型，其极常见参数选择见下表。

cl_device_type描述CL_DEVICE_TYPE_CPU作为宿主机处理器的OpenCL设备CL_DEVICE_TYPE_GPU作为GPU的OpenCL设备CL_DEVICE_TYPE_ACCELERATOROpenCL加速器(例如,IBM Cell Broaband)CL_DEVICE_TYPE_DEFAULT默认设备CL_DEVICE_TYPE_ALL与相应平台关联的所有OpenCL设备

1.3、创建上下文clCreateContext()

cl_context clCreateContext( const cl_context_properties *properties, // 此处包含平台信息 cl_uint num_devices, // 设备个数 const cl_device_id *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 );

第一参数着重说明一下，需要用确定的平台来构建，方法如下： 　　cl_context_properties props[3] = 　　{ CL_CONTEXT_PLATFORM , (cl_context_properties)platformIds, 0 }; 第四个参数那么复杂就直接忽略吧，因为调用时那个位置填NULL。使用中user_data也输入NULL。最后一个参数errcode_ret用来记录错误代码，正常时errcode_ret==CL_SUCCESS。

1.4、创建程序对象clCreateProgramWithSource()

cl_program clCreateProgramWithSource( cl_context context, cl_uint count, const char **string, const size_t *lengths, cl_int *errcode_ret );

context 创建程序对象的上下文 count 没看到过介绍 const strings 将这个参数的所有字符，构成了创建程序对象的完整源代码 lengths 这个参数可以设置为NULL，这种情况下，则认为字符串以null终止的 errcode_ret 如果非NULL，函数返回错误代码将由这个参数返回 一般执行时，先从.cl文件中读取代码，然后将保存代码的字符串传递给strings

1.5、构建clBuildProgram()

cl_int clBuildProgram( cl_program program, // 一个合法的程序对象 cl_uint num_devices, // 要构建程序对象的设备数 const cl_device_id *device_list, // 如果为空则为context所有关联设备创建 const char *options, void(CL_CALLBACK *pfn_notify)(cl_program program,void *user_data), void *user_data );

函数的第四、五、六个参数设置为NULL即可，前三个参数根据代码段注释传递即可。深入学习的话关于第四个参数options，可以参考《OpenCL编程指南》6.2.2或者查看文章开头给的参考网页。

1.6、创建内核clCreateKernel()

cl_kernel clCreateKernel( cl_program program, const char *kernel_name, cl_int *errcode_ret );

到这里就比较容易了，只介绍一下kernel_name：创建内核对象的内核函数名。这是程序源代码__kernel关键字后面的内核函数名。参数errcode_ret的作用和上面的API中意义相同。

1.7、篇外话之先创建上下文再获取设备

a、创建上下文clCreateContextFromType()b、从上下文中查询设备信息clGetContextInfo() cl_context clCreateContextFromType( const cl_context_properties *properties, cl_device_type device_type, void (CL_CALLBACK *pfn_notify)(const char *,const void *p,size_t, void *), void *user_data, cl_int *errcode_ret ); cl_int clCreatContextInfo( cl_Context context, cl_Context_info param_name, size_t param_value_size, void *param_value, size_t *param_value_size_ret )

和clCreateContext()类似，不过这里不需要事先指定设备列表，使用实例： 　　platform = platformIds[0]; 　　cl_context_properties props[3] = 　　{ CL_CONTEXT_PLATFORM,(cl_context_properties)platform, 0 }; 　　context = clCreateContextFromType( props, CL_DEVICE_TYPE_CPU, NULL, NULL, &errNum); 之后需要在上下文中查询设备clGetContextInfo()，获取到设备列表之后操作就和之前相同了，使用实例： 　　cl_uint numDevices ; 　　cl_device_id *deviceIds; 　　errNum = clGetContextInfo( context, CL_CONTEXT_DEVICES, 0, NULL, &numDevices); 　　deviceIds =(cl_device_id*)malloc(sizeof(cl_device_id)*numDevices); 　　errNum = clGetContextInfo( context, CL_CONTEXT_DEVICES, numDevices, deviceIds, NULL); 把这一个篇外话，是因为这有一个不可取之处，这样搞会有很奇怪的问题。首先在我的电脑上如果是用clCreateContextFromType创建上下文设备选择CL_DEVICE_TYPE_CPU；在调用clGetContextInfo获取的numDevices=4，这是我CPU的线程数，并不是CPU个数。这样本来我还能接受，因为只选择deviceIds[0]的话程序可以正常执行；但如果打印设备信息，deviceIds[0]的信息能正常获取并且printf出来，但是获取deviceIds[1]时就开始报错了，简直。。。。。 把这一个篇外话，只是为了在实践中理解上下文和平台设备之间的关系，并不是推荐使用。

二、OpenCL运行时API

这些API管理上下文来创建命令队列以及运行时发生的其他操作。包括创建读写缓冲区、设置执行内核等

2.1、设置内核参数clSetKernelArg()

cl_int clSetKernelArg( cl_kernel kernel, cl_uint arg_index, size_t *arg_size, const void *arg_value );

kernel 一个合法的内核对象；arg_index 内核参数的索引；arg_size 参数的大小；arg_value 传入内核函数的参数的一个指针。这里参数的意思看看就好，重要的是看例子学会使用。设置好kernel参数之后通过调用下文的clEnqueueNDRangeKernel来执行。 值得说明的是如果你要多次调用执行同一个内核，那么可以只设置一次内核。例如我写卷积神经网络，在程序开始设置内核参数，之后训练过程中只需把新的图像写入到图像缓冲区，直接调用clEnqueueNDRangeKernel执行内核就好。

2.2、执行内核clEnqueueNDRangeKernel()

cl_int clEnqueueNDRangeKernel( cl_command_queue commad_queue, cl_kernel kernel, cl_uint work_dim, const size_t *global_work_offset, const size_t *global_work_size, const size_t *local_work_size, cl_uint num_events_in_wait_list, const cl_event *events_in_wait_list, cl_event *event );

commend_queue 内核的执行需要提交到命令队列中 kernel 要执行的内核名称 work_dim 指定新建的work-item的纬度，这里我假设你已经知道了OpenCL的内核与工作项之间的关系。 global_work_offset 为work-item提供全局ID，该参数可以不从0开始，但一般设置为0或者NULL global_work_size 指定NDRange中每维work-item的数量，不可为空 local_work_size 指定workgroup中每维work-item的数量，可以设置为NULL让系统自动设置 num_events_in_wait_list、events_in_wait_list、event 这是OpenCL高级一点的操作，用于记录此事件或者需要等待的其他事件，可以用来规划不同事件执行顺序；如果不用可以分别设置为0、NULL、NULL。 需要注意的是global_work_size指向的数组大小要和work_dim相等，global_work_offset和local_work_size不为NULL时也是同样要求，不一样会发生什么我也没有测试，不过正常的程序这里该是一样的。

三、缓冲区的创建、写入和读取

3.1、flush和finish命令

这两个命令在命令队列中值两种不同类型的barrier操作。clFinish()函数阻塞直到命令队列中所有命令完成。clFlush()阻塞指导命令队列中的所有命令被移出队列，这意味着这些命令已经准备就绪但无法保证执行完毕。 cl_int clFlush(cl_command_queue command_queue); cl_int clFinish(cl_command_queue command_queue);

3.2、创建缓冲区clCreateBuffer()

cl_mem clCreateBuffer( cl_context context, cl_mem_flags flag, size_t size, void *host_ptr, cl_int *errcode_ret );

context 一个合法的上下文，为这个上下文分配缓冲区 size 所分配缓冲区的大小（字节数） host_ptr 这个指针在clCreateBuffer如何使用有flags参数确定。host_ptr指向的数据大小应大于等于size flags 用于指定关于缓冲区创建的分配和使用信息。其部分取值见下表。

cl_mem_flags描述CL_MEM_READ_WRITE指定内存对象将由内核读写；默认为此模式CL_MEM_WRITE_ONLY指定内存对象由内核写，但不能读。CL_MEM_READ_ONLY指定内存对象由内核读，但不能写。CL_MEM_USE_HOST_PTR只有当host_ptr为非NULL时，这个标志合法；使用host_ptr引用的内存作为内存对象的存储位CL_MEM_ALLOC_HOST_PTR指定缓冲区应当在宿主机可访问的内存中分配。不可与USE_HOST_PTR同时使用CL_MEM_COPY_HOST_PTR表示希望OpenCL实现分配内存对象的内存，并从host_ptr引用复制数据。

3.3、读写缓冲区

cl_int clEnqueueWriteBuffer( cl_command_queue commad_queue, cl_mem buffer, cl_bool blocking_write, size_t offset, size_t cb, void *ptr, cl_uint num_events_in_wait_list, const cl_event *events_in_wait_list, cl_event *event ); cl_int clEnqueueReadBuffer( cl_command_queue commad_queue, cl_mem buffer, cl_bool blocking_read, size_t offset, size_t cb, void *ptr, cl_uint num_events_in_wait_list, const cl_event *events_in_wait_list, cl_event *event );

command_queue 这是一个命令队列,读写命令将在这个队列中排队 buffer 一个合法的缓冲区对象（数据对这里读写） blocking_read 如果设置为CL_TRUE，则命令阻塞,直至ptr读写数据完成 offset 缓冲区对象的读写数据的起始偏移量（字节数） cb 对缓冲区读写的字节数 ptr 宿主机内存的一个指针，写入缓冲区的数据从哪里来 / 或者从缓冲区读数据写入哪里 关于读写的blocking_write的多一些说明： 　　如果blocking_write为CL_TRUE, 　　　　则OpenCL实现将复制ptr引用的数据，并在命令队列中对写操作进行排队。 　　　　在clEnqueueWriteBuffer调用返回后，由ptr指向的内存可以被应用程序重用。 　　如果blocking_write为CL_FALSE, 　　　　则OpenCL实现将使用ptr执行非阻塞写操作。 由于写是非阻塞的，实现可以立即返回。 　　　　ptr指向的内存在调用返回后不能被应用程序重用。 　　　　event参数返回一个事件对象，可以用来查询write命令的执行状态。 　　　　当写命令完成后，ptr指向的内存可以被应用程序重新使用

四、图像创建读写

图像类型的数据我还没有用过，不过好像挺有用的；在此只给出API的介绍。

4.1、图像格式

typedef struct _cl_image_format { cl_channel_order image_channel_order; cl_channel_type iamge_channel_date_type; } cl_image_format ;

The image format describes how the data will be stored in memory 使用示例: cl_image_format format; format.image_channel_order = CL_R; // single channel format.image_channel_data_type = CL_FLOAT; // float data type

4.2、图像创建

cl_mem clCreateImage2D( cl_context context, cl_mem_flags flags, const cl_image_format *image_format size_t image_with, size_t image_height, size_t image_row_pitch, void *host_ptr, cl_int *errcode_ret ); cl_mem clCreateImage3D( cl_context context, cl_mem_flags flags, const cl_image_format *image_format size_t image_with, size_t image_height, size_t image_depth, size_t image_row_pitch, size_t *image_slice_pitch, void *host_ptr, cl_int *errcode_ret );

context 创建图像对象的上下文 flags 其合法枚举由cl_mem_flags定义 image_format 描述通道次序和图像通道数据类型 image_with,image_height,image_depth 图像的长宽深 image_row_pitch 如果host_ptr不为NULL,这个值指定图像中各行的字节数;为0采取默认值 image_slice_pitch 如果host_ptr不为NULL,这个值指定图像中各个切片的字节数;为0采取默认值 host_ptr 内存中线性布局的图像缓冲区指针 errcode_ret 如果为非NULL,函数返回错误码 使用示例: cl_mem d_inputImage = clCreateImage2D(context, 0, &format, imageWidth, imageHeight, 0, NULL, &errNum);

4.3、图像读写

cl_int clEnqueueWriteImage( cl_command_queue commad_queue, cl_mem image, cl_bool blocking_read, const size_t origin[3], const size_t region[3], size_t row_pitch size_t slice_pitch, void *ptr, cl_uint num_events_in_wait_list const cl_event *event_wait_list, cl_event *event );

commad_queue 写入命令将放入这个队列 iamge 这是一个合法的图像对象 blocking_read 如果设置为CL_TRUE,则clEnqueueReadImage阻塞,直到数据读入ptr origin 要写入相对图像原点的(x,y,z)整数坐标,对于二维z=0 region 要写入区域的(宽,高,深),对于二维z=1 row_pitch 图像中各行字节数,默认为image_with*(byte_per_pixel) slice_pitch 三维图像中各切片的字节数image_height*row_pitch ptr 这个指针指向源数据的宿主机内存 使用示例: errNum = clEnqueueWriteImage(queue, d_inputImage, CL_FALSE, origin, region, 0, 0, inputImage, 0, NULL, NULL);

cl_int clEnqueueReadImage( cl_command_queue commad_queue, cl_mem image, cl_bool blocking_read, const size_t origin[3], const size_t region[3], size_t row_pitch size_t slice_pitch, void *ptr, cl_uint num_events_in_wait_list const cl_event *event_wait_list, cl_event *event );

commad_queue 读取命令将放入这个队列 iamge 这是一个合法的图像对象 blocking_read 如果设置为CL_TRUE,则clEnqueueReadImage阻塞,直到数据读入ptr origin 要读取相对原图像原点的(x,y,z)整数坐标,对于二维z=0 region 要读取区域的(宽,高,深),对于二维z=1 row_pitch 图像中各行字节数,默认为image_with*(byte_per_pixel) slice_pitch 三维图像中各切片的字节数image_height*row_pitch ptr 这个指针指向写入数据的宿主机内存 使用示例: errNum = clEnqueueReadImage(queue, d_outputImage, CL_TRUE, origin, region, 0, 0, outputImage, 0, NULL, NULL);

4.4、篇外话

1.要实现图像类数据的使用，好像还需要声明一个采样器（cl_sampler），具体怎么操作我还没有试验过。 2.在内核中读写图像好像需要使用固定的函数read_imagef()和write_imagef()；具体解释参考OpenCL Reference Pages =>OpenCL Compiler =>Built-in Functions => Image Functions =>read_imagef / write_imagef 。