将缓冲区写入设备时出现OpenCL访问冲突

Question

我在OpenCL中有一个项目。它是GPU上的矩阵分解。所有的工作正常，结果是好的。我看到的唯一情况是，当我连续多次执行程序（每秒大约一次）时，当我将初始缓冲区写入设备时，会遇到访问冲突。

它总是在写入它被卡住的缓冲区。我对OpenCL非常陌生，我想知道如果我退出程序时是否必须清除GPU中的内存？有时它在第一次运行时崩溃，但在2或3次尝试后成功。然后再次，有时立即成功，以及随后的运行。这只是非常随机的。失败的实际缓冲区写入也会不时发生变化。有时第三个缓冲区写入失败，有时是第四个。

我运行此程序的参数是工作组大小为7和70 * 70元素的矩阵。起初我认为这可能是因为我的矩阵对于GPU来说太大（GT650M有2GB），但是有时候也会使用一个矩阵为O.000的元素来运行。

直到缓冲区写入的代码如下所示。

任何帮助，非常感谢。 Ps：为了清楚起见，PRECISION是一个宏#define PRECISION float。

int main(int argc, char *argv[]) 
{ 
    //////////////////////////////////////////////////////////////////////////////////////////////////////////////// 
    //// INITIALIZATION PART /////////////////////////////////////////////////////////////////////////////////////// 
    //////////////////////////////////////////////////////////////////////////////////////////////////////////////// 
    try { 
     if (argc != 5) { 
      std::ostringstream oss; 
      oss << "Usage: " << argv[0] << " <kernel_file> <kernel_name> <workgroup_size> <array width>"; 
      throw std::runtime_error(oss.str()); 
     } 
     // Read in arguments. 
     std::string kernel_file(argv[1]); 
     std::string kernel_name(argv[2]); 
     unsigned int workgroup_size = atoi(argv[3]); 
     unsigned int array_dimension = atoi(argv[4]); 
     int total_matrix_length = array_dimension * array_dimension; 

     int total_workgroups = total_matrix_length/workgroup_size; 
     total_workgroups += total_matrix_length % workgroup_size == 0 ? 0 : 1; 

     // Print parameters 
     std::cout << "Workgroup size: " << workgroup_size  << std::endl; 
     std::cout << "Total workgroups: " << total_workgroups << std::endl; 
     std::cout << "Array dimension: " << array_dimension  << " x " << array_dimension << std::endl; 
     std::cout << "Total elements: " << total_matrix_length << std::endl; 


     // OpenCL initialization 
     std::vector<cl::Platform> platforms; 
     std::vector<cl::Device> devices; 
     cl::Platform::get(&platforms); 
     platforms[0].getDevices(CL_DEVICE_TYPE_GPU, &devices); 
     cl::Context context(devices); 
     cl::CommandQueue queue(context, devices[0], CL_QUEUE_PROFILING_ENABLE); 

     // Load the kernel source. 
     std::string file_text; 
     std::ifstream file_stream(kernel_file.c_str()); 
     if (!file_stream) { 
      std::ostringstream oss; 
      oss << "There is no file called " << kernel_file; 
      throw std::runtime_error(oss.str()); 
     } 
     file_text.assign(std::istreambuf_iterator<char>(file_stream), std::istreambuf_iterator<char>()); 

     // Compile the kernel source. 
     std::string source_code = file_text; 
     std::pair<const char *, size_t> source(source_code.c_str(), source_code.size()); 
     cl::Program::Sources sources; 
     sources.push_back(source); 
     cl::Program program(context, sources); 
     try { 
      program.build(devices); 
     } 
     catch (cl::Error& e) { 
      getchar(); 
      std::string msg; 
      program.getBuildInfo<std::string>(devices[0], CL_PROGRAM_BUILD_LOG, &msg); 
      std::cerr << "Your kernel failed to compile" << std::endl; 
      std::cerr << "-----------------------------" << std::endl; 
      std::cerr << msg; 
      throw(e); 
     } 
     //////////////////////////////////////////////////////////////////////////////////////////////////////////////// 
     //// CREATE RANDOM INPUT DATA ////////////////////////////////////////////////////////////////////////////////// 
     //////////////////////////////////////////////////////////////////////////////////////////////////////////////// 

     // Create matrix to work on. 
     // Create a random array. 
     int matrix_width   = sqrt(total_matrix_length); 
     PRECISION* random_matrix = new PRECISION[total_matrix_length]; 
     random_matrix   = randommatrix(total_matrix_length); 
     PRECISION* A    = new PRECISION[total_matrix_length]; 

     for (int i = 0; i < total_matrix_length; i++) 
      A[i] = random_matrix[i]; 

     PRECISION* L_SEQ = new PRECISION[total_matrix_length]; 
     PRECISION* U_SEQ = new PRECISION[total_matrix_length]; 
     PRECISION* P_SEQ = new PRECISION[total_matrix_length]; 

     // Do the sequential algorithm. 
     decompose(A, L_SEQ, U_SEQ, P_SEQ, matrix_width); 
     float* PA = multiply(P_SEQ, A, total_matrix_length); 
     float* LU = multiply(L_SEQ, U_SEQ, total_matrix_length); 
     std::cout << "PA = LU?" << std::endl; 
     bool eq = equalMatrices(PA, LU, total_matrix_length); 
     std::cout << eq << std::endl; 
     //////////////////////////////////////////////////////////////////////////////////////////////////////////////// 
     //// RUN AND SETUP KERNELS ///////////////////////////////////////////////////////////////////////////////////// 
     //////////////////////////////////////////////////////////////////////////////////////////////////////////////// 

     // Initialize arrays for GPU. 
     PRECISION* L_PAR = new PRECISION[total_matrix_length]; 
     PRECISION* U_PAR = new PRECISION[total_matrix_length]; 
     PRECISION* P_PAR = new PRECISION[total_matrix_length]; 

     PRECISION* ROW_IDX = new PRECISION[matrix_width]; 
     PRECISION* ROW_VAL = new PRECISION[matrix_width]; 
     // Write A to U and initialize P. 
     for (int i = 0; i < total_matrix_length; i++) 
      U_PAR[i] = A[i]; 
     // Initialize P_PAR. 
     for (int row = 0; row < matrix_width; row++) 
     { 
      for (int i = 0; i < matrix_width; i++) 
       IDX(P_PAR, row, i) = 0; 
      IDX(P_PAR, row, row) = 1; 
     } 
     // Allocate memory on the device 
     cl::Buffer P_BUFF(context, CL_MEM_READ_WRITE, total_matrix_length*sizeof(PRECISION)); 
     cl::Buffer L_BUFF(context, CL_MEM_READ_WRITE, total_matrix_length*sizeof(PRECISION)); 
     cl::Buffer U_BUFF(context, CL_MEM_READ_WRITE, total_matrix_length*sizeof(PRECISION)); 
     // Buffer to determine maximum row value. 
     cl::Buffer MAX_ROW_IDX_BUFF(context, CL_MEM_READ_WRITE, total_workgroups*sizeof(PRECISION)); 
     cl::Buffer MAX_ROW_VAL_BUFF(context, CL_MEM_READ_WRITE, total_workgroups*sizeof(PRECISION)); 

     // Create the actual kernels. 
     cl::Kernel kernel(program, kernel_name.c_str()); 

     std::string max_row_kernel_name = "max_row"; 
     cl::Kernel max_row(program, max_row_kernel_name.c_str()); 
     std::string swap_row_kernel_name = "swap_row"; 
     cl::Kernel swap_row(program, swap_row_kernel_name.c_str()); 

     // transfer source data from the host to the device 
     std::cout << "Writing buffers" << std::endl; 
     queue.enqueueWriteBuffer(P_BUFF, CL_TRUE, 0, total_matrix_length*sizeof(PRECISION), P_PAR); 
     queue.enqueueWriteBuffer(L_BUFF, CL_TRUE, 0, total_matrix_length*sizeof(PRECISION), L_PAR); 
     queue.enqueueWriteBuffer(U_BUFF, CL_TRUE, 0, total_matrix_length*sizeof(PRECISION), U_PAR); 

     queue.enqueueWriteBuffer(MAX_ROW_IDX_BUFF, CL_TRUE, 0, total_workgroups*sizeof(PRECISION), ROW_IDX); 
     queue.enqueueWriteBuffer(MAX_ROW_VAL_BUFF, CL_TRUE, 0, total_workgroups*sizeof(PRECISION), ROW_VAL); 

完整的错误，当我与调试器钩子，我得到的是这样的：

Unhandled exception at 0x55903CC0 (nvopencl.dll) in Project.exe: 
0xC0000005: Access violation reading location 0x0068F004. 

If there is a handler for this exception, the program may be safely continued. 

调试器显示我的下面，在命名空间cl功能：

cl_int enqueueWriteBuffer(
    const Buffer& buffer, 
    cl_bool blocking, 
    ::size_t offset, 
    ::size_t size, 
    const void* ptr, 
    const VECTOR_CLASS<Event>* events = NULL, 
    Event* event = NULL) const 
{ 
    return detail::errHandler(
     ::clEnqueueWriteBuffer(
      object_, buffer(), blocking, offset, size, 
      ptr, 
      (events != NULL) ? (cl_uint) events->size() : 0, 
      (events != NULL && events->size() > 0) ? (cl_event*) &events->front() : NULL, 
      (cl_event*) event), 
      __ENQUEUE_WRITE_BUFFER_ERR); 

编辑：完整源代码here。

Answer 1

看看这些行：

PRECISION* ROW_IDX = new PRECISION[matrix_width]; 
... 
cl::Buffer MAX_ROW_IDX_BUFF(context, CL_MEM_READ_WRITE, total_workgroups*sizeof(PRECISION)); 
... 
queue.enqueueWriteBuffer(MAX_ROW_IDX_BUFF, CL_TRUE, 0, total_workgroups*sizeof(PRECISION), ROW_IDX); 

所以，你想要写total_workgroups元素的缓冲区，但你的源阵列只与matrix_width要素分配。对于您提到的输入参数（工作组大小为7的70x70阵列），这将试图从70*4字节数组中读取700*4字节的数据字节 - 确定的内存访问冲突。

后来在您的代码中，您正在从同一个缓冲区读取相同的主机数组，这会破坏内存，并在我自己的系统上运行代码时导致其他崩溃和无法解释的行为。

Answer 2

仅仅因为排队缓冲区时发生错误，它不一定是原因。你可能已经修复了你的内存，并且由于入队过程而出现错误（很像CPU内存损坏，免费调用导致错误）。

您所有的CL函数都会返回错误代码，通过将它们与CL_SUCCESS进行比较来评估它们（OpenCL file, containing all error codes）。例如，如果内核调用确实损坏了内存，enqueueReadBuffer通常返回CL_INVALID_COMMAND_QUEUE。

从你对问题的描述中我假设你实际上反复启动了一个内核，但是我没有看到相应的代码。

最可能的原因是： 您在内核中的内存访问超出了界限并损坏了内存。 由于您不评估错误代码并继续执行程序，因此驱动程序迟早会报告错误（或者只是崩溃），但从这里开始，我们可能已经在处理未定义的行为，所以并不重要司机说。