Compute Kernel Metal - 如何检索结果和调试？

Question

我已经下载了苹果的 truedepth 流媒体示例，并正在尝试添加一个计算管道。我想我正在检索计算结果，但不确定，因为它们似乎都为零。

我是iOS开发的初学者，所以可能会有很多错误，请多多包涵！

管道设置：（我不太确定如何创建结果缓冲区，因为内核输出一个 float3）

int resultsCount = CVPixelBufferGetWidth(depthFrame) * CVPixelBufferGetHeight(depthFrame);

//because I will be output 3 floats for each value in depthframe
id<MTLBuffer> resultsBuffer = [self.device newBufferWithLength:(sizeof(float) * 3 * resultsCount) options:MTLResourceOptionCPUCacheModeDefault];


_threadgroupSize = MTLSizeMake(16, 16, 1);

// Calculate the number of rows and columns of threadgroups given the width of the input image
// Ensure that you cover the entire image (or more) so you process every pixel
_threadgroupCount.width  = (inTexture.width  + _threadgroupSize.width -  1) / _threadgroupSize.width;
_threadgroupCount.height = (inTexture.height + _threadgroupSize.height - 1) / _threadgroupSize.height;

// Since we're only dealing with a 2D data set, set depth to 1
_threadgroupCount.depth = 1;

id<MTLComputeCommandEncoder> computeEncoder = [commandBuffer computeCommandEncoder];

[computeEncoder setComputePipelineState:_computePipelineState];

[computeEncoder setTexture: inTexture atIndex:0];

[computeEncoder setBuffer:resultsBuffer offset:0 atIndex:1];

[computeEncoder setBytes:&intrinsics length:sizeof(intrinsics) atIndex:0];

[computeEncoder dispatchThreadgroups:_threadgroupCount
                       threadsPerThreadgroup:_threadgroupSize];

[computeEncoder endEncoding];


// Finalize rendering here & push the command buffer to the GPU
[commandBuffer commit];

//for testing
[commandBuffer waitUntilCompleted];

我添加了以下计算内核：

kernel void
calc(texture2d<float, access::read>  inTexture  [[texture(0)]],
                device float3 *resultsBuffer [[buffer(1)]],
                constant float3x3& cameraIntrinsics [[ buffer(0) ]],
                uint2 gid [[thread_position_in_grid]])
{

    float val = inTexture.read(gid).x * 1000.0f;

    float xrw = (gid.x - cameraIntrinsics[2][0]) * val / cameraIntrinsics[0][0];
    float yrw = (gid.y - cameraIntrinsics[2][1]) * val / cameraIntrinsics[1][1];

    int vertex_id = ((gid.y * inTexture.get_width()) + gid.x);

    resultsBuffer[vertex_id] = float3(xrw, yrw, val);

}

查看缓冲区结果的代码：（我尝试了两种不同的方式，目前都输出全零）

    void *output = [resultsBuffer contents];
    for (int i = 0; i < 10; ++i) {
        NSLog(@"value is %f", *(float *)(output) ); //= *(float *)(output + 4 * i);
    }

    NSData *data = [NSData dataWithBytesNoCopy:resultsBuffer.contents length:(sizeof(float) * 3 * resultsCount)freeWhenDone:NO];
    float *finalArray = new float [resultsCount * 3];
    [data getBytes:&finalArray[0] length:sizeof(finalArray)];
    for (int i = 0; i < 10; ++i) {
        NSLog(@"here is output %f", finalArray[i]);
    }

Answer 1

我在这里看到了几个问题，但它们都与您的 Metal 代码本身无关。

在您的第一个输出循环中，正如所写的那样，您只是将结果缓冲区的第一个元素打印了 10 次。第一个元素可能合法地为 0，导致您相信所有结果都为零。但是当我将第一个日志行更改为

NSLog(@"value is %f", ((float *)output)[i]);

在测试映像上运行内核时，我看到打印了不同的值。

另一个问题与您的getBytes:length: 通话有关。您想传递要复制的字节数，但sizeof(finalArray) 实际上是finalArray 指针 的大小，即4 个字节，而不是总大小它指向的缓冲区。这是 C 和 C++ 代码中极为常见的错误。

相反，您可以使用与分配空间时相同的字节数：

[data getBytes:&finalArray[0] length:(sizeof(float) * 3 * resultsCount)];

然后您应该会发现打印的值与上一步中相同（非零）。