如何在没有索引的情况下访问第 1 层参数缓冲区结构

Question

根据 Apple 的这个示例，无法通过指针索引 (https://developer.apple.com/documentation/metal/buffers/about_argument_buffers) 访问第 1 层参数缓冲区。如果不允许这样做，我怎样才能索引到我的参数缓冲区数组中的特定结构？

// Shader.metal
struct MyTexture {
    metal::texture2d<float, metal::access::sample> texture;
};

fragment half4 myFragment(VertexOut vert [[stage_in]],
                          ....,
                          constant int &count [[buffer(4)]],
                          constant MyTexture *textures [[buffer(5)]],
                          ....)
{
   for(int i = 0; i < count; i++) {
       MyTexture resource = textures[i];
       
       float4 color = resource.texture.sample(sampler, pos.xy);
       outputColor = mix(inputColor, color, 0.5); // <-- Causes error
   }
}

我得到的错误是使用此错误消息创建 MTLRenderPipelineState：

由于使用了间接参数 bufferbuffer(5)，所以内联所有函数：使用非零数组索引访问的参数缓冲区。

Answer 1

简短的回答：你不能。

您不能这样做的原因是第 1 层硬件只能使用常规绑定点模拟参数缓冲区。使用 tier2，您可以在那里绑定任意数量的纹理，因此驱动程序在绑定时无法知道它需要使用多少个插槽，并且硬件本身无法对其他 GPU 对象（如纹理和采样器）进行依赖读取.

Answer 2

第 1 层的解决方法是将指针传递给参数缓冲区内的实例，而不是整个缓冲区。

例如，看Material的用法。

// Argument-buffered resource
struct Material {
  metal::sampler sampler [[id(AB_MaterialSampler)]];
  metal::texture2d<float> base_color_texture [[id(AB_MaterialBaseColorTexture)]];
  metal::texture2d<float> normal_map [[id(AB_MaterialNormalMap)]];
  metal::texture2d<float> ao_metallic_roughness_map [[id(AB_MaterialAoMetallicRoughnessMap)]];
  float3 base_color_factor [[id(AB_MaterialBaseColorFactor)]];
  float metallic_factor [[id(AB_MaterialMetallicFactor)]];
  float roughness_factor [[id(AB_MaterialRoughnessFactor)]];
};

// GPU-driven rendering kernel
kernel void icb_frame_kernel(device IcbContainer& icb_container [[buffer(KB_IcbContainer)]],
                             constant VertexUniforms* vertex_uniforms [[buffer(KB_VertexUniforms)]],
                             constant FragmentUniforms* fragment_uniforms [[buffer(KB_FragmentUniforms)]],
                             device Mesh* meshes [[buffer(KB_Meshes)]],
                             constant Transform* transforms [[buffer(KB_Transforms)]],
                             device Material* materials [[buffer(KB_Materials)]],
                             constant ShadowMap* shadow_map [[buffer(KB_ShadowMap)]],
                             constant Ibl* ibl [[buffer(KB_Ibl)]],
                             constant Cubemap* cubemap [[buffer(KB_Cubemap)]],
                             device MTLIndirectCommandBufferExecutionRange& range [[buffer(KB_ExecutionRange)]],
                             const uint instance_id [[thread_position_in_grid]]) {
  device auto& mesh = meshes[instance_id];
  device auto* range_length = reinterpret_cast<device atomic_uint*>(&range.length);
  const auto index = atomic_fetch_add_explicit(range_length, 1, memory_order_relaxed);
  
  render_command cmd(icb_container.icb, index);
  cmd.set_render_pipeline_state(mesh.pipeline_state);
  cmd.set_vertex_buffer(mesh.vertex_buffer, VB_Vertices);
  cmd.set_vertex_buffer(vertex_uniforms, VB_VertexUniforms);
  cmd.set_vertex_buffer(transforms, VB_Transforms);
  cmd.set_fragment_buffer(fragment_uniforms, FB_FragmentUniforms);
  cmd.set_fragment_buffer(transforms, FB_Transforms);
  // Tier 1: use indexed access and pass pointer to instance
  cmd.set_fragment_buffer(&materials[instance_id], FB_Material);
  // Tier 2: pass entire buffer and use indexed access in fragment shader
  cmd.set_fragment_buffer(materials, FB_Material);
  cmd.set_fragment_buffer(shadow_map, FB_ShadowMap);
  cmd.set_fragment_buffer(ibl, FB_Ibl);
  cmd.set_fragment_buffer(cubemap, FB_Cubemap);
  
  if (mesh.is_uint16_index){
    constant auto* index_buffer = static_cast<constant ushort*>(mesh.index_buffer);
    cmd.draw_indexed_primitives(primitive_type::triangle, mesh.index_count, index_buffer, 1, 0, instance_id);
  } else {
    constant auto* index_buffer = static_cast<constant uint*>(mesh.index_buffer);
    cmd.draw_indexed_primitives(primitive_type::triangle, mesh.index_count, index_buffer, 1, 0, instance_id);
  }
}

// Tier 1
fragment half4 pbr_fragment(ProjectedVertex vert [[stage_in]],
                            constant FragmentUniforms& uniforms [[buffer(FB_FragmentUniforms)]],
                            constant Material& material [[buffer(FB_Material)]],
                            constant Ibl& ibl [[buffer(FB_Ibl), function_constant(HAS_IBL)]],
                            constant ShadowMap& shadow_map [[buffer(FB_ShadowMap), function_constant(HAS_SHADOW_MAP)]]
                            ) {
  // Use Material
}

// Tier 2
fragment half4 pbr_fragment(ProjectedVertex vert [[stage_in]],
                            constant FragmentUniforms& uniforms [[buffer(FB_FragmentUniforms)]],
                            device Material* materials [[buffer(FB_Material)]],
                            constant Ibl& ibl [[buffer(FB_Ibl), function_constant(HAS_IBL)]],
                            constant ShadowMap& shadow_map [[buffer(FB_ShadowMap), function_constant(HAS_SHADOW_MAP)]]
                            ) {
  // Use indexed Material
  const auto& material = materials[vert.instance_id];
}

为简洁起见，我没有时间编辑示例，但应该足够清楚。

旁注：Metal 规范建议在使用指针算法（索引访问）时使用device 地址空间。请参阅规范的第 61 页。