No.8 OpenCL 性能分析——内存带宽

测量时间

时钟和时间

在OpenCL 编程实践前面的文章中，为了比较 CPU 和 GPU 的性能，在统计函数的执行时间时使用了 glibc 中提供的时钟和时间相关函数。为了测得 func 函数的执行时间，在待测函数执行前，首先通过调用 system_time() 得到当前的时间戳 time_start，接着执行 func 函数，待其执行完后再次调用 system_time() 得到当前时间戳 time_end，使用 time_end 减去 time_start 就可以得到函数 func 的执行时间。下面是封装后的 system_time() 函数：

int64_t system_time()
{
    struct timespec t;

    clock_gettime(CLOCK_MONOTONIC, &t);
    return (int64_t)(t.tv_sec) * 1e9 + t.tv_nsec;
}

函数 clock_gettime() 的原型如下：

int clock_gettime(clockid_t clk_id, struct timespec *tp);

struct timespec {
    time_t   tv_sec;        /* seconds */
    long     tv_nsec;       /* nanoseconds */
};

参数描述：

• clk_id：表示时钟类型。CLOCK_MONOTONIC 为单调递增时钟，在系统启动后开始计时，其时间不能手动的被重新设置。在 Linux 内核中，CLOCK_MONOTONIC 时钟通过高精度定时器实现，使用 hrtimer_init() 来初始化；
• tp：指向结构体 timespec 的指针，返回当前的时间戳信息。

在 Linux 系统中，内核版本从 Linux 2.6.28 开始，增加了对 CLOCK_BOOTTIME 时钟的支持。该时钟和 CLOCK_MONOTONIC 比较类似，但是它包含了系统在休眠过程中的时间，这在记录的时间戳信息需要包含休眠时间时非常有用。关于时钟的精度可通过调用 clock_getres() 函数来获取，其原型如下：

int clock_getres(clockid_t clk_id, struct timespec *res);

下面是参数描述：

• clk_id：表示对应时钟的类型；
• res：返回该时钟的精度，以纳秒为单位。

在下面的示例代码中，为了更准确的计算函数 clEnqueueNDRangeKernel 的执行时间，在其前后分别调用了 clFinish() 函数。在 clEnqueueNDRangeKernel() 的前面调用 clFinish() 函数࿰