floating-point - NVIDIA GPU 的 OpenCL 中浮点值的原子加法？

Question

OpenCL 3.0 规范 does not seem to have用于对浮点值进行原子加法的内在函数/内置函数，仅适用于整数值(在 OpenCL 1.x 和 2.x 中似乎也是如此)。然而，CUDA has offered浮点原子现在有一段时间了:

float  atomicAdd(float*  address, float  val); // since Fermi
double atomicAdd(double* address, double val); // since Pascal
__half atomicAdd(__half *address, __half val); // ?

自然地，任何简单的原子操作都可以通过比较和交换来模拟，而这在 OpenCL 中是可用的。但我的问题是:

1. NVIDIA 是否以某种方式在 OpenCL 中公开了浮点原子？例如通过供应商扩展？使用编译指示？含蓄地？
2. 是否有比比较交换模拟更有效的机制，我可以考虑将其作为浮点原子的替代品？对于 NVIDIA GPU 还是一般？

Answer 1

native 浮点原子是 OpenCL 3.0 非常需要的扩展。截至目前，它们仍然不可用。

1. 唯一可能的方法是使用 inline PTX .
2. 没有。 implementation with atomic compare-exchange for FP32 and FP64目前是最先进的，没有更好的方法。

2022 年 6 月更新:Floating-point atomics are being added to the OpenCL 3.0 standard但硬件供应商的采用可能仍需要一些时间。