hmn
/
orca
5
Fork 2
Code Issues 21 Pull Requests 1 Packages Projects Releases Wiki Activity

fix some crashes in metal implementation when pathCount and eltCount are 0 #65

Merged
MartinFouilleul merged 1 commits from macos_metal_fixes into main 2023-08-28 10:16:52 +00:00
Conversation 2 Commits 1 Files Changed 1 +75 -82
1 changed files with 75 additions and 82 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

23
src/graphics/mtl_renderer.m
View File

@ -952,6 +952,11 @@ void oc_mtl_render_batch(oc_mtl_canvas_backend* backend,
int pathCount = backend->pathCount - backend->pathBatchStart;
int eltCount = backend->eltCount - backend->eltBatchStart;
if(!pathCount || !eltCount)
{
return;
}
//NOTE: update intermediate buffers sizes if needed
oc_mtl_grow_buffer_if_needed(backend, &backend->pathQueueBuffer, pathCount * sizeof(oc_mtl_path_queue));
@ -984,8 +989,6 @@ void oc_mtl_render_batch(oc_mtl_canvas_backend* backend,
[blitEncoder endEncoding];
//NOTE: path setup pass
if(pathCount > 0)
{
id<MTLComputeCommandEncoder> pathEncoder = [surface->commandBuffer computeCommandEncoder];
pathEncoder.label = @"path pass";
[pathEncoder setComputePipelineState:backend->pathPipeline];
@ -1006,18 +1009,15 @@ void oc_mtl_render_batch(oc_mtl_canvas_backend* backend,
[pathEncoder dispatchThreads:pathGridSize threadsPerThreadgroup:pathGroupSize];
[pathEncoder endEncoding];
}
int tileOpMax = [backend->tileOpBuffer length] / sizeof(oc_mtl_tile_op);
int segmentMax = [backend->segmentBuffer length] / sizeof(oc_mtl_segment);
//NOTE: segment setup pass
if(eltCount > 0)
{
id<MTLComputeCommandEncoder> segmentEncoder = [surface->commandBuffer computeCommandEncoder];
segmentEncoder.label = @"segment pass";
[segmentEncoder setComputePipelineState:backend->segmentPipeline];
int tileOpMax = [backend->tileOpBuffer length] / sizeof(oc_mtl_tile_op);
int segmentMax = [backend->segmentBuffer length] / sizeof(oc_mtl_segment);
[segmentEncoder setBytes:&eltCount length:sizeof(int) atIndex:0];
[segmentEncoder setBuffer:backend->elementBuffer[backend->bufferIndex] offset:elementBufferOffset atIndex:1];
[segmentEncoder setBuffer:backend->segmentCountBuffer offset:0 atIndex:2];
@ -1038,11 +1038,8 @@ void oc_mtl_render_batch(oc_mtl_canvas_backend* backend,
[segmentEncoder dispatchThreads:segmentGridSize threadsPerThreadgroup:segmentGroupSize];
[segmentEncoder endEncoding];
}
//NOTE: backprop pass
if(pathCount > 0)
{
id<MTLComputeCommandEncoder> backpropEncoder = [surface->commandBuffer computeCommandEncoder];
backpropEncoder.label = @"backprop pass";
[backpropEncoder setComputePipelineState:backend->backpropPipeline];
@ -1057,11 +1054,8 @@ void oc_mtl_render_batch(oc_mtl_canvas_backend* backend,
[backpropEncoder dispatchThreads:backpropGridSize threadsPerThreadgroup:backpropGroupSize];
[backpropEncoder endEncoding];
}
//NOTE: merge pass
if(pathCount > 0)
{
id<MTLComputeCommandEncoder> mergeEncoder = [surface->commandBuffer computeCommandEncoder];
mergeEncoder.label = @"merge pass";
[mergeEncoder setComputePipelineState:backend->mergePipeline];
@ -1085,7 +1079,6 @@ void oc_mtl_render_batch(oc_mtl_canvas_backend* backend,
[mergeEncoder dispatchThreads:mergeGridSize threadsPerThreadgroup:mergeGroupSize];
[mergeEncoder endEncoding];
}
//NOTE: raster pass
id<MTLComputeCommandEncoder> rasterEncoder = [surface->commandBuffer computeCommandEncoder];