metal - 如何将 float 组传递给 Metal 内核函数？

Question

我不想使用texture1d_array。我可以简单地传递一个 float 组吗？我将把它写入我的内核函数中。

Answer 1

为了写入内核函数内的 float 组，您需要向内核提供一个缓冲区参数。该参数应具有 device float * 类型，并具有 buffer 属性，指定它将占用哪个参数表槽:

kernel void my_kernel(device float *data [[buffer(0)]],
                      uint threadIndex [[thread_position_in_grid]])
{
    data[threadIndex] = /* calculate value for this element */;
}

要在您的应用代码中创建此类缓冲区，请请求您的 Metal 设备分配它:

let buffer = device.makeBuffer(length: MemoryLayout<Float>.stride * dataCount,
                               options: [])!

在 Mac 上，您可能需要使用 .storageModeManaged 选项创建缓冲区，该选项不会自动将您从内核写入的值同步回 CPU 可读内存。您可以使用 blit 编码器和 synchronize(resource:) 方法从 GPU 内存复制回来。在 iOS 上，不存在托管缓冲区，并且除了通常情况之外不需要同步(确保您永远不会从其他人正在写入的同一位置读取数据)。

当您准备好分派(dispatch)计算工作时，请将缓冲区绑定(bind)为计算命令编码器的参数:

computeCommandEncoder.setBuffer(buffer, offset: 0, index: 0)

调度任何大小的网格以完成工作。在命令缓冲区中对您可能需要执行的任何其他工作(包括同步命令)进行编码，提交命令缓冲区，并确保它在尝试读取缓冲区内容之前已完成。

要读取缓冲区的内容，请将缓冲区的内容转换为适当类型的 UnsafeMutableBufferPointer，这样您就可以像对待任何其他 Sequence 一样对待缓冲区>:

let data = UnsafeMutableBufferPointer<Float>(start: buffer.contents().assumingMemoryBound(to: Float.self),
                                             count:dataCount)
// iterate over elements of data or whatever...