2012-11-06 42 views
0

我试图编写一个OpenCL实现memchr,以帮助我了解OpenCL的工作方式。我打算做的是为每个工作项分配一大块内存进行搜索。然后,在每个工作项目中,它循环搜索该字符的块。在OpenCL中提前退出

尤其是如果缓冲区很大,我不希望其他线程在发现事件后继续搜索(假设任何给定缓冲区中只有一个字符出现)。

我卡在的是一个工作项目如何显示,主机和其他线程,当它已经找到了字符?

谢谢,

回答

1

你可以做到这一点的一种方法是使用全局标志变量。当你找到这个值的时候你自动将它设置为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 
} 

这可能比只运行整个内核(除非你是做了很多工作,在每次迭代)增加更多的开销。另外,如果每个线程只检查数组中的单个值,则此方法将不起作用。每个线程必须有多个工作迭代。

最后,我已经看到写入原子变量的实例不会立即提交,因此您需要检查此代码是否会在系统上死锁,因为写入没有提交。

+2

假设停止不需要是精确的,我可以跳过原子函数并非原子使用它们来节省开销吗? – tangrs

+4

严格来说,根据OpenCL规范,不能保证其他工作项目将不会在没有某种同步点的情况下看到“停止”指标。但是,如果您尝试设置并检查__global变量而不是原子,您可能会发现它在您的设备上可靠和/或性能更高。 –