Metal Indirect command buffer with multiple instances and multiple LOD

I have the following Metal shader, that computes an indirect command buffer for a set of different meshes/submeshes, with multiple instances for each.

struct ICBContainer {
    command_buffer icb [[id(0)]];
};

struct ICBMesh {
    // Mesh data
    constant float *vertices;
    constant float *uv;
    constant float *tangents;
    constant float *bitangents;
    // Submesh data
    constant uint *indices;
    constant float *materials;
};

kernel void model_encode(
    uint meshId [[thread_position_in_grid]],
    device ICBContainer *icbContainer [[buffer(BUF_ICB)]],
    constant ICBMesh *meshes [[buffer(BUF_MESHES)]],
    constant MTLDrawIndexedPrimitivesIndirectArguments *draw [[buffer(BUF_DRAW)]],
    constant InstanceData *instances [[buffer(BUF_INSTANCES)]],
    constant Uniforms &uniforms [[buffer(BUF_UNIFORMS)]])
{
    ICBMesh mesh = meshes[meshId];
    MTLDrawIndexedPrimitivesIndirectArguments drawArgs = draw[meshId];
    
    render_command cmd(icbContainer->icb, meshId);
    cmd.set_vertex_buffer(&uniforms, BUF_UNIFORMS);
    cmd.set_vertex_buffer(mesh.vertices, BUF_VERTEX);
    cmd.set_vertex_buffer(mesh.uv, BUF_UV);
    cmd.set_vertex_buffer(mesh.tangents, BUF_TANGENT);
    cmd.set_vertex_buffer(mesh.bitangents, BUF_BITANGENT);
    cmd.set_fragment_buffer(mesh.materials, BUF_MATERIAL);
    cmd.set_vertex_buffer(instances, BUF_INSTANCES);
    cmd.draw_indexed_primitives(
        primitive_type::triangle,
        drawArgs.indexCount,
        mesh.indices + drawArgs.indexStart,
        drawArgs.instanceCount,
        drawArgs.baseVertex,
        drawArgs.baseInstance);
}

It's working fine, but now I would like to provide different meshes at different LOD (0, 1, 2), but with the same list of instances, and make my compute shader calculate the LOD for each instance and filter the draw calls.

I added an argument (lod) to the ICBMesh structure, and the mesh buffer now contains all the different meshes for all different LODs.

struct ICBMesh {
    uint lod;
    // Mesh data
    constant float *vertices;
    constant float *uv;
    constant float *tangents;
    constant float *bitangents;
    // Submesh data
    constant uint *indices;
    constant float *materials;
};

Now, it's calling model_encode() for all LOD versions of my models, with the same list of instances. So, it is rendering all the LOD versions for each instances.

What I would like to do in my shader is calculate the LOD of all my instances (in function of the distance to camera), and draw only the instances whose LOD corresponds to the current mesh's LOD, without having to calculate instance LOD on the CPU and provide different buffers at every frame.

I guess it implies that I make a new array containing the filtered instances and pass this array to the render command and draw call. Something like this:

cmd.set_vertex_buffer(**filteredInstances**, BUF_INSTANCES);
cmd.draw_indexed_primitives(
    primitive_type::triangle,
    drawArgs.indexCount,
    mesh.indices + drawArgs.indexStart,
    **filteredInstancesCount**,
    drawArgs.baseVertex,
    **0**);

But how can I make this array of filtered instances without knowing its size in advance? And does this approach make sense in the first place?

Solution

The solution I have found is to create a buffer (uint *) for each LOD, filled by a new compute shader.

kernel void instance_filter(
uint instanceId [[thread_position_in_grid]],
constant CameraUniforms &camera [[buffer(BUF_CAMERA)]],
constant InstanceData *instances [[buffer(BUF_INSTANCES)]],
device atomic_uint *lod0 [[buffer(BUF_LOD0)]],
device atomic_uint *lod1 [[buffer(BUF_LOD1)]],
device atomic_uint *lod2 [[buffer(BUF_LOD2)]])
{...}

These buffers contain the number of instances (first position) and the list of instance ids I want to draw for each specific LOD, in a compact form, so that I can use them in my instanced draw calls:

kernel void model_encode(
uint meshId [[thread_position_in_grid]],
device ICBContainer *icbContainer [[buffer(BUF_ICB)]],
constant Uniforms &uniforms [[buffer(BUF_UNIFORMS)]],
constant ICBMesh *meshes [[buffer(BUF_MESHES)]],
constant MTLDrawIndexedPrimitivesIndirectArguments *draw [[buffer(BUF_DRAW)]],
constant InstanceData *instances [[buffer(BUF_INSTANCES)]],
constant uint *lod0 [[buffer(BUF_LOD0)]],
constant uint *lod1 [[buffer(BUF_LOD1)]],
constant uint *lod2 [[buffer(BUF_LOD2)]])
{
ICBMesh mesh = meshes[meshId];
MTLDrawIndexedPrimitivesIndirectArguments drawArgs = draw[meshId];
constant uint *indices = lod_select(mesh.lod, lod0, lod1, lod2);
uint numInstances = indices[0];
render_command cmd(icbContainer->icb, meshId);
cmd.set_vertex_buffer(&uniforms, BUF_UNIFORMS);
cmd.set_vertex_buffer(mesh.vertices, BUF_VERTEX);
cmd.set_vertex_buffer(mesh.uv, BUF_UV);
cmd.set_vertex_buffer(mesh.tangents, BUF_TANGENT);
cmd.set_vertex_buffer(mesh.bitangents, BUF_BITANGENT);
cmd.set_fragment_buffer(mesh.materials, BUF_MATERIAL);
cmd.set_vertex_buffer(instances, BUF_INSTANCES);
cmd.set_vertex_buffer(indices, BUF_INDICES);
cmd.draw_indexed_primitives(
    primitive_type::triangle,
    drawArgs.indexCount,
    mesh.indices + drawArgs.indexStart,
    numInstances,
    drawArgs.baseVertex,
    1); // ignore the first element (numInstances)
}

The instanced draw calls now use indices to my filtered list as instance ids. Then, in my vertex shader, I get the instance data from "instances" indirectly, by reading the actual instance id in the passed "indices" buffer.