Metal error when using discard: control reaches end of non-void fragment function

[almarklein]

In pygfx we generate wgsl shaders with a form of templating. One of the shaders we generate has a fragment function that ends with an unconditional discard:

[[stage(fragment)]]
 fn fs_main(varyings: Varyings) -> FragmentOutput {
     ...
    discard;
}

This used to work, but with the latest version of wgpu-native (v0.11.0.1, 19 dec, using Naga 8ffd6ba), this produces an error on MacOS:

thread '<unnamed>' panicked at 'Internal { stage: VERTEX, error: "Metal: program_source:149:1: error: control reaches end of non-void fragment function\n}\n^\n" }', src/device.rs:643:9
note: run with `RUST_BACKTRACE=1` environment variable to display a backtrace
fatal runtime error: failed to initiate panic, error 5

I was able to fix this by changing our wgsl shader to the following. It was necessary to both make the discard conditional, and to have a return after it:

[[stage(fragment)]]
 fn fs_main(varyings: Varyings) -> FragmentOutput {
     ...
    if (true) { discard; }
    ...
    return out;
}

The produced error also showed the Metal shader. I put it last because it's quite long:

Naga generated shader:
// language: metal2.2
#include <metal_stdlib>
#include <simd/simd.h>

struct DefaultConstructible {
    template<typename T>
    operator T() && {
        return T {};
    }
};
struct Struct_u_stdinfo {
    metal::float4x4 cam_transform;
    metal::float4x4 cam_transform_inv;
    metal::float4x4 projection_transform;
    metal::float4x4 projection_transform_inv;
    metal::float2 physical_size;
    metal::float2 logical_size;
    int flipped_winding;
};
struct Struct_u_wobject {
    metal::float4x4 world_transform;
    metal::float4x4 world_transform_inv;
    int id;
};
struct Struct_u_material {
    metal::float4 color_bottom_left;
    metal::float4 color_bottom_right;
    metal::float4 color_top_left;
    metal::float4 color_top_right;
    float opacity;
};
struct FragmentOutput {
    metal::float4 color;
};
struct VertexInput {
    metal::uint index;
};
struct Varyings {
    metal::float3 texcoord;
    metal::float4 position;
};
struct type_11 {
    metal::float2 inner[4];
};

bool check_clipping_planes(
    metal::float3 world_pos
) {
    return true;
}

void apply_clipping_planes(
    metal::float3 world_pos_1
) {
    return;
}

metal::float3 ndc_to_world_pos(
    metal::float4 ndc_pos,
    constant Struct_u_stdinfo& u_stdinfo
) {
    metal::float4x4 _e5 = u_stdinfo.cam_transform_inv;
    metal::float4x4 _e7 = u_stdinfo.projection_transform_inv;
    metal::float4x4 ndc_to_world = _e5 * _e7;
    metal::float4 world_pos_2 = ndc_to_world * ndc_pos;
    return world_pos_2.xyz / metal::float3(world_pos_2.w);
}

metal::uint4 pick_pack(
    metal::uint value,
    int bits,
    thread int& p_pick_bits_used
) {
    metal::uint v = metal::min(value, static_cast<uint>(metal::exp2(static_cast<float>(bits))));
    int _e10 = p_pick_bits_used;
    int _e11 = p_pick_bits_used;
    int _e14 = p_pick_bits_used;
    int _e17 = p_pick_bits_used;
    metal::int4 shift = metal::int4(_e10, _e11 - 16, _e14 - 32, _e17 - 48);
    int _e21 = p_pick_bits_used;
    p_pick_bits_used = _e21 + bits;
    metal::uint4 vv = metal::uint4(v);
    metal::bool4 selector1_ = metal::bool4(shift.x < 0, shift.y < 0, shift.z < 0, shift.w < 0);
    metal::uint4 pick_new = metal::select(vv << static_cast<metal::uint4>(shift), vv >> static_cast<metal::uint4>(-shift), selector1_);
    metal::uint4 mask = metal::uint4(65535u);
    metal::bool4 selector2_ = metal::bool4(metal::abs(shift.x) < 32, metal::abs(shift.y) < 32, metal::abs(shift.z) < 32, metal::abs(shift.w) < 32);
    return metal::select(metal::uint4(0u), pick_new & mask, selector2_);
}

FragmentOutput get_fragment_output(
    float depth,
    metal::float4 color
) {
    FragmentOutput out;
    if (color.w <= 0.0) {
        metal::discard_fragment();
    }
    out.color = metal::float4(color.xyz * color.w, color.w);
    FragmentOutput _e16 = out;
    return _e16;
}

struct vs_mainInput {
};
struct vs_mainOutput {
    metal::float3 texcoord [[user(loc0), center_perspective]];
    metal::float4 position [[position]];
};
vertex vs_mainOutput vs_main(
  metal::uint index [[vertex_id]]
) {
    const VertexInput in = { index };
    Varyings varyings;
    type_11 positions;
    for(int _i=0; _i<4; ++_i) positions.inner[_i] = type_11 {metal::float2(-1.0, -1.0), metal::float2(1.0, -1.0), metal::float2(-1.0, 1.0), metal::float2(1.0, 1.0)}.inner[_i];
    int _e21 = static_cast<int>(in.index);
    metal::float2 pos = metal::uint(_e21) < 4 ? positions.inner[_e21] : DefaultConstructible();
    varyings.position = metal::float4(pos, 0.9999998807907104, 1.0);
    varyings.texcoord = metal::float3((pos * 0.5) + metal::float2(0.5), 0.0);
    Varyings _e36 = varyings;
    const auto _tmp = _e36;
    return vs_mainOutput { _tmp.texcoord, _tmp.position };
}


struct fs_mainInput {
    metal::float3 texcoord [[user(loc0), center_perspective]];
};
struct fs_mainOutput {
    metal::float4 color [[color(0)]];
};
fragment fs_mainOutput fs_main(
  fs_mainInput varyings_3 [[stage_in]]
, metal::float4 position [[position]]
, constant Struct_u_material& u_material [[buffer(2)]]
) {
    const Varyings varyings_1 = { varyings_3.texcoord, position };
    metal::float4 final_color;
    metal::float2 f = varyings_1.texcoord.xy;
    metal::float4 _e9 = u_material.color_bottom_left;
    metal::float4 _e19 = u_material.color_bottom_right;
    metal::float4 _e28 = u_material.color_top_left;
    metal::float4 _e37 = u_material.color_top_right;
    final_color = ((((_e9 * (1.0 - f.x)) * (1.0 - f.y)) + ((_e19 * f.x) * (1.0 - f.y))) + ((_e28 * (1.0 - f.x)) * f.y)) + ((_e37 * f.x) * f.y);
    float _e45 = final_color.w;
    float _e47 = u_material.opacity;
    final_color.w = _e45 * _e47;
    metal::discard_fragment();
}

[kvark]

Filed gpuweb/gpuweb#2523 upstream to discuss

[dneto0]

I saw this from upstream. Thanks for this bug report. It's very helpful to see your real-life use case.