0
我试图编写一个OpenCL实现memchr
,以帮助我了解OpenCL的工作方式。我打算做的是为每个工作项分配一大块内存进行搜索。然后,在每个工作项目中,它循环搜索该字符的块。在OpenCL中提前退出
尤其是如果缓冲区很大,我不希望其他线程在发现事件后继续搜索(假设任何给定缓冲区中只有一个字符出现)。
我卡在的是一个工作项目如何显示,主机和其他线程,当它已经找到了字符?
谢谢,
我试图编写一个OpenCL实现memchr
,以帮助我了解OpenCL的工作方式。我打算做的是为每个工作项分配一大块内存进行搜索。然后,在每个工作项目中,它循环搜索该字符的块。在OpenCL中提前退出
尤其是如果缓冲区很大,我不希望其他线程在发现事件后继续搜索(假设任何给定缓冲区中只有一个字符出现)。
我卡在的是一个工作项目如何显示,主机和其他线程,当它已经找到了字符?
谢谢,
你可以做到这一点的一种方法是使用全局标志变量。当你找到这个值的时候你自动将它设置为1,而其他线程在他们工作时会检查这个值。
例如:
__kernel test(__global int* buffer, __global volatile int* flag)
{
int tid = get_global_id(0);
int sx = get_global_size(0);
int i = tid;
while(buffer[i] != 8) //Whatever value we're trying to find.
{
int stop = atomic_add(&flag, 0); //Read the atomic value
if(stop)
break;
i = i + sx;
}
atomic_xchg(&flag, 1); //Set the atomic value
}
这可能比只运行整个内核(除非你是做了很多工作,在每次迭代)增加更多的开销。另外,如果每个线程只检查数组中的单个值,则此方法将不起作用。每个线程必须有多个工作迭代。
最后,我已经看到写入原子变量的实例不会立即提交,因此您需要检查此代码是否会在系统上死锁,因为写入没有提交。
假设停止不需要是精确的,我可以跳过原子函数并非原子使用它们来节省开销吗? – tangrs
严格来说,根据OpenCL规范,不能保证其他工作项目将不会在没有某种同步点的情况下看到“停止”指标。但是,如果您尝试设置并检查__global变量而不是原子,您可能会发现它在您的设备上可靠和/或性能更高。 –