10.2 使用事件分析OpenCL代码

OpenCL命令队列支持64位的计时命令——使用clEnaueueXX()函数提交，比如:clEnqueueNDRangeKernel()。通常，命令入队都是异步的，并且开发者可以使用事件的方式对命令进行状态追踪。事件对象提供了一种方式来了解命令的执行过程。事件中记录了命令的很多相关信息，比如何时入队、何时提交到设备上、何时开始运行，以及何时执行完成。通过事件的信息获取函数——clGetEventProfilingInfo()，其能提供命令的相关计时信息：

使用事件对象显式的对OpenCL程序进行计，需要对对应的命令队列进行计时使能的操作。在创建命令队列的时候，需要设置CL_QUEUE_PROFILING_ENABLE标识。一旦命令命令队列创建完成，就无法在对事件计时的功能进行开启或关闭。

cl_int clGetEventProfilingInfo(
  cl_event event,
  cl_profling_info param_name,
  size_t param_value_size,
  void *param_value,
  size_t *param_value_size_ret)

第一个参数，event事件对象时必须给定的，第二个参数是一个枚举值，用来描述描述所要获取相应的时间信息。具体的值如表10.1所示。

表10.1 对应的命令状态可以用来获取OpenCL事件的时间戳

事件状态	param_value返回的信息
CL_PROFILING_COMMAND_QUEUE	使用一个64位的值对主机端将命令提交到命令队列的时间进行统计(单位：ns)
CL_PROFILING_COMMAND_SUBMIT	使用一个64位的值对命令从命令队列提交到相关的设备上的时间进行统计(单位：ns)
CL_PROFILING_COMMAND_START	使用一个64位的值对命令开始的时间进行记录(单位：ns)
CL_PROFILING_COMMAND_END	使用一个64位的值对命令完成的时间进行记录(单位：ns)
CL_PROFILING_COMMAND_COMPLETE	使用一个64位的值对命令及其相关子命令完成的时间进行记录(单位：ns)

如之前所述，OpenCL命令队列是异步工作的，因此函数在命令入队时就返回了。所以在对事件对象进行计时查询时，需要调用一次clFinish()，以同步相关任务，让队列中的所有任务都完成。下面一段简单的代码展示了，如何使用事件的方式对内核进行性能分析。

代码清单10.1 使用OpenCL事件获取内核的时间信息