OpenCL内核调试的最佳方法

Question

我有一个需要调试的OpenCL内核。

我在其中添加了一些printf，但这些printf并不有用，因为工作项是随机调度的，打印出来的值并不总是正确的。

我如何使我的内核中的工作项串行执行以进行调试?

以下是代码

__kernel
void SampleKernel( __global float4* gVtx,  __global float4* gColor,  
                 __global float4* gDst,
                 const int cNvtx, 
                 const int4 cRes )
                 {
                   printf("nVertex : %d ", cNvtx);

                   for(int i =0 ; i < 1; i+=4)
                   {

                   printf(" %f ",  gVtx[0].x);

                   printf(" %f ",  gVtx[0].y);

                   printf(" %f ",  gVtx[0].z);

                   printf(" %f ",  gVtx[0].w);

                   }

                 }

我还尝试在printf之前和之后放置调用barrier(CLK_LOCAL_MEM_FENCE | CLK_GLOBAL_MEM_FENCE);，但这并没有起到作用。 请问是否有办法使工作项执行串行化，以便于我可以打印和调试内核？或者有更好的OpenCL内核调试方法吗？我正在使用RX 580 AMD GPU。

- Paritosh Kulkarni

Answer 1

一些建议： 你可以使用全局ID和组ID来控制要打印哪个线程，当你打印时，也打印出线程和组ID。这将显著减少打印信息的复杂性，并使您更好地控制您可能需要的信息。

另一个提示是，请尽可能将多个打印合并为一个；例如，如果我们像下面这样使用print，那么这不是一个好的调试方法。

               printf(" %f ",  gVtx[0].x);

               printf(" %f ",  gVtx[0].y);

               printf(" %f ",  gVtx[0].z);

               printf(" %f ",  gVtx[0].w);

为避免来自其他线程的打印干扰，最好一次性将它们全部打印出来。

使用以上两个提示，调试内核可能会更容易处理。

- Robert Wang

---

谢谢这些提示。还有一件事，我能否为了调试目的而序列化执行？ - Paritosh Kulkarni

一种串行执行的方法是指定1x1的全局工作大小和1x1的工作组大小。 - Robert Wang