可以将___shfl_xor替换为在sm_21上运行？

-1

试图运行这样的：在一个NVS4200M，这是sm_21，不sm_35根据需要 https://github.com/Celebrandil/CudaSift 。在运行中提到的项目的唯一问题是此代码（cudaSiftD.cu:205）：可以将___shfl_xor替换为在sm_21上运行？

对（INT I = 1;我< = 16; I * = 2）总和+ = __shfl_xor（总和，一世）;

是否有可能的等效代码？

来源

2017-04-06 Alex Fliker

是的，如果你喜欢写它。 – talonmies

通过共享内存操作，您可以使用随机操作完成的任何操作都可以完成，这也允许进行线程间通信。我并不是说实现是相同的，只是有一个“可能的等价代码”使用共享内存。 –

@talonmies这个评论如何帮助OP？这是一个不平凡的问题，因为我不认为将内部函数作为cuda的一个简单特性来搅乱。 –

好，几乎任何CUDA内在的可以更换，所以我会解释为

你的问题可以被__shfl_xor对SM_21的GPU取代便宜？

而答案是：不是真的;你会受到惩罚。你最好的选择，因为@ RobertCrovella的意见建议是使用共享内存：

每个通道的数据写入到一个位置，在共享内存中（使这些连续4个字节大小的值，以避免bank conflicts）
执行某种同步（可能您必须__syncthreads()）
每个通道从共享内存位置读取其所需值已写入的通道。

我没拼出来的代码不采取乐趣离开你:-)

编辑：虽然洗牌的执行比较复杂，它是静止的，语义至少，对寄存器的操作;并且它不需要同步。所以共享内存的选择会更慢。

来源

2017-04-06 20:23:29 einpoklum

我不会考虑洗牌是一个时钟周期，原因有两个：1）在多处理器上每个周期有32个可发布的洗牌[http://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html ＃算术指令]，2）shuffle操作由管理共享内存的高速缓存执行。实质上，与共享内存相比，使用shuffle大约是性能的两倍 - 请参见[http：//on-demand.gputechconf。com/gtc/2013/presentations/S3174-Kepler-Shuffle-Tips-Tricks.pdf] –

@FlorentDUGUET：编辑以反映您的评论。您的链接不起作用，但我认为您的圆括号有一些错字。 – einpoklum

上述评论中的链接断开：http://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#arithmetic-instructions和http://on-demand.gputechconf.com/gtc/ 2013 /演示文稿/ S3174-Kepler-Shuffle-Tips-Tricks.pdf –

如果问题更多的是关于如何用与sm_21兼容的代码替换这段代码，您可能需要关注CUB，block-reduce部分here。其中一个模板参数是设备的体系结构。

该__CUDA_ARCH__宏可以帮助您选择最合适的实现，请参阅here。

来源

2017-04-07 09:49:11

可以将___shfl_xor替换为在sm_21上运行？

回答

相关问题