有没有一种方法可以让 gpu 跨多个线程同时更改一个变量？答案

【问题标题】：Is there a way where the gpu can change one variable simultaneously across multiple threads?有没有一种方法可以让 gpu 跨多个线程同时更改一个变量？
【发布时间】：2020-05-25 01:32:35
【问题描述】：

看看这段代码， main.swift：

import Metal

let device = MTLCreateSystemDefaultDevice()!
let lib = device.makeDefaultLibrary()!
let function = lib.makeFunction(name: "testout")!
let pso : MTLComputePipelineState
do {
    pso = try device.makeComputePipelineState(function: function)
}

let commandQueue = device.makeCommandQueue()!
// generate buffer
let buffer = device.makeBuffer(length: MemoryLayout<Float>.stride, options: .storageModeShared)!
// set the variable "a" in our testout.metal to the value -5
buffer.contents().assumingMemoryBound(to: Float.self).assign(repeating: -5, count: 1)

let commandBuffer = commandQueue.makeCommandBuffer()!
let computeEncoder = commandBuffer.makeComputeCommandEncoder()!
computeEncoder.setComputePipelineState(pso)
computeEncoder.setBuffer(buffer, offset: 0, index: 0)

// run it 3 times
let gridSize = MTLSizeMake(3, 1, 1)

computeEncoder.dispatchThreads(gridSize, threadsPerThreadgroup: gridSize)
computeEncoder.endEncoding()
commandBuffer.commit()
commandBuffer.waitUntilCompleted()

print(buffer.contents().assumingMemoryBound(to: Float.self).pointee)

testout.metal:

#include <metal_stdlib>
using namespace metal;

kernel void testout (device float* a) {
    a[0] += 2;
}

有没有可以输出1的函数/程序？甚至可能吗？（运行3次，原值为-5，metal增加a 2，-5 + (3 * 2) = 1）

任何答案将不胜感激:)

【问题讨论】：

您将使用原子操作。见here。
谢谢，但是你能把初始值设置为小于 0 的值吗？
如果你使用有符号类型（atomic_int 在着色器中，Int32 用于 Swift 中的缓冲区类型），那么可以。
谢谢，但是否有一段苹果文档说明 Int32 有效，而不是 Int 用于原子操作？如果没有，你是怎么知道的？（只是想更好地自己解决问题）
正如我在链接的答案中提到的，它没有记录原子类型的内存表示是什么。 Metal Shading Language spec 确实记录了 int 是 32 位类型。 Swift 的 Int 并不总是 32 位的。

标签： gpu metal

【解决方案1】：

感谢 Ken Thomases，我所要做的就是将 testout.metal 更改为：

kernel void testout (device atomic_int &a) {
    atomic_fetch_add_explicit(&a, 2, memory_order_relaxed);
}

在主脚本中，我不得不将缓冲区的数据类型更改为Int32。

【讨论】：