beignet 上的 OpenCL 超时不会引发错误？

Question

我运行以下（简化的）代码，它运行简化的内核几秒钟，然后检查结果。前 400,000 个左右的结果是正确的，然后接下来的结果都是零。内核应该将相同的值 (4228) 放入 450 万个元素的输出数组的每个元素中。看起来不知何故，某处某事超时，或未同步，但我有点困惑，因为我：

• 甚至调用 clFinish，只是为了确保
• 正在检查所有错误，没有返回错误

结果如下：

user@pear:~/git/machinelearning/prototyping/build$ ./testcltimeout 
out[442496] != 4228: 0

我期望发生的是：代码应该运行完成，没有错误。

上下文：运行于：

• 贝涅特，OpenCL 1.2
• Intel HD 4000 集成显卡

内核是：

kernel void test_read( const int one,  const int two, global int *out) {
    const int globalid = get_global_id(0);
    int sum = 0;
    int n = 0;
    while( n < 100000 ) {
        sum = (sum + one ) % 1357 * two;
        n++;
    }
    out[globalid] = sum;
}

测试代码（我已经尽可能地简化了……）

#include <iostream>
#include <sstream>
#include <stdexcept>
using namespace std;

#include "CL/cl.hpp"

template<typename T>
std::string toString(T val ) {
   std::ostringstream myostringstream;
   myostringstream << val;
   return myostringstream.str();
}

void checkError( cl_int error ) {
    if (error != CL_SUCCESS) {
       throw std::runtime_error( "Error: " + toString(error) );
    }
}

int main( int argc, char *argv[] ) {

     cl_int error;  

    cl_device_id *device_ids;

    cl_uint num_platforms;
    cl_uint num_devices;

    cl_platform_id platform_id;
    cl_device_id device;

    cl_context context;
    cl_command_queue queue;
    cl_program program;

    checkError( clGetPlatformIDs(1, &platform_id, &num_platforms) );
    checkError(  clGetDeviceIDs(platform_id, CL_DEVICE_TYPE_GPU, 1, &device, &num_devices) );
    device_ids = new cl_device_id[num_devices];
    checkError( clGetDeviceIDs(platform_id, CL_DEVICE_TYPE_GPU, num_devices, device_ids, &num_devices) );
    device = device_ids[0];
    context = clCreateContext(0, 1, &device, NULL, NULL, &error);
    checkError(error);
    queue = clCreateCommandQueue(context, device, 0, &error);
    checkError(error);

    string kernel_source = string( "kernel void test_read( const int one,  const int two, global int *out) {\n" ) +
    "    const int globalid = get_global_id(0);\n" +
    "    int sum = 0;\n" +
    "    int n = 0;\n" +
    "    while( n < 100000 ) {\n" +
    "        sum = (sum + one ) % 1357 * two;\n" +
    "        n++;\n" +
    "    }\n" +
    "    out[globalid] = sum;\n" +
    "}\n";
    const char *source_char = kernel_source.c_str();
    size_t src_size = strlen( source_char );
    program = clCreateProgramWithSource(context, 1, &source_char, &src_size, &error);
    checkError(error);

    checkError( clBuildProgram(program, 1, &device, 0, NULL, NULL) );

    cl_kernel kernel = clCreateKernel(program, "test_read", &error);
    checkError(error);

    const int N = 4500000;
    int *out = new int[N];
    if( out == 0 ) throw runtime_error("couldnt allocate array");

    int c1 = 3;
    int c2 = 7;
    checkError( clSetKernelArg(kernel, 0, sizeof(int), &c1 ) );
    checkError( clSetKernelArg(kernel, 1, sizeof(int), &c2 ) );
    cl_mem outbuffer = clCreateBuffer(context, CL_MEM_WRITE_ONLY, sizeof(int) * N, 0, &error);
    checkError(error);
    checkError( clSetKernelArg(kernel, 2, sizeof(cl_mem), &outbuffer) );

    size_t globalSize = N;
    size_t workgroupsize = 512;
    globalSize = ( ( globalSize + workgroupsize - 1 ) / workgroupsize ) * workgroupsize;
    checkError( clEnqueueNDRangeKernel( queue, kernel, 1, NULL, &globalSize, &workgroupsize, 0, NULL, NULL) );
    checkError( clFinish( queue ) );
    checkError( clEnqueueReadBuffer( queue, outbuffer, CL_TRUE, 0, sizeof(int) * N, out, 0, NULL, NULL) );    
    checkError( clFinish( queue ) );

    for( int i = 0; i < N; i++ ) {
       if( out[i] != 4228 ) {
           cout << "out[" << i << "] != 4228: " << out[i] << endl;
           exit(-1);
       }
    }

    return 0;
}

Answer 1

你的内核似乎很长。我怀疑您正在 TDR（超时）超时，Linux（Beignet）比 Windows 更安静地处理这个问题。因此，我有几个想法。

• 检查dmesg 以获取 TDR 消息。我没有使用 Beignet 或 Linux OpenCL 实现，但Beignet documentation 页面（在“已知问题”下）表明您可以通过dmesg 进行检查。

要检查 GPU 是否挂起，可以执行 dmesg 并检查是否 它有以下消息：[17909.175965] [drm:i915_hangcheck_hung] 错误 挂起检查计时器已过...如果是这样，则 GPU 挂起。通常，这意味着内核中有问题，因为它表明 OCL 内核大约有 6 秒甚至更长时间没有完成。

文档继续说，如果您确实知道内核只是需要更长的时间才能完成，您可以禁用超时检查，但警告您有机器挂起的风险。

• 在 Windows 上试用 Intel HD 4000 显卡。如果内核花费的时间超过几秒钟，它将超时并且驱动程序实际上会崩溃（但会自动重启）。

• 尝试使用 Intel OpenCL CPU 实现的内核（或任何其他没有 TRD 限制的内核）。检查正确性和运行时长（10 秒？10 分钟？）。我认为 CPU 实现不会超时。