0

我正在编写一个 openCL 内核来为益智游戏执行基于蛮力的 AI,但我的内核代码和/或它调用的辅助函数有问题。这是我的内核代码(我确信输入在这里正确传递): 60 是 clEnqueueNDRangeKernel 设置的全局工作大小。

内核的输入如下:

__global char * in //用于测试目的的虚拟输入

__global char * board_in, // 一个包含 60 个板的大型 char 数组

__global int * lookup, // 一个数组,我用来快速获取得分动作的分数

输出:

__global char * out, //用于测试的虚拟输出

__global int * score_out, //60 个分数的数组:每个棋盘一个

__global int * row_out, // 一个 60 行的数组:每个评估的板一个

__global int * col_out // 一个 60 列的数组: ...

__kernel void helloworld(__global char* in,
                    __global char* board_in,
                    __global int* lookup,
                    __global char* out, 
                    __global int * score_out,
                    __global int * row_out,
                    __global int * col_out)
{

    int num = get_global_id(0);
    char workingBoard[72];
    int scoreMat[64];
//set up the array for each thread to use
    for(int k=0; k< 72; k++)
    {
        workingBoard[k] = board_in[num*BOARDSIZE+k];
    }
// Make a copy of the score matrix for each thread to use
    for(int j=0; j<64; j++)
    {
        scoreMat[j] = lookup[j];
    }
    int s=0;
    int r=0;
    int c=0;
    findBestMove(workingBoard,scoreMat,&s,1,&r,&c);
    col_out[num] = ?????????
    score_out[num] = ???????????
    row_out[num] = ???????????????
}

函数 findBestMove 的工作原理是这样的(它经过了很好的测试。我在 CPU 实现中使用了一段时间):它需要一个 Board(字符数组)、一个得分查找数组、一个指向移动得分的指针、当前深度,以及指向行和列的指针。它应该设置分数,行和列。它调用我在同一个文档中定义的其他函数。

如果我在 CPU 上运行这个代码片段,我会得到正确的输出:

// workerBoard and lookuparr are set previous to this to be the same as what
//the kernel thread is supposed to have
int s=0;
int r=0;
int c=0;
findBestMove(workerBoard,lookuparr,&s,1,&r,&c);
cout<<s<<","<<r<<","<<c<<endl;

当我运行我的内核代码时,我不会让它通过函数调用。该函数与内核在同一个文档中定义,并且不使用动态内存、函数指针、递归或全局内存(在内核参数之外)。我确实使用了一些#define 语句。

我要设置???我的内核部分是 r、c 和 s,但如前所述,我没有到达那里。我是否犯了任何严重错误(注意:内核通过了我的代码检查器和 AMD 的内核分析器)。另外,我对 openCL 还很陌生,所以也欢迎任何提示。如果我可以提供更多信息来帮助回答这个问题,请告诉我!

4

1 回答 1

0

根据您的评论,问题似乎出在您的 findBestMove 函数中。顺便说一句,如果你有一个无限循环,看门狗会在某一时刻触发,你的驱动程序很可能会崩溃,导致黑屏或死机。

因此,我建议您在函数中注释所有代码,并为 r、s、c 变量分配一个选定的值,例如使用 get_global_id 函数处理这些特定变量的工作项 id。当然要换???和:

  col_out[num] = c;
  score_out[num] = s;
  row_out[num] = r;

如果你得到正确的值,开始调试你的函数,你会确定你的问题出在函数中。

由于您在这里询问了一些提示,我认为这是一个可以提高性能的提示(一旦您修复了错误:)):不要为 scoreMat 数组使用私有内存,而是使用本地内存。这样做可以避免让每个线程一遍又一遍地访问全局内存中的相同数据(这很慢)。要从全局内存中获取数据到本地内存,您可以使用async_work_group_copy函数。

所以在你的情况下,你会有这样的事情:

local int scoreMat[64];
event_t ev = async_work_group_copy(lookup, scoreMat, 64, 0); 
// Wait to make sure everything is copied   
wait_group_events (1, &ev);

您可能需要更改更多代码以考虑您现在使用的本地内存。基本上它的工作方式与全局方式相同(从访问的角度来看),但速度要快得多。

请注意,与您所拥有的不同之处在于只会制作一份副本,而不是 60 个(工作项的数量)。这一次,您从全局获取的数据也可以从工作组中的所有工作项访问。在每个工作项都有自己的副本之前。强调它在工作组中的事实也很重要。但由于您只使用 60 个工作项,您很可能只有一个工作组。

于 2013-07-27T21:55:39.653 回答