Add shadow snapshot shader

tests/snapshots.rs

@@ -182,3 +182,9 @@ fn convert_wgsl_skybox() {
 fn convert_wgsl_collatz() {
     convert_wgsl("collatz", Language::METAL | Language::SPIRV);
 }
+
+#[cfg(feature = "wgsl-in")]
+#[test]
+fn convert_wgsl_shadow() {
+    convert_wgsl("shadow", Language::METAL | Language::SPIRV);
+}

tests/snapshots/in/shadow.param.ron

@@ -0,0 +1,10 @@
+(
+    spv_flow_dump_prefix: "",
+    spv_capabilities: [ Shader ],
+    mtl_bindings: {
+        (stage: Fragment, group: 0, binding: 0): (buffer: Some(0), mutable: false),
+        (stage: Fragment, group: 0, binding: 1): (buffer: Some(1), mutable: false),
+        (stage: Fragment, group: 0, binding: 2): (texture: Some(0), mutable: false),
+        (stage: Fragment, group: 0, binding: 3): (sampler: Some(0), mutable: false),
+    }
+)

tests/snapshots/in/shadow.wgsl

@@ -0,0 +1,69 @@
+[[block]]
+struct Globals {
+    num_lights: vec4<u32>;
+};
+
+[[group(0), binding(0)]]
+var<uniform> u_globals: Globals;
+
+[[block]]
+struct Light {
+    proj: mat4x4<f32>;
+    pos: vec4<f32>;
+    color: vec4<f32>;
+};
+
+[[block]]
+struct Lights {
+    data: [[stride(96)]] array<Light>;
+};
+
+[[group(0), binding(1)]]
+var<storage> s_lights: [[access(read)]] Lights;
+[[group(0), binding(2)]]
+var t_shadow: texture_depth_2d_array;
+[[group(0), binding(3)]]
+var sampler_shadow: sampler;
+
+fn fetch_shadow(light_id: u32, homogeneous_coords: vec4<f32>) -> f32 {
+    if (homogeneous_coords.w <= 0.0) {
+        return 1.0;
+    }
+    const flip_correction: vec2<f32> = vec2<f32>(0.5, -0.5);
+    const proj_correction: f32 = 1.0 / homogeneous_coords.w;
+    const light_local: vec2<f32> = homogeneous_coords.xy * flip_correction * proj_correction + vec2<f32>(0.5, 0.5);
+    return textureSampleCompare(t_shadow, sampler_shadow, light_local, i32(light_id), homogeneous_coords.z * proj_correction);
+}
+
+[[location(0)]]
+var<in> in_normal_fs: vec3<f32>;
+[[location(1)]]
+var<in> in_position_fs: vec4<f32>;
+[[location(0)]]
+var<out> out_color_fs: vec4<f32>;
+
+const c_ambient: vec3<f32> = vec3<f32>(0.05, 0.05, 0.05);
+const c_max_lights: u32 = 10u;
+
+[[stage(fragment)]]
+fn fs_main() {
+    const normal: vec3<f32> = normalize(in_normal_fs);
+    // accumulate color
+    var color: vec3<f32> = c_ambient;
+    var i: u32 = 0u;
+    loop {
+        if (i >= min(u_globals.num_lights.x, c_max_lights)) {
+            break;
+        }
+        const light: Light = s_lights.data[i];
+        const shadow: f32 = fetch_shadow(i, light.proj * in_position_fs);
+        const light_dir: vec3<f32> = normalize(light.pos.xyz - in_position_fs.xyz);
+        const diffuse: f32 = max(0.0, dot(normal, light_dir));
+        color = color + shadow * diffuse * light.color.xyz;
+        continuing {
+            i = i + 1u;
+        }
+    }
+    // multiply the light by material color
+    out_color_fs = vec4<f32>(color, 1.0);
+}

tests/snapshots/snapshots__boids.msl.snap

@@ -51,7 +51,7 @@ kernel void main1(
     type5 cVelCount = 0;
     type1 pos1;
     type1 vel1;
-    type i = 0;
+    type i = 0u;
     if ((gl_GlobalInvocationID.x >= 1500)) {
         return ;
     }
@@ -63,7 +63,7 @@ kernel void main1(
     bool loop_init = true;
     while(true) {
         if (!loop_init) {
-            i = (i + 1);
+            i = (i + 1u);
         }
         loop_init = false;
         if ((i >= 1500)) {

tests/snapshots/snapshots__collatz.msl.snap

@@ -19,18 +19,18 @@ type1 collatz_iterations(
     type1 n_base
 ) {
     type1 n;
-    type1 i = 0;
+    type1 i = 0u;
     n = n_base;
     while(true) {
-        if ((n <= 1)) {
+        if ((n <= 1u)) {
             break;
         }
-        if (((n % 2) == 0)) {
+        if (((n % 2u) == 0u)) {
-            n = (n / 2);
+            n = (n / 2u);
         } else {
-            n = ((3 * n) + 1);
+            n = ((3u * n) + 1u);
         }
-        i = (i + 1);
+        i = (i + 1u);
     }
     return i;
 }

tests/snapshots/snapshots__shadow.msl.snap

@@ -0,0 +1,87 @@
+---
+source: tests/snapshots.rs
+expression: msl
+---
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+typedef metal::uint4 type;
+
+struct Globals {
+    type num_lights;
+};
+
+typedef metal::float4x4 type1;
+
+typedef metal::float4 type2;
+
+struct Light {
+    type1 proj;
+    type2 pos;
+    type2 color;
+};
+
+typedef Light type3[1];
+
+struct Lights {
+    type3 data;
+};
+
+typedef metal::depth2d_array<float, metal::access::sample> type4;
+
+typedef metal::sampler type5;
+
+typedef uint type6;
+
+typedef float type7;
+
+typedef metal::float2 type8;
+
+typedef metal::int2 type9;
+
+typedef int type10;
+
+typedef metal::float3 type11;
+
+type7 fetch_shadow(
+    type6 light_id,
+    type2 homogeneous_coords
+) {
+    if ((homogeneous_coords.w <= 0.0)) {
+        return 1.0;
+    }
+    return t_shadow.sample_compare(sampler_shadow, (((metal::float2(homogeneous_coords.x, homogeneous_coords.y) * metal::float2(0.5, -0.5)) * (1.0 / homogeneous_coords.w)) + metal::float2(0.5, 0.5)), static_cast<int>(light_id), (homogeneous_coords.z * (1.0 / homogeneous_coords.w)));
+}
+
+struct fs_mainInput {
+    type11 in_normal_fs [[user(loc0)]];
+    type2 in_position_fs [[user(loc1)]];
+};
+
+struct fs_mainOutput {
+    type2 out_color_fs [[color(0)]];
+};
+
+fragment fs_mainOutput fs_main(
+    fs_mainInput input [[stage_in]],
+    constant Globals& u_globals [[buffer(0)]],
+    constant Lights& s_lights [[buffer(1)]]
+) {
+    fs_mainOutput output;
+    type11 color1 = type11(0.05, 0.05, 0.05);
+    type6 i = 0u;
+    bool loop_init = true;
+    while(true) {
+        if (!loop_init) {
+            i = (i + 1u);
+        }
+        loop_init = false;
+        if ((i >= metal::min(u_globals.num_lights.x, 10u))) {
+            break;
+        }
+        color1 = (color1 + ((fetch_shadow(i, (s_lights.data[i].proj * input.in_position_fs)) * metal::max(0.0, metal::dot(metal::normalize(input.in_normal_fs), metal::normalize((metal::float3(s_lights.data[i].pos.x, s_lights.data[i].pos.y, s_lights.data[i].pos.z) - metal::float3(input.in_position_fs.x, input.in_position_fs.y, input.in_position_fs.z)))))) * metal::float3(s_lights.data[i].color.x, s_lights.data[i].color.y, s_lights.data[i].color.z)));
+    }
+    output.out_color_fs = metal::float4(color1, 1.0);
+    return output;
+}
+

tests/snapshots/snapshots__shadow.spvasm.snap

@@ -0,0 +1,273 @@
+---
+source: tests/snapshots.rs
+expression: dis
+---
+; SPIR-V
+; Version: 1.0
+; Generator: rspirv
+; Bound: 220
+OpCapability Shader
+%1 = OpExtInstImport "GLSL.std.450"
+OpMemoryModel Logical GLSL450
+OpEntryPoint Fragment %60 "fs_main" %112 %109 %216
+OpExecutionMode %60 OriginUpperLeft
+OpDecorate %18 DescriptorSet 0
+OpDecorate %18 Binding 2
+OpDecorate %23 DescriptorSet 0
+OpDecorate %23 Binding 3
+OpDecorate %72 Block
+OpMemberDecorate %72 0 Offset 0
+OpDecorate %71 DescriptorSet 0
+OpDecorate %71 Binding 0
+OpDecorate %97 BufferBlock
+OpMemberDecorate %97 0 Offset 0
+OpDecorate %98 ArrayStride 96
+OpDecorate %99 Block
+OpMemberDecorate %99 0 Offset 0
+OpMemberDecorate %99 0 MatrixStride 16
+OpDecorate %100 ColMajor
+OpMemberDecorate %99 1 Offset 64
+OpMemberDecorate %99 2 Offset 80
+OpDecorate %96 DescriptorSet 0
+OpDecorate %96 Binding 1
+OpDecorate %109 Location 1
+OpDecorate %112 Location 0
+OpDecorate %216 Location 0
+%2 = OpTypeFloat 32
+%4 = OpTypeInt 32 0
+%6 = OpTypeVector %2 4
+%8 = OpTypeFunction %2 %4 %6
+%12 = OpConstant  %2  0.0
+%13 = OpTypeBool
+%17 = OpConstant  %2  1.0
+%19 = OpTypeImage %2 2D 1 1 0 1 Unknown
+%20 = OpTypePointer UniformConstant %19
+%18 = OpVariable  %20  UniformConstant
+%22 = OpTypeSampledImage %19
+%24 = OpTypeSampler
+%25 = OpTypePointer UniformConstant %24
+%23 = OpVariable  %25  UniformConstant
+%30 = OpTypeVector %2 2
+%34 = OpConstant  %2  0.5
+%35 = OpConstant  %2  -0.5
+%43 = OpTypeVector %2 3
+%51 = OpConstant  %2  0.0
+%54 = OpConstant  %2  0.05
+%53 = OpConstantComposite  %43  %54 %54 %54
+%55 = OpTypePointer Function %43
+%57 = OpConstant  %4  0
+%58 = OpTypePointer Function %4
+%59 = OpTypeVoid
+%61 = OpTypeFunction %59
+%73 = OpTypeVector %4 4
+%72 = OpTypeStruct %73
+%74 = OpTypePointer Uniform %72
+%71 = OpVariable  %74  Uniform
+%75 = OpTypeInt 32 1
+%76 = OpConstant  %75  0
+%77 = OpTypePointer Uniform %73
+%78 = OpConstant  %75  0
+%79 = OpTypePointer Uniform %4
+%81 = OpConstant  %4  10
+%100 = OpTypeMatrix %6 4
+%99 = OpTypeStruct %100 %6 %6
+%98 = OpTypeRuntimeArray %99
+%97 = OpTypeStruct %98
+%101 = OpTypePointer Uniform %97
+%96 = OpVariable  %101  Uniform
+%102 = OpConstant  %75  0
+%103 = OpTypePointer Uniform %98
+%105 = OpTypePointer Uniform %99
+%106 = OpConstant  %75  0
+%107 = OpTypePointer Uniform %100
+%110 = OpTypePointer Input %6
+%109 = OpVariable  %110  Input
+%113 = OpTypePointer Input %43
+%112 = OpVariable  %113  Input
+%121 = OpConstant  %75  0
+%122 = OpTypePointer Uniform %98
+%124 = OpTypePointer Uniform %99
+%125 = OpConstant  %75  1
+%126 = OpTypePointer Uniform %6
+%127 = OpConstant  %75  0
+%128 = OpTypePointer Uniform %2
+%134 = OpConstant  %75  0
+%135 = OpTypePointer Uniform %98
+%137 = OpTypePointer Uniform %99
+%138 = OpConstant  %75  1
+%139 = OpTypePointer Uniform %6
+%140 = OpConstant  %75  1
+%141 = OpTypePointer Uniform %2
+%147 = OpConstant  %75  0
+%148 = OpTypePointer Uniform %98
+%150 = OpTypePointer Uniform %99
+%151 = OpConstant  %75  1
+%152 = OpTypePointer Uniform %6
+%153 = OpConstant  %75  2
+%154 = OpTypePointer Uniform %2
+%158 = OpConstant  %75  0
+%159 = OpTypePointer Input %2
+%162 = OpConstant  %75  1
+%163 = OpTypePointer Input %2
+%166 = OpConstant  %75  2
+%167 = OpTypePointer Input %2
+%177 = OpConstant  %75  0
+%178 = OpTypePointer Uniform %98
+%180 = OpTypePointer Uniform %99
+%181 = OpConstant  %75  2
+%182 = OpTypePointer Uniform %6
+%183 = OpConstant  %75  0
+%184 = OpTypePointer Uniform %2
+%190 = OpConstant  %75  0
+%191 = OpTypePointer Uniform %98
+%193 = OpTypePointer Uniform %99
+%194 = OpConstant  %75  2
+%195 = OpTypePointer Uniform %6
+%196 = OpConstant  %75  1
+%197 = OpTypePointer Uniform %2
+%203 = OpConstant  %75  0
+%204 = OpTypePointer Uniform %98
+%206 = OpTypePointer Uniform %99
+%207 = OpConstant  %75  2
+%208 = OpTypePointer Uniform %6
+%209 = OpConstant  %75  2
+%210 = OpTypePointer Uniform %2
+%215 = OpConstant  %4  1
+%217 = OpTypePointer Output %6
+%216 = OpVariable  %217  Output
+%7 = OpFunction  %2  None %8
+%3 = OpFunctionParameter  %4
+%5 = OpFunctionParameter  %6
+%9 = OpLabel
+%11 = OpCompositeExtract  %2  %5 3
+%10 = OpFOrdLessThanEqual  %13  %11 %12
+OpSelectionMerge %14 None
+OpBranchConditional %10 %15 %16
+%15 = OpLabel
+OpReturnValue %17
+%16 = OpLabel
+OpBranch %14
+%14 = OpLabel
+%21 = OpLoad  %19  %18
+%26 = OpLoad  %24  %23
+%31 = OpCompositeExtract  %2  %5 0
+%32 = OpCompositeExtract  %2  %5 1
+%33 = OpCompositeConstruct  %30  %31 %32
+%36 = OpCompositeConstruct  %30  %34 %35
+%29 = OpFMul  %30  %33 %36
+%38 = OpCompositeExtract  %2  %5 3
+%37 = OpFDiv  %2  %17 %38
+%28 = OpVectorTimesScalar  %30  %29 %37
+%39 = OpCompositeConstruct  %30  %34 %34
+%27 = OpFAdd  %30  %28 %39
+%40 = OpCompositeExtract  %2  %27 0
+%41 = OpCompositeExtract  %2  %27 1
+%42 = OpConvertUToF  %2  %3
+%44 = OpCompositeConstruct  %43  %40 %41 %42
+%45 = OpSampledImage  %22  %21 %26
+%48 = OpCompositeExtract  %2  %5 2
+%50 = OpCompositeExtract  %2  %5 3
+%49 = OpFDiv  %2  %17 %50
+%47 = OpFMul  %2  %48 %49
+%46 = OpImageSampleDrefExplicitLod  %2  %45 %44 %47 Lod %51
+OpReturnValue %46
+OpFunctionEnd
+%60 = OpFunction  %59  None %61
+%62 = OpLabel
+%52 = OpVariable  %55  Function %53
+%56 = OpVariable  %58  Function %57
+OpBranch %63
+%63 = OpLabel
+OpLoopMerge %64 %66 None
+OpBranch %65
+%65 = OpLabel
+%68 = OpLoad  %4  %56
+%70 = OpAccessChain  %77  %71 %76
+%69 = OpAccessChain  %79  %70 %78
+%80 = OpLoad  %4  %69
+%82 = OpExtInst  %4  %1 UMin %80 %81
+%67 = OpUGreaterThanEqual  %13  %68 %82
+OpSelectionMerge %83 None
+OpBranchConditional %67 %84 %85
+%84 = OpLabel
+OpBranch %64
+%85 = OpLabel
+OpBranch %83
+%83 = OpLabel
+%87 = OpLoad  %43  %52
+%91 = OpLoad  %4  %56
+%95 = OpAccessChain  %103  %96 %102
+%104 = OpLoad  %4  %56
+%94 = OpAccessChain  %105  %95 %104
+%93 = OpAccessChain  %107  %94 %106
+%108 = OpLoad  %100  %93
+%111 = OpLoad  %6  %109
+%92 = OpMatrixTimesVector  %6  %108 %111
+%90 = OpFunctionCall  %2  %7 %91 %92
+%114 = OpLoad  %43  %112
+%115 = OpExtInst  %43  %1 Normalize %114
+%120 = OpAccessChain  %122  %96 %121
+%123 = OpLoad  %4  %56
+%119 = OpAccessChain  %124  %120 %123
+%118 = OpAccessChain  %126  %119 %125
+%117 = OpAccessChain  %128  %118 %127
+%129 = OpLoad  %2  %117
+%133 = OpAccessChain  %135  %96 %134
+%136 = OpLoad  %4  %56
+%132 = OpAccessChain  %137  %133 %136
+%131 = OpAccessChain  %139  %132 %138
+%130 = OpAccessChain  %141  %131 %140
+%142 = OpLoad  %2  %130
+%146 = OpAccessChain  %148  %96 %147
+%149 = OpLoad  %4  %56
+%145 = OpAccessChain  %150  %146 %149
+%144 = OpAccessChain  %152  %145 %151
+%143 = OpAccessChain  %154  %144 %153
+%155 = OpLoad  %2  %143
+%156 = OpCompositeConstruct  %43  %129 %142 %155
+%157 = OpAccessChain  %159  %109 %158
+%160 = OpLoad  %2  %157
+%161 = OpAccessChain  %163  %109 %162
+%164 = OpLoad  %2  %161
+%165 = OpAccessChain  %167  %109 %166
+%168 = OpLoad  %2  %165
+%169 = OpCompositeConstruct  %43  %160 %164 %168
+%116 = OpFSub  %43  %156 %169
+%170 = OpExtInst  %43  %1 Normalize %116
+%171 = OpDot  %2  %115 %170
+%172 = OpExtInst  %2  %1 FMax %12 %171
+%89 = OpFMul  %2  %90 %172
+%176 = OpAccessChain  %178  %96 %177
+%179 = OpLoad  %4  %56
+%175 = OpAccessChain  %180  %176 %179
+%174 = OpAccessChain  %182  %175 %181
+%173 = OpAccessChain  %184  %174 %183
+%185 = OpLoad  %2  %173
+%189 = OpAccessChain  %191  %96 %190
+%192 = OpLoad  %4  %56
+%188 = OpAccessChain  %193  %189 %192
+%187 = OpAccessChain  %195  %188 %194
+%186 = OpAccessChain  %197  %187 %196
+%198 = OpLoad  %2  %186
+%202 = OpAccessChain  %204  %96 %203
+%205 = OpLoad  %4  %56
+%201 = OpAccessChain  %206  %202 %205
+%200 = OpAccessChain  %208  %201 %207
+%199 = OpAccessChain  %210  %200 %209
+%211 = OpLoad  %2  %199
+%212 = OpCompositeConstruct  %43  %185 %198 %211
+%88 = OpVectorTimesScalar  %43  %212 %89
+%86 = OpFAdd  %43  %87 %88
+OpStore %52 %86
+OpBranch %66
+%66 = OpLabel
+%214 = OpLoad  %4  %56
+%213 = OpIAdd  %4  %214 %215
+OpStore %56 %213
+OpBranch %63
+%64 = OpLabel
+%218 = OpLoad  %43  %52
+%219 = OpCompositeConstruct  %6  %218 %17
+OpStore %216 %219
+OpReturn
+OpFunctionEnd