金属内核在新款 MacBook Pro（2016 年末）GPU 上无法正常运行

Question

我正在研究使用 Swift 和 Metal 在 GPU 上进行图像处理的 macOS 项目。上周，我收到了我的新 15 英寸 MacBook Pro（2016 年末）并注意到我的代码有些奇怪：应该写入纹理的内核似乎没有这样做......

经过大量挖掘，我发现问题与 Metal（AMD Radeon Pro 455 或 Intel(R) HD Graphics 530）使用哪个 GPU 进行计算有关。

使用 MTLCopyAllDevices() returns 代表 Radeon 和 Intel GPU 的设备数组初始化 MTLDevice（而 MTLCreateSystemDefaultDevice() returns 默认设备是 Radeon）。在任何情况下，代码在 Intel GPU 上都按预期工作，但在 Radeon GPU 上却不是这样。

举个例子。

首先，这是一个简单的内核，它采用输入纹理并将其颜色复制到输出纹理：

    kernel void passthrough(texture2d<uint, access::read> inTexture [[texture(0)]],
                            texture2d<uint, access::write> outTexture [[texture(1)]],
                            uint2 gid [[thread_position_in_grid]])
    {
        uint4 out = inTexture.read(gid);
        outTexture.write(out, gid);
    }

我为了使用这个内核，我使用了这段代码：

    let devices = MTLCopyAllDevices()
    for device in devices {
        print(device.name!) // [0] -> "AMD Radeon Pro 455", [1] -> "Intel(R) HD Graphics 530"
    }

    let device = devices[0] 
    let library = device.newDefaultLibrary()
    let commandQueue = device.makeCommandQueue()

    let passthroughKernelFunction = library!.makeFunction(name: "passthrough")

    let cps = try! device.makeComputePipelineState(function: passthroughKernelFunction!)

    let commandBuffer = commandQueue.makeCommandBuffer()
    let commandEncoder = commandBuffer.makeComputeCommandEncoder()

    commandEncoder.setComputePipelineState(cps)

    // Texture setup
    let width = 16
    let height = 16
    let byteCount = height*width*4
    let bytesPerRow = width*4
    let region = MTLRegionMake2D(0, 0, width, height)
    let textureDescriptor = MTLTextureDescriptor.texture2DDescriptor(pixelFormat: .rgba8Uint, width: width, height: height, mipmapped: false)

    // inTexture
    var inData = [UInt8](repeating: 255, count: Int(byteCount))
    let inTexture = device.makeTexture(descriptor: textureDescriptor)
    inTexture.replace(region: region, mipmapLevel: 0, withBytes: &inData, bytesPerRow: bytesPerRow)

    // outTexture
    var outData = [UInt8](repeating: 128, count: Int(byteCount))
    let outTexture = device.makeTexture(descriptor: textureDescriptor)
    outTexture.replace(region: region, mipmapLevel: 0, withBytes: &outData, bytesPerRow: bytesPerRow)

    commandEncoder.setTexture(inTexture, at: 0)
    commandEncoder.setTexture(outTexture, at: 1)
    commandEncoder.dispatchThreadgroups(MTLSize(width: 1,height: 1,depth: 1), threadsPerThreadgroup: MTLSize(width: width, height: height, depth: 1))

    commandEncoder.endEncoding()
    commandBuffer.commit()
    commandBuffer.waitUntilCompleted()

    // Get the data back from the GPU
    outTexture.getBytes(&outData, bytesPerRow: bytesPerRow, from: region , mipmapLevel: 0)

    // Validation
    // outData should be exactly the same as inData 
    for (i,outElement) in outData.enumerated() {
        if outElement != inData[i] {
            print("Dest: \(outElement) != Src: \(inData[i]) at \(i))")
        }
    }

当运行此代码与 let device = devices[0] (Radeon GPU) 时，outTexture 永远不会写入（我的假设），因此 outData 保持不变。另一方面，当运行此代码与 let device = devices[1] (Intel GPU) 时，一切都按预期工作，并且 outData 更新为 inData 中的值。

Answer 1

我认为每当 GPU 写入 MTLStorageModeManaged 资源（例如纹理）然后您想要从 CPU 读取该资源（例如使用 getBytes()）时，您需要使用 blit 编码器同步它。尝试将以下内容放在 commandBuffer.commit() 行上方：

let blitEncoder = commandBuffer.makeBlitCommandEncoder()
blitEncoder.synchronize(outTexture)
blitEncoder.endEncoding()

如果集成 GPU 上没有这个，你可能会逃脱，因为 GPU 正在使用系统内存作为资源，没有什么可以同步的。

金属内核在新款 MacBook Pro（2016 年末）GPU 上无法正常运行

Metal kernels not behaving properly on the new MacBook Pro (late 2016) GPUs

macos

kernel

gpu

swift

metal