CUDA可以做argmax吗？答案

【问题标题】：Can CUDA do argmax?CUDA可以做argmax吗？
【发布时间】：2011-04-18 23:24:15
【问题描述】：

问题说明了一切；

假设每个线程都在做类似的事情

value=blockDim.x*blockIdx.x+threadIdx.x;
result=f(value);

其中 f 是一个设备函数，通过添加 atomicMax() 调用很容易找到最大结果，但是您如何找出该值是什么？

【问题讨论】：

你有没有研究过归约内核？无论如何，它们可能比 atomicMax 更快。
基本上 f() 正在做一大堆工作，最后我得到了 f() 的输出值，但我想要的是 f 的输入，它给出了最高的输出.我正在想办法解决这个问题，但我们会看到......
是的，我想通了，我只是太慢了。
当你说“最大结果”时，你问的是什么范围？线程、块还是网格？每个人的最佳解决方案可能并不相同。
你可能已经看到了，但是...stackoverflow.com/questions/5720267/…

标签： c cuda mapreduce gpgpu

【解决方案1】：

这有意义吗？只需添加一个 if 语句，将最大结果与线程的结果进行比较。如果匹配，则保存线程的值。

value=blockDim.x*blockIdx.x+threadIdx.x;
result=f(value);
atomicMax(max,result);

if result==*max:
    max_value = value;

或者，如果多个线程具有最大结果，也许您需要指定行为...例如采用最低线程：

value=blockDim.x*blockIdx.x+threadIdx.x;
result=f(value);
atomicMax(max,result);

if result==*max:
    atomicMin(max_value,value);

也就是说，如果您要从每个线程中找到最大结果，您将需要使用归约而不是 atomicMax。如果我理解正确， atomicMax 函数基本上是串行执行的，而减少将主要是并行执行的。当您使用减少时，您可以手动跟踪值以及结果 - 这就是我所做的。 （虽然也许上面的 if 语句方法在归约结束时也可以工作。我可能不得不在我的代码中尝试它......）

【讨论】：

我不太了解实现您所说的最佳方法。我看过使用推力，但从我的阅读来看，没有简单的方法可以有条件地使用最大值（即，如果这个值大于那个值，将我的 id 提供给 max 模板）还有更多提示吗？ :D
我认为我的代码完全符合您的要求。也许如果你把它放在一个真正的内核的上下文中，我可以提供更多帮助。
我在考虑您对缩减内核的评论，但谢谢！（由于内核安排的特殊问题，没有完全以相同的方式实现它，但或多或少与你的相同）
在我的简化中，我有两个共享内存数组：values 和 results。在每次比较中，我都会比较结果，并保存最大结果和产生它的值。所以在归约结束时，results[0] 具有最大结果，values[0] 具有产生它的值。