diff --git a/naga/src/back/glsl/mod.rs b/naga/src/back/glsl/mod.rs index 54f2b2e07..e4c2d3ab7 100644 --- a/naga/src/back/glsl/mod.rs +++ b/naga/src/back/glsl/mod.rs @@ -1259,7 +1259,7 @@ impl<'a, W: Write> Writer<'a, W> { if global.space.initializable() && is_value_init_supported(self.module, global.ty) { write!(self.out, " = ")?; if let Some(init) = global.init { - self.write_const_expr(init)?; + self.write_const_expr(init, &self.module.global_expressions)?; } else { self.write_zero_init_value(global.ty)?; } @@ -1904,7 +1904,7 @@ impl<'a, W: Write> Writer<'a, W> { self.write_array_size(base, size)?; } write!(self.out, " = ")?; - self.write_const_expr(constant.init)?; + self.write_const_expr(constant.init, &self.module.global_expressions)?; writeln!(self.out, ";")?; Ok(()) } @@ -2654,12 +2654,16 @@ impl<'a, W: Write> Writer<'a, W> { /// /// [`Expression`]: crate::Expression /// [`Module`]: crate::Module - fn write_const_expr(&mut self, expr: Handle<crate::Expression>) -> BackendResult { + fn write_const_expr( + &mut self, + expr: Handle<crate::Expression>, + arena: &crate::Arena<crate::Expression>, + ) -> BackendResult { self.write_possibly_const_expr( expr, - &self.module.global_expressions, + arena, |expr| &self.info[expr], - |writer, expr| writer.write_const_expr(expr), + |writer, expr| writer.write_const_expr(expr, arena), ) } @@ -2726,7 +2730,7 @@ impl<'a, W: Write> Writer<'a, W> { if constant.name.is_some() { write!(self.out, "{}", self.names[&NameKey::Constant(handle)])?; } else { - self.write_const_expr(constant.init)?; + self.write_const_expr(constant.init, &self.module.global_expressions)?; } } Expression::ZeroValue(ty) => { @@ -3033,7 +3037,7 @@ impl<'a, W: Write> Writer<'a, W> { if tex_1d_hack { write!(self.out, "ivec2(")?; } - self.write_const_expr(constant)?; + self.write_const_expr(constant, ctx.expressions)?; if tex_1d_hack { write!(self.out, ", 0)")?; } diff --git a/naga/src/back/hlsl/writer.rs b/naga/src/back/hlsl/writer.rs index 2ed1371a1..dae4d1ee0 100644 --- a/naga/src/back/hlsl/writer.rs +++ b/naga/src/back/hlsl/writer.rs @@ -1002,7 +1002,7 @@ impl<'a, W: fmt::Write> super::Writer<'a, W> { if global.space == crate::AddressSpace::Private { write!(self.out, " = ")?; if let Some(init) = global.init { - self.write_const_expression(module, init)?; + self.write_const_expression(module, init, &module.global_expressions)?; } else { self.write_default_init(module, global.ty)?; } @@ -1112,7 +1112,7 @@ impl<'a, W: fmt::Write> super::Writer<'a, W> { self.write_array_size(module, base, size)?; } write!(self.out, " = ")?; - self.write_const_expression(module, constant.init)?; + self.write_const_expression(module, constant.init, &module.global_expressions)?; writeln!(self.out, ";")?; Ok(()) } @@ -2609,13 +2609,11 @@ impl<'a, W: fmt::Write> super::Writer<'a, W> { &mut self, module: &Module, expr: Handle<crate::Expression>, + arena: &crate::Arena<crate::Expression>, ) -> BackendResult { - self.write_possibly_const_expression( - module, - expr, - &module.global_expressions, - |writer, expr| writer.write_const_expression(module, expr), - ) + self.write_possibly_const_expression(module, expr, arena, |writer, expr| { + writer.write_const_expression(module, expr, arena) + }) } fn write_possibly_const_expression<E>( @@ -2655,7 +2653,7 @@ impl<'a, W: fmt::Write> super::Writer<'a, W> { if constant.name.is_some() { write!(self.out, "{}", self.names[&NameKey::Constant(handle)])?; } else { - self.write_const_expression(module, constant.init)?; + self.write_const_expression(module, constant.init, &module.global_expressions)?; } } Expression::ZeroValue(ty) => { @@ -3178,7 +3176,7 @@ impl<'a, W: fmt::Write> super::Writer<'a, W> { if let Some(offset) = offset { write!(self.out, ", ")?; write!(self.out, "int2(")?; // work around https://github.com/microsoft/DirectXShaderCompiler/issues/5082#issuecomment-1540147807 - self.write_const_expression(module, offset)?; + self.write_const_expression(module, offset, func_ctx.expressions)?; write!(self.out, ")")?; } diff --git a/naga/src/back/msl/writer.rs b/naga/src/back/msl/writer.rs index c799df743..981ba78c1 100644 --- a/naga/src/back/msl/writer.rs +++ b/naga/src/back/msl/writer.rs @@ -1429,15 +1429,16 @@ impl<W: Write> Writer<W> { expr_handle: Handle<crate::Expression>, module: &crate::Module, mod_info: &valid::ModuleInfo, + arena: &crate::Arena<crate::Expression>, ) -> BackendResult { self.put_possibly_const_expression( expr_handle, - &module.global_expressions, + arena, module, mod_info, &(module, mod_info), |&(_, mod_info), expr| &mod_info[expr], - |writer, &(module, _), expr| writer.put_const_expression(expr, module, mod_info), + |writer, &(module, _), expr| writer.put_const_expression(expr, module, mod_info, arena), ) } @@ -1498,7 +1499,12 @@ impl<W: Write> Writer<W> { if constant.name.is_some() { write!(self.out, "{}", self.names[&NameKey::Constant(handle)])?; } else { - self.put_const_expression(constant.init, module, mod_info)?; + self.put_const_expression( + constant.init, + module, + mod_info, + &module.global_expressions, + )?; } } crate::Expression::ZeroValue(ty) => { @@ -1723,7 +1729,7 @@ impl<W: Write> Writer<W> { if let Some(offset) = offset { write!(self.out, ", ")?; - self.put_const_expression(offset, context.module, context.mod_info)?; + self.put_expression(offset, context, true)?; } match gather { @@ -4178,7 +4184,7 @@ template <typename A> }; let name = &self.names[&NameKey::Constant(handle)]; write!(self.out, "constant {ty_name} {name} = ")?; - self.put_const_expression(constant.init, module, mod_info)?; + self.put_const_expression(constant.init, module, mod_info, &module.global_expressions)?; writeln!(self.out, ";")?; } @@ -6149,7 +6155,7 @@ template <typename A> } if let Some(value) = var.init { write!(self.out, " = ")?; - self.put_const_expression(value, module, mod_info)?; + self.put_const_expression(value, module, mod_info, &module.global_expressions)?; } writeln!(self.out)?; } @@ -6354,7 +6360,12 @@ template <typename A> match var.init { Some(value) => { write!(self.out, " = ")?; - self.put_const_expression(value, module, mod_info)?; + self.put_const_expression( + value, + module, + mod_info, + &module.global_expressions, + )?; writeln!(self.out, ";")?; } None => { diff --git a/naga/src/back/spv/image.rs b/naga/src/back/spv/image.rs index 8285ca482..c554be423 100644 --- a/naga/src/back/spv/image.rs +++ b/naga/src/back/spv/image.rs @@ -996,7 +996,7 @@ impl BlockContext<'_> { }; if let Some(offset_const) = offset { - let offset_id = self.writer.constant_ids[offset_const]; + let offset_id = self.cached[offset_const]; main_instruction.add_operand(offset_id); } diff --git a/naga/src/back/wgsl/writer.rs b/naga/src/back/wgsl/writer.rs index 7e049b0ba..0f051621b 100644 --- a/naga/src/back/wgsl/writer.rs +++ b/naga/src/back/wgsl/writer.rs @@ -1225,13 +1225,11 @@ impl<W: Write> Writer<W> { &mut self, module: &Module, expr: Handle<crate::Expression>, + arena: &crate::Arena<crate::Expression>, ) -> BackendResult { - self.write_possibly_const_expression( - module, - expr, - &module.global_expressions, - |writer, expr| writer.write_const_expression(module, expr), - ) + self.write_possibly_const_expression(module, expr, arena, |writer, expr| { + writer.write_const_expression(module, expr, arena) + }) } fn write_possibly_const_expression<E>( @@ -1284,7 +1282,7 @@ impl<W: Write> Writer<W> { if constant.name.is_some() { write!(self.out, "{}", self.names[&NameKey::Constant(handle)])?; } else { - self.write_const_expression(module, constant.init)?; + self.write_const_expression(module, constant.init, &module.global_expressions)?; } } Expression::ZeroValue(ty) => { @@ -1480,7 +1478,7 @@ impl<W: Write> Writer<W> { if let Some(offset) = offset { write!(self.out, ", ")?; - self.write_const_expression(module, offset)?; + self.write_const_expression(module, offset, func_ctx.expressions)?; } write!(self.out, ")")?; @@ -1529,7 +1527,7 @@ impl<W: Write> Writer<W> { if let Some(offset) = offset { write!(self.out, ", ")?; - self.write_const_expression(module, offset)?; + self.write_const_expression(module, offset, func_ctx.expressions)?; } write!(self.out, ")")?; @@ -1840,7 +1838,7 @@ impl<W: Write> Writer<W> { // Write initializer if let Some(init) = global.init { write!(self.out, " = ")?; - self.write_const_expression(module, init)?; + self.write_const_expression(module, init, &module.global_expressions)?; } // End with semicolon @@ -1864,7 +1862,7 @@ impl<W: Write> Writer<W> { self.write_type(module, module.constants[handle].ty)?; write!(self.out, " = ")?; let init = module.constants[handle].init; - self.write_const_expression(module, init)?; + self.write_const_expression(module, init, &module.global_expressions)?; writeln!(self.out, ";")?; Ok(()) diff --git a/naga/src/compact/expressions.rs b/naga/src/compact/expressions.rs index d1b800a93..b0a858786 100644 --- a/naga/src/compact/expressions.rs +++ b/naga/src/compact/expressions.rs @@ -150,10 +150,7 @@ impl ExpressionTracer<'_> { self.expressions_used .insert_iter([image, sampler, coordinate]); self.expressions_used.insert_iter(array_index); - match self.global_expressions_used { - Some(ref mut used) => used.insert_iter(offset), - None => self.expressions_used.insert_iter(offset), - } + self.expressions_used.insert_iter(offset); use crate::SampleLevel as Sl; match *level { Sl::Auto | Sl::Zero => {} @@ -324,9 +321,7 @@ impl ModuleMap { adjust(sampler); adjust(coordinate); operand_map.adjust_option(array_index); - if let Some(ref mut offset) = *offset { - self.global_expressions.adjust(offset); - } + operand_map.adjust_option(offset); self.adjust_sample_level(level, operand_map); operand_map.adjust_option(depth_ref); } diff --git a/naga/src/front/glsl/builtins.rs b/naga/src/front/glsl/builtins.rs index 502da8a20..f73004aa9 100644 --- a/naga/src/front/glsl/builtins.rs +++ b/naga/src/front/glsl/builtins.rs @@ -1700,13 +1700,7 @@ impl MacroCall { true => { let offset_arg = args[num_args]; num_args += 1; - match ctx.lift_up_const_expression(offset_arg) { - Ok(v) => Some(v), - Err(e) => { - frontend.errors.push(e); - None - } - } + Some(offset_arg) } false => None, }; diff --git a/naga/src/front/wgsl/lower/mod.rs b/naga/src/front/wgsl/lower/mod.rs index 96032f826..a61ee1d17 100644 --- a/naga/src/front/wgsl/lower/mod.rs +++ b/naga/src/front/wgsl/lower/mod.rs @@ -3110,7 +3110,7 @@ impl<'source, 'temp> Lowerer<'source, 'temp> { let offset = args .next() - .map(|arg| self.expression(arg, &mut ctx.as_global().as_const())) + .map(|arg| self.expression(arg, &mut ctx.as_const())) .ok() .transpose()?; diff --git a/naga/src/lib.rs b/naga/src/lib.rs index afb9f8d0e..835de8392 100644 --- a/naga/src/lib.rs +++ b/naga/src/lib.rs @@ -1522,7 +1522,7 @@ pub enum Expression { gather: Option<SwizzleComponent>, coordinate: Handle<Expression>, array_index: Option<Handle<Expression>>, - /// This refers to an expression in [`Module::global_expressions`]. + /// This must be a const-expression. offset: Option<Handle<Expression>>, level: SampleLevel, depth_ref: Option<Handle<Expression>>, @@ -2216,9 +2216,6 @@ pub struct Function { /// - Various expressions hold [`Type`] handles, and [`Type`]s may refer to /// global expressions, for things like array lengths. /// - /// - [`Expression::ImageSample::offset`] refers to an expression in - /// [`Module::global_expressions`]. - /// /// An [`Expression`] must occur before all other [`Expression`]s that use /// its value. /// diff --git a/naga/src/valid/analyzer.rs b/naga/src/valid/analyzer.rs index 5a3d6ebc3..335a5b8a4 100644 --- a/naga/src/valid/analyzer.rs +++ b/naga/src/valid/analyzer.rs @@ -658,7 +658,7 @@ impl FunctionInfo { gather: _, coordinate, array_index, - offset: _, + offset, level, depth_ref, } => { @@ -685,6 +685,7 @@ impl FunctionInfo { Sl::Gradient { x, y } => self.add_ref(x).or(self.add_ref(y)), }; let dref_nur = depth_ref.and_then(|h| self.add_ref(h)); + let offset_nur = offset.and_then(|h| self.add_ref(h)); Uniformity { non_uniform_result: self .add_ref(image) @@ -692,7 +693,8 @@ impl FunctionInfo { .or(self.add_ref(coordinate)) .or(array_nur) .or(level_nur) - .or(dref_nur), + .or(dref_nur) + .or(offset_nur), requirements: if level.implicit_derivatives() { UniformityRequirements::IMPLICIT_LEVEL } else { diff --git a/naga/src/valid/expression.rs b/naga/src/valid/expression.rs index ec16c6d95..9d9e4d2f1 100644 --- a/naga/src/valid/expression.rs +++ b/naga/src/valid/expression.rs @@ -242,7 +242,7 @@ impl super::Validator { module: &crate::Module, info: &FunctionInfo, mod_info: &ModuleInfo, - global_expr_kind: &crate::proc::ExpressionKindTracker, + expr_kind: &crate::proc::ExpressionKindTracker, ) -> Result<ShaderStages, ExpressionError> { use crate::{Expression as E, Scalar as Sc, ScalarKind as Sk, TypeInner as Ti}; @@ -489,11 +489,11 @@ impl super::Validator { // check constant offset if let Some(const_expr) = offset { - if !global_expr_kind.is_const(const_expr) { + if !expr_kind.is_const(const_expr) { return Err(ExpressionError::InvalidSampleOffsetExprType); } - match *mod_info[const_expr].inner_with(&module.types) { + match resolver[const_expr] { Ti::Scalar(Sc { kind: Sk::Sint, .. }) if num_components == 1 => {} Ti::Vector { size, diff --git a/naga/src/valid/function.rs b/naga/src/valid/function.rs index b6ff92bbd..d00a28e10 100644 --- a/naga/src/valid/function.rs +++ b/naga/src/valid/function.rs @@ -1651,7 +1651,6 @@ impl super::Validator { module: &crate::Module, mod_info: &ModuleInfo, entry_point: bool, - global_expr_kind: &crate::proc::ExpressionKindTracker, ) -> Result<FunctionInfo, WithSpan<FunctionError>> { let mut info = mod_info.process_function(fun, module, self.flags, self.capabilities)?; @@ -1740,7 +1739,7 @@ impl super::Validator { module, &info, mod_info, - global_expr_kind, + &local_expr_kind, ) { Ok(stages) => info.available_stages &= stages, Err(source) => { diff --git a/naga/src/valid/handles.rs b/naga/src/valid/handles.rs index 62f5d00bb..86285c281 100644 --- a/naga/src/valid/handles.rs +++ b/naga/src/valid/handles.rs @@ -209,7 +209,6 @@ impl super::Validator { handle_and_expr, constants, overrides, - global_expressions, types, local_variables, global_variables, @@ -386,7 +385,6 @@ impl super::Validator { (handle, expression): (Handle<crate::Expression>, &crate::Expression), constants: &Arena<crate::Constant>, overrides: &Arena<crate::Override>, - global_expressions: &Arena<crate::Expression>, types: &UniqueArena<crate::Type>, local_variables: &Arena<crate::LocalVariable>, global_variables: &Arena<crate::GlobalVariable>, @@ -396,8 +394,6 @@ impl super::Validator { ) -> Result<(), InvalidHandleError> { let validate_constant = |handle| Self::validate_constant_handle(handle, constants); let validate_override = |handle| Self::validate_override_handle(handle, overrides); - let validate_const_expr = - |handle| Self::validate_expression_handle(handle, global_expressions); let validate_type = |handle| Self::validate_type_handle(handle, types); match *expression { @@ -447,15 +443,12 @@ impl super::Validator { level, depth_ref, } => { - if let Some(offset) = offset { - validate_const_expr(offset)?; - } - handle .check_dep(image)? .check_dep(sampler)? .check_dep(coordinate)? - .check_dep_opt(array_index)?; + .check_dep_opt(array_index)? + .check_dep_opt(offset)?; match level { crate::SampleLevel::Auto | crate::SampleLevel::Zero => (), diff --git a/naga/src/valid/interface.rs b/naga/src/valid/interface.rs index d182d9575..86869ae29 100644 --- a/naga/src/valid/interface.rs +++ b/naga/src/valid/interface.rs @@ -621,7 +621,6 @@ impl super::Validator { ep: &crate::EntryPoint, module: &crate::Module, mod_info: &ModuleInfo, - global_expr_kind: &crate::proc::ExpressionKindTracker, ) -> Result<FunctionInfo, WithSpan<EntryPointError>> { if ep.early_depth_test.is_some() { let required = Capabilities::EARLY_DEPTH_TEST; @@ -650,7 +649,7 @@ impl super::Validator { } let mut info = self - .validate_function(&ep.function, module, mod_info, true, global_expr_kind) + .validate_function(&ep.function, module, mod_info, true) .map_err(WithSpan::into_other)?; { diff --git a/naga/src/valid/mod.rs b/naga/src/valid/mod.rs index a2cec05c1..197579316 100644 --- a/naga/src/valid/mod.rs +++ b/naga/src/valid/mod.rs @@ -713,7 +713,7 @@ impl Validator { } for (handle, fun) in module.functions.iter() { - match self.validate_function(fun, module, &mod_info, false, &global_expr_kind) { + match self.validate_function(fun, module, &mod_info, false) { Ok(info) => mod_info.functions.push(info), Err(error) => { return Err(error.and_then(|source| { @@ -739,7 +739,7 @@ impl Validator { .with_span()); // TODO: keep some EP span information? } - match self.validate_entry_point(ep, module, &mod_info, &global_expr_kind) { + match self.validate_entry_point(ep, module, &mod_info) { Ok(info) => mod_info.entry_points.push(info), Err(error) => { return Err(error.and_then(|source| { diff --git a/naga/tests/in/wgsl/image.wgsl b/naga/tests/in/wgsl/image.wgsl index 53f9a925c..f159bbea9 100644 --- a/naga/tests/in/wgsl/image.wgsl +++ b/naga/tests/in/wgsl/image.wgsl @@ -113,26 +113,27 @@ var sampler_reg: sampler; @fragment fn texture_sample() -> @location(0) vec4<f32> { - let tc = vec2<f32>(0.5); - let tc3 = vec3<f32>(0.5); + const tc = vec2<f32>(0.5); + const tc3 = vec3<f32>(0.5); + const offset = vec2<i32>(3, 1); let level = 2.3; var a: vec4<f32>; a += textureSample(image_1d, sampler_reg, tc.x); a += textureSample(image_2d, sampler_reg, tc); a += textureSample(image_2d, sampler_reg, tc, vec2<i32>(3, 1)); a += textureSampleLevel(image_2d, sampler_reg, tc, level); - a += textureSampleLevel(image_2d, sampler_reg, tc, level, vec2<i32>(3, 1)); - a += textureSampleBias(image_2d, sampler_reg, tc, 2.0, vec2<i32>(3, 1)); + a += textureSampleLevel(image_2d, sampler_reg, tc, level, offset); + a += textureSampleBias(image_2d, sampler_reg, tc, 2.0, offset); a += textureSample(image_2d_array, sampler_reg, tc, 0u); - a += textureSample(image_2d_array, sampler_reg, tc, 0u, vec2<i32>(3, 1)); + a += textureSample(image_2d_array, sampler_reg, tc, 0u, offset); a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0u, level); - a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0u, level, vec2<i32>(3, 1)); - a += textureSampleBias(image_2d_array, sampler_reg, tc, 0u, 2.0, vec2<i32>(3, 1)); + a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0u, level, offset); + a += textureSampleBias(image_2d_array, sampler_reg, tc, 0u, 2.0, offset); a += textureSample(image_2d_array, sampler_reg, tc, 0); - a += textureSample(image_2d_array, sampler_reg, tc, 0, vec2<i32>(3, 1)); + a += textureSample(image_2d_array, sampler_reg, tc, 0, offset); a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0, level); - a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0, level, vec2<i32>(3, 1)); - a += textureSampleBias(image_2d_array, sampler_reg, tc, 0, 2.0, vec2<i32>(3, 1)); + a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0, level, offset); + a += textureSampleBias(image_2d_array, sampler_reg, tc, 0, 2.0, offset); a += textureSample(image_cube_array, sampler_reg, tc3, 0u); a += textureSampleLevel(image_cube_array, sampler_reg, tc3, 0u, level); a += textureSampleBias(image_cube_array, sampler_reg, tc3, 0u, 2.0); diff --git a/naga/tests/out/glsl/image.texture_sample.Fragment.glsl b/naga/tests/out/glsl/image.texture_sample.Fragment.glsl index 3657b2ea1..fc0d0db38 100644 --- a/naga/tests/out/glsl/image.texture_sample.Fragment.glsl +++ b/naga/tests/out/glsl/image.texture_sample.Fragment.glsl @@ -13,74 +13,75 @@ void main() { vec4 a = vec4(0.0); vec2 tc = vec2(0.5); vec3 tc3_ = vec3(0.5); - vec4 _e9 = texture(_group_0_binding_0_fs, tc.x); - vec4 _e10 = a; - a = (_e10 + _e9); - vec4 _e14 = texture(_group_0_binding_1_fs, vec2(tc)); - vec4 _e15 = a; - a = (_e15 + _e14); - vec4 _e19 = textureOffset(_group_0_binding_1_fs, vec2(tc), ivec2(3, 1)); - vec4 _e20 = a; - a = (_e20 + _e19); - vec4 _e24 = textureLod(_group_0_binding_1_fs, vec2(tc), 2.3); + ivec2 offset = ivec2(3, 1); + vec4 _e11 = texture(_group_0_binding_0_fs, 0.5); + vec4 _e12 = a; + a = (_e12 + _e11); + vec4 _e16 = texture(_group_0_binding_1_fs, vec2(tc)); + vec4 _e17 = a; + a = (_e17 + _e16); + vec4 _e24 = textureOffset(_group_0_binding_1_fs, vec2(tc), ivec2(3, 1)); vec4 _e25 = a; a = (_e25 + _e24); - vec4 _e29 = textureLodOffset(_group_0_binding_1_fs, vec2(tc), 2.3, ivec2(3, 1)); + vec4 _e29 = textureLod(_group_0_binding_1_fs, vec2(tc), 2.3); vec4 _e30 = a; a = (_e30 + _e29); - vec4 _e35 = textureOffset(_group_0_binding_1_fs, vec2(tc), ivec2(3, 1), 2.0); - vec4 _e36 = a; - a = (_e36 + _e35); - vec4 _e41 = texture(_group_0_binding_4_fs, vec3(tc, 0u)); - vec4 _e42 = a; - a = (_e42 + _e41); - vec4 _e47 = textureOffset(_group_0_binding_4_fs, vec3(tc, 0u), ivec2(3, 1)); - vec4 _e48 = a; - a = (_e48 + _e47); - vec4 _e53 = textureLod(_group_0_binding_4_fs, vec3(tc, 0u), 2.3); - vec4 _e54 = a; - a = (_e54 + _e53); - vec4 _e59 = textureLodOffset(_group_0_binding_4_fs, vec3(tc, 0u), 2.3, ivec2(3, 1)); - vec4 _e60 = a; - a = (_e60 + _e59); - vec4 _e66 = textureOffset(_group_0_binding_4_fs, vec3(tc, 0u), ivec2(3, 1), 2.0); - vec4 _e67 = a; - a = (_e67 + _e66); - vec4 _e72 = texture(_group_0_binding_4_fs, vec3(tc, 0)); - vec4 _e73 = a; - a = (_e73 + _e72); - vec4 _e78 = textureOffset(_group_0_binding_4_fs, vec3(tc, 0), ivec2(3, 1)); - vec4 _e79 = a; - a = (_e79 + _e78); - vec4 _e84 = textureLod(_group_0_binding_4_fs, vec3(tc, 0), 2.3); - vec4 _e85 = a; - a = (_e85 + _e84); - vec4 _e90 = textureLodOffset(_group_0_binding_4_fs, vec3(tc, 0), 2.3, ivec2(3, 1)); - vec4 _e91 = a; - a = (_e91 + _e90); - vec4 _e97 = textureOffset(_group_0_binding_4_fs, vec3(tc, 0), ivec2(3, 1), 2.0); - vec4 _e98 = a; - a = (_e98 + _e97); - vec4 _e103 = texture(_group_0_binding_6_fs, vec4(tc3_, 0u)); - vec4 _e104 = a; - a = (_e104 + _e103); - vec4 _e109 = textureLod(_group_0_binding_6_fs, vec4(tc3_, 0u), 2.3); - vec4 _e110 = a; - a = (_e110 + _e109); - vec4 _e116 = texture(_group_0_binding_6_fs, vec4(tc3_, 0u), 2.0); - vec4 _e117 = a; - a = (_e117 + _e116); - vec4 _e122 = texture(_group_0_binding_6_fs, vec4(tc3_, 0)); - vec4 _e123 = a; - a = (_e123 + _e122); - vec4 _e128 = textureLod(_group_0_binding_6_fs, vec4(tc3_, 0), 2.3); - vec4 _e129 = a; - a = (_e129 + _e128); - vec4 _e135 = texture(_group_0_binding_6_fs, vec4(tc3_, 0), 2.0); - vec4 _e136 = a; - a = (_e136 + _e135); - vec4 _e138 = a; - _fs2p_location0 = _e138; + vec4 _e34 = textureLodOffset(_group_0_binding_1_fs, vec2(tc), 2.3, ivec2(3, 1)); + vec4 _e35 = a; + a = (_e35 + _e34); + vec4 _e40 = textureOffset(_group_0_binding_1_fs, vec2(tc), ivec2(3, 1), 2.0); + vec4 _e41 = a; + a = (_e41 + _e40); + vec4 _e46 = texture(_group_0_binding_4_fs, vec3(tc, 0u)); + vec4 _e47 = a; + a = (_e47 + _e46); + vec4 _e52 = textureOffset(_group_0_binding_4_fs, vec3(tc, 0u), ivec2(3, 1)); + vec4 _e53 = a; + a = (_e53 + _e52); + vec4 _e58 = textureLod(_group_0_binding_4_fs, vec3(tc, 0u), 2.3); + vec4 _e59 = a; + a = (_e59 + _e58); + vec4 _e64 = textureLodOffset(_group_0_binding_4_fs, vec3(tc, 0u), 2.3, ivec2(3, 1)); + vec4 _e65 = a; + a = (_e65 + _e64); + vec4 _e71 = textureOffset(_group_0_binding_4_fs, vec3(tc, 0u), ivec2(3, 1), 2.0); + vec4 _e72 = a; + a = (_e72 + _e71); + vec4 _e77 = texture(_group_0_binding_4_fs, vec3(tc, 0)); + vec4 _e78 = a; + a = (_e78 + _e77); + vec4 _e83 = textureOffset(_group_0_binding_4_fs, vec3(tc, 0), ivec2(3, 1)); + vec4 _e84 = a; + a = (_e84 + _e83); + vec4 _e89 = textureLod(_group_0_binding_4_fs, vec3(tc, 0), 2.3); + vec4 _e90 = a; + a = (_e90 + _e89); + vec4 _e95 = textureLodOffset(_group_0_binding_4_fs, vec3(tc, 0), 2.3, ivec2(3, 1)); + vec4 _e96 = a; + a = (_e96 + _e95); + vec4 _e102 = textureOffset(_group_0_binding_4_fs, vec3(tc, 0), ivec2(3, 1), 2.0); + vec4 _e103 = a; + a = (_e103 + _e102); + vec4 _e108 = texture(_group_0_binding_6_fs, vec4(tc3_, 0u)); + vec4 _e109 = a; + a = (_e109 + _e108); + vec4 _e114 = textureLod(_group_0_binding_6_fs, vec4(tc3_, 0u), 2.3); + vec4 _e115 = a; + a = (_e115 + _e114); + vec4 _e121 = texture(_group_0_binding_6_fs, vec4(tc3_, 0u), 2.0); + vec4 _e122 = a; + a = (_e122 + _e121); + vec4 _e127 = texture(_group_0_binding_6_fs, vec4(tc3_, 0)); + vec4 _e128 = a; + a = (_e128 + _e127); + vec4 _e133 = textureLod(_group_0_binding_6_fs, vec4(tc3_, 0), 2.3); + vec4 _e134 = a; + a = (_e134 + _e133); + vec4 _e140 = texture(_group_0_binding_6_fs, vec4(tc3_, 0), 2.0); + vec4 _e141 = a; + a = (_e141 + _e140); + vec4 _e143 = a; + _fs2p_location0 = _e143; return; } diff --git a/naga/tests/out/hlsl/image.hlsl b/naga/tests/out/hlsl/image.hlsl index 5d13dd5a6..5319a1ca3 100644 --- a/naga/tests/out/hlsl/image.hlsl +++ b/naga/tests/out/hlsl/image.hlsl @@ -256,74 +256,75 @@ float4 texture_sample() : SV_Target0 float2 tc = (0.5).xx; float3 tc3_ = (0.5).xxx; - float4 _e9 = image_1d.Sample(sampler_reg, tc.x); - float4 _e10 = a; - a = (_e10 + _e9); - float4 _e14 = image_2d.Sample(sampler_reg, tc); - float4 _e15 = a; - a = (_e15 + _e14); - float4 _e19 = image_2d.Sample(sampler_reg, tc, int2(int2(int(3), int(1)))); - float4 _e20 = a; - a = (_e20 + _e19); - float4 _e24 = image_2d.SampleLevel(sampler_reg, tc, 2.3); + int2 offset = int2(int(3), int(1)); + float4 _e11 = image_1d.Sample(sampler_reg, 0.5); + float4 _e12 = a; + a = (_e12 + _e11); + float4 _e16 = image_2d.Sample(sampler_reg, tc); + float4 _e17 = a; + a = (_e17 + _e16); + float4 _e24 = image_2d.Sample(sampler_reg, tc, int2(int2(int(3), int(1)))); float4 _e25 = a; a = (_e25 + _e24); - float4 _e29 = image_2d.SampleLevel(sampler_reg, tc, 2.3, int2(int2(int(3), int(1)))); + float4 _e29 = image_2d.SampleLevel(sampler_reg, tc, 2.3); float4 _e30 = a; a = (_e30 + _e29); - float4 _e35 = image_2d.SampleBias(sampler_reg, tc, 2.0, int2(int2(int(3), int(1)))); - float4 _e36 = a; - a = (_e36 + _e35); - float4 _e41 = image_2d_array.Sample(sampler_reg, float3(tc, 0u)); - float4 _e42 = a; - a = (_e42 + _e41); - float4 _e47 = image_2d_array.Sample(sampler_reg, float3(tc, 0u), int2(int2(int(3), int(1)))); - float4 _e48 = a; - a = (_e48 + _e47); - float4 _e53 = image_2d_array.SampleLevel(sampler_reg, float3(tc, 0u), 2.3); - float4 _e54 = a; - a = (_e54 + _e53); - float4 _e59 = image_2d_array.SampleLevel(sampler_reg, float3(tc, 0u), 2.3, int2(int2(int(3), int(1)))); - float4 _e60 = a; - a = (_e60 + _e59); - float4 _e66 = image_2d_array.SampleBias(sampler_reg, float3(tc, 0u), 2.0, int2(int2(int(3), int(1)))); - float4 _e67 = a; - a = (_e67 + _e66); - float4 _e72 = image_2d_array.Sample(sampler_reg, float3(tc, int(0))); - float4 _e73 = a; - a = (_e73 + _e72); - float4 _e78 = image_2d_array.Sample(sampler_reg, float3(tc, int(0)), int2(int2(int(3), int(1)))); - float4 _e79 = a; - a = (_e79 + _e78); - float4 _e84 = image_2d_array.SampleLevel(sampler_reg, float3(tc, int(0)), 2.3); - float4 _e85 = a; - a = (_e85 + _e84); - float4 _e90 = image_2d_array.SampleLevel(sampler_reg, float3(tc, int(0)), 2.3, int2(int2(int(3), int(1)))); - float4 _e91 = a; - a = (_e91 + _e90); - float4 _e97 = image_2d_array.SampleBias(sampler_reg, float3(tc, int(0)), 2.0, int2(int2(int(3), int(1)))); - float4 _e98 = a; - a = (_e98 + _e97); - float4 _e103 = image_cube_array.Sample(sampler_reg, float4(tc3_, 0u)); - float4 _e104 = a; - a = (_e104 + _e103); - float4 _e109 = image_cube_array.SampleLevel(sampler_reg, float4(tc3_, 0u), 2.3); - float4 _e110 = a; - a = (_e110 + _e109); - float4 _e116 = image_cube_array.SampleBias(sampler_reg, float4(tc3_, 0u), 2.0); - float4 _e117 = a; - a = (_e117 + _e116); - float4 _e122 = image_cube_array.Sample(sampler_reg, float4(tc3_, int(0))); - float4 _e123 = a; - a = (_e123 + _e122); - float4 _e128 = image_cube_array.SampleLevel(sampler_reg, float4(tc3_, int(0)), 2.3); - float4 _e129 = a; - a = (_e129 + _e128); - float4 _e135 = image_cube_array.SampleBias(sampler_reg, float4(tc3_, int(0)), 2.0); - float4 _e136 = a; - a = (_e136 + _e135); - float4 _e138 = a; - return _e138; + float4 _e34 = image_2d.SampleLevel(sampler_reg, tc, 2.3, int2(int2(int(3), int(1)))); + float4 _e35 = a; + a = (_e35 + _e34); + float4 _e40 = image_2d.SampleBias(sampler_reg, tc, 2.0, int2(int2(int(3), int(1)))); + float4 _e41 = a; + a = (_e41 + _e40); + float4 _e46 = image_2d_array.Sample(sampler_reg, float3(tc, 0u)); + float4 _e47 = a; + a = (_e47 + _e46); + float4 _e52 = image_2d_array.Sample(sampler_reg, float3(tc, 0u), int2(int2(int(3), int(1)))); + float4 _e53 = a; + a = (_e53 + _e52); + float4 _e58 = image_2d_array.SampleLevel(sampler_reg, float3(tc, 0u), 2.3); + float4 _e59 = a; + a = (_e59 + _e58); + float4 _e64 = image_2d_array.SampleLevel(sampler_reg, float3(tc, 0u), 2.3, int2(int2(int(3), int(1)))); + float4 _e65 = a; + a = (_e65 + _e64); + float4 _e71 = image_2d_array.SampleBias(sampler_reg, float3(tc, 0u), 2.0, int2(int2(int(3), int(1)))); + float4 _e72 = a; + a = (_e72 + _e71); + float4 _e77 = image_2d_array.Sample(sampler_reg, float3(tc, int(0))); + float4 _e78 = a; + a = (_e78 + _e77); + float4 _e83 = image_2d_array.Sample(sampler_reg, float3(tc, int(0)), int2(int2(int(3), int(1)))); + float4 _e84 = a; + a = (_e84 + _e83); + float4 _e89 = image_2d_array.SampleLevel(sampler_reg, float3(tc, int(0)), 2.3); + float4 _e90 = a; + a = (_e90 + _e89); + float4 _e95 = image_2d_array.SampleLevel(sampler_reg, float3(tc, int(0)), 2.3, int2(int2(int(3), int(1)))); + float4 _e96 = a; + a = (_e96 + _e95); + float4 _e102 = image_2d_array.SampleBias(sampler_reg, float3(tc, int(0)), 2.0, int2(int2(int(3), int(1)))); + float4 _e103 = a; + a = (_e103 + _e102); + float4 _e108 = image_cube_array.Sample(sampler_reg, float4(tc3_, 0u)); + float4 _e109 = a; + a = (_e109 + _e108); + float4 _e114 = image_cube_array.SampleLevel(sampler_reg, float4(tc3_, 0u), 2.3); + float4 _e115 = a; + a = (_e115 + _e114); + float4 _e121 = image_cube_array.SampleBias(sampler_reg, float4(tc3_, 0u), 2.0); + float4 _e122 = a; + a = (_e122 + _e121); + float4 _e127 = image_cube_array.Sample(sampler_reg, float4(tc3_, int(0))); + float4 _e128 = a; + a = (_e128 + _e127); + float4 _e133 = image_cube_array.SampleLevel(sampler_reg, float4(tc3_, int(0)), 2.3); + float4 _e134 = a; + a = (_e134 + _e133); + float4 _e140 = image_cube_array.SampleBias(sampler_reg, float4(tc3_, int(0)), 2.0); + float4 _e141 = a; + a = (_e141 + _e140); + float4 _e143 = a; + return _e143; } float texture_sample_comparison() : SV_Target0 diff --git a/naga/tests/out/msl/image.msl b/naga/tests/out/msl/image.msl index 8dc223089..920d20f6c 100644 --- a/naga/tests/out/msl/image.msl +++ b/naga/tests/out/msl/image.msl @@ -126,74 +126,75 @@ fragment texture_sampleOutput texture_sample( metal::float4 a = {}; metal::float2 tc = metal::float2(0.5); metal::float3 tc3_ = metal::float3(0.5); - metal::float4 _e9 = image_1d.sample(sampler_reg, tc.x); - metal::float4 _e10 = a; - a = _e10 + _e9; - metal::float4 _e14 = image_2d.sample(sampler_reg, tc); - metal::float4 _e15 = a; - a = _e15 + _e14; - metal::float4 _e19 = image_2d.sample(sampler_reg, tc, metal::int2(3, 1)); - metal::float4 _e20 = a; - a = _e20 + _e19; - metal::float4 _e24 = image_2d.sample(sampler_reg, tc, metal::level(2.3)); + metal::int2 offset = metal::int2(3, 1); + metal::float4 _e11 = image_1d.sample(sampler_reg, 0.5); + metal::float4 _e12 = a; + a = _e12 + _e11; + metal::float4 _e16 = image_2d.sample(sampler_reg, tc); + metal::float4 _e17 = a; + a = _e17 + _e16; + metal::float4 _e24 = image_2d.sample(sampler_reg, tc, metal::int2(3, 1)); metal::float4 _e25 = a; a = _e25 + _e24; - metal::float4 _e29 = image_2d.sample(sampler_reg, tc, metal::level(2.3), metal::int2(3, 1)); + metal::float4 _e29 = image_2d.sample(sampler_reg, tc, metal::level(2.3)); metal::float4 _e30 = a; a = _e30 + _e29; - metal::float4 _e35 = image_2d.sample(sampler_reg, tc, metal::bias(2.0), metal::int2(3, 1)); - metal::float4 _e36 = a; - a = _e36 + _e35; - metal::float4 _e41 = image_2d_array.sample(sampler_reg, tc, 0u); - metal::float4 _e42 = a; - a = _e42 + _e41; - metal::float4 _e47 = image_2d_array.sample(sampler_reg, tc, 0u, metal::int2(3, 1)); - metal::float4 _e48 = a; - a = _e48 + _e47; - metal::float4 _e53 = image_2d_array.sample(sampler_reg, tc, 0u, metal::level(2.3)); - metal::float4 _e54 = a; - a = _e54 + _e53; - metal::float4 _e59 = image_2d_array.sample(sampler_reg, tc, 0u, metal::level(2.3), metal::int2(3, 1)); - metal::float4 _e60 = a; - a = _e60 + _e59; - metal::float4 _e66 = image_2d_array.sample(sampler_reg, tc, 0u, metal::bias(2.0), metal::int2(3, 1)); - metal::float4 _e67 = a; - a = _e67 + _e66; - metal::float4 _e72 = image_2d_array.sample(sampler_reg, tc, 0); - metal::float4 _e73 = a; - a = _e73 + _e72; - metal::float4 _e78 = image_2d_array.sample(sampler_reg, tc, 0, metal::int2(3, 1)); - metal::float4 _e79 = a; - a = _e79 + _e78; - metal::float4 _e84 = image_2d_array.sample(sampler_reg, tc, 0, metal::level(2.3)); - metal::float4 _e85 = a; - a = _e85 + _e84; - metal::float4 _e90 = image_2d_array.sample(sampler_reg, tc, 0, metal::level(2.3), metal::int2(3, 1)); - metal::float4 _e91 = a; - a = _e91 + _e90; - metal::float4 _e97 = image_2d_array.sample(sampler_reg, tc, 0, metal::bias(2.0), metal::int2(3, 1)); - metal::float4 _e98 = a; - a = _e98 + _e97; - metal::float4 _e103 = image_cube_array.sample(sampler_reg, tc3_, 0u); - metal::float4 _e104 = a; - a = _e104 + _e103; - metal::float4 _e109 = image_cube_array.sample(sampler_reg, tc3_, 0u, metal::level(2.3)); - metal::float4 _e110 = a; - a = _e110 + _e109; - metal::float4 _e116 = image_cube_array.sample(sampler_reg, tc3_, 0u, metal::bias(2.0)); - metal::float4 _e117 = a; - a = _e117 + _e116; - metal::float4 _e122 = image_cube_array.sample(sampler_reg, tc3_, 0); - metal::float4 _e123 = a; - a = _e123 + _e122; - metal::float4 _e128 = image_cube_array.sample(sampler_reg, tc3_, 0, metal::level(2.3)); - metal::float4 _e129 = a; - a = _e129 + _e128; - metal::float4 _e135 = image_cube_array.sample(sampler_reg, tc3_, 0, metal::bias(2.0)); - metal::float4 _e136 = a; - a = _e136 + _e135; - metal::float4 _e138 = a; - return texture_sampleOutput { _e138 }; + metal::float4 _e34 = image_2d.sample(sampler_reg, tc, metal::level(2.3), offset); + metal::float4 _e35 = a; + a = _e35 + _e34; + metal::float4 _e40 = image_2d.sample(sampler_reg, tc, metal::bias(2.0), offset); + metal::float4 _e41 = a; + a = _e41 + _e40; + metal::float4 _e46 = image_2d_array.sample(sampler_reg, tc, 0u); + metal::float4 _e47 = a; + a = _e47 + _e46; + metal::float4 _e52 = image_2d_array.sample(sampler_reg, tc, 0u, offset); + metal::float4 _e53 = a; + a = _e53 + _e52; + metal::float4 _e58 = image_2d_array.sample(sampler_reg, tc, 0u, metal::level(2.3)); + metal::float4 _e59 = a; + a = _e59 + _e58; + metal::float4 _e64 = image_2d_array.sample(sampler_reg, tc, 0u, metal::level(2.3), offset); + metal::float4 _e65 = a; + a = _e65 + _e64; + metal::float4 _e71 = image_2d_array.sample(sampler_reg, tc, 0u, metal::bias(2.0), offset); + metal::float4 _e72 = a; + a = _e72 + _e71; + metal::float4 _e77 = image_2d_array.sample(sampler_reg, tc, 0); + metal::float4 _e78 = a; + a = _e78 + _e77; + metal::float4 _e83 = image_2d_array.sample(sampler_reg, tc, 0, offset); + metal::float4 _e84 = a; + a = _e84 + _e83; + metal::float4 _e89 = image_2d_array.sample(sampler_reg, tc, 0, metal::level(2.3)); + metal::float4 _e90 = a; + a = _e90 + _e89; + metal::float4 _e95 = image_2d_array.sample(sampler_reg, tc, 0, metal::level(2.3), offset); + metal::float4 _e96 = a; + a = _e96 + _e95; + metal::float4 _e102 = image_2d_array.sample(sampler_reg, tc, 0, metal::bias(2.0), offset); + metal::float4 _e103 = a; + a = _e103 + _e102; + metal::float4 _e108 = image_cube_array.sample(sampler_reg, tc3_, 0u); + metal::float4 _e109 = a; + a = _e109 + _e108; + metal::float4 _e114 = image_cube_array.sample(sampler_reg, tc3_, 0u, metal::level(2.3)); + metal::float4 _e115 = a; + a = _e115 + _e114; + metal::float4 _e121 = image_cube_array.sample(sampler_reg, tc3_, 0u, metal::bias(2.0)); + metal::float4 _e122 = a; + a = _e122 + _e121; + metal::float4 _e127 = image_cube_array.sample(sampler_reg, tc3_, 0); + metal::float4 _e128 = a; + a = _e128 + _e127; + metal::float4 _e133 = image_cube_array.sample(sampler_reg, tc3_, 0, metal::level(2.3)); + metal::float4 _e134 = a; + a = _e134 + _e133; + metal::float4 _e140 = image_cube_array.sample(sampler_reg, tc3_, 0, metal::bias(2.0)); + metal::float4 _e141 = a; + a = _e141 + _e140; + metal::float4 _e143 = a; + return texture_sampleOutput { _e143 }; } diff --git a/naga/tests/out/spv/image.spvasm b/naga/tests/out/spv/image.spvasm index 04f9e6978..152ede488 100644 --- a/naga/tests/out/spv/image.spvasm +++ b/naga/tests/out/spv/image.spvasm @@ -1,7 +1,7 @@ ; SPIR-V ; Version: 1.1 ; Generator: rspirv -; Bound: 548 +; Bound: 547 OpCapability Shader OpCapability Image1D OpCapability Sampled1D @@ -9,20 +9,20 @@ OpCapability SampledCubeArray OpCapability ImageQuery %1 = OpExtInstImport "GLSL.std.450" OpMemoryModel Logical GLSL450 -OpEntryPoint GLCompute %98 "main" %95 -OpEntryPoint GLCompute %192 "depth_load" %190 -OpEntryPoint Vertex %212 "queries" %210 -OpEntryPoint Vertex %264 "levels_queries" %263 -OpEntryPoint Fragment %295 "texture_sample" %294 -OpEntryPoint Fragment %441 "texture_sample_comparison" %439 -OpEntryPoint Fragment %497 "gather" %496 -OpEntryPoint Fragment %531 "depth_no_comparison" %530 -OpExecutionMode %98 LocalSize 16 1 1 -OpExecutionMode %192 LocalSize 16 1 1 -OpExecutionMode %295 OriginUpperLeft -OpExecutionMode %441 OriginUpperLeft -OpExecutionMode %497 OriginUpperLeft -OpExecutionMode %531 OriginUpperLeft +OpEntryPoint GLCompute %96 "main" %93 +OpEntryPoint GLCompute %190 "depth_load" %188 +OpEntryPoint Vertex %210 "queries" %208 +OpEntryPoint Vertex %262 "levels_queries" %261 +OpEntryPoint Fragment %293 "texture_sample" %292 +OpEntryPoint Fragment %440 "texture_sample_comparison" %438 +OpEntryPoint Fragment %496 "gather" %495 +OpEntryPoint Fragment %530 "depth_no_comparison" %529 +OpExecutionMode %96 LocalSize 16 1 1 +OpExecutionMode %190 LocalSize 16 1 1 +OpExecutionMode %293 OriginUpperLeft +OpExecutionMode %440 OriginUpperLeft +OpExecutionMode %496 OriginUpperLeft +OpExecutionMode %530 OriginUpperLeft %3 = OpString "image.wgsl" OpSource Unknown 0 %3 "@group(0) @binding(0) var image_mipmapped_src: texture_2d<u32>; @@ -139,26 +139,27 @@ var sampler_reg: sampler; @fragment fn texture_sample() -> @location(0) vec4<f32> { - let tc = vec2<f32>(0.5); - let tc3 = vec3<f32>(0.5); + const tc = vec2<f32>(0.5); + const tc3 = vec3<f32>(0.5); + const offset = vec2<i32>(3, 1); let level = 2.3; var a: vec4<f32>; a += textureSample(image_1d, sampler_reg, tc.x); a += textureSample(image_2d, sampler_reg, tc); a += textureSample(image_2d, sampler_reg, tc, vec2<i32>(3, 1)); a += textureSampleLevel(image_2d, sampler_reg, tc, level); - a += textureSampleLevel(image_2d, sampler_reg, tc, level, vec2<i32>(3, 1)); - a += textureSampleBias(image_2d, sampler_reg, tc, 2.0, vec2<i32>(3, 1)); + a += textureSampleLevel(image_2d, sampler_reg, tc, level, offset); + a += textureSampleBias(image_2d, sampler_reg, tc, 2.0, offset); a += textureSample(image_2d_array, sampler_reg, tc, 0u); - a += textureSample(image_2d_array, sampler_reg, tc, 0u, vec2<i32>(3, 1)); + a += textureSample(image_2d_array, sampler_reg, tc, 0u, offset); a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0u, level); - a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0u, level, vec2<i32>(3, 1)); - a += textureSampleBias(image_2d_array, sampler_reg, tc, 0u, 2.0, vec2<i32>(3, 1)); + a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0u, level, offset); + a += textureSampleBias(image_2d_array, sampler_reg, tc, 0u, 2.0, offset); a += textureSample(image_2d_array, sampler_reg, tc, 0); - a += textureSample(image_2d_array, sampler_reg, tc, 0, vec2<i32>(3, 1)); + a += textureSample(image_2d_array, sampler_reg, tc, 0, offset); a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0, level); - a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0, level, vec2<i32>(3, 1)); - a += textureSampleBias(image_2d_array, sampler_reg, tc, 0, 2.0, vec2<i32>(3, 1)); + a += textureSampleLevel(image_2d_array, sampler_reg, tc, 0, level, offset); + a += textureSampleBias(image_2d_array, sampler_reg, tc, 0, 2.0, offset); a += textureSample(image_cube_array, sampler_reg, tc3, 0u); a += textureSampleLevel(image_cube_array, sampler_reg, tc3, 0u, level); a += textureSampleBias(image_cube_array, sampler_reg, tc3, 0u, 2.0); @@ -220,98 +221,98 @@ fn depth_no_comparison() -> @location(0) vec4<f32> { return s2d + s2d_gather + s2d_level; } " -OpName %32 "image_mipmapped_src" -OpName %34 "image_multisampled_src" -OpName %36 "image_depth_multisampled_src" -OpName %38 "image_storage_src" -OpName %40 "image_array_src" -OpName %42 "image_dup_src" -OpName %44 "image_1d_src" -OpName %46 "image_dst" -OpName %47 "image_1d" -OpName %49 "image_2d" -OpName %51 "image_2d_u32" -OpName %52 "image_2d_i32" -OpName %54 "image_2d_array" -OpName %56 "image_cube" -OpName %58 "image_cube_array" -OpName %60 "image_3d" -OpName %62 "image_aa" -OpName %64 "sampler_reg" -OpName %66 "sampler_cmp" -OpName %67 "image_2d_depth" -OpName %69 "image_2d_array_depth" -OpName %71 "image_cube_depth" -OpName %73 "naga_mod" -OpName %75 "lhs" -OpName %76 "rhs" -OpName %95 "local_id" -OpName %98 "main" -OpName %190 "local_id" -OpName %192 "depth_load" -OpName %212 "queries" -OpName %264 "levels_queries" -OpName %295 "texture_sample" +OpName %29 "image_mipmapped_src" +OpName %31 "image_multisampled_src" +OpName %33 "image_depth_multisampled_src" +OpName %35 "image_storage_src" +OpName %37 "image_array_src" +OpName %39 "image_dup_src" +OpName %41 "image_1d_src" +OpName %43 "image_dst" +OpName %44 "image_1d" +OpName %46 "image_2d" +OpName %48 "image_2d_u32" +OpName %49 "image_2d_i32" +OpName %51 "image_2d_array" +OpName %53 "image_cube" +OpName %55 "image_cube_array" +OpName %57 "image_3d" +OpName %59 "image_aa" +OpName %61 "sampler_reg" +OpName %63 "sampler_cmp" +OpName %64 "image_2d_depth" +OpName %66 "image_2d_array_depth" +OpName %68 "image_cube_depth" +OpName %70 "naga_mod" +OpName %72 "lhs" +OpName %73 "rhs" +OpName %93 "local_id" +OpName %96 "main" +OpName %188 "local_id" +OpName %190 "depth_load" +OpName %210 "queries" +OpName %262 "levels_queries" +OpName %293 "texture_sample" OpName %308 "a" -OpName %441 "texture_sample_comparison" -OpName %446 "a" -OpName %497 "gather" -OpName %531 "depth_no_comparison" -OpDecorate %32 DescriptorSet 0 -OpDecorate %32 Binding 0 -OpDecorate %34 DescriptorSet 0 -OpDecorate %34 Binding 3 -OpDecorate %36 DescriptorSet 0 -OpDecorate %36 Binding 4 -OpDecorate %38 NonWritable -OpDecorate %38 DescriptorSet 0 -OpDecorate %38 Binding 1 -OpDecorate %40 DescriptorSet 0 -OpDecorate %40 Binding 5 -OpDecorate %42 NonWritable -OpDecorate %42 DescriptorSet 0 -OpDecorate %42 Binding 6 +OpName %440 "texture_sample_comparison" +OpName %445 "a" +OpName %496 "gather" +OpName %530 "depth_no_comparison" +OpDecorate %29 DescriptorSet 0 +OpDecorate %29 Binding 0 +OpDecorate %31 DescriptorSet 0 +OpDecorate %31 Binding 3 +OpDecorate %33 DescriptorSet 0 +OpDecorate %33 Binding 4 +OpDecorate %35 NonWritable +OpDecorate %35 DescriptorSet 0 +OpDecorate %35 Binding 1 +OpDecorate %37 DescriptorSet 0 +OpDecorate %37 Binding 5 +OpDecorate %39 NonWritable +OpDecorate %39 DescriptorSet 0 +OpDecorate %39 Binding 6 +OpDecorate %41 DescriptorSet 0 +OpDecorate %41 Binding 7 +OpDecorate %43 NonReadable +OpDecorate %43 DescriptorSet 0 +OpDecorate %43 Binding 2 OpDecorate %44 DescriptorSet 0 -OpDecorate %44 Binding 7 -OpDecorate %46 NonReadable +OpDecorate %44 Binding 0 OpDecorate %46 DescriptorSet 0 -OpDecorate %46 Binding 2 -OpDecorate %47 DescriptorSet 0 -OpDecorate %47 Binding 0 +OpDecorate %46 Binding 1 +OpDecorate %48 DescriptorSet 0 +OpDecorate %48 Binding 2 OpDecorate %49 DescriptorSet 0 -OpDecorate %49 Binding 1 +OpDecorate %49 Binding 3 OpDecorate %51 DescriptorSet 0 -OpDecorate %51 Binding 2 -OpDecorate %52 DescriptorSet 0 -OpDecorate %52 Binding 3 -OpDecorate %54 DescriptorSet 0 -OpDecorate %54 Binding 4 -OpDecorate %56 DescriptorSet 0 -OpDecorate %56 Binding 5 -OpDecorate %58 DescriptorSet 0 -OpDecorate %58 Binding 6 -OpDecorate %60 DescriptorSet 0 -OpDecorate %60 Binding 7 -OpDecorate %62 DescriptorSet 0 -OpDecorate %62 Binding 8 +OpDecorate %51 Binding 4 +OpDecorate %53 DescriptorSet 0 +OpDecorate %53 Binding 5 +OpDecorate %55 DescriptorSet 0 +OpDecorate %55 Binding 6 +OpDecorate %57 DescriptorSet 0 +OpDecorate %57 Binding 7 +OpDecorate %59 DescriptorSet 0 +OpDecorate %59 Binding 8 +OpDecorate %61 DescriptorSet 1 +OpDecorate %61 Binding 0 +OpDecorate %63 DescriptorSet 1 +OpDecorate %63 Binding 1 OpDecorate %64 DescriptorSet 1 -OpDecorate %64 Binding 0 +OpDecorate %64 Binding 2 OpDecorate %66 DescriptorSet 1 -OpDecorate %66 Binding 1 -OpDecorate %67 DescriptorSet 1 -OpDecorate %67 Binding 2 -OpDecorate %69 DescriptorSet 1 -OpDecorate %69 Binding 3 -OpDecorate %71 DescriptorSet 1 -OpDecorate %71 Binding 4 -OpDecorate %95 BuiltIn LocalInvocationId -OpDecorate %190 BuiltIn LocalInvocationId -OpDecorate %210 BuiltIn Position -OpDecorate %263 BuiltIn Position -OpDecorate %294 Location 0 -OpDecorate %439 Location 0 -OpDecorate %496 Location 0 -OpDecorate %530 Location 0 +OpDecorate %66 Binding 3 +OpDecorate %68 DescriptorSet 1 +OpDecorate %68 Binding 4 +OpDecorate %93 BuiltIn LocalInvocationId +OpDecorate %188 BuiltIn LocalInvocationId +OpDecorate %208 BuiltIn Position +OpDecorate %261 BuiltIn Position +OpDecorate %292 Location 0 +OpDecorate %438 Location 0 +OpDecorate %495 Location 0 +OpDecorate %529 Location 0 %2 = OpTypeVoid %5 = OpTypeInt 32 0 %4 = OpTypeImage %5 2D 0 0 0 1 Unknown @@ -338,721 +339,721 @@ OpDecorate %530 Location 0 %26 = OpTypeImage %8 2D 1 0 0 1 Unknown %27 = OpTypeImage %8 2D 1 1 0 1 Unknown %28 = OpTypeImage %8 Cube 1 0 0 1 Unknown -%29 = OpConstant %15 3 -%30 = OpConstant %15 1 -%31 = OpConstantComposite %14 %29 %30 -%33 = OpTypePointer UniformConstant %4 -%32 = OpVariable %33 UniformConstant -%35 = OpTypePointer UniformConstant %6 -%34 = OpVariable %35 UniformConstant -%37 = OpTypePointer UniformConstant %7 -%36 = OpVariable %37 UniformConstant -%39 = OpTypePointer UniformConstant %9 -%38 = OpVariable %39 UniformConstant -%41 = OpTypePointer UniformConstant %10 -%40 = OpVariable %41 UniformConstant -%43 = OpTypePointer UniformConstant %11 -%42 = OpVariable %43 UniformConstant -%45 = OpTypePointer UniformConstant %12 +%30 = OpTypePointer UniformConstant %4 +%29 = OpVariable %30 UniformConstant +%32 = OpTypePointer UniformConstant %6 +%31 = OpVariable %32 UniformConstant +%34 = OpTypePointer UniformConstant %7 +%33 = OpVariable %34 UniformConstant +%36 = OpTypePointer UniformConstant %9 +%35 = OpVariable %36 UniformConstant +%38 = OpTypePointer UniformConstant %10 +%37 = OpVariable %38 UniformConstant +%40 = OpTypePointer UniformConstant %11 +%39 = OpVariable %40 UniformConstant +%42 = OpTypePointer UniformConstant %12 +%41 = OpVariable %42 UniformConstant +%43 = OpVariable %40 UniformConstant +%45 = OpTypePointer UniformConstant %16 %44 = OpVariable %45 UniformConstant -%46 = OpVariable %43 UniformConstant -%48 = OpTypePointer UniformConstant %16 -%47 = OpVariable %48 UniformConstant -%50 = OpTypePointer UniformConstant %17 +%47 = OpTypePointer UniformConstant %17 +%46 = OpVariable %47 UniformConstant +%48 = OpVariable %30 UniformConstant +%50 = OpTypePointer UniformConstant %18 %49 = OpVariable %50 UniformConstant -%51 = OpVariable %33 UniformConstant -%53 = OpTypePointer UniformConstant %18 -%52 = OpVariable %53 UniformConstant -%55 = OpTypePointer UniformConstant %19 -%54 = OpVariable %55 UniformConstant -%57 = OpTypePointer UniformConstant %20 -%56 = OpVariable %57 UniformConstant -%59 = OpTypePointer UniformConstant %21 -%58 = OpVariable %59 UniformConstant -%61 = OpTypePointer UniformConstant %22 -%60 = OpVariable %61 UniformConstant -%63 = OpTypePointer UniformConstant %23 -%62 = OpVariable %63 UniformConstant -%65 = OpTypePointer UniformConstant %25 +%52 = OpTypePointer UniformConstant %19 +%51 = OpVariable %52 UniformConstant +%54 = OpTypePointer UniformConstant %20 +%53 = OpVariable %54 UniformConstant +%56 = OpTypePointer UniformConstant %21 +%55 = OpVariable %56 UniformConstant +%58 = OpTypePointer UniformConstant %22 +%57 = OpVariable %58 UniformConstant +%60 = OpTypePointer UniformConstant %23 +%59 = OpVariable %60 UniformConstant +%62 = OpTypePointer UniformConstant %25 +%61 = OpVariable %62 UniformConstant +%63 = OpVariable %62 UniformConstant +%65 = OpTypePointer UniformConstant %26 %64 = OpVariable %65 UniformConstant -%66 = OpVariable %65 UniformConstant -%68 = OpTypePointer UniformConstant %26 -%67 = OpVariable %68 UniformConstant -%70 = OpTypePointer UniformConstant %27 -%69 = OpVariable %70 UniformConstant -%72 = OpTypePointer UniformConstant %28 -%71 = OpVariable %72 UniformConstant -%74 = OpTypeFunction %14 %14 %14 -%79 = OpTypeBool -%78 = OpTypeVector %79 2 -%80 = OpConstant %15 0 -%81 = OpConstantComposite %14 %80 %80 -%83 = OpConstant %15 -2147483648 -%84 = OpConstant %15 -1 -%85 = OpConstantComposite %14 %83 %83 -%86 = OpConstantComposite %14 %84 %84 -%91 = OpConstantComposite %14 %30 %30 -%96 = OpTypePointer Input %13 -%95 = OpVariable %96 Input -%99 = OpTypeFunction %2 -%106 = OpConstant %15 10 -%107 = OpConstant %15 20 -%108 = OpConstantComposite %14 %106 %107 -%110 = OpTypeVector %5 2 -%118 = OpTypeVector %5 4 -%132 = OpTypeVector %15 3 -%190 = OpVariable %96 Input -%211 = OpTypePointer Output %24 -%210 = OpVariable %211 Output -%221 = OpConstant %5 0 -%263 = OpVariable %211 Output -%294 = OpVariable %211 Output -%301 = OpConstant %8 0.5 -%302 = OpTypeVector %8 2 -%303 = OpConstantComposite %302 %301 %301 -%304 = OpTypeVector %8 3 -%305 = OpConstantComposite %304 %301 %301 %301 +%67 = OpTypePointer UniformConstant %27 +%66 = OpVariable %67 UniformConstant +%69 = OpTypePointer UniformConstant %28 +%68 = OpVariable %69 UniformConstant +%71 = OpTypeFunction %14 %14 %14 +%76 = OpTypeBool +%75 = OpTypeVector %76 2 +%77 = OpConstant %15 0 +%78 = OpConstantComposite %14 %77 %77 +%80 = OpConstant %15 -2147483648 +%81 = OpConstant %15 -1 +%82 = OpConstantComposite %14 %80 %80 +%83 = OpConstantComposite %14 %81 %81 +%88 = OpConstant %15 1 +%89 = OpConstantComposite %14 %88 %88 +%94 = OpTypePointer Input %13 +%93 = OpVariable %94 Input +%97 = OpTypeFunction %2 +%104 = OpConstant %15 10 +%105 = OpConstant %15 20 +%106 = OpConstantComposite %14 %104 %105 +%108 = OpTypeVector %5 2 +%116 = OpTypeVector %5 4 +%130 = OpTypeVector %15 3 +%188 = OpVariable %94 Input +%209 = OpTypePointer Output %24 +%208 = OpVariable %209 Output +%219 = OpConstant %5 0 +%261 = OpVariable %209 Output +%292 = OpVariable %209 Output +%299 = OpConstant %8 0.5 +%300 = OpTypeVector %8 2 +%301 = OpConstantComposite %300 %299 %299 +%302 = OpTypeVector %8 3 +%303 = OpConstantComposite %302 %299 %299 %299 +%304 = OpConstant %15 3 +%305 = OpConstantComposite %14 %304 %88 %306 = OpConstant %8 2.3 %307 = OpConstant %8 2.0 %309 = OpTypePointer Function %24 %310 = OpConstantNull %24 -%313 = OpTypeSampledImage %16 -%318 = OpTypeSampledImage %17 -%339 = OpTypeSampledImage %19 -%400 = OpTypeSampledImage %21 -%440 = OpTypePointer Output %8 -%439 = OpVariable %440 Output -%447 = OpTypePointer Function %8 -%448 = OpConstantNull %8 -%450 = OpTypeSampledImage %26 -%455 = OpTypeSampledImage %27 -%468 = OpTypeSampledImage %28 -%475 = OpConstant %8 0.0 -%496 = OpVariable %211 Output -%507 = OpConstant %5 1 -%510 = OpConstant %5 3 -%515 = OpTypeSampledImage %4 -%518 = OpTypeVector %15 4 -%519 = OpTypeSampledImage %18 -%530 = OpVariable %211 Output -%73 = OpFunction %14 None %74 -%75 = OpFunctionParameter %14 -%76 = OpFunctionParameter %14 -%77 = OpLabel -%82 = OpIEqual %78 %76 %81 -%87 = OpIEqual %78 %75 %85 -%88 = OpIEqual %78 %76 %86 -%89 = OpLogicalAnd %78 %87 %88 -%90 = OpLogicalOr %78 %82 %89 -%92 = OpSelect %14 %90 %91 %76 -%93 = OpSRem %14 %75 %92 -OpReturnValue %93 +%312 = OpTypeSampledImage %16 +%317 = OpTypeSampledImage %17 +%338 = OpTypeSampledImage %19 +%399 = OpTypeSampledImage %21 +%439 = OpTypePointer Output %8 +%438 = OpVariable %439 Output +%446 = OpTypePointer Function %8 +%447 = OpConstantNull %8 +%449 = OpTypeSampledImage %26 +%454 = OpTypeSampledImage %27 +%467 = OpTypeSampledImage %28 +%474 = OpConstant %8 0.0 +%495 = OpVariable %209 Output +%506 = OpConstant %5 1 +%509 = OpConstant %5 3 +%514 = OpTypeSampledImage %4 +%517 = OpTypeVector %15 4 +%518 = OpTypeSampledImage %18 +%529 = OpVariable %209 Output +%70 = OpFunction %14 None %71 +%72 = OpFunctionParameter %14 +%73 = OpFunctionParameter %14 +%74 = OpLabel +%79 = OpIEqual %75 %73 %78 +%84 = OpIEqual %75 %72 %82 +%85 = OpIEqual %75 %73 %83 +%86 = OpLogicalAnd %75 %84 %85 +%87 = OpLogicalOr %75 %79 %86 +%90 = OpSelect %14 %87 %89 %73 +%91 = OpSRem %14 %72 %90 +OpReturnValue %91 OpFunctionEnd -%98 = OpFunction %2 None %99 -%94 = OpLabel -%97 = OpLoad %13 %95 -%100 = OpLoad %4 %32 -%101 = OpLoad %6 %34 -%102 = OpLoad %9 %38 -%103 = OpLoad %10 %40 -%104 = OpLoad %12 %44 -%105 = OpLoad %11 %46 -OpBranch %109 -%109 = OpLabel +%96 = OpFunction %2 None %97 +%92 = OpLabel +%95 = OpLoad %13 %93 +%98 = OpLoad %4 %29 +%99 = OpLoad %6 %31 +%100 = OpLoad %9 %35 +%101 = OpLoad %10 %37 +%102 = OpLoad %12 %41 +%103 = OpLoad %11 %43 +OpBranch %107 +%107 = OpLabel OpLine %3 20 15 -%111 = OpImageQuerySize %110 %102 +%109 = OpImageQuerySize %108 %100 OpLine %3 21 15 -%112 = OpVectorShuffle %110 %97 %97 0 1 -%113 = OpIMul %110 %111 %112 -%114 = OpBitcast %14 %113 +%110 = OpVectorShuffle %108 %95 %95 0 1 +%111 = OpIMul %108 %109 %110 +%112 = OpBitcast %14 %111 OpLine %3 21 15 -%115 = OpFunctionCall %14 %73 %114 %108 +%113 = OpFunctionCall %14 %70 %112 %106 OpLine %3 23 18 -%116 = OpCompositeExtract %5 %97 2 -%117 = OpBitcast %15 %116 -%119 = OpImageFetch %118 %100 %115 Lod %117 +%114 = OpCompositeExtract %5 %95 2 +%115 = OpBitcast %15 %114 +%117 = OpImageFetch %116 %98 %113 Lod %115 OpLine %3 25 20 -%120 = OpCompositeExtract %5 %97 2 -%122 = OpImageFetch %118 %100 %115 Lod %120 +%118 = OpCompositeExtract %5 %95 2 +%120 = OpImageFetch %116 %98 %113 Lod %118 OpLine %3 26 18 -%123 = OpCompositeExtract %5 %97 2 -%124 = OpBitcast %15 %123 -%125 = OpImageFetch %118 %101 %115 Sample %124 +%121 = OpCompositeExtract %5 %95 2 +%122 = OpBitcast %15 %121 +%123 = OpImageFetch %116 %99 %113 Sample %122 OpLine %3 27 18 -%126 = OpImageRead %118 %102 %115 +%124 = OpImageRead %116 %100 %113 OpLine %3 28 52 -%127 = OpCompositeExtract %5 %97 2 -%128 = OpCompositeExtract %5 %97 2 -%129 = OpBitcast %15 %128 +%125 = OpCompositeExtract %5 %95 2 +%126 = OpCompositeExtract %5 %95 2 +%127 = OpBitcast %15 %126 OpLine %3 28 18 -%130 = OpIAdd %15 %129 %30 -%131 = OpBitcast %15 %127 -%133 = OpCompositeConstruct %132 %115 %131 -%134 = OpImageFetch %118 %103 %133 Lod %130 +%128 = OpIAdd %15 %127 %88 +%129 = OpBitcast %15 %125 +%131 = OpCompositeConstruct %130 %113 %129 +%132 = OpImageFetch %116 %101 %131 Lod %128 OpLine %3 29 52 -%135 = OpCompositeExtract %5 %97 2 +%133 = OpCompositeExtract %5 %95 2 +%134 = OpBitcast %15 %133 +%135 = OpCompositeExtract %5 %95 2 %136 = OpBitcast %15 %135 -%137 = OpCompositeExtract %5 %97 2 -%138 = OpBitcast %15 %137 OpLine %3 29 18 -%139 = OpIAdd %15 %138 %30 -%140 = OpCompositeConstruct %132 %115 %136 -%141 = OpImageFetch %118 %103 %140 Lod %139 +%137 = OpIAdd %15 %136 %88 +%138 = OpCompositeConstruct %130 %113 %134 +%139 = OpImageFetch %116 %101 %138 Lod %137 OpLine %3 30 18 -%142 = OpCompositeExtract %5 %97 0 +%140 = OpCompositeExtract %5 %95 0 +%141 = OpBitcast %15 %140 +%142 = OpCompositeExtract %5 %95 2 %143 = OpBitcast %15 %142 -%144 = OpCompositeExtract %5 %97 2 -%145 = OpBitcast %15 %144 -%146 = OpImageFetch %118 %104 %143 Lod %145 +%144 = OpImageFetch %116 %102 %141 Lod %143 OpLine %3 32 19 -%147 = OpBitcast %110 %115 -%148 = OpCompositeExtract %5 %97 2 -%149 = OpBitcast %15 %148 -%150 = OpImageFetch %118 %100 %147 Lod %149 +%145 = OpBitcast %108 %113 +%146 = OpCompositeExtract %5 %95 2 +%147 = OpBitcast %15 %146 +%148 = OpImageFetch %116 %98 %145 Lod %147 OpLine %3 33 19 -%151 = OpBitcast %110 %115 -%152 = OpCompositeExtract %5 %97 2 -%153 = OpBitcast %15 %152 -%154 = OpImageFetch %118 %101 %151 Sample %153 +%149 = OpBitcast %108 %113 +%150 = OpCompositeExtract %5 %95 2 +%151 = OpBitcast %15 %150 +%152 = OpImageFetch %116 %99 %149 Sample %151 OpLine %3 34 19 -%155 = OpBitcast %110 %115 -%156 = OpImageRead %118 %102 %155 +%153 = OpBitcast %108 %113 +%154 = OpImageRead %116 %100 %153 OpLine %3 35 48 -%157 = OpBitcast %110 %115 -%158 = OpCompositeExtract %5 %97 2 -%159 = OpCompositeExtract %5 %97 2 -%160 = OpBitcast %15 %159 +%155 = OpBitcast %108 %113 +%156 = OpCompositeExtract %5 %95 2 +%157 = OpCompositeExtract %5 %95 2 +%158 = OpBitcast %15 %157 OpLine %3 35 19 -%161 = OpIAdd %15 %160 %30 -%162 = OpCompositeConstruct %13 %157 %158 -%163 = OpImageFetch %118 %103 %162 Lod %161 +%159 = OpIAdd %15 %158 %88 +%160 = OpCompositeConstruct %13 %155 %156 +%161 = OpImageFetch %116 %101 %160 Lod %159 OpLine %3 36 48 -%164 = OpBitcast %110 %115 -%165 = OpCompositeExtract %5 %97 2 +%162 = OpBitcast %108 %113 +%163 = OpCompositeExtract %5 %95 2 +%164 = OpBitcast %15 %163 +%165 = OpCompositeExtract %5 %95 2 %166 = OpBitcast %15 %165 -%167 = OpCompositeExtract %5 %97 2 -%168 = OpBitcast %15 %167 OpLine %3 36 19 -%169 = OpIAdd %15 %168 %30 -%170 = OpBitcast %5 %166 -%171 = OpCompositeConstruct %13 %164 %170 -%172 = OpImageFetch %118 %103 %171 Lod %169 +%167 = OpIAdd %15 %166 %88 +%168 = OpBitcast %5 %164 +%169 = OpCompositeConstruct %13 %162 %168 +%170 = OpImageFetch %116 %101 %169 Lod %167 OpLine %3 37 19 -%173 = OpCompositeExtract %5 %97 0 -%175 = OpCompositeExtract %5 %97 2 -%176 = OpBitcast %15 %175 -%177 = OpImageFetch %118 %104 %173 Lod %176 +%171 = OpCompositeExtract %5 %95 0 +%173 = OpCompositeExtract %5 %95 2 +%174 = OpBitcast %15 %173 +%175 = OpImageFetch %116 %102 %171 Lod %174 OpLine %3 39 29 -%178 = OpCompositeExtract %15 %115 0 -%179 = OpIAdd %118 %119 %125 -%180 = OpIAdd %118 %179 %126 -%181 = OpIAdd %118 %180 %134 -%182 = OpIAdd %118 %181 %141 +%176 = OpCompositeExtract %15 %113 0 +%177 = OpIAdd %116 %117 %123 +%178 = OpIAdd %116 %177 %124 +%179 = OpIAdd %116 %178 %132 +%180 = OpIAdd %116 %179 %139 OpLine %3 39 5 -OpImageWrite %105 %178 %182 +OpImageWrite %103 %176 %180 OpLine %3 41 29 -%183 = OpCompositeExtract %15 %115 0 -%184 = OpBitcast %5 %183 -%185 = OpIAdd %118 %150 %154 -%186 = OpIAdd %118 %185 %156 -%187 = OpIAdd %118 %186 %163 -%188 = OpIAdd %118 %187 %172 +%181 = OpCompositeExtract %15 %113 0 +%182 = OpBitcast %5 %181 +%183 = OpIAdd %116 %148 %152 +%184 = OpIAdd %116 %183 %154 +%185 = OpIAdd %116 %184 %161 +%186 = OpIAdd %116 %185 %170 OpLine %3 41 5 -OpImageWrite %105 %184 %188 +OpImageWrite %103 %182 %186 OpReturn OpFunctionEnd -%192 = OpFunction %2 None %99 -%189 = OpLabel -%191 = OpLoad %13 %190 -%193 = OpLoad %7 %36 -%194 = OpLoad %9 %38 -%195 = OpLoad %11 %46 -OpBranch %196 -%196 = OpLabel +%190 = OpFunction %2 None %97 +%187 = OpLabel +%189 = OpLoad %13 %188 +%191 = OpLoad %7 %33 +%192 = OpLoad %9 %35 +%193 = OpLoad %11 %43 +OpBranch %194 +%194 = OpLabel OpLine %3 46 26 -%197 = OpImageQuerySize %110 %194 +%195 = OpImageQuerySize %108 %192 OpLine %3 47 27 -%198 = OpVectorShuffle %110 %191 %191 0 1 -%199 = OpIMul %110 %197 %198 -%200 = OpBitcast %14 %199 +%196 = OpVectorShuffle %108 %189 %189 0 1 +%197 = OpIMul %108 %195 %196 +%198 = OpBitcast %14 %197 OpLine %3 47 27 -%201 = OpFunctionCall %14 %73 %200 %108 +%199 = OpFunctionCall %14 %70 %198 %106 OpLine %3 48 20 -%202 = OpCompositeExtract %5 %191 2 -%203 = OpBitcast %15 %202 -%204 = OpImageFetch %24 %193 %201 Sample %203 -%205 = OpCompositeExtract %8 %204 0 +%200 = OpCompositeExtract %5 %189 2 +%201 = OpBitcast %15 %200 +%202 = OpImageFetch %24 %191 %199 Sample %201 +%203 = OpCompositeExtract %8 %202 0 OpLine %3 49 29 -%206 = OpCompositeExtract %15 %201 0 -%207 = OpConvertFToU %5 %205 -%208 = OpCompositeConstruct %118 %207 %207 %207 %207 +%204 = OpCompositeExtract %15 %199 0 +%205 = OpConvertFToU %5 %203 +%206 = OpCompositeConstruct %116 %205 %205 %205 %205 OpLine %3 49 5 -OpImageWrite %195 %206 %208 +OpImageWrite %193 %204 %206 OpReturn OpFunctionEnd -%212 = OpFunction %2 None %99 -%209 = OpLabel -%213 = OpLoad %16 %47 -%214 = OpLoad %17 %49 -%215 = OpLoad %19 %54 -%216 = OpLoad %20 %56 -%217 = OpLoad %21 %58 -%218 = OpLoad %22 %60 -%219 = OpLoad %23 %62 -OpBranch %220 -%220 = OpLabel +%210 = OpFunction %2 None %97 +%207 = OpLabel +%211 = OpLoad %16 %44 +%212 = OpLoad %17 %46 +%213 = OpLoad %19 %51 +%214 = OpLoad %20 %53 +%215 = OpLoad %21 %55 +%216 = OpLoad %22 %57 +%217 = OpLoad %23 %59 +OpBranch %218 +%218 = OpLabel OpLine %3 74 18 -%222 = OpImageQuerySizeLod %5 %213 %221 +%220 = OpImageQuerySizeLod %5 %211 %219 OpLine %3 75 22 -%223 = OpBitcast %15 %222 -%224 = OpImageQuerySizeLod %5 %213 %223 +%221 = OpBitcast %15 %220 +%222 = OpImageQuerySizeLod %5 %211 %221 OpLine %3 76 18 -%225 = OpImageQuerySizeLod %110 %214 %221 +%223 = OpImageQuerySizeLod %108 %212 %219 OpLine %3 77 22 -%226 = OpImageQuerySizeLod %110 %214 %30 +%224 = OpImageQuerySizeLod %108 %212 %88 OpLine %3 78 24 -%227 = OpImageQuerySizeLod %13 %215 %221 -%228 = OpVectorShuffle %110 %227 %227 0 1 +%225 = OpImageQuerySizeLod %13 %213 %219 +%226 = OpVectorShuffle %108 %225 %225 0 1 OpLine %3 79 28 -%229 = OpImageQuerySizeLod %13 %215 %30 -%230 = OpVectorShuffle %110 %229 %229 0 1 +%227 = OpImageQuerySizeLod %13 %213 %88 +%228 = OpVectorShuffle %108 %227 %227 0 1 OpLine %3 80 20 -%231 = OpImageQuerySizeLod %110 %216 %221 +%229 = OpImageQuerySizeLod %108 %214 %219 OpLine %3 81 24 -%232 = OpImageQuerySizeLod %110 %216 %30 +%230 = OpImageQuerySizeLod %108 %214 %88 OpLine %3 82 26 -%233 = OpImageQuerySizeLod %13 %217 %221 -%234 = OpVectorShuffle %110 %233 %233 0 0 +%231 = OpImageQuerySizeLod %13 %215 %219 +%232 = OpVectorShuffle %108 %231 %231 0 0 OpLine %3 83 30 -%235 = OpImageQuerySizeLod %13 %217 %30 -%236 = OpVectorShuffle %110 %235 %235 0 0 +%233 = OpImageQuerySizeLod %13 %215 %88 +%234 = OpVectorShuffle %108 %233 %233 0 0 OpLine %3 84 18 -%237 = OpImageQuerySizeLod %13 %218 %221 +%235 = OpImageQuerySizeLod %13 %216 %219 OpLine %3 85 22 -%238 = OpImageQuerySizeLod %13 %218 %30 +%236 = OpImageQuerySizeLod %13 %216 %88 OpLine %3 86 21 -%239 = OpImageQuerySize %110 %219 +%237 = OpImageQuerySize %108 %217 OpLine %3 88 15 -%240 = OpCompositeExtract %5 %225 1 -%241 = OpIAdd %5 %222 %240 +%238 = OpCompositeExtract %5 %223 1 +%239 = OpIAdd %5 %220 %238 +%240 = OpCompositeExtract %5 %224 1 +%241 = OpIAdd %5 %239 %240 %242 = OpCompositeExtract %5 %226 1 %243 = OpIAdd %5 %241 %242 %244 = OpCompositeExtract %5 %228 1 %245 = OpIAdd %5 %243 %244 -%246 = OpCompositeExtract %5 %230 1 +%246 = OpCompositeExtract %5 %229 1 %247 = OpIAdd %5 %245 %246 -%248 = OpCompositeExtract %5 %231 1 +%248 = OpCompositeExtract %5 %230 1 %249 = OpIAdd %5 %247 %248 %250 = OpCompositeExtract %5 %232 1 %251 = OpIAdd %5 %249 %250 %252 = OpCompositeExtract %5 %234 1 %253 = OpIAdd %5 %251 %252 -%254 = OpCompositeExtract %5 %236 1 +%254 = OpCompositeExtract %5 %235 2 %255 = OpIAdd %5 %253 %254 -%256 = OpCompositeExtract %5 %237 2 +%256 = OpCompositeExtract %5 %236 2 %257 = OpIAdd %5 %255 %256 -%258 = OpCompositeExtract %5 %238 2 -%259 = OpIAdd %5 %257 %258 OpLine %3 91 12 -%260 = OpConvertUToF %8 %259 -%261 = OpCompositeConstruct %24 %260 %260 %260 %260 -OpStore %210 %261 +%258 = OpConvertUToF %8 %257 +%259 = OpCompositeConstruct %24 %258 %258 %258 %258 +OpStore %208 %259 OpReturn OpFunctionEnd -%264 = OpFunction %2 None %99 -%262 = OpLabel -%265 = OpLoad %17 %49 -%266 = OpLoad %19 %54 -%267 = OpLoad %20 %56 -%268 = OpLoad %21 %58 -%269 = OpLoad %22 %60 -%270 = OpLoad %23 %62 -OpBranch %271 -%271 = OpLabel +%262 = OpFunction %2 None %97 +%260 = OpLabel +%263 = OpLoad %17 %46 +%264 = OpLoad %19 %51 +%265 = OpLoad %20 %53 +%266 = OpLoad %21 %55 +%267 = OpLoad %22 %57 +%268 = OpLoad %23 %59 +OpBranch %269 +%269 = OpLabel OpLine %3 96 25 -%272 = OpImageQueryLevels %5 %265 +%270 = OpImageQueryLevels %5 %263 OpLine %3 97 25 -%273 = OpImageQuerySizeLod %13 %266 %221 -%274 = OpCompositeExtract %5 %273 2 +%271 = OpImageQuerySizeLod %13 %264 %219 +%272 = OpCompositeExtract %5 %271 2 OpLine %3 98 31 -%275 = OpImageQueryLevels %5 %266 +%273 = OpImageQueryLevels %5 %264 OpLine %3 99 31 -%276 = OpImageQuerySizeLod %13 %266 %221 -%277 = OpCompositeExtract %5 %276 2 +%274 = OpImageQuerySizeLod %13 %264 %219 +%275 = OpCompositeExtract %5 %274 2 OpLine %3 100 27 -%278 = OpImageQueryLevels %5 %267 +%276 = OpImageQueryLevels %5 %265 OpLine %3 101 33 -%279 = OpImageQueryLevels %5 %268 +%277 = OpImageQueryLevels %5 %266 OpLine %3 102 27 -%280 = OpImageQuerySizeLod %13 %268 %221 -%281 = OpCompositeExtract %5 %280 2 +%278 = OpImageQuerySizeLod %13 %266 %219 +%279 = OpCompositeExtract %5 %278 2 OpLine %3 103 25 -%282 = OpImageQueryLevels %5 %269 +%280 = OpImageQueryLevels %5 %267 OpLine %3 104 26 -%283 = OpImageQuerySamples %5 %270 +%281 = OpImageQuerySamples %5 %268 OpLine %3 106 15 -%284 = OpIAdd %5 %274 %281 -%285 = OpIAdd %5 %284 %283 -%286 = OpIAdd %5 %285 %272 -%287 = OpIAdd %5 %286 %275 -%288 = OpIAdd %5 %287 %282 -%289 = OpIAdd %5 %288 %278 -%290 = OpIAdd %5 %289 %279 +%282 = OpIAdd %5 %272 %279 +%283 = OpIAdd %5 %282 %281 +%284 = OpIAdd %5 %283 %270 +%285 = OpIAdd %5 %284 %273 +%286 = OpIAdd %5 %285 %280 +%287 = OpIAdd %5 %286 %276 +%288 = OpIAdd %5 %287 %277 OpLine %3 108 12 -%291 = OpConvertUToF %8 %290 -%292 = OpCompositeConstruct %24 %291 %291 %291 %291 -OpStore %263 %292 +%289 = OpConvertUToF %8 %288 +%290 = OpCompositeConstruct %24 %289 %289 %289 %289 +OpStore %261 %290 OpReturn OpFunctionEnd -%295 = OpFunction %2 None %99 -%293 = OpLabel +%293 = OpFunction %2 None %97 +%291 = OpLabel %308 = OpVariable %309 Function %310 -%296 = OpLoad %16 %47 -%297 = OpLoad %17 %49 -%298 = OpLoad %19 %54 -%299 = OpLoad %21 %58 -%300 = OpLoad %25 %64 +%294 = OpLoad %16 %44 +%295 = OpLoad %17 %46 +%296 = OpLoad %19 %51 +%297 = OpLoad %21 %55 +%298 = OpLoad %25 %61 OpBranch %311 %311 = OpLabel -OpLine %3 116 14 -OpLine %3 117 15 -OpLine %3 120 5 -%312 = OpCompositeExtract %8 %303 0 -%314 = OpSampledImage %313 %296 %300 -%315 = OpImageSampleImplicitLod %24 %314 %312 -%316 = OpLoad %24 %308 -%317 = OpFAdd %24 %316 %315 -OpLine %3 120 5 -OpStore %308 %317 +OpLine %3 116 16 +OpLine %3 117 17 +OpLine %3 118 20 OpLine %3 121 5 -%319 = OpSampledImage %318 %297 %300 -%320 = OpImageSampleImplicitLod %24 %319 %303 -%321 = OpLoad %24 %308 -%322 = OpFAdd %24 %321 %320 +%313 = OpSampledImage %312 %294 %298 +%314 = OpImageSampleImplicitLod %24 %313 %299 +%315 = OpLoad %24 %308 +%316 = OpFAdd %24 %315 %314 OpLine %3 121 5 -OpStore %308 %322 +OpStore %308 %316 OpLine %3 122 5 -%323 = OpSampledImage %318 %297 %300 -%324 = OpImageSampleImplicitLod %24 %323 %303 ConstOffset %31 -%325 = OpLoad %24 %308 -%326 = OpFAdd %24 %325 %324 +%318 = OpSampledImage %317 %295 %298 +%319 = OpImageSampleImplicitLod %24 %318 %301 +%320 = OpLoad %24 %308 +%321 = OpFAdd %24 %320 %319 OpLine %3 122 5 -OpStore %308 %326 +OpStore %308 %321 OpLine %3 123 5 -%327 = OpSampledImage %318 %297 %300 -%328 = OpImageSampleExplicitLod %24 %327 %303 Lod %306 -%329 = OpLoad %24 %308 -%330 = OpFAdd %24 %329 %328 +%322 = OpSampledImage %317 %295 %298 +%323 = OpImageSampleImplicitLod %24 %322 %301 ConstOffset %305 +%324 = OpLoad %24 %308 +%325 = OpFAdd %24 %324 %323 OpLine %3 123 5 -OpStore %308 %330 +OpStore %308 %325 OpLine %3 124 5 -%331 = OpSampledImage %318 %297 %300 -%332 = OpImageSampleExplicitLod %24 %331 %303 Lod|ConstOffset %306 %31 -%333 = OpLoad %24 %308 -%334 = OpFAdd %24 %333 %332 +%326 = OpSampledImage %317 %295 %298 +%327 = OpImageSampleExplicitLod %24 %326 %301 Lod %306 +%328 = OpLoad %24 %308 +%329 = OpFAdd %24 %328 %327 OpLine %3 124 5 -OpStore %308 %334 +OpStore %308 %329 OpLine %3 125 5 -%335 = OpSampledImage %318 %297 %300 -%336 = OpImageSampleImplicitLod %24 %335 %303 Bias|ConstOffset %307 %31 -%337 = OpLoad %24 %308 -%338 = OpFAdd %24 %337 %336 +%330 = OpSampledImage %317 %295 %298 +%331 = OpImageSampleExplicitLod %24 %330 %301 Lod|ConstOffset %306 %305 +%332 = OpLoad %24 %308 +%333 = OpFAdd %24 %332 %331 OpLine %3 125 5 -OpStore %308 %338 +OpStore %308 %333 OpLine %3 126 5 -%340 = OpConvertUToF %8 %221 -%341 = OpCompositeConstruct %304 %303 %340 -%342 = OpSampledImage %339 %298 %300 -%343 = OpImageSampleImplicitLod %24 %342 %341 -%344 = OpLoad %24 %308 -%345 = OpFAdd %24 %344 %343 +%334 = OpSampledImage %317 %295 %298 +%335 = OpImageSampleImplicitLod %24 %334 %301 Bias|ConstOffset %307 %305 +%336 = OpLoad %24 %308 +%337 = OpFAdd %24 %336 %335 OpLine %3 126 5 -OpStore %308 %345 +OpStore %308 %337 OpLine %3 127 5 -%346 = OpConvertUToF %8 %221 -%347 = OpCompositeConstruct %304 %303 %346 -%348 = OpSampledImage %339 %298 %300 -%349 = OpImageSampleImplicitLod %24 %348 %347 ConstOffset %31 -%350 = OpLoad %24 %308 -%351 = OpFAdd %24 %350 %349 +%339 = OpConvertUToF %8 %219 +%340 = OpCompositeConstruct %302 %301 %339 +%341 = OpSampledImage %338 %296 %298 +%342 = OpImageSampleImplicitLod %24 %341 %340 +%343 = OpLoad %24 %308 +%344 = OpFAdd %24 %343 %342 OpLine %3 127 5 -OpStore %308 %351 +OpStore %308 %344 OpLine %3 128 5 -%352 = OpConvertUToF %8 %221 -%353 = OpCompositeConstruct %304 %303 %352 -%354 = OpSampledImage %339 %298 %300 -%355 = OpImageSampleExplicitLod %24 %354 %353 Lod %306 -%356 = OpLoad %24 %308 -%357 = OpFAdd %24 %356 %355 +%345 = OpConvertUToF %8 %219 +%346 = OpCompositeConstruct %302 %301 %345 +%347 = OpSampledImage %338 %296 %298 +%348 = OpImageSampleImplicitLod %24 %347 %346 ConstOffset %305 +%349 = OpLoad %24 %308 +%350 = OpFAdd %24 %349 %348 OpLine %3 128 5 -OpStore %308 %357 +OpStore %308 %350 OpLine %3 129 5 -%358 = OpConvertUToF %8 %221 -%359 = OpCompositeConstruct %304 %303 %358 -%360 = OpSampledImage %339 %298 %300 -%361 = OpImageSampleExplicitLod %24 %360 %359 Lod|ConstOffset %306 %31 -%362 = OpLoad %24 %308 -%363 = OpFAdd %24 %362 %361 +%351 = OpConvertUToF %8 %219 +%352 = OpCompositeConstruct %302 %301 %351 +%353 = OpSampledImage %338 %296 %298 +%354 = OpImageSampleExplicitLod %24 %353 %352 Lod %306 +%355 = OpLoad %24 %308 +%356 = OpFAdd %24 %355 %354 OpLine %3 129 5 -OpStore %308 %363 +OpStore %308 %356 OpLine %3 130 5 -%364 = OpConvertUToF %8 %221 -%365 = OpCompositeConstruct %304 %303 %364 -%366 = OpSampledImage %339 %298 %300 -%367 = OpImageSampleImplicitLod %24 %366 %365 Bias|ConstOffset %307 %31 -%368 = OpLoad %24 %308 -%369 = OpFAdd %24 %368 %367 +%357 = OpConvertUToF %8 %219 +%358 = OpCompositeConstruct %302 %301 %357 +%359 = OpSampledImage %338 %296 %298 +%360 = OpImageSampleExplicitLod %24 %359 %358 Lod|ConstOffset %306 %305 +%361 = OpLoad %24 %308 +%362 = OpFAdd %24 %361 %360 OpLine %3 130 5 -OpStore %308 %369 +OpStore %308 %362 OpLine %3 131 5 -%370 = OpConvertSToF %8 %80 -%371 = OpCompositeConstruct %304 %303 %370 -%372 = OpSampledImage %339 %298 %300 -%373 = OpImageSampleImplicitLod %24 %372 %371 -%374 = OpLoad %24 %308 -%375 = OpFAdd %24 %374 %373 +%363 = OpConvertUToF %8 %219 +%364 = OpCompositeConstruct %302 %301 %363 +%365 = OpSampledImage %338 %296 %298 +%366 = OpImageSampleImplicitLod %24 %365 %364 Bias|ConstOffset %307 %305 +%367 = OpLoad %24 %308 +%368 = OpFAdd %24 %367 %366 OpLine %3 131 5 -OpStore %308 %375 +OpStore %308 %368 OpLine %3 132 5 -%376 = OpConvertSToF %8 %80 -%377 = OpCompositeConstruct %304 %303 %376 -%378 = OpSampledImage %339 %298 %300 -%379 = OpImageSampleImplicitLod %24 %378 %377 ConstOffset %31 -%380 = OpLoad %24 %308 -%381 = OpFAdd %24 %380 %379 +%369 = OpConvertSToF %8 %77 +%370 = OpCompositeConstruct %302 %301 %369 +%371 = OpSampledImage %338 %296 %298 +%372 = OpImageSampleImplicitLod %24 %371 %370 +%373 = OpLoad %24 %308 +%374 = OpFAdd %24 %373 %372 OpLine %3 132 5 -OpStore %308 %381 +OpStore %308 %374 OpLine %3 133 5 -%382 = OpConvertSToF %8 %80 -%383 = OpCompositeConstruct %304 %303 %382 -%384 = OpSampledImage %339 %298 %300 -%385 = OpImageSampleExplicitLod %24 %384 %383 Lod %306 -%386 = OpLoad %24 %308 -%387 = OpFAdd %24 %386 %385 +%375 = OpConvertSToF %8 %77 +%376 = OpCompositeConstruct %302 %301 %375 +%377 = OpSampledImage %338 %296 %298 +%378 = OpImageSampleImplicitLod %24 %377 %376 ConstOffset %305 +%379 = OpLoad %24 %308 +%380 = OpFAdd %24 %379 %378 OpLine %3 133 5 -OpStore %308 %387 +OpStore %308 %380 OpLine %3 134 5 -%388 = OpConvertSToF %8 %80 -%389 = OpCompositeConstruct %304 %303 %388 -%390 = OpSampledImage %339 %298 %300 -%391 = OpImageSampleExplicitLod %24 %390 %389 Lod|ConstOffset %306 %31 -%392 = OpLoad %24 %308 -%393 = OpFAdd %24 %392 %391 +%381 = OpConvertSToF %8 %77 +%382 = OpCompositeConstruct %302 %301 %381 +%383 = OpSampledImage %338 %296 %298 +%384 = OpImageSampleExplicitLod %24 %383 %382 Lod %306 +%385 = OpLoad %24 %308 +%386 = OpFAdd %24 %385 %384 OpLine %3 134 5 -OpStore %308 %393 +OpStore %308 %386 OpLine %3 135 5 -%394 = OpConvertSToF %8 %80 -%395 = OpCompositeConstruct %304 %303 %394 -%396 = OpSampledImage %339 %298 %300 -%397 = OpImageSampleImplicitLod %24 %396 %395 Bias|ConstOffset %307 %31 -%398 = OpLoad %24 %308 -%399 = OpFAdd %24 %398 %397 +%387 = OpConvertSToF %8 %77 +%388 = OpCompositeConstruct %302 %301 %387 +%389 = OpSampledImage %338 %296 %298 +%390 = OpImageSampleExplicitLod %24 %389 %388 Lod|ConstOffset %306 %305 +%391 = OpLoad %24 %308 +%392 = OpFAdd %24 %391 %390 OpLine %3 135 5 -OpStore %308 %399 +OpStore %308 %392 OpLine %3 136 5 -%401 = OpConvertUToF %8 %221 -%402 = OpCompositeConstruct %24 %305 %401 -%403 = OpSampledImage %400 %299 %300 -%404 = OpImageSampleImplicitLod %24 %403 %402 -%405 = OpLoad %24 %308 -%406 = OpFAdd %24 %405 %404 +%393 = OpConvertSToF %8 %77 +%394 = OpCompositeConstruct %302 %301 %393 +%395 = OpSampledImage %338 %296 %298 +%396 = OpImageSampleImplicitLod %24 %395 %394 Bias|ConstOffset %307 %305 +%397 = OpLoad %24 %308 +%398 = OpFAdd %24 %397 %396 OpLine %3 136 5 -OpStore %308 %406 +OpStore %308 %398 OpLine %3 137 5 -%407 = OpConvertUToF %8 %221 -%408 = OpCompositeConstruct %24 %305 %407 -%409 = OpSampledImage %400 %299 %300 -%410 = OpImageSampleExplicitLod %24 %409 %408 Lod %306 -%411 = OpLoad %24 %308 -%412 = OpFAdd %24 %411 %410 +%400 = OpConvertUToF %8 %219 +%401 = OpCompositeConstruct %24 %303 %400 +%402 = OpSampledImage %399 %297 %298 +%403 = OpImageSampleImplicitLod %24 %402 %401 +%404 = OpLoad %24 %308 +%405 = OpFAdd %24 %404 %403 OpLine %3 137 5 -OpStore %308 %412 +OpStore %308 %405 OpLine %3 138 5 -%413 = OpConvertUToF %8 %221 -%414 = OpCompositeConstruct %24 %305 %413 -%415 = OpSampledImage %400 %299 %300 -%416 = OpImageSampleImplicitLod %24 %415 %414 Bias %307 -%417 = OpLoad %24 %308 -%418 = OpFAdd %24 %417 %416 +%406 = OpConvertUToF %8 %219 +%407 = OpCompositeConstruct %24 %303 %406 +%408 = OpSampledImage %399 %297 %298 +%409 = OpImageSampleExplicitLod %24 %408 %407 Lod %306 +%410 = OpLoad %24 %308 +%411 = OpFAdd %24 %410 %409 OpLine %3 138 5 -OpStore %308 %418 +OpStore %308 %411 OpLine %3 139 5 -%419 = OpConvertSToF %8 %80 -%420 = OpCompositeConstruct %24 %305 %419 -%421 = OpSampledImage %400 %299 %300 -%422 = OpImageSampleImplicitLod %24 %421 %420 -%423 = OpLoad %24 %308 -%424 = OpFAdd %24 %423 %422 +%412 = OpConvertUToF %8 %219 +%413 = OpCompositeConstruct %24 %303 %412 +%414 = OpSampledImage %399 %297 %298 +%415 = OpImageSampleImplicitLod %24 %414 %413 Bias %307 +%416 = OpLoad %24 %308 +%417 = OpFAdd %24 %416 %415 OpLine %3 139 5 -OpStore %308 %424 +OpStore %308 %417 OpLine %3 140 5 -%425 = OpConvertSToF %8 %80 -%426 = OpCompositeConstruct %24 %305 %425 -%427 = OpSampledImage %400 %299 %300 -%428 = OpImageSampleExplicitLod %24 %427 %426 Lod %306 -%429 = OpLoad %24 %308 -%430 = OpFAdd %24 %429 %428 +%418 = OpConvertSToF %8 %77 +%419 = OpCompositeConstruct %24 %303 %418 +%420 = OpSampledImage %399 %297 %298 +%421 = OpImageSampleImplicitLod %24 %420 %419 +%422 = OpLoad %24 %308 +%423 = OpFAdd %24 %422 %421 OpLine %3 140 5 -OpStore %308 %430 +OpStore %308 %423 OpLine %3 141 5 -%431 = OpConvertSToF %8 %80 -%432 = OpCompositeConstruct %24 %305 %431 -%433 = OpSampledImage %400 %299 %300 -%434 = OpImageSampleImplicitLod %24 %433 %432 Bias %307 -%435 = OpLoad %24 %308 -%436 = OpFAdd %24 %435 %434 +%424 = OpConvertSToF %8 %77 +%425 = OpCompositeConstruct %24 %303 %424 +%426 = OpSampledImage %399 %297 %298 +%427 = OpImageSampleExplicitLod %24 %426 %425 Lod %306 +%428 = OpLoad %24 %308 +%429 = OpFAdd %24 %428 %427 OpLine %3 141 5 -OpStore %308 %436 +OpStore %308 %429 +OpLine %3 142 5 +%430 = OpConvertSToF %8 %77 +%431 = OpCompositeConstruct %24 %303 %430 +%432 = OpSampledImage %399 %297 %298 +%433 = OpImageSampleImplicitLod %24 %432 %431 Bias %307 +%434 = OpLoad %24 %308 +%435 = OpFAdd %24 %434 %433 +OpLine %3 142 5 +OpStore %308 %435 OpLine %3 1 1 -%437 = OpLoad %24 %308 -OpStore %294 %437 +%436 = OpLoad %24 %308 +OpStore %292 %436 OpReturn OpFunctionEnd -%441 = OpFunction %2 None %99 -%438 = OpLabel -%446 = OpVariable %447 Function %448 -%442 = OpLoad %25 %66 -%443 = OpLoad %26 %67 -%444 = OpLoad %27 %69 -%445 = OpLoad %28 %71 -OpBranch %449 -%449 = OpLabel -OpLine %3 156 14 -OpLine %3 157 15 -OpLine %3 160 5 -%451 = OpSampledImage %450 %443 %442 -%452 = OpImageSampleDrefImplicitLod %8 %451 %303 %301 -%453 = OpLoad %8 %446 -%454 = OpFAdd %8 %453 %452 -OpLine %3 160 5 -OpStore %446 %454 +%440 = OpFunction %2 None %97 +%437 = OpLabel +%445 = OpVariable %446 Function %447 +%441 = OpLoad %25 %63 +%442 = OpLoad %26 %64 +%443 = OpLoad %27 %66 +%444 = OpLoad %28 %68 +OpBranch %448 +%448 = OpLabel +OpLine %3 157 14 +OpLine %3 158 15 OpLine %3 161 5 -%456 = OpConvertUToF %8 %221 -%457 = OpCompositeConstruct %304 %303 %456 -%458 = OpSampledImage %455 %444 %442 -%459 = OpImageSampleDrefImplicitLod %8 %458 %457 %301 -%460 = OpLoad %8 %446 -%461 = OpFAdd %8 %460 %459 +%450 = OpSampledImage %449 %442 %441 +%451 = OpImageSampleDrefImplicitLod %8 %450 %301 %299 +%452 = OpLoad %8 %445 +%453 = OpFAdd %8 %452 %451 OpLine %3 161 5 -OpStore %446 %461 +OpStore %445 %453 OpLine %3 162 5 -%462 = OpConvertSToF %8 %80 -%463 = OpCompositeConstruct %304 %303 %462 -%464 = OpSampledImage %455 %444 %442 -%465 = OpImageSampleDrefImplicitLod %8 %464 %463 %301 -%466 = OpLoad %8 %446 -%467 = OpFAdd %8 %466 %465 +%455 = OpConvertUToF %8 %219 +%456 = OpCompositeConstruct %302 %301 %455 +%457 = OpSampledImage %454 %443 %441 +%458 = OpImageSampleDrefImplicitLod %8 %457 %456 %299 +%459 = OpLoad %8 %445 +%460 = OpFAdd %8 %459 %458 OpLine %3 162 5 -OpStore %446 %467 +OpStore %445 %460 OpLine %3 163 5 -%469 = OpSampledImage %468 %445 %442 -%470 = OpImageSampleDrefImplicitLod %8 %469 %305 %301 -%471 = OpLoad %8 %446 -%472 = OpFAdd %8 %471 %470 +%461 = OpConvertSToF %8 %77 +%462 = OpCompositeConstruct %302 %301 %461 +%463 = OpSampledImage %454 %443 %441 +%464 = OpImageSampleDrefImplicitLod %8 %463 %462 %299 +%465 = OpLoad %8 %445 +%466 = OpFAdd %8 %465 %464 OpLine %3 163 5 -OpStore %446 %472 +OpStore %445 %466 OpLine %3 164 5 -%473 = OpSampledImage %450 %443 %442 -%474 = OpImageSampleDrefExplicitLod %8 %473 %303 %301 Lod %475 -%476 = OpLoad %8 %446 -%477 = OpFAdd %8 %476 %474 +%468 = OpSampledImage %467 %444 %441 +%469 = OpImageSampleDrefImplicitLod %8 %468 %303 %299 +%470 = OpLoad %8 %445 +%471 = OpFAdd %8 %470 %469 OpLine %3 164 5 -OpStore %446 %477 +OpStore %445 %471 OpLine %3 165 5 -%478 = OpConvertUToF %8 %221 -%479 = OpCompositeConstruct %304 %303 %478 -%480 = OpSampledImage %455 %444 %442 -%481 = OpImageSampleDrefExplicitLod %8 %480 %479 %301 Lod %475 -%482 = OpLoad %8 %446 -%483 = OpFAdd %8 %482 %481 +%472 = OpSampledImage %449 %442 %441 +%473 = OpImageSampleDrefExplicitLod %8 %472 %301 %299 Lod %474 +%475 = OpLoad %8 %445 +%476 = OpFAdd %8 %475 %473 OpLine %3 165 5 -OpStore %446 %483 +OpStore %445 %476 OpLine %3 166 5 -%484 = OpConvertSToF %8 %80 -%485 = OpCompositeConstruct %304 %303 %484 -%486 = OpSampledImage %455 %444 %442 -%487 = OpImageSampleDrefExplicitLod %8 %486 %485 %301 Lod %475 -%488 = OpLoad %8 %446 -%489 = OpFAdd %8 %488 %487 +%477 = OpConvertUToF %8 %219 +%478 = OpCompositeConstruct %302 %301 %477 +%479 = OpSampledImage %454 %443 %441 +%480 = OpImageSampleDrefExplicitLod %8 %479 %478 %299 Lod %474 +%481 = OpLoad %8 %445 +%482 = OpFAdd %8 %481 %480 OpLine %3 166 5 -OpStore %446 %489 +OpStore %445 %482 OpLine %3 167 5 -%490 = OpSampledImage %468 %445 %442 -%491 = OpImageSampleDrefExplicitLod %8 %490 %305 %301 Lod %475 -%492 = OpLoad %8 %446 -%493 = OpFAdd %8 %492 %491 +%483 = OpConvertSToF %8 %77 +%484 = OpCompositeConstruct %302 %301 %483 +%485 = OpSampledImage %454 %443 %441 +%486 = OpImageSampleDrefExplicitLod %8 %485 %484 %299 Lod %474 +%487 = OpLoad %8 %445 +%488 = OpFAdd %8 %487 %486 OpLine %3 167 5 -OpStore %446 %493 +OpStore %445 %488 +OpLine %3 168 5 +%489 = OpSampledImage %467 %444 %441 +%490 = OpImageSampleDrefExplicitLod %8 %489 %303 %299 Lod %474 +%491 = OpLoad %8 %445 +%492 = OpFAdd %8 %491 %490 +OpLine %3 168 5 +OpStore %445 %492 OpLine %3 1 1 -%494 = OpLoad %8 %446 -OpStore %439 %494 +%493 = OpLoad %8 %445 +OpStore %438 %493 OpReturn OpFunctionEnd -%497 = OpFunction %2 None %99 -%495 = OpLabel -%498 = OpLoad %17 %49 -%499 = OpLoad %4 %51 -%500 = OpLoad %18 %52 -%501 = OpLoad %25 %64 -%502 = OpLoad %25 %66 -%503 = OpLoad %26 %67 -OpBranch %504 -%504 = OpLabel -OpLine %3 173 14 -OpLine %3 175 15 -%505 = OpSampledImage %318 %498 %501 -%506 = OpImageGather %24 %505 %303 %507 -OpLine %3 176 22 -%508 = OpSampledImage %318 %498 %501 -%509 = OpImageGather %24 %508 %303 %510 ConstOffset %31 -OpLine %3 177 21 -%511 = OpSampledImage %450 %503 %502 -%512 = OpImageDrefGather %24 %511 %303 %301 -OpLine %3 178 28 -%513 = OpSampledImage %450 %503 %502 -%514 = OpImageDrefGather %24 %513 %303 %301 ConstOffset %31 -OpLine %3 180 13 -%516 = OpSampledImage %515 %499 %501 -%517 = OpImageGather %118 %516 %303 %221 +%496 = OpFunction %2 None %97 +%494 = OpLabel +%497 = OpLoad %17 %46 +%498 = OpLoad %4 %48 +%499 = OpLoad %18 %49 +%500 = OpLoad %25 %61 +%501 = OpLoad %25 %63 +%502 = OpLoad %26 %64 +OpBranch %503 +%503 = OpLabel +OpLine %3 174 14 +OpLine %3 176 15 +%504 = OpSampledImage %317 %497 %500 +%505 = OpImageGather %24 %504 %301 %506 +OpLine %3 177 22 +%507 = OpSampledImage %317 %497 %500 +%508 = OpImageGather %24 %507 %301 %509 ConstOffset %305 +OpLine %3 178 21 +%510 = OpSampledImage %449 %502 %501 +%511 = OpImageDrefGather %24 %510 %301 %299 +OpLine %3 179 28 +%512 = OpSampledImage %449 %502 %501 +%513 = OpImageDrefGather %24 %512 %301 %299 ConstOffset %305 OpLine %3 181 13 -%520 = OpSampledImage %519 %500 %501 -%521 = OpImageGather %518 %520 %303 %221 +%515 = OpSampledImage %514 %498 %500 +%516 = OpImageGather %116 %515 %301 %219 OpLine %3 182 13 -%522 = OpConvertUToF %24 %517 -%523 = OpConvertSToF %24 %521 -%524 = OpFAdd %24 %522 %523 -OpLine %3 184 12 -%525 = OpFAdd %24 %506 %509 -%526 = OpFAdd %24 %525 %512 -%527 = OpFAdd %24 %526 %514 -%528 = OpFAdd %24 %527 %524 -OpStore %496 %528 +%519 = OpSampledImage %518 %499 %500 +%520 = OpImageGather %517 %519 %301 %219 +OpLine %3 183 13 +%521 = OpConvertUToF %24 %516 +%522 = OpConvertSToF %24 %520 +%523 = OpFAdd %24 %521 %522 +OpLine %3 185 12 +%524 = OpFAdd %24 %505 %508 +%525 = OpFAdd %24 %524 %511 +%526 = OpFAdd %24 %525 %513 +%527 = OpFAdd %24 %526 %523 +OpStore %495 %527 OpReturn OpFunctionEnd -%531 = OpFunction %2 None %99 -%529 = OpLabel -%532 = OpLoad %25 %64 -%533 = OpLoad %26 %67 -OpBranch %534 -%534 = OpLabel -OpLine %3 189 14 -OpLine %3 191 15 -%535 = OpSampledImage %450 %533 %532 -%536 = OpImageSampleImplicitLod %24 %535 %303 -%537 = OpCompositeExtract %8 %536 0 -OpLine %3 192 22 -%538 = OpSampledImage %450 %533 %532 -%539 = OpImageGather %24 %538 %303 %221 -OpLine %3 193 21 -%540 = OpSampledImage %450 %533 %532 -%542 = OpConvertSToF %8 %30 -%541 = OpImageSampleExplicitLod %24 %540 %303 Lod %542 -%543 = OpCompositeExtract %8 %541 0 -OpLine %3 191 15 -%544 = OpCompositeConstruct %24 %537 %537 %537 %537 -%545 = OpFAdd %24 %544 %539 -%546 = OpCompositeConstruct %24 %543 %543 %543 %543 -%547 = OpFAdd %24 %545 %546 -OpStore %530 %547 +%530 = OpFunction %2 None %97 +%528 = OpLabel +%531 = OpLoad %25 %61 +%532 = OpLoad %26 %64 +OpBranch %533 +%533 = OpLabel +OpLine %3 190 14 +OpLine %3 192 15 +%534 = OpSampledImage %449 %532 %531 +%535 = OpImageSampleImplicitLod %24 %534 %301 +%536 = OpCompositeExtract %8 %535 0 +OpLine %3 193 22 +%537 = OpSampledImage %449 %532 %531 +%538 = OpImageGather %24 %537 %301 %219 +OpLine %3 194 21 +%539 = OpSampledImage %449 %532 %531 +%541 = OpConvertSToF %8 %88 +%540 = OpImageSampleExplicitLod %24 %539 %301 Lod %541 +%542 = OpCompositeExtract %8 %540 0 +OpLine %3 192 15 +%543 = OpCompositeConstruct %24 %536 %536 %536 %536 +%544 = OpFAdd %24 %543 %538 +%545 = OpCompositeConstruct %24 %542 %542 %542 %542 +%546 = OpFAdd %24 %544 %545 +OpStore %529 %546 OpReturn OpFunctionEnd \ No newline at end of file diff --git a/naga/tests/out/wgsl/image.wgsl b/naga/tests/out/wgsl/image.wgsl index 6789ebc3e..c3ac10f54 100644 --- a/naga/tests/out/wgsl/image.wgsl +++ b/naga/tests/out/wgsl/image.wgsl @@ -114,74 +114,75 @@ fn texture_sample() -> @location(0) vec4<f32> { const tc = vec2(0.5f); const tc3_ = vec3(0.5f); - let _e9 = textureSample(image_1d, sampler_reg, tc.x); - let _e10 = a; - a = (_e10 + _e9); - let _e14 = textureSample(image_2d, sampler_reg, tc); - let _e15 = a; - a = (_e15 + _e14); - let _e19 = textureSample(image_2d, sampler_reg, tc, vec2<i32>(3i, 1i)); - let _e20 = a; - a = (_e20 + _e19); - let _e24 = textureSampleLevel(image_2d, sampler_reg, tc, 2.3f); + const offset = vec2<i32>(3i, 1i); + let _e11 = textureSample(image_1d, sampler_reg, 0.5f); + let _e12 = a; + a = (_e12 + _e11); + let _e16 = textureSample(image_2d, sampler_reg, tc); + let _e17 = a; + a = (_e17 + _e16); + let _e24 = textureSample(image_2d, sampler_reg, tc, vec2<i32>(3i, 1i)); let _e25 = a; a = (_e25 + _e24); - let _e29 = textureSampleLevel(image_2d, sampler_reg, tc, 2.3f, vec2<i32>(3i, 1i)); + let _e29 = textureSampleLevel(image_2d, sampler_reg, tc, 2.3f); let _e30 = a; a = (_e30 + _e29); - let _e35 = textureSampleBias(image_2d, sampler_reg, tc, 2f, vec2<i32>(3i, 1i)); - let _e36 = a; - a = (_e36 + _e35); - let _e41 = textureSample(image_2d_array, sampler_reg, tc, 0u); - let _e42 = a; - a = (_e42 + _e41); - let _e47 = textureSample(image_2d_array, sampler_reg, tc, 0u, vec2<i32>(3i, 1i)); - let _e48 = a; - a = (_e48 + _e47); - let _e53 = textureSampleLevel(image_2d_array, sampler_reg, tc, 0u, 2.3f); - let _e54 = a; - a = (_e54 + _e53); - let _e59 = textureSampleLevel(image_2d_array, sampler_reg, tc, 0u, 2.3f, vec2<i32>(3i, 1i)); - let _e60 = a; - a = (_e60 + _e59); - let _e66 = textureSampleBias(image_2d_array, sampler_reg, tc, 0u, 2f, vec2<i32>(3i, 1i)); - let _e67 = a; - a = (_e67 + _e66); - let _e72 = textureSample(image_2d_array, sampler_reg, tc, 0i); - let _e73 = a; - a = (_e73 + _e72); - let _e78 = textureSample(image_2d_array, sampler_reg, tc, 0i, vec2<i32>(3i, 1i)); - let _e79 = a; - a = (_e79 + _e78); - let _e84 = textureSampleLevel(image_2d_array, sampler_reg, tc, 0i, 2.3f); - let _e85 = a; - a = (_e85 + _e84); - let _e90 = textureSampleLevel(image_2d_array, sampler_reg, tc, 0i, 2.3f, vec2<i32>(3i, 1i)); - let _e91 = a; - a = (_e91 + _e90); - let _e97 = textureSampleBias(image_2d_array, sampler_reg, tc, 0i, 2f, vec2<i32>(3i, 1i)); - let _e98 = a; - a = (_e98 + _e97); - let _e103 = textureSample(image_cube_array, sampler_reg, tc3_, 0u); - let _e104 = a; - a = (_e104 + _e103); - let _e109 = textureSampleLevel(image_cube_array, sampler_reg, tc3_, 0u, 2.3f); - let _e110 = a; - a = (_e110 + _e109); - let _e116 = textureSampleBias(image_cube_array, sampler_reg, tc3_, 0u, 2f); - let _e117 = a; - a = (_e117 + _e116); - let _e122 = textureSample(image_cube_array, sampler_reg, tc3_, 0i); - let _e123 = a; - a = (_e123 + _e122); - let _e128 = textureSampleLevel(image_cube_array, sampler_reg, tc3_, 0i, 2.3f); - let _e129 = a; - a = (_e129 + _e128); - let _e135 = textureSampleBias(image_cube_array, sampler_reg, tc3_, 0i, 2f); - let _e136 = a; - a = (_e136 + _e135); - let _e138 = a; - return _e138; + let _e34 = textureSampleLevel(image_2d, sampler_reg, tc, 2.3f, vec2<i32>(3i, 1i)); + let _e35 = a; + a = (_e35 + _e34); + let _e40 = textureSampleBias(image_2d, sampler_reg, tc, 2f, vec2<i32>(3i, 1i)); + let _e41 = a; + a = (_e41 + _e40); + let _e46 = textureSample(image_2d_array, sampler_reg, tc, 0u); + let _e47 = a; + a = (_e47 + _e46); + let _e52 = textureSample(image_2d_array, sampler_reg, tc, 0u, vec2<i32>(3i, 1i)); + let _e53 = a; + a = (_e53 + _e52); + let _e58 = textureSampleLevel(image_2d_array, sampler_reg, tc, 0u, 2.3f); + let _e59 = a; + a = (_e59 + _e58); + let _e64 = textureSampleLevel(image_2d_array, sampler_reg, tc, 0u, 2.3f, vec2<i32>(3i, 1i)); + let _e65 = a; + a = (_e65 + _e64); + let _e71 = textureSampleBias(image_2d_array, sampler_reg, tc, 0u, 2f, vec2<i32>(3i, 1i)); + let _e72 = a; + a = (_e72 + _e71); + let _e77 = textureSample(image_2d_array, sampler_reg, tc, 0i); + let _e78 = a; + a = (_e78 + _e77); + let _e83 = textureSample(image_2d_array, sampler_reg, tc, 0i, vec2<i32>(3i, 1i)); + let _e84 = a; + a = (_e84 + _e83); + let _e89 = textureSampleLevel(image_2d_array, sampler_reg, tc, 0i, 2.3f); + let _e90 = a; + a = (_e90 + _e89); + let _e95 = textureSampleLevel(image_2d_array, sampler_reg, tc, 0i, 2.3f, vec2<i32>(3i, 1i)); + let _e96 = a; + a = (_e96 + _e95); + let _e102 = textureSampleBias(image_2d_array, sampler_reg, tc, 0i, 2f, vec2<i32>(3i, 1i)); + let _e103 = a; + a = (_e103 + _e102); + let _e108 = textureSample(image_cube_array, sampler_reg, tc3_, 0u); + let _e109 = a; + a = (_e109 + _e108); + let _e114 = textureSampleLevel(image_cube_array, sampler_reg, tc3_, 0u, 2.3f); + let _e115 = a; + a = (_e115 + _e114); + let _e121 = textureSampleBias(image_cube_array, sampler_reg, tc3_, 0u, 2f); + let _e122 = a; + a = (_e122 + _e121); + let _e127 = textureSample(image_cube_array, sampler_reg, tc3_, 0i); + let _e128 = a; + a = (_e128 + _e127); + let _e133 = textureSampleLevel(image_cube_array, sampler_reg, tc3_, 0i, 2.3f); + let _e134 = a; + a = (_e134 + _e133); + let _e140 = textureSampleBias(image_cube_array, sampler_reg, tc3_, 0i, 2f); + let _e141 = a; + a = (_e141 + _e140); + let _e143 = a; + return _e143; } @fragment diff --git a/naga/tests/out/wgsl/samplers.frag.wgsl b/naga/tests/out/wgsl/samplers.frag.wgsl index 24469b593..950c4967c 100644 --- a/naga/tests/out/wgsl/samplers.frag.wgsl +++ b/naga/tests/out/wgsl/samplers.frag.wgsl @@ -53,71 +53,71 @@ fn testTex1D(coord: f32) { let _e17 = textureSampleGrad(tex1D, samp, _e14, 4f, 4f); c = _e17; let _e18 = coord_1; - let _e21 = textureSampleGrad(tex1D, samp, _e18, 4f, 4f, 5i); - c = _e21; - let _e22 = coord_1; - let _e24 = textureSampleLevel(tex1D, samp, _e22, 3f); - c = _e24; - let _e25 = coord_1; - let _e27 = textureSampleLevel(tex1D, samp, _e25, 3f, 5i); - c = _e27; - let _e28 = coord_1; - let _e29 = textureSample(tex1D, samp, _e28, 5i); + let _e22 = textureSampleGrad(tex1D, samp, _e18, 4f, 4f, 5i); + c = _e22; + let _e23 = coord_1; + let _e25 = textureSampleLevel(tex1D, samp, _e23, 3f); + c = _e25; + let _e26 = coord_1; + let _e29 = textureSampleLevel(tex1D, samp, _e26, 3f, 5i); c = _e29; let _e30 = coord_1; - let _e32 = vec2<f32>(_e30, 6f); - let _e36 = textureSample(tex1D, samp, (_e32.x / _e32.y)); - c = _e36; - let _e37 = coord_1; - let _e41 = vec4<f32>(_e37, 0f, 0f, 6f); - let _e47 = textureSample(tex1D, samp, (_e41.xyz / vec3(_e41.w)).x); - c = _e47; - let _e48 = coord_1; - let _e50 = vec2<f32>(_e48, 6f); - let _e56 = textureSampleGrad(tex1D, samp, (_e50.x / _e50.y), 4f, 4f); - c = _e56; - let _e57 = coord_1; - let _e61 = vec4<f32>(_e57, 0f, 0f, 6f); - let _e69 = textureSampleGrad(tex1D, samp, (_e61.xyz / vec3(_e61.w)).x, 4f, 4f); - c = _e69; - let _e70 = coord_1; - let _e72 = vec2<f32>(_e70, 6f); - let _e78 = textureSampleGrad(tex1D, samp, (_e72.x / _e72.y), 4f, 4f, 5i); - c = _e78; - let _e79 = coord_1; - let _e83 = vec4<f32>(_e79, 0f, 0f, 6f); - let _e91 = textureSampleGrad(tex1D, samp, (_e83.xyz / vec3(_e83.w)).x, 4f, 4f, 5i); - c = _e91; - let _e92 = coord_1; - let _e94 = vec2<f32>(_e92, 6f); - let _e99 = textureSampleLevel(tex1D, samp, (_e94.x / _e94.y), 3f); - c = _e99; - let _e100 = coord_1; - let _e104 = vec4<f32>(_e100, 0f, 0f, 6f); - let _e111 = textureSampleLevel(tex1D, samp, (_e104.xyz / vec3(_e104.w)).x, 3f); - c = _e111; - let _e112 = coord_1; - let _e114 = vec2<f32>(_e112, 6f); - let _e119 = textureSampleLevel(tex1D, samp, (_e114.x / _e114.y), 3f, 5i); - c = _e119; - let _e120 = coord_1; - let _e124 = vec4<f32>(_e120, 0f, 0f, 6f); - let _e131 = textureSampleLevel(tex1D, samp, (_e124.xyz / vec3(_e124.w)).x, 3f, 5i); - c = _e131; - let _e132 = coord_1; - let _e134 = vec2<f32>(_e132, 6f); - let _e138 = textureSample(tex1D, samp, (_e134.x / _e134.y), 5i); + let _e32 = textureSample(tex1D, samp, _e30, 5i); + c = _e32; + let _e33 = coord_1; + let _e35 = vec2<f32>(_e33, 6f); + let _e39 = textureSample(tex1D, samp, (_e35.x / _e35.y)); + c = _e39; + let _e40 = coord_1; + let _e44 = vec4<f32>(_e40, 0f, 0f, 6f); + let _e50 = textureSample(tex1D, samp, (_e44.xyz / vec3(_e44.w)).x); + c = _e50; + let _e51 = coord_1; + let _e53 = vec2<f32>(_e51, 6f); + let _e59 = textureSampleGrad(tex1D, samp, (_e53.x / _e53.y), 4f, 4f); + c = _e59; + let _e60 = coord_1; + let _e64 = vec4<f32>(_e60, 0f, 0f, 6f); + let _e72 = textureSampleGrad(tex1D, samp, (_e64.xyz / vec3(_e64.w)).x, 4f, 4f); + c = _e72; + let _e73 = coord_1; + let _e75 = vec2<f32>(_e73, 6f); + let _e82 = textureSampleGrad(tex1D, samp, (_e75.x / _e75.y), 4f, 4f, 5i); + c = _e82; + let _e83 = coord_1; + let _e87 = vec4<f32>(_e83, 0f, 0f, 6f); + let _e96 = textureSampleGrad(tex1D, samp, (_e87.xyz / vec3(_e87.w)).x, 4f, 4f, 5i); + c = _e96; + let _e97 = coord_1; + let _e99 = vec2<f32>(_e97, 6f); + let _e104 = textureSampleLevel(tex1D, samp, (_e99.x / _e99.y), 3f); + c = _e104; + let _e105 = coord_1; + let _e109 = vec4<f32>(_e105, 0f, 0f, 6f); + let _e116 = textureSampleLevel(tex1D, samp, (_e109.xyz / vec3(_e109.w)).x, 3f); + c = _e116; + let _e117 = coord_1; + let _e119 = vec2<f32>(_e117, 6f); + let _e125 = textureSampleLevel(tex1D, samp, (_e119.x / _e119.y), 3f, 5i); + c = _e125; + let _e126 = coord_1; + let _e130 = vec4<f32>(_e126, 0f, 0f, 6f); + let _e138 = textureSampleLevel(tex1D, samp, (_e130.xyz / vec3(_e130.w)).x, 3f, 5i); c = _e138; let _e139 = coord_1; - let _e143 = vec4<f32>(_e139, 0f, 0f, 6f); - let _e149 = textureSample(tex1D, samp, (_e143.xyz / vec3(_e143.w)).x, 5i); - c = _e149; - let _e150 = coord_1; - let _e153 = textureLoad(tex1D, i32(_e150), 3i); - c = _e153; - let _e154 = coord_1; - let _e157 = textureLoad(tex1D, i32(_e154), 3i); - c = _e157; + let _e141 = vec2<f32>(_e139, 6f); + let _e146 = textureSample(tex1D, samp, (_e141.x / _e141.y), 5i); + c = _e146; + let _e147 = coord_1; + let _e151 = vec4<f32>(_e147, 0f, 0f, 6f); + let _e158 = textureSample(tex1D, samp, (_e151.xyz / vec3(_e151.w)).x, 5i); + c = _e158; + let _e159 = coord_1; + let _e162 = textureLoad(tex1D, i32(_e159), 3i); + c = _e162; + let _e163 = coord_1; + let _e166 = textureLoad(tex1D, i32(_e163), 3i); + c = _e166; return; } @@ -140,25 +140,25 @@ fn testTex1DArray(coord_2: vec2<f32>) { let _e25 = textureSampleGrad(tex1DArray, samp, _e19.x, i32(_e19.y), 4f, 4f); c_1 = _e25; let _e26 = coord_3; - let _e32 = textureSampleGrad(tex1DArray, samp, _e26.x, i32(_e26.y), 4f, 4f, 5i); - c_1 = _e32; - let _e33 = coord_3; - let _e38 = textureSampleLevel(tex1DArray, samp, _e33.x, i32(_e33.y), 3f); - c_1 = _e38; - let _e39 = coord_3; - let _e44 = textureSampleLevel(tex1DArray, samp, _e39.x, i32(_e39.y), 3f, 5i); - c_1 = _e44; - let _e45 = coord_3; - let _e49 = textureSample(tex1DArray, samp, _e45.x, i32(_e45.y), 5i); - c_1 = _e49; - let _e50 = coord_3; - let _e51 = vec2<i32>(_e50); - let _e55 = textureLoad(tex1DArray, _e51.x, _e51.y, 3i); - c_1 = _e55; - let _e56 = coord_3; - let _e57 = vec2<i32>(_e56); - let _e61 = textureLoad(tex1DArray, _e57.x, _e57.y, 3i); - c_1 = _e61; + let _e33 = textureSampleGrad(tex1DArray, samp, _e26.x, i32(_e26.y), 4f, 4f, 5i); + c_1 = _e33; + let _e34 = coord_3; + let _e39 = textureSampleLevel(tex1DArray, samp, _e34.x, i32(_e34.y), 3f); + c_1 = _e39; + let _e40 = coord_3; + let _e46 = textureSampleLevel(tex1DArray, samp, _e40.x, i32(_e40.y), 3f, 5i); + c_1 = _e46; + let _e47 = coord_3; + let _e52 = textureSample(tex1DArray, samp, _e47.x, i32(_e47.y), 5i); + c_1 = _e52; + let _e53 = coord_3; + let _e54 = vec2<i32>(_e53); + let _e58 = textureLoad(tex1DArray, _e54.x, _e54.y, 3i); + c_1 = _e58; + let _e59 = coord_3; + let _e60 = vec2<i32>(_e59); + let _e64 = textureLoad(tex1DArray, _e60.x, _e60.y, 3i); + c_1 = _e64; return; } @@ -183,102 +183,102 @@ fn testTex2D(coord_4: vec2<f32>) { let _e24 = textureSampleGrad(tex2D, samp, _e19, vec2(4f), vec2(4f)); c_2 = _e24; let _e25 = coord_5; - let _e30 = textureSampleGrad(tex2D, samp, _e25, vec2(4f), vec2(4f), vec2(5i)); - c_2 = _e30; - let _e31 = coord_5; - let _e33 = textureSampleLevel(tex2D, samp, _e31, 3f); - c_2 = _e33; - let _e34 = coord_5; - let _e36 = textureSampleLevel(tex2D, samp, _e34, 3f, vec2(5i)); - c_2 = _e36; - let _e37 = coord_5; - let _e38 = textureSample(tex2D, samp, _e37, vec2(5i)); - c_2 = _e38; - let _e39 = coord_5; - let _e41 = textureSampleBias(tex2D, samp, _e39, 2f, vec2(5i)); - c_2 = _e41; - let _e42 = coord_5; - let _e46 = vec3<f32>(_e42.x, _e42.y, 6f); - let _e51 = textureSample(tex2D, samp, (_e46.xy / vec2(_e46.z))); - c_2 = _e51; - let _e52 = coord_5; - let _e57 = vec4<f32>(_e52.x, _e52.y, 0f, 6f); - let _e63 = textureSample(tex2D, samp, (_e57.xyz / vec3(_e57.w)).xy); - c_2 = _e63; - let _e64 = coord_5; - let _e68 = vec3<f32>(_e64.x, _e64.y, 6f); - let _e74 = textureSampleBias(tex2D, samp, (_e68.xy / vec2(_e68.z)), 2f); - c_2 = _e74; - let _e75 = coord_5; - let _e80 = vec4<f32>(_e75.x, _e75.y, 0f, 6f); - let _e87 = textureSampleBias(tex2D, samp, (_e80.xyz / vec3(_e80.w)).xy, 2f); - c_2 = _e87; - let _e88 = coord_5; - let _e92 = vec3<f32>(_e88.x, _e88.y, 6f); - let _e101 = textureSampleGrad(tex2D, samp, (_e92.xy / vec2(_e92.z)), vec2(4f), vec2(4f)); - c_2 = _e101; - let _e102 = coord_5; - let _e107 = vec4<f32>(_e102.x, _e102.y, 0f, 6f); - let _e117 = textureSampleGrad(tex2D, samp, (_e107.xyz / vec3(_e107.w)).xy, vec2(4f), vec2(4f)); - c_2 = _e117; - let _e118 = coord_5; - let _e122 = vec3<f32>(_e118.x, _e118.y, 6f); - let _e131 = textureSampleGrad(tex2D, samp, (_e122.xy / vec2(_e122.z)), vec2(4f), vec2(4f), vec2(5i)); - c_2 = _e131; - let _e132 = coord_5; - let _e137 = vec4<f32>(_e132.x, _e132.y, 0f, 6f); - let _e147 = textureSampleGrad(tex2D, samp, (_e137.xyz / vec3(_e137.w)).xy, vec2(4f), vec2(4f), vec2(5i)); - c_2 = _e147; - let _e148 = coord_5; - let _e152 = vec3<f32>(_e148.x, _e148.y, 6f); - let _e158 = textureSampleLevel(tex2D, samp, (_e152.xy / vec2(_e152.z)), 3f); - c_2 = _e158; - let _e159 = coord_5; - let _e164 = vec4<f32>(_e159.x, _e159.y, 0f, 6f); - let _e171 = textureSampleLevel(tex2D, samp, (_e164.xyz / vec3(_e164.w)).xy, 3f); - c_2 = _e171; - let _e172 = coord_5; - let _e176 = vec3<f32>(_e172.x, _e172.y, 6f); - let _e182 = textureSampleLevel(tex2D, samp, (_e176.xy / vec2(_e176.z)), 3f, vec2(5i)); - c_2 = _e182; - let _e183 = coord_5; - let _e188 = vec4<f32>(_e183.x, _e183.y, 0f, 6f); - let _e195 = textureSampleLevel(tex2D, samp, (_e188.xyz / vec3(_e188.w)).xy, 3f, vec2(5i)); - c_2 = _e195; - let _e196 = coord_5; - let _e200 = vec3<f32>(_e196.x, _e196.y, 6f); - let _e205 = textureSample(tex2D, samp, (_e200.xy / vec2(_e200.z)), vec2(5i)); - c_2 = _e205; - let _e206 = coord_5; - let _e211 = vec4<f32>(_e206.x, _e206.y, 0f, 6f); - let _e217 = textureSample(tex2D, samp, (_e211.xyz / vec3(_e211.w)).xy, vec2(5i)); - c_2 = _e217; - let _e218 = coord_5; - let _e222 = vec3<f32>(_e218.x, _e218.y, 6f); - let _e228 = textureSampleBias(tex2D, samp, (_e222.xy / vec2(_e222.z)), 2f, vec2(5i)); - c_2 = _e228; - let _e229 = coord_5; - let _e234 = vec4<f32>(_e229.x, _e229.y, 0f, 6f); - let _e241 = textureSampleBias(tex2D, samp, (_e234.xyz / vec3(_e234.w)).xy, 2f, vec2(5i)); - c_2 = _e241; - let _e242 = coord_5; - let _e245 = textureLoad(tex2D, vec2<i32>(_e242), 3i); - c_2 = _e245; - let _e246 = coord_5; - let _e249 = textureLoad(utex2D, vec2<i32>(_e246), 3i); - c_2 = vec4<f32>(_e249); + let _e32 = textureSampleGrad(tex2D, samp, _e25, vec2(4f), vec2(4f), vec2(5i)); + c_2 = _e32; + let _e33 = coord_5; + let _e35 = textureSampleLevel(tex2D, samp, _e33, 3f); + c_2 = _e35; + let _e36 = coord_5; + let _e40 = textureSampleLevel(tex2D, samp, _e36, 3f, vec2(5i)); + c_2 = _e40; + let _e41 = coord_5; + let _e44 = textureSample(tex2D, samp, _e41, vec2(5i)); + c_2 = _e44; + let _e45 = coord_5; + let _e49 = textureSampleBias(tex2D, samp, _e45, 2f, vec2(5i)); + c_2 = _e49; + let _e50 = coord_5; + let _e54 = vec3<f32>(_e50.x, _e50.y, 6f); + let _e59 = textureSample(tex2D, samp, (_e54.xy / vec2(_e54.z))); + c_2 = _e59; + let _e60 = coord_5; + let _e65 = vec4<f32>(_e60.x, _e60.y, 0f, 6f); + let _e71 = textureSample(tex2D, samp, (_e65.xyz / vec3(_e65.w)).xy); + c_2 = _e71; + let _e72 = coord_5; + let _e76 = vec3<f32>(_e72.x, _e72.y, 6f); + let _e82 = textureSampleBias(tex2D, samp, (_e76.xy / vec2(_e76.z)), 2f); + c_2 = _e82; + let _e83 = coord_5; + let _e88 = vec4<f32>(_e83.x, _e83.y, 0f, 6f); + let _e95 = textureSampleBias(tex2D, samp, (_e88.xyz / vec3(_e88.w)).xy, 2f); + c_2 = _e95; + let _e96 = coord_5; + let _e100 = vec3<f32>(_e96.x, _e96.y, 6f); + let _e109 = textureSampleGrad(tex2D, samp, (_e100.xy / vec2(_e100.z)), vec2(4f), vec2(4f)); + c_2 = _e109; + let _e110 = coord_5; + let _e115 = vec4<f32>(_e110.x, _e110.y, 0f, 6f); + let _e125 = textureSampleGrad(tex2D, samp, (_e115.xyz / vec3(_e115.w)).xy, vec2(4f), vec2(4f)); + c_2 = _e125; + let _e126 = coord_5; + let _e130 = vec3<f32>(_e126.x, _e126.y, 6f); + let _e141 = textureSampleGrad(tex2D, samp, (_e130.xy / vec2(_e130.z)), vec2(4f), vec2(4f), vec2(5i)); + c_2 = _e141; + let _e142 = coord_5; + let _e147 = vec4<f32>(_e142.x, _e142.y, 0f, 6f); + let _e159 = textureSampleGrad(tex2D, samp, (_e147.xyz / vec3(_e147.w)).xy, vec2(4f), vec2(4f), vec2(5i)); + c_2 = _e159; + let _e160 = coord_5; + let _e164 = vec3<f32>(_e160.x, _e160.y, 6f); + let _e170 = textureSampleLevel(tex2D, samp, (_e164.xy / vec2(_e164.z)), 3f); + c_2 = _e170; + let _e171 = coord_5; + let _e176 = vec4<f32>(_e171.x, _e171.y, 0f, 6f); + let _e183 = textureSampleLevel(tex2D, samp, (_e176.xyz / vec3(_e176.w)).xy, 3f); + c_2 = _e183; + let _e184 = coord_5; + let _e188 = vec3<f32>(_e184.x, _e184.y, 6f); + let _e196 = textureSampleLevel(tex2D, samp, (_e188.xy / vec2(_e188.z)), 3f, vec2(5i)); + c_2 = _e196; + let _e197 = coord_5; + let _e202 = vec4<f32>(_e197.x, _e197.y, 0f, 6f); + let _e211 = textureSampleLevel(tex2D, samp, (_e202.xyz / vec3(_e202.w)).xy, 3f, vec2(5i)); + c_2 = _e211; + let _e212 = coord_5; + let _e216 = vec3<f32>(_e212.x, _e212.y, 6f); + let _e223 = textureSample(tex2D, samp, (_e216.xy / vec2(_e216.z)), vec2(5i)); + c_2 = _e223; + let _e224 = coord_5; + let _e229 = vec4<f32>(_e224.x, _e224.y, 0f, 6f); + let _e237 = textureSample(tex2D, samp, (_e229.xyz / vec3(_e229.w)).xy, vec2(5i)); + c_2 = _e237; + let _e238 = coord_5; + let _e242 = vec3<f32>(_e238.x, _e238.y, 6f); + let _e250 = textureSampleBias(tex2D, samp, (_e242.xy / vec2(_e242.z)), 2f, vec2(5i)); + c_2 = _e250; let _e251 = coord_5; - let _e254 = textureLoad(itex2D, vec2<i32>(_e251), 3i); - c_2 = vec4<f32>(_e254); - let _e256 = coord_5; - let _e259 = textureLoad(tex2D, vec2<i32>(_e256), 3i); - c_2 = _e259; - let _e260 = coord_5; - let _e263 = textureLoad(utex2D, vec2<i32>(_e260), 3i); - c_2 = vec4<f32>(_e263); - let _e265 = coord_5; - let _e268 = textureLoad(itex2D, vec2<i32>(_e265), 3i); - c_2 = vec4<f32>(_e268); + let _e256 = vec4<f32>(_e251.x, _e251.y, 0f, 6f); + let _e265 = textureSampleBias(tex2D, samp, (_e256.xyz / vec3(_e256.w)).xy, 2f, vec2(5i)); + c_2 = _e265; + let _e266 = coord_5; + let _e269 = textureLoad(tex2D, vec2<i32>(_e266), 3i); + c_2 = _e269; + let _e270 = coord_5; + let _e273 = textureLoad(utex2D, vec2<i32>(_e270), 3i); + c_2 = vec4<f32>(_e273); + let _e275 = coord_5; + let _e278 = textureLoad(itex2D, vec2<i32>(_e275), 3i); + c_2 = vec4<f32>(_e278); + let _e280 = coord_5; + let _e283 = textureLoad(tex2D, vec2<i32>(_e280), 3i); + c_2 = _e283; + let _e284 = coord_5; + let _e287 = textureLoad(utex2D, vec2<i32>(_e284), 3i); + c_2 = vec4<f32>(_e287); + let _e289 = coord_5; + let _e292 = textureLoad(itex2D, vec2<i32>(_e289), 3i); + c_2 = vec4<f32>(_e292); return; } @@ -303,50 +303,50 @@ fn testTex2DShadow(coord_6: vec2<f32>) { d = _e27; let _e28 = coord_7; let _e32 = vec3<f32>(_e28.x, _e28.y, 1f); - let _e35 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e32.xy, _e32.z, vec2(5i)); - d = _e35; - let _e36 = coord_7; - let _e40 = vec3<f32>(_e36.x, _e36.y, 1f); - let _e43 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e40.xy, _e40.z); - d = _e43; - let _e44 = coord_7; - let _e48 = vec3<f32>(_e44.x, _e44.y, 1f); - let _e51 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e48.xy, _e48.z, vec2(5i)); - d = _e51; - let _e52 = coord_7; - let _e56 = vec3<f32>(_e52.x, _e52.y, 1f); - let _e59 = textureSampleCompare(tex2DShadow, sampShadow, _e56.xy, _e56.z, vec2(5i)); - d = _e59; - let _e60 = coord_7; - let _e65 = vec4<f32>(_e60.x, _e60.y, 1f, 6f); - let _e69 = (_e65.xyz / vec3(_e65.w)); - let _e72 = textureSampleCompare(tex2DShadow, sampShadow, _e69.xy, _e69.z); - d = _e72; - let _e73 = coord_7; - let _e78 = vec4<f32>(_e73.x, _e73.y, 1f, 6f); - let _e82 = (_e78.xyz / vec3(_e78.w)); - let _e85 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e82.xy, _e82.z); - d = _e85; - let _e86 = coord_7; - let _e91 = vec4<f32>(_e86.x, _e86.y, 1f, 6f); - let _e95 = (_e91.xyz / vec3(_e91.w)); - let _e98 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e95.xy, _e95.z, vec2(5i)); - d = _e98; - let _e99 = coord_7; - let _e104 = vec4<f32>(_e99.x, _e99.y, 1f, 6f); - let _e108 = (_e104.xyz / vec3(_e104.w)); - let _e111 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e108.xy, _e108.z); - d = _e111; - let _e112 = coord_7; - let _e117 = vec4<f32>(_e112.x, _e112.y, 1f, 6f); - let _e121 = (_e117.xyz / vec3(_e117.w)); - let _e124 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e121.xy, _e121.z, vec2(5i)); - d = _e124; - let _e125 = coord_7; - let _e130 = vec4<f32>(_e125.x, _e125.y, 1f, 6f); - let _e134 = (_e130.xyz / vec3(_e130.w)); - let _e137 = textureSampleCompare(tex2DShadow, sampShadow, _e134.xy, _e134.z, vec2(5i)); - d = _e137; + let _e37 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e32.xy, _e32.z, vec2(5i)); + d = _e37; + let _e38 = coord_7; + let _e42 = vec3<f32>(_e38.x, _e38.y, 1f); + let _e45 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e42.xy, _e42.z); + d = _e45; + let _e46 = coord_7; + let _e50 = vec3<f32>(_e46.x, _e46.y, 1f); + let _e55 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e50.xy, _e50.z, vec2(5i)); + d = _e55; + let _e56 = coord_7; + let _e60 = vec3<f32>(_e56.x, _e56.y, 1f); + let _e65 = textureSampleCompare(tex2DShadow, sampShadow, _e60.xy, _e60.z, vec2(5i)); + d = _e65; + let _e66 = coord_7; + let _e71 = vec4<f32>(_e66.x, _e66.y, 1f, 6f); + let _e75 = (_e71.xyz / vec3(_e71.w)); + let _e78 = textureSampleCompare(tex2DShadow, sampShadow, _e75.xy, _e75.z); + d = _e78; + let _e79 = coord_7; + let _e84 = vec4<f32>(_e79.x, _e79.y, 1f, 6f); + let _e88 = (_e84.xyz / vec3(_e84.w)); + let _e91 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e88.xy, _e88.z); + d = _e91; + let _e92 = coord_7; + let _e97 = vec4<f32>(_e92.x, _e92.y, 1f, 6f); + let _e103 = (_e97.xyz / vec3(_e97.w)); + let _e106 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e103.xy, _e103.z, vec2(5i)); + d = _e106; + let _e107 = coord_7; + let _e112 = vec4<f32>(_e107.x, _e107.y, 1f, 6f); + let _e116 = (_e112.xyz / vec3(_e112.w)); + let _e119 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e116.xy, _e116.z); + d = _e119; + let _e120 = coord_7; + let _e125 = vec4<f32>(_e120.x, _e120.y, 1f, 6f); + let _e131 = (_e125.xyz / vec3(_e125.w)); + let _e134 = textureSampleCompareLevel(tex2DShadow, sampShadow, _e131.xy, _e131.z, vec2(5i)); + d = _e134; + let _e135 = coord_7; + let _e140 = vec4<f32>(_e135.x, _e135.y, 1f, 6f); + let _e146 = (_e140.xyz / vec3(_e140.w)); + let _e149 = textureSampleCompare(tex2DShadow, sampShadow, _e146.xy, _e146.z, vec2(5i)); + d = _e149; return; } @@ -372,28 +372,28 @@ fn testTex2DArray(coord_8: vec3<f32>) { let _e35 = textureSampleGrad(tex2DArray, samp, _e27.xy, i32(_e27.z), vec2(4f), vec2(4f)); c_3 = _e35; let _e36 = coord_9; - let _e44 = textureSampleGrad(tex2DArray, samp, _e36.xy, i32(_e36.z), vec2(4f), vec2(4f), vec2(5i)); - c_3 = _e44; - let _e45 = coord_9; - let _e50 = textureSampleLevel(tex2DArray, samp, _e45.xy, i32(_e45.z), 3f); - c_3 = _e50; - let _e51 = coord_9; - let _e56 = textureSampleLevel(tex2DArray, samp, _e51.xy, i32(_e51.z), 3f, vec2(5i)); - c_3 = _e56; - let _e57 = coord_9; - let _e61 = textureSample(tex2DArray, samp, _e57.xy, i32(_e57.z), vec2(5i)); - c_3 = _e61; - let _e62 = coord_9; - let _e67 = textureSampleBias(tex2DArray, samp, _e62.xy, i32(_e62.z), 2f, vec2(5i)); + let _e46 = textureSampleGrad(tex2DArray, samp, _e36.xy, i32(_e36.z), vec2(4f), vec2(4f), vec2(5i)); + c_3 = _e46; + let _e47 = coord_9; + let _e52 = textureSampleLevel(tex2DArray, samp, _e47.xy, i32(_e47.z), 3f); + c_3 = _e52; + let _e53 = coord_9; + let _e60 = textureSampleLevel(tex2DArray, samp, _e53.xy, i32(_e53.z), 3f, vec2(5i)); + c_3 = _e60; + let _e61 = coord_9; + let _e67 = textureSample(tex2DArray, samp, _e61.xy, i32(_e61.z), vec2(5i)); c_3 = _e67; let _e68 = coord_9; - let _e69 = vec3<i32>(_e68); - let _e73 = textureLoad(tex2DArray, _e69.xy, _e69.z, 3i); - c_3 = _e73; - let _e74 = coord_9; - let _e75 = vec3<i32>(_e74); - let _e79 = textureLoad(tex2DArray, _e75.xy, _e75.z, 3i); - c_3 = _e79; + let _e75 = textureSampleBias(tex2DArray, samp, _e68.xy, i32(_e68.z), 2f, vec2(5i)); + c_3 = _e75; + let _e76 = coord_9; + let _e77 = vec3<i32>(_e76); + let _e81 = textureLoad(tex2DArray, _e77.xy, _e77.z, 3i); + c_3 = _e81; + let _e82 = coord_9; + let _e83 = vec3<i32>(_e82); + let _e87 = textureLoad(tex2DArray, _e83.xy, _e83.z, 3i); + c_3 = _e87; return; } @@ -419,12 +419,12 @@ fn testTex2DArrayShadow(coord_10: vec3<f32>) { d_1 = _e37; let _e38 = coord_11; let _e43 = vec4<f32>(_e38.x, _e38.y, _e38.z, 1f); - let _e48 = textureSampleCompareLevel(tex2DArrayShadow, sampShadow, _e43.xy, i32(_e43.z), _e43.w, vec2(5i)); - d_1 = _e48; - let _e49 = coord_11; - let _e54 = vec4<f32>(_e49.x, _e49.y, _e49.z, 1f); - let _e59 = textureSampleCompare(tex2DArrayShadow, sampShadow, _e54.xy, i32(_e54.z), _e54.w, vec2(5i)); - d_1 = _e59; + let _e50 = textureSampleCompareLevel(tex2DArrayShadow, sampShadow, _e43.xy, i32(_e43.z), _e43.w, vec2(5i)); + d_1 = _e50; + let _e51 = coord_11; + let _e56 = vec4<f32>(_e51.x, _e51.y, _e51.z, 1f); + let _e63 = textureSampleCompare(tex2DArrayShadow, sampShadow, _e56.xy, i32(_e56.z), _e56.w, vec2(5i)); + d_1 = _e63; return; } @@ -548,52 +548,52 @@ fn testTex3D(coord_20: vec3<f32>) { c_6 = _e39; let _e40 = coord_21; let _e45 = vec4<f32>(_e40.x, _e40.y, _e40.z, 6f); - let _e50 = textureSample(tex3D, samp, (_e45.xyz / vec3(_e45.w)), vec3(5i)); - c_6 = _e50; - let _e51 = coord_21; - let _e56 = vec4<f32>(_e51.x, _e51.y, _e51.z, 6f); - let _e62 = textureSampleBias(tex3D, samp, (_e56.xyz / vec3(_e56.w)), 2f, vec3(5i)); - c_6 = _e62; - let _e63 = coord_21; - let _e68 = vec4<f32>(_e63.x, _e63.y, _e63.z, 6f); - let _e74 = textureSampleLevel(tex3D, samp, (_e68.xyz / vec3(_e68.w)), 3f); - c_6 = _e74; - let _e75 = coord_21; - let _e80 = vec4<f32>(_e75.x, _e75.y, _e75.z, 6f); - let _e86 = textureSampleLevel(tex3D, samp, (_e80.xyz / vec3(_e80.w)), 3f, vec3(5i)); - c_6 = _e86; - let _e87 = coord_21; - let _e92 = vec4<f32>(_e87.x, _e87.y, _e87.z, 6f); - let _e101 = textureSampleGrad(tex3D, samp, (_e92.xyz / vec3(_e92.w)), vec3(4f), vec3(4f)); - c_6 = _e101; - let _e102 = coord_21; - let _e107 = vec4<f32>(_e102.x, _e102.y, _e102.z, 6f); - let _e116 = textureSampleGrad(tex3D, samp, (_e107.xyz / vec3(_e107.w)), vec3(4f), vec3(4f), vec3(5i)); - c_6 = _e116; - let _e117 = coord_21; - let _e122 = textureSampleGrad(tex3D, samp, _e117, vec3(4f), vec3(4f)); - c_6 = _e122; - let _e123 = coord_21; - let _e128 = textureSampleGrad(tex3D, samp, _e123, vec3(4f), vec3(4f), vec3(5i)); - c_6 = _e128; - let _e129 = coord_21; - let _e131 = textureSampleLevel(tex3D, samp, _e129, 3f); - c_6 = _e131; - let _e132 = coord_21; - let _e134 = textureSampleLevel(tex3D, samp, _e132, 3f, vec3(5i)); - c_6 = _e134; - let _e135 = coord_21; - let _e136 = textureSample(tex3D, samp, _e135, vec3(5i)); - c_6 = _e136; - let _e137 = coord_21; - let _e139 = textureSampleBias(tex3D, samp, _e137, 2f, vec3(5i)); - c_6 = _e139; - let _e140 = coord_21; - let _e143 = textureLoad(tex3D, vec3<i32>(_e140), 3i); - c_6 = _e143; - let _e144 = coord_21; - let _e147 = textureLoad(tex3D, vec3<i32>(_e144), 3i); - c_6 = _e147; + let _e52 = textureSample(tex3D, samp, (_e45.xyz / vec3(_e45.w)), vec3(5i)); + c_6 = _e52; + let _e53 = coord_21; + let _e58 = vec4<f32>(_e53.x, _e53.y, _e53.z, 6f); + let _e66 = textureSampleBias(tex3D, samp, (_e58.xyz / vec3(_e58.w)), 2f, vec3(5i)); + c_6 = _e66; + let _e67 = coord_21; + let _e72 = vec4<f32>(_e67.x, _e67.y, _e67.z, 6f); + let _e78 = textureSampleLevel(tex3D, samp, (_e72.xyz / vec3(_e72.w)), 3f); + c_6 = _e78; + let _e79 = coord_21; + let _e84 = vec4<f32>(_e79.x, _e79.y, _e79.z, 6f); + let _e92 = textureSampleLevel(tex3D, samp, (_e84.xyz / vec3(_e84.w)), 3f, vec3(5i)); + c_6 = _e92; + let _e93 = coord_21; + let _e98 = vec4<f32>(_e93.x, _e93.y, _e93.z, 6f); + let _e107 = textureSampleGrad(tex3D, samp, (_e98.xyz / vec3(_e98.w)), vec3(4f), vec3(4f)); + c_6 = _e107; + let _e108 = coord_21; + let _e113 = vec4<f32>(_e108.x, _e108.y, _e108.z, 6f); + let _e124 = textureSampleGrad(tex3D, samp, (_e113.xyz / vec3(_e113.w)), vec3(4f), vec3(4f), vec3(5i)); + c_6 = _e124; + let _e125 = coord_21; + let _e130 = textureSampleGrad(tex3D, samp, _e125, vec3(4f), vec3(4f)); + c_6 = _e130; + let _e131 = coord_21; + let _e138 = textureSampleGrad(tex3D, samp, _e131, vec3(4f), vec3(4f), vec3(5i)); + c_6 = _e138; + let _e139 = coord_21; + let _e141 = textureSampleLevel(tex3D, samp, _e139, 3f); + c_6 = _e141; + let _e142 = coord_21; + let _e146 = textureSampleLevel(tex3D, samp, _e142, 3f, vec3(5i)); + c_6 = _e146; + let _e147 = coord_21; + let _e150 = textureSample(tex3D, samp, _e147, vec3(5i)); + c_6 = _e150; + let _e151 = coord_21; + let _e155 = textureSampleBias(tex3D, samp, _e151, 2f, vec3(5i)); + c_6 = _e155; + let _e156 = coord_21; + let _e159 = textureLoad(tex3D, vec3<i32>(_e156), 3i); + c_6 = _e159; + let _e160 = coord_21; + let _e163 = textureLoad(tex3D, vec3<i32>(_e160), 3i); + c_6 = _e163; return; }