cuda - CUDA 共享内存中的并行扫描

Question

我有一个问题，每个线程 block (一维)必须对共享内存内的一个数组进行扫描，并执行几个其他任务。 (该数组最多有 1024 个元素。)

有没有支持这种操作的好库？

我检查了 Thrust 和 CudPP 但它们都只在数据最初位于全局内存中时才工作，这不是我想要的，因为我不想为很少的工作启动一个新内核并将数据复制回全局内存，因为它有相当大的开销。

如果没有支持此类操作的公共(public)库，那么如果我只想编写一次此操作并在几个不同的问题中使用它，您有什么建议？

我的第一个想法是编写一个简单的设备函数来运行扫描操作，但是是否可以从不是内核函数的设备函数访问共享内存并执行 __syncthreads() 操作？
我的另一个想法是在宏中编写整个函数，然后它会起作用，因为预处理器将代码复制到内核代码，但编写如此复杂的宏并不是使用宏的最佳方式。

PS.:我有流式多处理器版本 2.1，所以我不能从另一个启动新内核。

Answer 1

Is there any good library which support this type of operation?

If there is no public library which support this type of operations, than what do you suggest if I want to write this operation only once and use it in several different problems?

有一个名为 CUB 的库 - 代表 CUDA UnBound。你找到它 here .它提供扫描功能。
如果您想自己编写类似的东西，CUB 可能有助于了解它是如何完成的。

My first idea is to write a simple host function what runs the scan opertaion, but is it possible to accessing shared memory and doing __syncthreads() operations from a device function which is not the kernel function?

我不确定“运行扫描操作的主机功能”是什么意思。
您可以从主机代码中做的唯一一件事就是启动内核和内存副本。除非从拥有相关共享内存的 block 中调用设备函数，否则答案是否定的。
共享内存是按 block 分配的，并且只能从 block 内部访问。