c++ - Opencl 独立队列和 openmp 部分中的 clFinish

Question

我正在制作一个 opencl 包装器，它为每个缓冲区使用不同的命令队列。

所有读/写操作都是非阻塞的。

问题是，其中一个读/写操作未发出(从分析器看来)。 CodeXL 探查器显示漏洞。我在图片上做了标记。

问题:是什么导致了这个问题？我检查了 cl 错误，但都给出了 CL_SUCCESS。

c++ opencl 1.2程序流程:

  (from main thread)
   issue write buffer 0
   issue write buffer 1
   issue write buffer N

   (from many threads by openmp body)
   clFinish(queue0)
   clFinish(queue1)
   clFinish(queue2)

   (from main thread)
   issue kernel 0
   issue kernel 1
   issue kernel N
   clFinish(kernelQueue)

  (from main thread)
   issue read buffer 0
   issue read buffer 1
   issue read buffer N

   (from many threads by openmp body)
   clFinish(queue0)
   clFinish(queue1)
   clFinish(queue2) ----> has some holes!!!
   clFinish(queueN)

这是主线程的样子:

void Air::update_air_and_h_gpu(void)
{
    if (airMode != 4) 
    {




            /*  update air*/

        gpu->WriteToClFrom2DCPP("pv",&pv[0][0]);    
        gpu->WriteToClFrom2DCPP("vx",&vx[0][0]);    
        gpu->WriteToClFrom2DCPP("vy",&vy[0][0]);    
        gpu->WriteToClFrom2DCPP("fvx",&fvx[0][0]);  
        gpu->WriteToClFrom2DCPP("fvy",&fvy[0][0]);      
        gpu->WriteToClFrom2DCPP("bmap_blockair",&bmap_blockair[0][0]);  
        gpu->WriteToClFrom2DCPP("bmap",&bmap[0][0]);    
        gpu->WriteToClFromCPP("kernelArr",&kernel[0]);


            /* update heat*/    
        gpu->WriteToClFrom2DCPP("hv",&hv[0][0]);    
        gpu->WriteToClFrom2DCPP("ambientAirTemp",&(ambientAirTemp));    
        gpu->WriteToClFrom2DCPP("gravityMode",&(sim.gravityMode));  
        gpu->WriteToClFrom2DCPP("bmap_blockairh",&bmap_blockairh[0][0]);


        gpu->syncW();

        /* update air*/
        gpu->Compute("UpdateAirReduceEdge0");
        gpu->Compute("UpdateAirReduceEdge1");
        gpu->Compute("UpdateAirClearVelWall");
        gpu->Compute("UpdateAirPressAdjVel");
        gpu->Compute("UpdateAirVelAdjPress");
        gpu->Compute("UpdateAirBigLoop");
        gpu->Compute("UpdateAirBigLoop2");

        /* update heat*/
        gpu->Compute("UpdateAirHeatLoop0");
        gpu->Compute("UpdateAirHeatLoop1");
        gpu->Compute("UpdateAirHeatLastLoop");
        gpu->Compute("UpdateAirHeatLastLoop2");

        gpu->sync();

        /*update air*/
        gpu->ReadFromClTo2DCPP("pv",&pv[0][0]);
        gpu->ReadFromClTo2DCPP("vx",&vx[0][0]);
        gpu->ReadFromClTo2DCPP("vy",&vy[0][0]);

        /* update heat*/
        gpu->ReadFromClTo2DCPP("hv",&hv[0][0]);


        gpu->syncR();



    }

}

以下是 gpu->syncR() 和 gpu->syncW 是如何通过 openmp 完成的(也在选项中激活):

void syncR()
{
    omp_set_num_threads(2);  //tried 8 but holes still exist
    #pragma omp parallel for
    for(int i=0;i<cqR.size();i++)
    {
        cqR[i].finish();
    }
}

void syncW()
{
    omp_set_num_threads(2); //tried 8 but holes still exist
    #pragma omp parallel for
    for(int i=0;i<cqW.size();i++)
    {

        cqW[i].finish();
    }
}

仔细看看:

请看右侧的蓝色条，其中一个有时会丢失。

设备:HD7870主机:FX8150

Opencl 1.2 C++ 绑定(bind)。

Answer 1

恕我直言，为每个内存对象使用单独的命令队列看起来不是最佳实践。在这种情况下，您很少在每个队列的范围内发出命令。命令无法重新排序、放入传送带等。简而言之，您无法从海量数据传输中受益，而且您正在将 CPU 资源浪费在多个队列处理上。我的建议是使用多个具有 OpenMP 线程安全访问的命令队列:

cl_command_queue 
    q_DtoH = clCreateCommandQueue(...),
    q_DtoD = clCreateCommandQueue(...)
    ...;

#pragma omp parallel for
for(;;)
{
    #pragma omp critical
    clEnqueueNDWriteBuffer(q_DtoH, ...); //Or whatever else

    ....
}

理想情况下，尝试使用具有事件同步的无序队列进行数据传输。此外，对象引用计数对于调试目的很方便(假设您的命令队列使用相同的上下文，您可以检查它的引用计数)。

并且，在您的代码片段中:

/* update heat*/
....
gpu->Compute("UpdateAirHeatLoop0");
gpu->sync();

/*update air*/
gpu->ReadFromClTo2DCPP("pv",&pv[0][0]);
...

为什么不使用相同的命令队列来发出内核执行和数据传输？这将使您摆脱同步点。