gpu是否可以跨多个线程同时更改一个变量？

Question

看看这段代码，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，金属将a增加2，-5-（3 * 2）= 1）

任何答案将不胜感激：）

Answer 1

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

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

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