diff --git a/naga/tests/in/operators.wgsl b/naga/tests/in/operators.wgsl index 40d1ce8ead..293b84ab11 100644 --- a/naga/tests/in/operators.wgsl +++ b/naga/tests/in/operators.wgsl @@ -21,9 +21,9 @@ fn builtins() -> vec4 { return vec4(vec4(s1) + v_i32_zero) + s2 + m1 + m2 + b1 + b2; } -fn splat() -> vec4 { - let a = (1.0 + vec2(2.0) - 3.0) / 4.0; - let b = vec4(5) % 2; +fn splat(m: f32, n: i32) -> vec4 { + let a = (2.0 + vec2(m) - 4.0) / 8.0; + let b = vec4(n) % 2; return a.xyxy + vec4(b); } @@ -280,9 +280,9 @@ fn assignment() { } @compute @workgroup_size(1) -fn main() { +fn main(@builtin(workgroup_id) id: vec3) { builtins(); - splat(); + splat(f32(id.x), i32(id.y)); bool_cast(v_f32_one.xyz); logical(); diff --git a/naga/tests/out/glsl/operators.main.Compute.glsl b/naga/tests/out/glsl/operators.main.Compute.glsl index 83dd56f8a4..006bce205e 100644 --- a/naga/tests/out/glsl/operators.main.Compute.glsl +++ b/naga/tests/out/glsl/operators.main.Compute.glsl @@ -23,9 +23,9 @@ vec4 builtins() { return (((((vec4((ivec4(s1_) + v_i32_zero)) + s2_) + m1_) + m2_) + vec4(b1_)) + b2_); } -vec4 splat() { - vec2 a_2 = (((vec2(1.0) + vec2(2.0)) - vec2(3.0)) / vec2(4.0)); - ivec4 b = (ivec4(5) % ivec4(2)); +vec4 splat(float m, int n) { + vec2 a_2 = (((vec2(2.0) + vec2(m)) - vec2(4.0)) / vec2(8.0)); + ivec4 b = (ivec4(n) % ivec4(2)); return (a_2.xyxy + vec4(b)); } @@ -247,9 +247,10 @@ void negation_avoids_prefix_decrement() { } void main() { - vec4 _e0 = builtins(); - vec4 _e1 = splat(); - vec3 _e6 = bool_cast(vec3(1.0, 1.0, 1.0)); + uvec3 id = gl_WorkGroupID; + vec4 _e1 = builtins(); + vec4 _e6 = splat(float(id.x), int(id.y)); + vec3 _e11 = bool_cast(vec3(1.0, 1.0, 1.0)); logical(); arithmetic(); bit(); diff --git a/naga/tests/out/hlsl/operators.hlsl b/naga/tests/out/hlsl/operators.hlsl index 6d18d07ed6..58ec5a170d 100644 --- a/naga/tests/out/hlsl/operators.hlsl +++ b/naga/tests/out/hlsl/operators.hlsl @@ -16,10 +16,10 @@ float4 builtins() return (((((float4(((s1_).xxxx + v_i32_zero)) + s2_) + m1_) + m2_) + (b1_).xxxx) + b2_); } -float4 splat() +float4 splat(float m, int n) { - float2 a_2 = ((((1.0).xx + (2.0).xx) - (3.0).xx) / (4.0).xx); - int4 b = ((5).xxxx % (2).xxxx); + float2 a_2 = ((((2.0).xx + (m).xx) - (4.0).xx) / (8.0).xx); + int4 b = ((n).xxxx % (2).xxxx); return (a_2.xyxy + float4(b)); } @@ -251,11 +251,11 @@ void negation_avoids_prefix_decrement() } [numthreads(1, 1, 1)] -void main() +void main(uint3 id : SV_GroupID) { - const float4 _e0 = builtins(); - const float4 _e1 = splat(); - const float3 _e6 = bool_cast(float3(1.0, 1.0, 1.0)); + const float4 _e1 = builtins(); + const float4 _e6 = splat(float(id.x), int(id.y)); + const float3 _e11 = bool_cast(float3(1.0, 1.0, 1.0)); logical(); arithmetic(); bit(); diff --git a/naga/tests/out/msl/operators.msl b/naga/tests/out/msl/operators.msl index 960a87f01c..85fba28c33 100644 --- a/naga/tests/out/msl/operators.msl +++ b/naga/tests/out/msl/operators.msl @@ -23,9 +23,11 @@ metal::float4 builtins( } metal::float4 splat( + float m, + int n ) { - metal::float2 a_2 = ((metal::float2(1.0) + metal::float2(2.0)) - metal::float2(3.0)) / metal::float2(4.0); - metal::int4 b = metal::int4(5) % metal::int4(2); + metal::float2 a_2 = ((metal::float2(2.0) + metal::float2(m)) - metal::float2(4.0)) / metal::float2(8.0); + metal::int4 b = metal::int4(n) % metal::int4(2); return a_2.xyxy + static_cast(b); } @@ -255,11 +257,14 @@ void negation_avoids_prefix_decrement( int p7_ = -(-(-(-(-(1))))); } +struct main_Input { +}; kernel void main_( + metal::uint3 id [[threadgroup_position_in_grid]] ) { - metal::float4 _e0 = builtins(); - metal::float4 _e1 = splat(); - metal::float3 _e6 = bool_cast(metal::float3(1.0, 1.0, 1.0)); + metal::float4 _e1 = builtins(); + metal::float4 _e6 = splat(static_cast(id.x), static_cast(id.y)); + metal::float3 _e11 = bool_cast(metal::float3(1.0, 1.0, 1.0)); logical(); arithmetic(); bit(); diff --git a/naga/tests/out/spv/operators.spvasm b/naga/tests/out/spv/operators.spvasm index 89bc4ccf28..974623bc70 100644 --- a/naga/tests/out/spv/operators.spvasm +++ b/naga/tests/out/spv/operators.spvasm @@ -1,12 +1,13 @@ ; SPIR-V ; Version: 1.1 ; Generator: rspirv -; Bound: 377 +; Bound: 389 OpCapability Shader %1 = OpExtInstImport "GLSL.std.450" OpMemoryModel Logical GLSL450 -OpEntryPoint GLCompute %366 "main" -OpExecutionMode %366 LocalSize 1 1 1 +OpEntryPoint GLCompute %374 "main" %371 +OpExecutionMode %374 LocalSize 1 1 1 +OpDecorate %371 BuiltIn WorkgroupId %2 = OpTypeVoid %4 = OpTypeFloat 32 %3 = OpTypeVector %4 4 @@ -20,418 +21,430 @@ OpExecutionMode %366 LocalSize 1 1 1 %12 = OpTypeMatrix %10 4 %13 = OpTypeMatrix %3 3 %14 = OpTypeVector %6 3 -%15 = OpConstant %4 1.0 -%16 = OpConstantComposite %3 %15 %15 %15 %15 -%17 = OpConstant %4 0.0 +%16 = OpTypeInt 32 0 +%15 = OpTypeVector %16 3 +%17 = OpConstant %4 1.0 %18 = OpConstantComposite %3 %17 %17 %17 %17 -%19 = OpConstant %4 0.5 +%19 = OpConstant %4 0.0 %20 = OpConstantComposite %3 %19 %19 %19 %19 -%21 = OpConstant %6 1 -%22 = OpConstantComposite %5 %21 %21 %21 %21 -%25 = OpTypeFunction %3 -%26 = OpConstantTrue %8 -%27 = OpConstant %6 0 -%28 = OpConstantFalse %8 -%29 = OpConstantComposite %7 %28 %28 %28 %28 -%30 = OpConstant %4 0.1 -%31 = OpConstantComposite %5 %27 %27 %27 %27 -%53 = OpConstant %4 2.0 -%54 = OpConstantComposite %9 %53 %53 -%55 = OpConstantComposite %9 %15 %15 -%56 = OpConstant %4 3.0 -%57 = OpConstantComposite %9 %56 %56 -%58 = OpConstant %4 4.0 +%21 = OpConstant %4 0.5 +%22 = OpConstantComposite %3 %21 %21 %21 %21 +%23 = OpConstant %6 1 +%24 = OpConstantComposite %5 %23 %23 %23 %23 +%27 = OpTypeFunction %3 +%28 = OpConstantTrue %8 +%29 = OpConstant %6 0 +%30 = OpConstantFalse %8 +%31 = OpConstantComposite %7 %30 %30 %30 %30 +%32 = OpConstant %4 0.1 +%33 = OpConstantComposite %5 %29 %29 %29 %29 +%57 = OpTypeFunction %3 %4 %6 +%58 = OpConstant %4 2.0 %59 = OpConstantComposite %9 %58 %58 -%60 = OpConstant %6 5 -%61 = OpConstantComposite %5 %60 %60 %60 %60 -%62 = OpConstant %6 2 -%63 = OpConstantComposite %5 %62 %62 %62 %62 -%74 = OpTypeFunction %9 -%76 = OpTypePointer Function %9 -%88 = OpTypeFunction %10 %10 -%90 = OpTypeVector %8 3 -%91 = OpConstantComposite %10 %17 %17 %17 -%93 = OpConstantComposite %10 %15 %15 %15 -%97 = OpTypeFunction %2 -%98 = OpTypeVector %8 2 -%99 = OpConstantComposite %98 %26 %26 -%100 = OpConstantComposite %90 %26 %26 %26 -%101 = OpConstantComposite %90 %28 %28 %28 -%102 = OpConstantComposite %7 %26 %26 %26 %26 -%103 = OpConstantComposite %7 %28 %28 %28 %28 -%115 = OpTypeInt 32 0 -%116 = OpConstant %115 1 -%117 = OpConstant %115 2 -%118 = OpTypeVector %6 2 -%119 = OpConstantComposite %118 %21 %21 -%120 = OpConstantComposite %118 %62 %62 -%121 = OpTypeVector %115 3 -%122 = OpConstantComposite %121 %117 %117 %117 -%123 = OpConstantComposite %121 %116 %116 %116 -%124 = OpConstantComposite %3 %53 %53 %53 %53 -%125 = OpConstantComposite %3 %15 %15 %15 %15 -%126 = OpTypeVector %115 2 -%127 = OpConstantComposite %126 %117 %117 -%128 = OpConstantComposite %126 %116 %116 -%129 = OpConstantNull %11 -%130 = OpConstantNull %12 -%131 = OpConstantComposite %10 %53 %53 %53 -%132 = OpConstantNull %13 -%296 = OpConstantNull %14 -%298 = OpTypePointer Function %6 -%299 = OpConstantNull %6 -%301 = OpTypePointer Function %14 -%329 = OpTypePointer Function %6 -%367 = OpConstantComposite %10 %15 %15 %15 -%24 = OpFunction %3 None %25 -%23 = OpLabel -OpBranch %32 -%32 = OpLabel -%33 = OpSelect %6 %26 %21 %27 -%35 = OpCompositeConstruct %7 %26 %26 %26 %26 -%34 = OpSelect %3 %35 %16 %18 -%36 = OpSelect %3 %29 %18 %16 -%37 = OpExtInst %3 %1 FMix %18 %16 %20 -%39 = OpCompositeConstruct %3 %30 %30 %30 %30 -%38 = OpExtInst %3 %1 FMix %18 %16 %39 -%40 = OpBitcast %4 %21 -%41 = OpBitcast %3 %22 -%42 = OpCompositeConstruct %5 %33 %33 %33 %33 -%43 = OpIAdd %5 %42 %31 -%44 = OpConvertSToF %3 %43 -%45 = OpFAdd %3 %44 %34 -%46 = OpFAdd %3 %45 %37 -%47 = OpFAdd %3 %46 %38 -%48 = OpCompositeConstruct %3 %40 %40 %40 %40 -%49 = OpFAdd %3 %47 %48 -%50 = OpFAdd %3 %49 %41 -OpReturnValue %50 +%60 = OpConstant %4 4.0 +%61 = OpConstantComposite %9 %60 %60 +%62 = OpConstant %4 8.0 +%63 = OpConstantComposite %9 %62 %62 +%64 = OpConstant %6 2 +%65 = OpConstantComposite %5 %64 %64 %64 %64 +%78 = OpTypeFunction %9 +%79 = OpConstantComposite %9 %17 %17 +%80 = OpConstant %4 3.0 +%81 = OpConstantComposite %9 %80 %80 +%83 = OpTypePointer Function %9 +%95 = OpTypeFunction %10 %10 +%97 = OpTypeVector %8 3 +%98 = OpConstantComposite %10 %19 %19 %19 +%100 = OpConstantComposite %10 %17 %17 %17 +%104 = OpTypeFunction %2 +%105 = OpTypeVector %8 2 +%106 = OpConstantComposite %105 %28 %28 +%107 = OpConstantComposite %97 %28 %28 %28 +%108 = OpConstantComposite %97 %30 %30 %30 +%109 = OpConstantComposite %7 %28 %28 %28 %28 +%110 = OpConstantComposite %7 %30 %30 %30 %30 +%122 = OpConstant %16 1 +%123 = OpConstant %16 2 +%124 = OpTypeVector %6 2 +%125 = OpConstantComposite %124 %23 %23 +%126 = OpConstantComposite %124 %64 %64 +%127 = OpConstantComposite %15 %123 %123 %123 +%128 = OpConstantComposite %15 %122 %122 %122 +%129 = OpConstantComposite %3 %58 %58 %58 %58 +%130 = OpConstantComposite %3 %17 %17 %17 %17 +%131 = OpTypeVector %16 2 +%132 = OpConstantComposite %131 %123 %123 +%133 = OpConstantComposite %131 %122 %122 +%134 = OpConstantNull %11 +%135 = OpConstantNull %12 +%136 = OpConstantComposite %10 %58 %58 %58 +%137 = OpConstantNull %13 +%301 = OpConstantNull %14 +%303 = OpTypePointer Function %6 +%304 = OpConstantNull %6 +%306 = OpTypePointer Function %14 +%334 = OpTypePointer Function %6 +%372 = OpTypePointer Input %15 +%371 = OpVariable %372 Input +%375 = OpConstantComposite %10 %17 %17 %17 +%26 = OpFunction %3 None %27 +%25 = OpLabel +OpBranch %34 +%34 = OpLabel +%35 = OpSelect %6 %28 %23 %29 +%37 = OpCompositeConstruct %7 %28 %28 %28 %28 +%36 = OpSelect %3 %37 %18 %20 +%38 = OpSelect %3 %31 %20 %18 +%39 = OpExtInst %3 %1 FMix %20 %18 %22 +%41 = OpCompositeConstruct %3 %32 %32 %32 %32 +%40 = OpExtInst %3 %1 FMix %20 %18 %41 +%42 = OpBitcast %4 %23 +%43 = OpBitcast %3 %24 +%44 = OpCompositeConstruct %5 %35 %35 %35 %35 +%45 = OpIAdd %5 %44 %33 +%46 = OpConvertSToF %3 %45 +%47 = OpFAdd %3 %46 %36 +%48 = OpFAdd %3 %47 %39 +%49 = OpFAdd %3 %48 %40 +%50 = OpCompositeConstruct %3 %42 %42 %42 %42 +%51 = OpFAdd %3 %49 %50 +%52 = OpFAdd %3 %51 %43 +OpReturnValue %52 OpFunctionEnd -%52 = OpFunction %3 None %25 -%51 = OpLabel -OpBranch %64 -%64 = OpLabel -%65 = OpFAdd %9 %55 %54 -%66 = OpFSub %9 %65 %57 -%67 = OpFDiv %9 %66 %59 -%68 = OpSRem %5 %61 %63 -%69 = OpVectorShuffle %3 %67 %67 0 1 0 1 -%70 = OpConvertSToF %3 %68 -%71 = OpFAdd %3 %69 %70 -OpReturnValue %71 +%56 = OpFunction %3 None %57 +%54 = OpFunctionParameter %4 +%55 = OpFunctionParameter %6 +%53 = OpLabel +OpBranch %66 +%66 = OpLabel +%67 = OpCompositeConstruct %9 %54 %54 +%68 = OpFAdd %9 %59 %67 +%69 = OpFSub %9 %68 %61 +%70 = OpFDiv %9 %69 %63 +%71 = OpCompositeConstruct %5 %55 %55 %55 %55 +%72 = OpSRem %5 %71 %65 +%73 = OpVectorShuffle %3 %70 %70 0 1 0 1 +%74 = OpConvertSToF %3 %72 +%75 = OpFAdd %3 %73 %74 +OpReturnValue %75 OpFunctionEnd -%73 = OpFunction %9 None %74 -%72 = OpLabel -%75 = OpVariable %76 Function %54 -OpBranch %77 -%77 = OpLabel -%78 = OpLoad %9 %75 -%79 = OpFAdd %9 %78 %55 -OpStore %75 %79 -%80 = OpLoad %9 %75 -%81 = OpFSub %9 %80 %57 -OpStore %75 %81 -%82 = OpLoad %9 %75 -%83 = OpFDiv %9 %82 %59 -OpStore %75 %83 -%84 = OpLoad %9 %75 -OpReturnValue %84 +%77 = OpFunction %9 None %78 +%76 = OpLabel +%82 = OpVariable %83 Function %59 +OpBranch %84 +%84 = OpLabel +%85 = OpLoad %9 %82 +%86 = OpFAdd %9 %85 %79 +OpStore %82 %86 +%87 = OpLoad %9 %82 +%88 = OpFSub %9 %87 %81 +OpStore %82 %88 +%89 = OpLoad %9 %82 +%90 = OpFDiv %9 %89 %61 +OpStore %82 %90 +%91 = OpLoad %9 %82 +OpReturnValue %91 OpFunctionEnd -%87 = OpFunction %10 None %88 -%86 = OpFunctionParameter %10 -%85 = OpLabel -OpBranch %89 -%89 = OpLabel -%92 = OpFUnordNotEqual %90 %86 %91 -%94 = OpSelect %10 %92 %93 %91 -OpReturnValue %94 +%94 = OpFunction %10 None %95 +%93 = OpFunctionParameter %10 +%92 = OpLabel +OpBranch %96 +%96 = OpLabel +%99 = OpFUnordNotEqual %97 %93 %98 +%101 = OpSelect %10 %99 %100 %98 +OpReturnValue %101 OpFunctionEnd -%96 = OpFunction %2 None %97 -%95 = OpLabel -OpBranch %104 -%104 = OpLabel -%105 = OpLogicalNot %8 %26 -%106 = OpLogicalNot %98 %99 -%107 = OpLogicalOr %8 %26 %28 -%108 = OpLogicalAnd %8 %26 %28 -%109 = OpLogicalOr %8 %26 %28 -%110 = OpLogicalOr %90 %100 %101 -%111 = OpLogicalAnd %8 %26 %28 -%112 = OpLogicalAnd %7 %102 %103 +%103 = OpFunction %2 None %104 +%102 = OpLabel +OpBranch %111 +%111 = OpLabel +%112 = OpLogicalNot %8 %28 +%113 = OpLogicalNot %105 %106 +%114 = OpLogicalOr %8 %28 %30 +%115 = OpLogicalAnd %8 %28 %30 +%116 = OpLogicalOr %8 %28 %30 +%117 = OpLogicalOr %97 %107 %108 +%118 = OpLogicalAnd %8 %28 %30 +%119 = OpLogicalAnd %7 %109 %110 OpReturn OpFunctionEnd -%114 = OpFunction %2 None %97 -%113 = OpLabel -OpBranch %133 -%133 = OpLabel -%134 = OpFNegate %4 %15 -%135 = OpSNegate %118 %119 -%136 = OpFNegate %9 %55 -%137 = OpIAdd %6 %62 %21 -%138 = OpIAdd %115 %117 %116 -%139 = OpFAdd %4 %53 %15 -%140 = OpIAdd %118 %120 %119 -%141 = OpIAdd %121 %122 %123 -%142 = OpFAdd %3 %124 %125 -%143 = OpISub %6 %62 %21 -%144 = OpISub %115 %117 %116 -%145 = OpFSub %4 %53 %15 -%146 = OpISub %118 %120 %119 -%147 = OpISub %121 %122 %123 -%148 = OpFSub %3 %124 %125 -%149 = OpIMul %6 %62 %21 -%150 = OpIMul %115 %117 %116 -%151 = OpFMul %4 %53 %15 -%152 = OpIMul %118 %120 %119 -%153 = OpIMul %121 %122 %123 -%154 = OpFMul %3 %124 %125 -%155 = OpSDiv %6 %62 %21 -%156 = OpUDiv %115 %117 %116 -%157 = OpFDiv %4 %53 %15 -%158 = OpSDiv %118 %120 %119 -%159 = OpUDiv %121 %122 %123 -%160 = OpFDiv %3 %124 %125 -%161 = OpSRem %6 %62 %21 -%162 = OpUMod %115 %117 %116 -%163 = OpFRem %4 %53 %15 -%164 = OpSRem %118 %120 %119 -%165 = OpUMod %121 %122 %123 -%166 = OpFRem %3 %124 %125 -OpBranch %167 -%167 = OpLabel -%169 = OpIAdd %118 %120 %119 -%170 = OpIAdd %118 %120 %119 -%171 = OpIAdd %126 %127 %128 -%172 = OpIAdd %126 %127 %128 -%173 = OpFAdd %9 %54 %55 -%174 = OpFAdd %9 %54 %55 -%175 = OpISub %118 %120 %119 -%176 = OpISub %118 %120 %119 -%177 = OpISub %126 %127 %128 -%178 = OpISub %126 %127 %128 -%179 = OpFSub %9 %54 %55 -%180 = OpFSub %9 %54 %55 -%182 = OpCompositeConstruct %118 %21 %21 -%181 = OpIMul %118 %120 %182 -%184 = OpCompositeConstruct %118 %62 %62 -%183 = OpIMul %118 %119 %184 -%186 = OpCompositeConstruct %126 %116 %116 -%185 = OpIMul %126 %127 %186 -%188 = OpCompositeConstruct %126 %117 %117 -%187 = OpIMul %126 %128 %188 -%189 = OpVectorTimesScalar %9 %54 %15 -%190 = OpVectorTimesScalar %9 %55 %53 -%191 = OpSDiv %118 %120 %119 -%192 = OpSDiv %118 %120 %119 -%193 = OpUDiv %126 %127 %128 -%194 = OpUDiv %126 %127 %128 -%195 = OpFDiv %9 %54 %55 -%196 = OpFDiv %9 %54 %55 -%197 = OpSRem %118 %120 %119 -%198 = OpSRem %118 %120 %119 -%199 = OpUMod %126 %127 %128 -%200 = OpUMod %126 %127 %128 -%201 = OpFRem %9 %54 %55 -%202 = OpFRem %9 %54 %55 -OpBranch %168 -%168 = OpLabel -%204 = OpCompositeExtract %10 %129 0 -%205 = OpCompositeExtract %10 %129 0 -%206 = OpFAdd %10 %204 %205 -%207 = OpCompositeExtract %10 %129 1 -%208 = OpCompositeExtract %10 %129 1 -%209 = OpFAdd %10 %207 %208 -%210 = OpCompositeExtract %10 %129 2 -%211 = OpCompositeExtract %10 %129 2 -%212 = OpFAdd %10 %210 %211 -%203 = OpCompositeConstruct %11 %206 %209 %212 -%214 = OpCompositeExtract %10 %129 0 -%215 = OpCompositeExtract %10 %129 0 -%216 = OpFSub %10 %214 %215 -%217 = OpCompositeExtract %10 %129 1 -%218 = OpCompositeExtract %10 %129 1 -%219 = OpFSub %10 %217 %218 -%220 = OpCompositeExtract %10 %129 2 -%221 = OpCompositeExtract %10 %129 2 -%222 = OpFSub %10 %220 %221 -%213 = OpCompositeConstruct %11 %216 %219 %222 -%223 = OpMatrixTimesScalar %11 %129 %15 -%224 = OpMatrixTimesScalar %11 %129 %53 -%225 = OpMatrixTimesVector %10 %130 %125 -%226 = OpVectorTimesMatrix %3 %131 %130 -%227 = OpMatrixTimesMatrix %11 %130 %132 +%121 = OpFunction %2 None %104 +%120 = OpLabel +OpBranch %138 +%138 = OpLabel +%139 = OpFNegate %4 %17 +%140 = OpSNegate %124 %125 +%141 = OpFNegate %9 %79 +%142 = OpIAdd %6 %64 %23 +%143 = OpIAdd %16 %123 %122 +%144 = OpFAdd %4 %58 %17 +%145 = OpIAdd %124 %126 %125 +%146 = OpIAdd %15 %127 %128 +%147 = OpFAdd %3 %129 %130 +%148 = OpISub %6 %64 %23 +%149 = OpISub %16 %123 %122 +%150 = OpFSub %4 %58 %17 +%151 = OpISub %124 %126 %125 +%152 = OpISub %15 %127 %128 +%153 = OpFSub %3 %129 %130 +%154 = OpIMul %6 %64 %23 +%155 = OpIMul %16 %123 %122 +%156 = OpFMul %4 %58 %17 +%157 = OpIMul %124 %126 %125 +%158 = OpIMul %15 %127 %128 +%159 = OpFMul %3 %129 %130 +%160 = OpSDiv %6 %64 %23 +%161 = OpUDiv %16 %123 %122 +%162 = OpFDiv %4 %58 %17 +%163 = OpSDiv %124 %126 %125 +%164 = OpUDiv %15 %127 %128 +%165 = OpFDiv %3 %129 %130 +%166 = OpSRem %6 %64 %23 +%167 = OpUMod %16 %123 %122 +%168 = OpFRem %4 %58 %17 +%169 = OpSRem %124 %126 %125 +%170 = OpUMod %15 %127 %128 +%171 = OpFRem %3 %129 %130 +OpBranch %172 +%172 = OpLabel +%174 = OpIAdd %124 %126 %125 +%175 = OpIAdd %124 %126 %125 +%176 = OpIAdd %131 %132 %133 +%177 = OpIAdd %131 %132 %133 +%178 = OpFAdd %9 %59 %79 +%179 = OpFAdd %9 %59 %79 +%180 = OpISub %124 %126 %125 +%181 = OpISub %124 %126 %125 +%182 = OpISub %131 %132 %133 +%183 = OpISub %131 %132 %133 +%184 = OpFSub %9 %59 %79 +%185 = OpFSub %9 %59 %79 +%187 = OpCompositeConstruct %124 %23 %23 +%186 = OpIMul %124 %126 %187 +%189 = OpCompositeConstruct %124 %64 %64 +%188 = OpIMul %124 %125 %189 +%191 = OpCompositeConstruct %131 %122 %122 +%190 = OpIMul %131 %132 %191 +%193 = OpCompositeConstruct %131 %123 %123 +%192 = OpIMul %131 %133 %193 +%194 = OpVectorTimesScalar %9 %59 %17 +%195 = OpVectorTimesScalar %9 %79 %58 +%196 = OpSDiv %124 %126 %125 +%197 = OpSDiv %124 %126 %125 +%198 = OpUDiv %131 %132 %133 +%199 = OpUDiv %131 %132 %133 +%200 = OpFDiv %9 %59 %79 +%201 = OpFDiv %9 %59 %79 +%202 = OpSRem %124 %126 %125 +%203 = OpSRem %124 %126 %125 +%204 = OpUMod %131 %132 %133 +%205 = OpUMod %131 %132 %133 +%206 = OpFRem %9 %59 %79 +%207 = OpFRem %9 %59 %79 +OpBranch %173 +%173 = OpLabel +%209 = OpCompositeExtract %10 %134 0 +%210 = OpCompositeExtract %10 %134 0 +%211 = OpFAdd %10 %209 %210 +%212 = OpCompositeExtract %10 %134 1 +%213 = OpCompositeExtract %10 %134 1 +%214 = OpFAdd %10 %212 %213 +%215 = OpCompositeExtract %10 %134 2 +%216 = OpCompositeExtract %10 %134 2 +%217 = OpFAdd %10 %215 %216 +%208 = OpCompositeConstruct %11 %211 %214 %217 +%219 = OpCompositeExtract %10 %134 0 +%220 = OpCompositeExtract %10 %134 0 +%221 = OpFSub %10 %219 %220 +%222 = OpCompositeExtract %10 %134 1 +%223 = OpCompositeExtract %10 %134 1 +%224 = OpFSub %10 %222 %223 +%225 = OpCompositeExtract %10 %134 2 +%226 = OpCompositeExtract %10 %134 2 +%227 = OpFSub %10 %225 %226 +%218 = OpCompositeConstruct %11 %221 %224 %227 +%228 = OpMatrixTimesScalar %11 %134 %17 +%229 = OpMatrixTimesScalar %11 %134 %58 +%230 = OpMatrixTimesVector %10 %135 %130 +%231 = OpVectorTimesMatrix %3 %136 %135 +%232 = OpMatrixTimesMatrix %11 %135 %137 OpReturn OpFunctionEnd -%229 = OpFunction %2 None %97 -%228 = OpLabel -OpBranch %230 -%230 = OpLabel -%231 = OpNot %6 %21 -%232 = OpNot %115 %116 -%233 = OpNot %118 %119 -%234 = OpNot %121 %123 -%235 = OpBitwiseOr %6 %62 %21 -%236 = OpBitwiseOr %115 %117 %116 -%237 = OpBitwiseOr %118 %120 %119 -%238 = OpBitwiseOr %121 %122 %123 -%239 = OpBitwiseAnd %6 %62 %21 -%240 = OpBitwiseAnd %115 %117 %116 -%241 = OpBitwiseAnd %118 %120 %119 -%242 = OpBitwiseAnd %121 %122 %123 -%243 = OpBitwiseXor %6 %62 %21 -%244 = OpBitwiseXor %115 %117 %116 -%245 = OpBitwiseXor %118 %120 %119 -%246 = OpBitwiseXor %121 %122 %123 -%247 = OpShiftLeftLogical %6 %62 %116 -%248 = OpShiftLeftLogical %115 %117 %116 -%249 = OpShiftLeftLogical %118 %120 %128 -%250 = OpShiftLeftLogical %121 %122 %123 -%251 = OpShiftRightArithmetic %6 %62 %116 -%252 = OpShiftRightLogical %115 %117 %116 -%253 = OpShiftRightArithmetic %118 %120 %128 -%254 = OpShiftRightLogical %121 %122 %123 +%234 = OpFunction %2 None %104 +%233 = OpLabel +OpBranch %235 +%235 = OpLabel +%236 = OpNot %6 %23 +%237 = OpNot %16 %122 +%238 = OpNot %124 %125 +%239 = OpNot %15 %128 +%240 = OpBitwiseOr %6 %64 %23 +%241 = OpBitwiseOr %16 %123 %122 +%242 = OpBitwiseOr %124 %126 %125 +%243 = OpBitwiseOr %15 %127 %128 +%244 = OpBitwiseAnd %6 %64 %23 +%245 = OpBitwiseAnd %16 %123 %122 +%246 = OpBitwiseAnd %124 %126 %125 +%247 = OpBitwiseAnd %15 %127 %128 +%248 = OpBitwiseXor %6 %64 %23 +%249 = OpBitwiseXor %16 %123 %122 +%250 = OpBitwiseXor %124 %126 %125 +%251 = OpBitwiseXor %15 %127 %128 +%252 = OpShiftLeftLogical %6 %64 %122 +%253 = OpShiftLeftLogical %16 %123 %122 +%254 = OpShiftLeftLogical %124 %126 %133 +%255 = OpShiftLeftLogical %15 %127 %128 +%256 = OpShiftRightArithmetic %6 %64 %122 +%257 = OpShiftRightLogical %16 %123 %122 +%258 = OpShiftRightArithmetic %124 %126 %133 +%259 = OpShiftRightLogical %15 %127 %128 OpReturn OpFunctionEnd -%256 = OpFunction %2 None %97 -%255 = OpLabel -OpBranch %257 -%257 = OpLabel -%258 = OpIEqual %8 %62 %21 -%259 = OpIEqual %8 %117 %116 -%260 = OpFOrdEqual %8 %53 %15 -%261 = OpIEqual %98 %120 %119 -%262 = OpIEqual %90 %122 %123 -%263 = OpFOrdEqual %7 %124 %125 -%264 = OpINotEqual %8 %62 %21 -%265 = OpINotEqual %8 %117 %116 -%266 = OpFOrdNotEqual %8 %53 %15 -%267 = OpINotEqual %98 %120 %119 -%268 = OpINotEqual %90 %122 %123 -%269 = OpFOrdNotEqual %7 %124 %125 -%270 = OpSLessThan %8 %62 %21 -%271 = OpULessThan %8 %117 %116 -%272 = OpFOrdLessThan %8 %53 %15 -%273 = OpSLessThan %98 %120 %119 -%274 = OpULessThan %90 %122 %123 -%275 = OpFOrdLessThan %7 %124 %125 -%276 = OpSLessThanEqual %8 %62 %21 -%277 = OpULessThanEqual %8 %117 %116 -%278 = OpFOrdLessThanEqual %8 %53 %15 -%279 = OpSLessThanEqual %98 %120 %119 -%280 = OpULessThanEqual %90 %122 %123 -%281 = OpFOrdLessThanEqual %7 %124 %125 -%282 = OpSGreaterThan %8 %62 %21 -%283 = OpUGreaterThan %8 %117 %116 -%284 = OpFOrdGreaterThan %8 %53 %15 -%285 = OpSGreaterThan %98 %120 %119 -%286 = OpUGreaterThan %90 %122 %123 -%287 = OpFOrdGreaterThan %7 %124 %125 -%288 = OpSGreaterThanEqual %8 %62 %21 -%289 = OpUGreaterThanEqual %8 %117 %116 -%290 = OpFOrdGreaterThanEqual %8 %53 %15 -%291 = OpSGreaterThanEqual %98 %120 %119 -%292 = OpUGreaterThanEqual %90 %122 %123 -%293 = OpFOrdGreaterThanEqual %7 %124 %125 +%261 = OpFunction %2 None %104 +%260 = OpLabel +OpBranch %262 +%262 = OpLabel +%263 = OpIEqual %8 %64 %23 +%264 = OpIEqual %8 %123 %122 +%265 = OpFOrdEqual %8 %58 %17 +%266 = OpIEqual %105 %126 %125 +%267 = OpIEqual %97 %127 %128 +%268 = OpFOrdEqual %7 %129 %130 +%269 = OpINotEqual %8 %64 %23 +%270 = OpINotEqual %8 %123 %122 +%271 = OpFOrdNotEqual %8 %58 %17 +%272 = OpINotEqual %105 %126 %125 +%273 = OpINotEqual %97 %127 %128 +%274 = OpFOrdNotEqual %7 %129 %130 +%275 = OpSLessThan %8 %64 %23 +%276 = OpULessThan %8 %123 %122 +%277 = OpFOrdLessThan %8 %58 %17 +%278 = OpSLessThan %105 %126 %125 +%279 = OpULessThan %97 %127 %128 +%280 = OpFOrdLessThan %7 %129 %130 +%281 = OpSLessThanEqual %8 %64 %23 +%282 = OpULessThanEqual %8 %123 %122 +%283 = OpFOrdLessThanEqual %8 %58 %17 +%284 = OpSLessThanEqual %105 %126 %125 +%285 = OpULessThanEqual %97 %127 %128 +%286 = OpFOrdLessThanEqual %7 %129 %130 +%287 = OpSGreaterThan %8 %64 %23 +%288 = OpUGreaterThan %8 %123 %122 +%289 = OpFOrdGreaterThan %8 %58 %17 +%290 = OpSGreaterThan %105 %126 %125 +%291 = OpUGreaterThan %97 %127 %128 +%292 = OpFOrdGreaterThan %7 %129 %130 +%293 = OpSGreaterThanEqual %8 %64 %23 +%294 = OpUGreaterThanEqual %8 %123 %122 +%295 = OpFOrdGreaterThanEqual %8 %58 %17 +%296 = OpSGreaterThanEqual %105 %126 %125 +%297 = OpUGreaterThanEqual %97 %127 %128 +%298 = OpFOrdGreaterThanEqual %7 %129 %130 OpReturn OpFunctionEnd -%295 = OpFunction %2 None %97 -%294 = OpLabel -%297 = OpVariable %298 Function %299 -%300 = OpVariable %301 Function %296 -OpBranch %302 -%302 = OpLabel -OpStore %297 %21 -%303 = OpLoad %6 %297 -%304 = OpIAdd %6 %303 %21 -OpStore %297 %304 -%305 = OpLoad %6 %297 -%306 = OpISub %6 %305 %21 -OpStore %297 %306 -%307 = OpLoad %6 %297 -%308 = OpLoad %6 %297 -%309 = OpIMul %6 %308 %307 -OpStore %297 %309 -%310 = OpLoad %6 %297 -%311 = OpLoad %6 %297 -%312 = OpSDiv %6 %311 %310 -OpStore %297 %312 -%313 = OpLoad %6 %297 -%314 = OpSRem %6 %313 %21 -OpStore %297 %314 -%315 = OpLoad %6 %297 -%316 = OpBitwiseAnd %6 %315 %27 -OpStore %297 %316 -%317 = OpLoad %6 %297 -%318 = OpBitwiseOr %6 %317 %27 -OpStore %297 %318 -%319 = OpLoad %6 %297 -%320 = OpBitwiseXor %6 %319 %27 -OpStore %297 %320 -%321 = OpLoad %6 %297 -%322 = OpShiftLeftLogical %6 %321 %117 -OpStore %297 %322 -%323 = OpLoad %6 %297 -%324 = OpShiftRightArithmetic %6 %323 %116 -OpStore %297 %324 -%325 = OpLoad %6 %297 -%326 = OpIAdd %6 %325 %21 -OpStore %297 %326 -%327 = OpLoad %6 %297 -%328 = OpISub %6 %327 %21 -OpStore %297 %328 -%330 = OpAccessChain %329 %300 %21 -%331 = OpLoad %6 %330 -%332 = OpIAdd %6 %331 %21 -%333 = OpAccessChain %329 %300 %21 -OpStore %333 %332 -%334 = OpAccessChain %329 %300 %21 -%335 = OpLoad %6 %334 -%336 = OpISub %6 %335 %21 -%337 = OpAccessChain %329 %300 %21 -OpStore %337 %336 +%300 = OpFunction %2 None %104 +%299 = OpLabel +%302 = OpVariable %303 Function %304 +%305 = OpVariable %306 Function %301 +OpBranch %307 +%307 = OpLabel +OpStore %302 %23 +%308 = OpLoad %6 %302 +%309 = OpIAdd %6 %308 %23 +OpStore %302 %309 +%310 = OpLoad %6 %302 +%311 = OpISub %6 %310 %23 +OpStore %302 %311 +%312 = OpLoad %6 %302 +%313 = OpLoad %6 %302 +%314 = OpIMul %6 %313 %312 +OpStore %302 %314 +%315 = OpLoad %6 %302 +%316 = OpLoad %6 %302 +%317 = OpSDiv %6 %316 %315 +OpStore %302 %317 +%318 = OpLoad %6 %302 +%319 = OpSRem %6 %318 %23 +OpStore %302 %319 +%320 = OpLoad %6 %302 +%321 = OpBitwiseAnd %6 %320 %29 +OpStore %302 %321 +%322 = OpLoad %6 %302 +%323 = OpBitwiseOr %6 %322 %29 +OpStore %302 %323 +%324 = OpLoad %6 %302 +%325 = OpBitwiseXor %6 %324 %29 +OpStore %302 %325 +%326 = OpLoad %6 %302 +%327 = OpShiftLeftLogical %6 %326 %123 +OpStore %302 %327 +%328 = OpLoad %6 %302 +%329 = OpShiftRightArithmetic %6 %328 %122 +OpStore %302 %329 +%330 = OpLoad %6 %302 +%331 = OpIAdd %6 %330 %23 +OpStore %302 %331 +%332 = OpLoad %6 %302 +%333 = OpISub %6 %332 %23 +OpStore %302 %333 +%335 = OpAccessChain %334 %305 %23 +%336 = OpLoad %6 %335 +%337 = OpIAdd %6 %336 %23 +%338 = OpAccessChain %334 %305 %23 +OpStore %338 %337 +%339 = OpAccessChain %334 %305 %23 +%340 = OpLoad %6 %339 +%341 = OpISub %6 %340 %23 +%342 = OpAccessChain %334 %305 %23 +OpStore %342 %341 OpReturn OpFunctionEnd -%339 = OpFunction %2 None %97 -%338 = OpLabel -OpBranch %340 -%340 = OpLabel -%341 = OpSNegate %6 %21 -%342 = OpSNegate %6 %21 -%343 = OpSNegate %6 %342 -%344 = OpSNegate %6 %21 -%345 = OpSNegate %6 %344 -%346 = OpSNegate %6 %21 -%347 = OpSNegate %6 %346 -%348 = OpSNegate %6 %21 -%349 = OpSNegate %6 %348 +%344 = OpFunction %2 None %104 +%343 = OpLabel +OpBranch %345 +%345 = OpLabel +%346 = OpSNegate %6 %23 +%347 = OpSNegate %6 %23 +%348 = OpSNegate %6 %347 +%349 = OpSNegate %6 %23 %350 = OpSNegate %6 %349 -%351 = OpSNegate %6 %21 +%351 = OpSNegate %6 %23 %352 = OpSNegate %6 %351 -%353 = OpSNegate %6 %352 +%353 = OpSNegate %6 %23 %354 = OpSNegate %6 %353 -%355 = OpSNegate %6 %21 -%356 = OpSNegate %6 %355 +%355 = OpSNegate %6 %354 +%356 = OpSNegate %6 %23 %357 = OpSNegate %6 %356 %358 = OpSNegate %6 %357 %359 = OpSNegate %6 %358 -%360 = OpSNegate %6 %21 +%360 = OpSNegate %6 %23 %361 = OpSNegate %6 %360 %362 = OpSNegate %6 %361 %363 = OpSNegate %6 %362 %364 = OpSNegate %6 %363 +%365 = OpSNegate %6 %23 +%366 = OpSNegate %6 %365 +%367 = OpSNegate %6 %366 +%368 = OpSNegate %6 %367 +%369 = OpSNegate %6 %368 OpReturn OpFunctionEnd -%366 = OpFunction %2 None %97 -%365 = OpLabel -OpBranch %368 -%368 = OpLabel -%369 = OpFunctionCall %3 %24 -%370 = OpFunctionCall %3 %52 -%371 = OpFunctionCall %10 %87 %367 -%372 = OpFunctionCall %2 %96 -%373 = OpFunctionCall %2 %114 -%374 = OpFunctionCall %2 %229 -%375 = OpFunctionCall %2 %256 -%376 = OpFunctionCall %2 %295 +%374 = OpFunction %2 None %104 +%370 = OpLabel +%373 = OpLoad %15 %371 +OpBranch %376 +%376 = OpLabel +%377 = OpFunctionCall %3 %26 +%378 = OpCompositeExtract %16 %373 0 +%379 = OpConvertUToF %4 %378 +%380 = OpCompositeExtract %16 %373 1 +%381 = OpBitcast %6 %380 +%382 = OpFunctionCall %3 %56 %379 %381 +%383 = OpFunctionCall %10 %94 %375 +%384 = OpFunctionCall %2 %103 +%385 = OpFunctionCall %2 %121 +%386 = OpFunctionCall %2 %234 +%387 = OpFunctionCall %2 %261 +%388 = OpFunctionCall %2 %300 OpReturn OpFunctionEnd \ No newline at end of file diff --git a/naga/tests/out/wgsl/operators.wgsl b/naga/tests/out/wgsl/operators.wgsl index 95621e32e2..a107f2b501 100644 --- a/naga/tests/out/wgsl/operators.wgsl +++ b/naga/tests/out/wgsl/operators.wgsl @@ -15,9 +15,9 @@ fn builtins() -> vec4 { return (((((vec4((vec4(s1_) + v_i32_zero)) + s2_) + m1_) + m2_) + vec4(b1_)) + b2_); } -fn splat() -> vec4 { - let a_2 = (((vec2(1.0) + vec2(2.0)) - vec2(3.0)) / vec2(4.0)); - let b = (vec4(5) % vec4(2)); +fn splat(m: f32, n: i32) -> vec4 { + let a_2 = (((vec2(2.0) + vec2(m)) - vec2(4.0)) / vec2(8.0)); + let b = (vec4(n) % vec4(2)); return (a_2.xyxy + vec4(b)); } @@ -241,10 +241,10 @@ fn negation_avoids_prefix_decrement() { } @compute @workgroup_size(1, 1, 1) -fn main() { - let _e0 = builtins(); - let _e1 = splat(); - let _e6 = bool_cast(vec3(1.0, 1.0, 1.0)); +fn main(@builtin(workgroup_id) id: vec3) { + let _e1 = builtins(); + let _e6 = splat(f32(id.x), i32(id.y)); + let _e11 = bool_cast(vec3(1.0, 1.0, 1.0)); logical(); arithmetic(); bit();