將緩衝區寫入設備時出現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。

從你對問題的描述中我假設你實際上反覆啓動了一個內核，但是我沒有看到相應的代碼。

最可能的原因是： 您在內核中的內存訪問超出了界限並損壞了內存。 由於您不評估錯誤代碼並繼續執行程序，因此驅動程序遲早會報告錯誤（或者只是崩潰），但從這裏開始，我們可能已經在處理未定義的行爲，所以並不重要司機說。