OpenCL的kernel

最终阶段是通过从cl_program中抽取出kernel，获取在设备上执行的的cl_kernel对象。从程序中抽取kernel类似于从动态库中导出函数。程序导出的kernel名称要求在程序对象中编译。kernel名称和程序对象传递给clCreateKernel()，如果程序对象有效并且指定的kernel已被发现，则会有kernel对象返回。

与常规C程序中函数调用不同，kernel在被执行前还需要一些步骤。我们不能简单地提供一些参数就可以调用kernel。

kernel的执行需要通过一个入队函数来进行分发。由于C语言的语法和kernel参数是持久的(因此我们不必重复设置它们来构建调度的参数列表)，我们必须分别使用函数clSetKernelArg()指定每个kernel参数，该函数接受一个kernel对象，参数索引、大小和指针。当kernel执行时，这些信息作为参数传输给设备。然后，运行时将kernel参数列表中的类型信息拆箱成(类似于类型转换)适当的类型。设置kernel参数的过程，将在本章结尾处提供的源代码清单中步骤9进行详细描述。

所需要的内存对象全部都传输到设备且kernel参数设置之后，通过调用clEnqueueNDRangeKernel()来开始执行kernel：

看看这个函数的签名。必须先指定一个命令队列，以便发现目标设备。

类似地，kernel对象识别出将要执行的代码。有四个字段与work-item新建相关， work_dim参数指定新建work-item的维度(1，2，3)。global_work_size参数指定NDRange中每维work-item的数量，local_work_size指定 workgroup中每维work-item的数量。可以使用参数global_work_offset为work-item提供全局的ID，该参数不从0开始。和所有clEnqueue命令一样，需要提供event_wait_list。非NULL值时，运行时将保证所有相应的事件在kernel开始执行之前完成。clEnqueueNDRangeKernel()调用是异步的：当命令进入命令队列中后立即返回，甚至可能在kernel已经开始执行之前。clWaitForEvents()或clFinish()可以阻塞直到kernel完成。在本章结尾处提供的源代码清单中，步骤10显示了work-item的配置，向量相加kernel入队的命令则出现在步骤11中。

至此，我们已经介绍了运行一个完整的OpenCL程序所需要的所有主机端API命令。