写入 MTLTexture 会导致致命错误

Question

给定一个MTLTexture，定义如下。

// Create device.
id<MTLDevice> dev = MTLCreateDefaultSystemDevice();

// Size of texture.
const unsigned int W = 640;
const unsigned int H = 480;

// Define texture.
MTLTextureDescriptor *desc = [[MTLTextureDescriptor alloc] init];
desc.pixelFormat = MTLPixelFormatBGRA8Unorm;
desc.width = W;
desc.height = H;

// Create texture.
id<MTLTexture> tex = [device newTextureWithDescriptor:desc];

据我了解，此时我应该有一个在 desc 中定义的纹理，分配在设备 dev 上并可通过 [=16] 访问=].

现在，给定另一个纹理 tex2（已知已分配和可访问）和定义如下的 Metal 计算内核。

kernel void foo(texture2d<float, access::read> in [[texture(0)]],
                texture2d<float, access::write> out [[texture(1)]],
                uint2 pix [[thread_position_in_grid]]) {
    // Out of bounds check.
    if (pix.x >= out.get_width() || pix.y >= out.get_height()) {
        return;
    }

    // Do some processing on the input texture.
    // ... All fine up to here.

    // Write out a pixel to the output buffer.
    const float4 p = abc; // abc is computed above.
    out.write(p, pix);
}

据我了解，当像素p写出到out时，p的值将被转换为符合[=16=的像素格式], 在这种情况下 MTLPixelFormatBGRA8Unorm.

但是，如下启动内核时，p写入out的行（上面定义为tex）触发严重错误（SIGABRT).

// Create a Metal library.
id<MTLLibrary> lib = [dev newDefaultLibrary];

// Load the kernel.
id<MTLFunction> kernel = [lib newFunctionWithName:@"foo"];

// Create a pipeline state.
id<MTLComputePipelineState> pipelineState = [dev newComputePipelineStateWithFunction:kernel error:NULL];

// Create a command queue.
id<MTLCommandQueue> cmdQueue = [dev newCommandQueue];

// Create command buffer.
id<MTLCommandBuffer> cmdBuff = [cmdQueue commandBuffer];

// Create compute encoder.
id<MTLComputeCommandEncoder> enc = [cmdBuff computeCommandEncoder];

// Set the pipeline state.
[enc setComputePipelineState:pipelineState];

// Set the input textures (tex2 is read only in the kernel, as above).
[enc setTexture:tex2 atIndex:0];
[enc setTexture:tex atIndex:1];

// 2D launch configuration.
const MTLSize groupDim = MTLSizeMake(16, 16, 1);
const MTLSize gridDim = MTLSizeMake((int)ceil((float)(W / (float)groupDim.width)),
                                    (int)ceil((float)(H / (float)groupDim.height)),
                                    1);

// Launch kernel.
[enc dispatchThreadgroups:gridDim threadsPerThreadgroup:groupDim];
[enc endEncoding];
[enc commit];
[cmdBuff waitUntilCompleted];

我的问题是，在上述情况下，我对如何分配 MTLTexture 的理解是否正确？或者，上面的示例是否只是在我需要单独分配的某些纹理周围定义了一个 wrapper？

Answer 1

以上纹理分配和计算内核启动是正确的。在进一步挖掘文档后，缺少的部分是 MTLTextureDescriptor 的 usage 属性。在documentation中说明如下。

The default value for this property is MTLTextureUsageShaderRead.

因此，在问题中给出的示例中，需要对 MTLTextureDescriptor 进行以下额外的属性赋值。

desc.usage = MTLTextureUsageShaderWrite;

写入 MTLTexture 会导致致命错误

Writing to an MTLTexture causes a fatal error

textures

objective-c

compute-shader

ios

metal