[naga wgsl-in] Apply automatic conversions to sampling arguments.

Properly apply WGSL's automatic conversions to the arguments
to texture sampling functions.

Introduce helper function `Lowerer::expression_with_leaf_scalar`.

Although this can't affect behavior, use
`Lowerer::expression_for_abstract` for the `image` and `sampler`
arguments, simply because we want to move away from
`Lower::expression`'s automatic concretization, and move towards
having callers say explicitly what sort of conversions they need.

Although this can't affect behavior, use
`Lowerer::expression_with_leaf_scalar` for the `offset` argument, so
that the code spells out that this requires an `i32` value, rather
than depending on blind concretization giving it that.

Continue to use `Lowerer::expression` for `gather` and `array_index`,
since those happen to behave correctly with blind concretization, and
can be cleaned up later.

Fixes #7427.

naga/src/front/wgsl/lower/mod.rs

@@ -2025,6 +2025,16 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {
         ctx.apply_load_rule(expr)
     }
 
+    fn expression_with_leaf_scalar(
+        &mut self,
+        expr: Handle<ast::Expression<'source>>,
+        scalar: ir::Scalar,
+        ctx: &mut ExpressionContext<'source, '_, '_>,
+    ) -> Result<'source, Handle<ir::Expression>> {
+        let unconverted = self.expression_for_abstract(expr, ctx)?;
+        ctx.try_automatic_conversion_for_leaf_scalar(unconverted, scalar, Span::default())
+    }
+
     fn expression_for_reference(
         &mut self,
         expr: Handle<ast::Expression<'source>>,
@@ -3272,7 +3282,7 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {
         ) -> Result<'source, (Handle<ir::Expression>, Span)> {
             let image = args.next()?;
             let image_span = ctx.ast_expressions.get_span(image);
-            let image = lowerer.expression(image, ctx)?;
+            let image = lowerer.expression_for_abstract(image, ctx)?;
             Ok((image, image_span))
         }
 
@@ -3316,11 +3326,11 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {
             }
         };
 
-        let sampler = self.expression(args.next()?, ctx)?;
+        let sampler = self.expression_for_abstract(args.next()?, ctx)?;
 
-        let coordinate = self.expression(args.next()?, ctx)?;
+        let coordinate = self.expression_with_leaf_scalar(args.next()?, ir::Scalar::F32, ctx)?;
 
-        let (_, arrayed) = ctx.image_data(image, image_span)?;
+        let (class, arrayed) = ctx.image_data(image, image_span)?;
         let array_index = arrayed
             .then(|| self.expression(args.next()?, ctx))
             .transpose()?;
@@ -3333,7 +3343,8 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {
                 depth_ref = None;
             }
             Texture::GatherCompare => {
-                let reference = self.expression(args.next()?, ctx)?;
+                let reference =
+                    self.expression_with_leaf_scalar(args.next()?, ir::Scalar::F32, ctx)?;
                 level = ir::SampleLevel::Zero;
                 depth_ref = Some(reference);
             }
@@ -3343,28 +3354,44 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {
                 depth_ref = None;
             }
             Texture::SampleBias => {
-                let bias = self.expression(args.next()?, ctx)?;
+                let bias = self.expression_with_leaf_scalar(args.next()?, ir::Scalar::F32, ctx)?;
                 level = ir::SampleLevel::Bias(bias);
                 depth_ref = None;
             }
             Texture::SampleCompare => {
-                let reference = self.expression(args.next()?, ctx)?;
+                let reference =
+                    self.expression_with_leaf_scalar(args.next()?, ir::Scalar::F32, ctx)?;
                 level = ir::SampleLevel::Auto;
                 depth_ref = Some(reference);
             }
             Texture::SampleCompareLevel => {
-                let reference = self.expression(args.next()?, ctx)?;
+                let reference =
+                    self.expression_with_leaf_scalar(args.next()?, ir::Scalar::F32, ctx)?;
                 level = ir::SampleLevel::Zero;
                 depth_ref = Some(reference);
             }
             Texture::SampleGrad => {
-                let x = self.expression(args.next()?, ctx)?;
-                let y = self.expression(args.next()?, ctx)?;
+                let x = self.expression_with_leaf_scalar(args.next()?, ir::Scalar::F32, ctx)?;
+                let y = self.expression_with_leaf_scalar(args.next()?, ir::Scalar::F32, ctx)?;
                 level = ir::SampleLevel::Gradient { x, y };
                 depth_ref = None;
             }
             Texture::SampleLevel => {
-                let exact = self.expression(args.next()?, ctx)?;
+                let exact = match class {
+                    // When applied to depth textures, `textureSampleLevel`'s
+                    // `level` argument is an `i32` or `u32`.
+                    ir::ImageClass::Depth { .. } => self.expression(args.next()?, ctx)?,
+
+                    // When applied to other sampled types, its `level` argument
+                    // is an `f32`.
+                    ir::ImageClass::Sampled { .. } => {
+                        self.expression_with_leaf_scalar(args.next()?, ir::Scalar::F32, ctx)?
+                    }
+
+                    // Sampling `Storage` textures isn't allowed at all. Let the
+                    // validator report the error.
+                    ir::ImageClass::Storage { .. } => self.expression(args.next()?, ctx)?,
+                };
                 level = ir::SampleLevel::Exact(exact);
                 depth_ref = None;
             }
@@ -3372,7 +3399,7 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {
 
         let offset = args
             .next()
-            .map(|arg| self.expression(arg, &mut ctx.as_const()))
+            .map(|arg| self.expression_with_leaf_scalar(arg, ir::Scalar::I32, &mut ctx.as_const()))
             .ok()
             .transpose()?;
 

naga/tests/in/wgsl/abstract-types-texture.toml

@@ -0,0 +1 @@
+targets = "METAL"

naga/tests/in/wgsl/abstract-types-texture.wgsl

@@ -0,0 +1,20 @@
+@group(0) @binding(0) var t: texture_2d<f32>;
+@group(0) @binding(1) var s: sampler;
+
+fn color() {
+  _ = textureSample(t, s, vec2(1,2));
+  _ = textureSample(t, s, vec2(1,2), vec2(3,4));
+  _ = textureSampleLevel(t, s, vec2(1,2), 0);
+  _ = textureSampleLevel(t, s, vec2(1,2), 0.0);
+  _ = textureSampleGrad(t, s, vec2(1,2), vec2(3,4), vec2(5,6));
+  _ = textureSampleBias(t, s, vec2(1,2), 1);
+}
+
+@group(0) @binding(2) var d: texture_depth_2d;
+@group(0) @binding(3) var c: sampler_comparison;
+
+fn depth() {
+  _ = textureSampleLevel(d, s, vec2(1,2), 1i);
+  _ = textureSampleCompare(d, c, vec2(1,2), 0);
+  _ = textureGatherCompare(d, c, vec2(1,2), 0);
+}

naga/tests/out/msl/wgsl-abstract-types-texture.msl

@@ -0,0 +1,30 @@
+// language: metal1.0
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using metal::uint;
+
+
+void color(
+    metal::texture2d<float, metal::access::sample> t,
+    metal::sampler s
+) {
+    metal::float4 phony = t.sample(s, metal::float2(1.0, 2.0));
+    metal::float4 phony_1 = t.sample(s, metal::float2(1.0, 2.0), metal::int2(3, 4));
+    metal::float4 phony_2 = t.sample(s, metal::float2(1.0, 2.0), metal::level(0.0));
+    metal::float4 phony_3 = t.sample(s, metal::float2(1.0, 2.0), metal::level(0.0));
+    metal::float4 phony_4 = t.sample(s, metal::float2(1.0, 2.0), metal::gradient2d(metal::float2(3.0, 4.0), metal::float2(5.0, 6.0)));
+    metal::float4 phony_5 = t.sample(s, metal::float2(1.0, 2.0), metal::bias(1.0));
+    return;
+}
+
+void depth(
+    metal::sampler s,
+    metal::depth2d<float, metal::access::sample> d,
+    metal::sampler c
+) {
+    float phony_6 = d.sample(s, metal::float2(1.0, 2.0), metal::level(1));
+    float phony_7 = d.sample_compare(c, metal::float2(1.0, 2.0), 0.0);
+    metal::float4 phony_8 = d.gather_compare(c, metal::float2(1.0, 2.0), 0.0);
+    return;
+}