我很困惑为什么我的内核着色器不起作用。
我有真正的原始RGBA32像素缓冲区(inBuffer)，我将其发送到内核着色器。 我还有一个接收MTLTexture，我将它的用法设置为在其RGBA8Norm描述符中MTLTextureUsageRenderTarget。
然后我这样发送编码...
id<MTLLibrary> library = [_device newDefaultLibrary];
id<MTLFunction> kernelFunction = [library newFunctionWithName:@"stripe_Kernel"];
id<MTLComputePipelineState> pipeline = [_device newComputePipelineStateWithFunction:kernelFunction error:&error];
id<MTLCommandQueue> commandQueue = [_device newCommandQueue];
MTLTextureDescriptor *textureDescription = [MTLTextureDescriptor texture2DDescriptorWithPixelFormat:MTLPixelFormatRGBA8Unorm
width:outputSize.width
height:outputSize.height
mipmapped:NO];
[textureDescription setUsage:MTLTextureUsageRenderTarget];
id<MTLTexture> metalTexture = [_device newTextureWithDescriptor:textureDescription];
MTLSize threadgroupCounts = MTLSizeMake(8, 8, 1);
MTLSize threadgroups = MTLSizeMake([metalTexture width] / threadgroupCounts.width,
[metalTexture height] / threadgroupCounts.height, 1);
...
id<MTLBuffer> metalBuffer = [_device newBufferWithBytesNoCopy:inBuffer
length:inputByteCount
options:MTLResourceStorageModeShared
deallocator:nil];
[commandEncoder setComputePipelineState:pipeline];
[commandEncoder setTexture:metalTexture atIndex:0];
[commandEncoder setBuffer:metalBuffer offset:0 atIndex:0];
[commandEncoder setBytes:&imageW length:sizeof(ushort) atIndex:1];
[commandEncoder setBytes:&imageH length:sizeof(ushort) atIndex:2];
[commandEncoder dispatchThreadgroups:threadgroups threadsPerThreadgroup:threadgroupCounts];
[commandEncoder endEncoding];
[commandBuffer commit];
[commandBuffer waitUntilCompleted];
目的是获取大小为mxn的原始图像，并将其打包成 2048x896 的纹理。 这是我的内核着色器：
kernel void stripe_Kernel(texture2d<float, access::write> outTexture [[ texture(0) ]],
device const float *inBuffer [[ buffer(0) ]],
device const ushort * imageWidth [[ buffer(1) ]],
device const ushort * imageHeight [[ buffer(2) ]],
uint2 gid [[ thread_position_in_grid ]])
{
const ushort imageW = *imageWidth;
const ushort imageH = *imageHeight;
const uint32_t textureW = outTexture.get_width(); // eg. 2048
uint32_t posX = gid.x; // eg. 0...2047
uint32_t posY = gid.y; // eg. 0...895
uint32_t sourceX = ((int)(posY/imageH)*textureW + posX) % imageW;
uint32_t sourceY = (int)(posY% imageH);
const uint32_t ptr = (sourceX + sourceY* imageW);
float pixel = inBuffer[ptr];
outTexture.write(pixel, gid);
}
后来我抓住了那个纹理缓冲区并将其转换为CVPixelBuffer：
MTLRegion region = MTLRegionMake2D(0, 0, (int)outputSize.width, (int)outputSize.height);
// lock buffers, copy texture over
CVPixelBufferLockBaseAddress(outBuffer, 0);
void *pixelData = CVPixelBufferGetBaseAddress(outBuffer);
[metalTexture getBytes:CVPixelBufferGetBaseAddress(outBuffer)
bytesPerRow:CVPixelBufferGetBytesPerRow(outBuffer)
fromRegion:region
mipmapLevel:0];
CVPixelBufferUnlockBaseAddress(outBuffer, 0);
我的问题是，我的CVPixelBuffer总是空的(已分配但为零)。 在配备 Radeon M395 GPU 的 iMac 17,1 上运行。
我什至将不透明的红色像素撞到内核着色器的输出纹理中。 不过，我什至没有看到红色。
更新：我对这个问题的解决方案是完全放弃使用MTLTextures(我什至尝试了与MTLBlitCommandEncoder同步的纹理) - 没有骰子。
我最终将 MTLBuffers 用于输入"纹理"和输出"纹理"，并在内核着色器中重新设计了数学。 我的输出缓冲区现在是一个预先分配的、锁定的 CVPixelBuffer，无论如何我最终都想要的。
首先，使用MTLTextureUsage.renderTarget，我收到错误"validateComputeFunctionArguments：825：失败的断言'函数写入纹理(outTexture[0])，其用法(0x04)未指定MTLTextureUsageShaderWrite (0x02)'"，所以它可能应该是MTLTextureUsage.shaderWrite。
出于某种原因，如果我使用 gfxSwitch 强制英特尔 GPU，纹理的回读会返回正确的数据，但对于 Radeon，无论"textureDesc.resourceOptions = MTLResourceOptions.storageModeXXX"标志，它始终为零。
英特尔和Radeon 460对我有用的是创建一个MTLBuffer并使用它而不是纹理。不过，您必须计算指数。如果您不使用 mip 映射或带有浮点索引的采样，切换到缓冲区应该没什么大不了的，对吧？
让 texBuffer = 设备？。makeBuffer(长度：4 * 宽度 * 高度，选项：MTLResourceOptions.storageModeShared)
变量结果 = [浮点型](重复：0，计数：宽度 * 高度 * 4) let data = NSData(bytesNoCopy： texBuffer！.内容()， 长度： 4 * 宽度 * 高度， freeWhenDone： false) data.getBytes(&result， length： 4 * width * height)
我认为创建由 MTLBuffer 支持的纹理会起作用，但 api 仅在 OSX 10.13 中。
编辑：正如Ken Thomases所指出的那样，在Metal内核上也有类似的讨论，在新的MacBook Pro(2016年底)GPU上表现不佳
我使用此线程第一篇文章中的方法和着色器制作了一个示例应用程序，并且链接线程的修复程序对我有用。这是应用程序代码的链接，以防有人想要一个可重现的示例。 https://gist.github.com/astarasikov/9e4f58e540a6ff066806d37eb5b2af29