# 3.6。支持的原子操作
> 原文: [http://numba.pydata.org/numba-doc/latest/cuda/intrinsics.html](http://numba.pydata.org/numba-doc/latest/cuda/intrinsics.html)
Numba 提供了对`numba.cuda.atomic`类中 CUDA 支持的一些原子操作的访问。
目前实施的内容如下:
```py
class numba.cuda.atomic
```
用于原子操作的命名空间
```py
class add(ary, idx, val)
```
执行原子 ary [idx] + = val。仅在 int32,float32 和 float64 操作数上受支持。
返回索引位置的旧值,就像它以原子方式加载一样。
```py
class compare_and_swap(ary, old, val)
```
如果当前值与`old`匹配,则有条件地将`val`分配给 1D 数组`ary`的第一个元素。
返回当前值,就像它以原子方式加载一样。
```py
class max(ary, idx, val)
```
执行原子 ary [idx] = max(ary [idx],val)。 NaN 被视为缺失值,因此 max(NaN,n)== max(n,NaN)== n。请注意,这与 Python 和 Numpy 行为不同,其中当 a 或 b 是 NaN 时,max(a,b)始终为 a。
仅在 int32,int64,uint32,uint64,float32,float64 操作数上受支持。
返回索引位置的旧值,就像它以原子方式加载一样。
```py
class min(ary, idx, val)
```
执行原子 ary [idx] = min(ary [idx],val)。 NaN 被视为缺失值,因此 min(NaN,n)== min(n,NaN)== n。请注意,这与 Python 和 Numpy 行为不同,其中 min(a,b)始终是 a 或 b 是 NaN 时的行为。
仅在 int32,int64,uint32,uint64,float32,float64 操作数上受支持。
## 3.6.1。示例
以下代码演示了如何使用 [`numba.cuda.atomic.max`](../cuda-reference/kernel.html#numba.cuda.atomic.max "numba.cuda.atomic.max") 查找数组中的最大值。请注意,在这种情况下,这不是找到最大值的最有效方法,但它是一个例子:
```py
from numba import cuda
import numpy as np
@cuda.jit
def max_example(result, values):
"""Find the maximum value in values and store in result[0]"""
tid = cuda.threadIdx.x
bid = cuda.blockIdx.x
bdim = cuda.blockDim.x
i = (bid * bdim) + tid
cuda.atomic.max(result, 0, values[i])
arr = np.random.rand(16384)
result = np.zeros(1, dtype=np.float64)
max_example[256,64](result, arr)
print(result[0]) # Found using cuda.atomic.max
print(max(arr)) # Print max(arr) for comparision (should be equal!)
```
使用索引的元组元组支持多维数组:
```py
@cuda.jit
def max_example_3d(result, values):
"""
Find the maximum value in values and store in result[0].
Both result and values are 3d arrays.
"""
i, j, k = cuda.grid(3)
# Atomically store to result[0,1,2] from values[i, j, k]
cuda.atomic.max(result, (0, 1, 2), values[i, j, k])
arr = np.random.rand(1000).reshape(10,10,10)
result = np.zeros((3, 3, 3), dtype=np.float64)
max_example_3d[(2, 2, 2), (5, 5, 5)](result, arr)
print(result[0, 1, 2], '==', np.max(arr))
```
- 1. 用户手册
- 1.1。 Numba 的约 5 分钟指南
- 1.2。概述
- 1.3。安装
- 1.4。使用@jit 编译 Python 代码
- 1.5。使用@generated_jit 进行灵活的专业化
- 1.6。创建 Numpy 通用函数
- 1.7。用@jitclass 编译 python 类
- 1.8。使用@cfunc 创建 C 回调
- 1.9。提前编译代码
- 1.10。使用@jit 自动并行化
- 1.11。使用@stencil装饰器
- 1.12。从 JIT 代码 中回调到 Python 解释器
- 1.13。性能提示
- 1.14。线程层
- 1.15。故障排除和提示
- 1.16。常见问题
- 1.17。示例
- 1.18。会谈和教程
- 2. 参考手册
- 2.1。类型和签名
- 2.2。即时编译
- 2.3。提前编译
- 2.4。公用事业
- 2.5。环境变量
- 2.6。支持的 Python 功能
- 2.7。支持的 NumPy 功能
- 2.8。与 Python 语义的偏差
- 2.9。浮点陷阱
- 2.10。 Python 2.7 寿命终止计划
- 3. 用于 CUDA GPU 的 Numba
- 3.1。概述
- 3.2。编写 CUDA 内核
- 3.3。内存管理
- 3.4。编写设备功能
- 3.5。 CUDA Python 中支持的 Python 功能
- 3.6。支持的原子操作
- 3.7。随机数生成
- 3.8。设备管理
- 3.10。示例
- 3.11。使用 CUDA 模拟器 调试 CUDA Python
- 3.12。 GPU 减少
- 3.13。 CUDA Ufuncs 和广义 Ufuncs
- 3.14。共享 CUDA 内存
- 3.15。 CUDA 阵列接口
- 3.16。 CUDA 常见问题
- 4. CUDA Python 参考
- 4.1。 CUDA 主机 API
- 4.2。 CUDA 内核 API
- 4.3。内存管理
- 5. 用于 AMD ROC GPU 的 Numba
- 5.1。概述
- 5.2。编写 HSA 内核
- 5.3。内存管理
- 5.4。编写设备功能
- 5.5。支持的原子操作
- 5.6。代理商
- 5.7。 ROC Ufuncs 和广义 Ufuncs
- 5.8。示例
- 6. 扩展 Numba
- 6.1。高级扩展 API
- 6.2。低级扩展 API
- 6.3。示例:间隔类型
- 7. 开发者手册
- 7.1。贡献给 Numba
- 7.2。 Numba 建筑
- 7.3。多态调度
- 7.4。关于发电机的注意事项
- 7.5。关于 Numba Runtime 的注意事项
- 7.6。使用 Numba Rewrite Pass 获得乐趣和优化
- 7.7。实时变量分析
- 7.8。上市
- 7.9。模板注释
- 7.10。关于自定义管道的注意事项
- 7.11。环境对象
- 7.12。哈希 的注意事项
- 7.13。 Numba 项目路线图
- 8. Numba 增强建议
- 9. 术语表