diff --git a/naga/src/back/wgsl/writer.rs b/naga/src/back/wgsl/writer.rs index af4923c65a..a43c320553 100644 --- a/naga/src/back/wgsl/writer.rs +++ b/naga/src/back/wgsl/writer.rs @@ -1139,12 +1139,12 @@ impl Writer { crate::Literal::Bool(value) => write!(self.out, "{value}")?, crate::Literal::F64(value) => write!(self.out, "{value:?}lf")?, crate::Literal::I64(value) => { - // `-9223372036854775808li` is not valid WGSL. Nor can we use the AbstractInt - // trick above, as AbstractInt also cannot represent `9223372036854775808`. - // The most negative `i64` value can only be expressed in WGSL using - // subtracting 1 from the second most negative value. + // `-9223372036854775808li` is not valid WGSL. Nor can we simply use the + // AbstractInt trick above, as AbstractInt also cannot represent + // `9223372036854775808`. Instead construct the second most negative + // AbstractInt, subtract one from it, then cast to i64. if value == i64::MIN { - write!(self.out, "{}li - 1li", value + 1)?; + write!(self.out, "i64({} - 1)", value + 1)?; } else { write!(self.out, "{value}li")?; } diff --git a/naga/tests/in/wgsl/int64.wgsl b/naga/tests/in/wgsl/int64.wgsl index f9766cfdc7..86ce73261c 100644 --- a/naga/tests/in/wgsl/int64.wgsl +++ b/naga/tests/in/wgsl/int64.wgsl @@ -56,7 +56,7 @@ fn int64_function(x: i64) -> i64 { val += bitcast>(input_uniform.val_u64_3).z; val += bitcast>(input_uniform.val_u64_4).w; // Most negative i64 - val += -9223372036854775807li - 1li; + val += i64(-9223372036854775807 - 1); // Reading/writing to a uniform/storage buffer output.val_i64 = input_uniform.val_i64 + input_storage.val_i64; diff --git a/naga/tests/out/hlsl/wgsl-int64.hlsl b/naga/tests/out/hlsl/wgsl-int64.hlsl index ef13f9c623..6705e77e71 100644 --- a/naga/tests/out/hlsl/wgsl-int64.hlsl +++ b/naga/tests/out/hlsl/wgsl-int64.hlsl @@ -99,51 +99,51 @@ int64_t int64_function(int64_t x) uint64_t4 _e71 = input_uniform.val_u64_4_; int64_t _e74 = val; val = (_e74 + _e71.w); - int64_t _e79 = val; - val = (_e79 + (-9223372036854775807L - 1L)); - int64_t _e85 = input_uniform.val_i64_; - int64_t _e88 = input_storage.Load(128); - output.Store(128, (_e85 + _e88)); - int64_t2 _e94 = input_uniform.val_i64_2_; - int64_t2 _e97 = input_storage.Load(144); - output.Store(144, (_e94 + _e97)); - int64_t3 _e103 = input_uniform.val_i64_3_; - int64_t3 _e106 = input_storage.Load(160); - output.Store(160, (_e103 + _e106)); - int64_t4 _e112 = input_uniform.val_i64_4_; - int64_t4 _e115 = input_storage.Load(192); - output.Store(192, (_e112 + _e115)); - int64_t _e121[2] = Constructarray2_int64_t_(input_arrays.Load(16+0), input_arrays.Load(16+8)); + int64_t _e77 = val; + val = (_e77 + -9223372036854775808L); + int64_t _e83 = input_uniform.val_i64_; + int64_t _e86 = input_storage.Load(128); + output.Store(128, (_e83 + _e86)); + int64_t2 _e92 = input_uniform.val_i64_2_; + int64_t2 _e95 = input_storage.Load(144); + output.Store(144, (_e92 + _e95)); + int64_t3 _e101 = input_uniform.val_i64_3_; + int64_t3 _e104 = input_storage.Load(160); + output.Store(160, (_e101 + _e104)); + int64_t4 _e110 = input_uniform.val_i64_4_; + int64_t4 _e113 = input_storage.Load(192); + output.Store(192, (_e110 + _e113)); + int64_t _e119[2] = Constructarray2_int64_t_(input_arrays.Load(16+0), input_arrays.Load(16+8)); { - int64_t _value2[2] = _e121; + int64_t _value2[2] = _e119; output_arrays.Store(16+0, _value2[0]); output_arrays.Store(16+8, _value2[1]); } + int64_t _e120 = val; int64_t _e122 = val; + val = (_e122 + abs(_e120)); int64_t _e124 = val; - val = (_e124 + abs(_e122)); + int64_t _e125 = val; int64_t _e126 = val; - int64_t _e127 = val; int64_t _e128 = val; + val = (_e128 + clamp(_e124, _e125, _e126)); int64_t _e130 = val; - val = (_e130 + clamp(_e126, _e127, _e128)); int64_t _e132 = val; - int64_t _e134 = val; + int64_t _e135 = val; + val = (_e135 + dot((_e130).xx, (_e132).xx)); int64_t _e137 = val; - val = (_e137 + dot((_e132).xx, (_e134).xx)); - int64_t _e139 = val; + int64_t _e138 = val; int64_t _e140 = val; + val = (_e140 + max(_e137, _e138)); int64_t _e142 = val; - val = (_e142 + max(_e139, _e140)); - int64_t _e144 = val; + int64_t _e143 = val; int64_t _e145 = val; + val = (_e145 + min(_e142, _e143)); int64_t _e147 = val; - val = (_e147 + min(_e144, _e145)); int64_t _e149 = val; + val = (_e149 + sign(_e147)); int64_t _e151 = val; - val = (_e151 + sign(_e149)); - int64_t _e153 = val; - return _e153; + return _e151; } uint64_t naga_f2u64(float value) { diff --git a/naga/tests/out/msl/wgsl-int64.msl b/naga/tests/out/msl/wgsl-int64.msl index e026cac919..ed5d3da2e4 100644 --- a/naga/tests/out/msl/wgsl-int64.msl +++ b/naga/tests/out/msl/wgsl-int64.msl @@ -83,49 +83,49 @@ long int64_function( metal::ulong4 _e71 = input_uniform.val_u64_4_; long _e74 = val; val = as_type(as_type(_e74) + as_type(as_type(_e71).w)); - long _e79 = val; - val = as_type(as_type(_e79) + as_type(as_type(as_type(-9223372036854775807L) - as_type(1L)))); - long _e85 = input_uniform.val_i64_; - long _e88 = input_storage.val_i64_; - output.val_i64_ = as_type(as_type(_e85) + as_type(_e88)); - metal::long2 _e94 = input_uniform.val_i64_2_; - metal::long2 _e97 = input_storage.val_i64_2_; - output.val_i64_2_ = as_type(as_type(_e94) + as_type(_e97)); - metal::long3 _e103 = input_uniform.val_i64_3_; - metal::long3 _e106 = input_storage.val_i64_3_; - output.val_i64_3_ = as_type(as_type(_e103) + as_type(_e106)); - metal::long4 _e112 = input_uniform.val_i64_4_; - metal::long4 _e115 = input_storage.val_i64_4_; - output.val_i64_4_ = as_type(as_type(_e112) + as_type(_e115)); - type_12 _e121 = input_arrays.val_i64_array_2_; - output_arrays.val_i64_array_2_ = _e121; + long _e77 = val; + val = as_type(as_type(_e77) + as_type(-9223372036854775808L)); + long _e83 = input_uniform.val_i64_; + long _e86 = input_storage.val_i64_; + output.val_i64_ = as_type(as_type(_e83) + as_type(_e86)); + metal::long2 _e92 = input_uniform.val_i64_2_; + metal::long2 _e95 = input_storage.val_i64_2_; + output.val_i64_2_ = as_type(as_type(_e92) + as_type(_e95)); + metal::long3 _e101 = input_uniform.val_i64_3_; + metal::long3 _e104 = input_storage.val_i64_3_; + output.val_i64_3_ = as_type(as_type(_e101) + as_type(_e104)); + metal::long4 _e110 = input_uniform.val_i64_4_; + metal::long4 _e113 = input_storage.val_i64_4_; + output.val_i64_4_ = as_type(as_type(_e110) + as_type(_e113)); + type_12 _e119 = input_arrays.val_i64_array_2_; + output_arrays.val_i64_array_2_ = _e119; + long _e120 = val; long _e122 = val; + val = as_type(as_type(_e122) + as_type(naga_abs(_e120))); long _e124 = val; - val = as_type(as_type(_e124) + as_type(naga_abs(_e122))); + long _e125 = val; long _e126 = val; - long _e127 = val; long _e128 = val; + val = as_type(as_type(_e128) + as_type(metal::clamp(_e124, _e125, _e126))); long _e130 = val; - val = as_type(as_type(_e130) + as_type(metal::clamp(_e126, _e127, _e128))); + metal::long2 _e131 = metal::long2(_e130); long _e132 = val; metal::long2 _e133 = metal::long2(_e132); - long _e134 = val; - metal::long2 _e135 = metal::long2(_e134); + long _e135 = val; + val = as_type(as_type(_e135) + as_type(( + _e131.x * _e133.x + _e131.y * _e133.y))); long _e137 = val; - val = as_type(as_type(_e137) + as_type(( + _e133.x * _e135.x + _e133.y * _e135.y))); - long _e139 = val; + long _e138 = val; long _e140 = val; + val = as_type(as_type(_e140) + as_type(metal::max(_e137, _e138))); long _e142 = val; - val = as_type(as_type(_e142) + as_type(metal::max(_e139, _e140))); - long _e144 = val; + long _e143 = val; long _e145 = val; + val = as_type(as_type(_e145) + as_type(metal::min(_e142, _e143))); long _e147 = val; - val = as_type(as_type(_e147) + as_type(metal::min(_e144, _e145))); long _e149 = val; + val = as_type(as_type(_e149) + as_type(metal::select(metal::select(long(-1), long(1), (_e147 > 0)), long(0), (_e147 == 0)))); long _e151 = val; - val = as_type(as_type(_e151) + as_type(metal::select(metal::select(long(-1), long(1), (_e149 > 0)), long(0), (_e149 == 0)))); - long _e153 = val; - return _e153; + return _e151; } ulong naga_f2u64(float value) { diff --git a/naga/tests/out/spv/wgsl-int64.spvasm b/naga/tests/out/spv/wgsl-int64.spvasm index 377247a215..883eb157f5 100644 --- a/naga/tests/out/spv/wgsl-int64.spvasm +++ b/naga/tests/out/spv/wgsl-int64.spvasm @@ -93,23 +93,24 @@ OpMemberDecorate %36 0 Offset 0 %53 = OpConstant %3 1002003004005006 %54 = OpConstant %3 -9223372036854775807 %55 = OpConstant %3 5 -%57 = OpTypePointer Function %3 -%67 = OpTypePointer Uniform %5 -%76 = OpTypePointer Uniform %6 -%77 = OpConstant %5 1 -%86 = OpTypePointer Uniform %7 -%92 = OpConstant %7 -9.223372e18 -%93 = OpConstant %7 9.2233715e18 -%98 = OpTypePointer Uniform %3 -%99 = OpConstant %5 7 -%106 = OpTypePointer Uniform %4 -%107 = OpConstant %5 3 -%113 = OpTypePointer Uniform %8 -%114 = OpConstant %5 4 -%121 = OpTypePointer Uniform %9 -%122 = OpConstant %5 5 -%129 = OpTypePointer Uniform %10 -%130 = OpConstant %5 6 +%56 = OpConstant %3 -9223372036854775808 +%58 = OpTypePointer Function %3 +%68 = OpTypePointer Uniform %5 +%77 = OpTypePointer Uniform %6 +%78 = OpConstant %5 1 +%87 = OpTypePointer Uniform %7 +%93 = OpConstant %7 -9.223372e18 +%94 = OpConstant %7 9.2233715e18 +%99 = OpTypePointer Uniform %3 +%100 = OpConstant %5 7 +%107 = OpTypePointer Uniform %4 +%108 = OpConstant %5 3 +%114 = OpTypePointer Uniform %8 +%115 = OpConstant %5 4 +%122 = OpTypePointer Uniform %9 +%123 = OpConstant %5 5 +%130 = OpTypePointer Uniform %10 +%131 = OpConstant %5 6 %140 = OpTypePointer StorageBuffer %3 %147 = OpTypePointer StorageBuffer %11 %148 = OpTypePointer Uniform %11 @@ -142,96 +143,95 @@ OpMemberDecorate %36 0 Offset 0 %40 = OpFunction %3 None %41 %39 = OpFunctionParameter %3 %38 = OpLabel -%56 = OpVariable %57 Function %51 +%57 = OpVariable %58 Function %51 %44 = OpAccessChain %42 %23 %43 %46 = OpAccessChain %45 %26 %43 %48 = OpAccessChain %47 %29 %43 %49 = OpAccessChain %45 %32 %43 %50 = OpAccessChain %47 %35 %43 -OpBranch %58 -%58 = OpLabel -%59 = OpISub %3 %52 %53 -%60 = OpIAdd %3 %59 %54 -%61 = OpLoad %3 %56 -%62 = OpIAdd %3 %61 %60 -OpStore %56 %62 -%63 = OpLoad %3 %56 -%64 = OpIAdd %3 %63 %55 -%65 = OpLoad %3 %56 -%66 = OpIAdd %3 %65 %64 -OpStore %56 %66 -%68 = OpAccessChain %67 %44 %43 -%69 = OpLoad %5 %68 -%70 = OpLoad %3 %56 -%71 = OpUConvert %5 %70 -%72 = OpIAdd %5 %69 %71 -%73 = OpSConvert %3 %72 -%74 = OpLoad %3 %56 -%75 = OpIAdd %3 %74 %73 -OpStore %56 %75 -%78 = OpAccessChain %76 %44 %77 -%79 = OpLoad %6 %78 -%80 = OpLoad %3 %56 -%81 = OpSConvert %6 %80 -%82 = OpIAdd %6 %79 %81 -%83 = OpSConvert %3 %82 -%84 = OpLoad %3 %56 -%85 = OpIAdd %3 %84 %83 -OpStore %56 %85 -%87 = OpAccessChain %86 %44 %16 -%88 = OpLoad %7 %87 -%89 = OpLoad %3 %56 -%90 = OpConvertSToF %7 %89 -%91 = OpFAdd %7 %88 %90 -%94 = OpExtInst %7 %1 FClamp %91 %92 %93 -%95 = OpConvertFToS %3 %94 -%96 = OpLoad %3 %56 -%97 = OpIAdd %3 %96 %95 -OpStore %56 %97 -%100 = OpAccessChain %98 %44 %99 -%101 = OpLoad %3 %100 -%102 = OpCompositeConstruct %12 %101 %101 %101 -%103 = OpCompositeExtract %3 %102 2 -%104 = OpLoad %3 %56 -%105 = OpIAdd %3 %104 %103 -OpStore %56 %105 -%108 = OpAccessChain %106 %44 %107 -%109 = OpLoad %4 %108 -%110 = OpBitcast %3 %109 -%111 = OpLoad %3 %56 -%112 = OpIAdd %3 %111 %110 -OpStore %56 %112 -%115 = OpAccessChain %113 %44 %114 -%116 = OpLoad %8 %115 -%117 = OpBitcast %11 %116 -%118 = OpCompositeExtract %3 %117 1 -%119 = OpLoad %3 %56 -%120 = OpIAdd %3 %119 %118 -OpStore %56 %120 -%123 = OpAccessChain %121 %44 %122 -%124 = OpLoad %9 %123 -%125 = OpBitcast %12 %124 -%126 = OpCompositeExtract %3 %125 2 -%127 = OpLoad %3 %56 -%128 = OpIAdd %3 %127 %126 -OpStore %56 %128 -%131 = OpAccessChain %129 %44 %130 -%132 = OpLoad %10 %131 -%133 = OpBitcast %13 %132 -%134 = OpCompositeExtract %3 %133 3 -%135 = OpLoad %3 %56 -%136 = OpIAdd %3 %135 %134 -OpStore %56 %136 -%137 = OpISub %3 %54 %19 -%138 = OpLoad %3 %56 -%139 = OpIAdd %3 %138 %137 -OpStore %56 %139 -%141 = OpAccessChain %98 %44 %99 +OpBranch %59 +%59 = OpLabel +%60 = OpISub %3 %52 %53 +%61 = OpIAdd %3 %60 %54 +%62 = OpLoad %3 %57 +%63 = OpIAdd %3 %62 %61 +OpStore %57 %63 +%64 = OpLoad %3 %57 +%65 = OpIAdd %3 %64 %55 +%66 = OpLoad %3 %57 +%67 = OpIAdd %3 %66 %65 +OpStore %57 %67 +%69 = OpAccessChain %68 %44 %43 +%70 = OpLoad %5 %69 +%71 = OpLoad %3 %57 +%72 = OpUConvert %5 %71 +%73 = OpIAdd %5 %70 %72 +%74 = OpSConvert %3 %73 +%75 = OpLoad %3 %57 +%76 = OpIAdd %3 %75 %74 +OpStore %57 %76 +%79 = OpAccessChain %77 %44 %78 +%80 = OpLoad %6 %79 +%81 = OpLoad %3 %57 +%82 = OpSConvert %6 %81 +%83 = OpIAdd %6 %80 %82 +%84 = OpSConvert %3 %83 +%85 = OpLoad %3 %57 +%86 = OpIAdd %3 %85 %84 +OpStore %57 %86 +%88 = OpAccessChain %87 %44 %16 +%89 = OpLoad %7 %88 +%90 = OpLoad %3 %57 +%91 = OpConvertSToF %7 %90 +%92 = OpFAdd %7 %89 %91 +%95 = OpExtInst %7 %1 FClamp %92 %93 %94 +%96 = OpConvertFToS %3 %95 +%97 = OpLoad %3 %57 +%98 = OpIAdd %3 %97 %96 +OpStore %57 %98 +%101 = OpAccessChain %99 %44 %100 +%102 = OpLoad %3 %101 +%103 = OpCompositeConstruct %12 %102 %102 %102 +%104 = OpCompositeExtract %3 %103 2 +%105 = OpLoad %3 %57 +%106 = OpIAdd %3 %105 %104 +OpStore %57 %106 +%109 = OpAccessChain %107 %44 %108 +%110 = OpLoad %4 %109 +%111 = OpBitcast %3 %110 +%112 = OpLoad %3 %57 +%113 = OpIAdd %3 %112 %111 +OpStore %57 %113 +%116 = OpAccessChain %114 %44 %115 +%117 = OpLoad %8 %116 +%118 = OpBitcast %11 %117 +%119 = OpCompositeExtract %3 %118 1 +%120 = OpLoad %3 %57 +%121 = OpIAdd %3 %120 %119 +OpStore %57 %121 +%124 = OpAccessChain %122 %44 %123 +%125 = OpLoad %9 %124 +%126 = OpBitcast %12 %125 +%127 = OpCompositeExtract %3 %126 2 +%128 = OpLoad %3 %57 +%129 = OpIAdd %3 %128 %127 +OpStore %57 %129 +%132 = OpAccessChain %130 %44 %131 +%133 = OpLoad %10 %132 +%134 = OpBitcast %13 %133 +%135 = OpCompositeExtract %3 %134 3 +%136 = OpLoad %3 %57 +%137 = OpIAdd %3 %136 %135 +OpStore %57 %137 +%138 = OpLoad %3 %57 +%139 = OpIAdd %3 %138 %56 +OpStore %57 %139 +%141 = OpAccessChain %99 %44 %100 %142 = OpLoad %3 %141 -%143 = OpAccessChain %140 %46 %99 +%143 = OpAccessChain %140 %46 %100 %144 = OpLoad %3 %143 %145 = OpIAdd %3 %142 %144 -%146 = OpAccessChain %140 %49 %99 +%146 = OpAccessChain %140 %49 %100 OpStore %146 %145 %150 = OpAccessChain %148 %44 %149 %151 = OpLoad %11 %150 @@ -254,26 +254,26 @@ OpStore %164 %163 %172 = OpIAdd %13 %169 %171 %173 = OpAccessChain %165 %49 %167 OpStore %173 %172 -%175 = OpAccessChain %174 %48 %77 +%175 = OpAccessChain %174 %48 %78 %176 = OpLoad %17 %175 -%177 = OpAccessChain %174 %50 %77 +%177 = OpAccessChain %174 %50 %78 OpStore %177 %176 -%178 = OpLoad %3 %56 +%178 = OpLoad %3 %57 %179 = OpExtInst %3 %1 SAbs %178 -%180 = OpLoad %3 %56 +%180 = OpLoad %3 %57 %181 = OpIAdd %3 %180 %179 -OpStore %56 %181 -%182 = OpLoad %3 %56 -%183 = OpLoad %3 %56 -%184 = OpLoad %3 %56 +OpStore %57 %181 +%182 = OpLoad %3 %57 +%183 = OpLoad %3 %57 +%184 = OpLoad %3 %57 %186 = OpExtInst %3 %1 SMax %182 %183 %185 = OpExtInst %3 %1 SMin %186 %184 -%187 = OpLoad %3 %56 +%187 = OpLoad %3 %57 %188 = OpIAdd %3 %187 %185 -OpStore %56 %188 -%189 = OpLoad %3 %56 +OpStore %57 %188 +%189 = OpLoad %3 %57 %190 = OpCompositeConstruct %11 %189 %189 -%191 = OpLoad %3 %56 +%191 = OpLoad %3 %57 %192 = OpCompositeConstruct %11 %191 %191 %195 = OpCompositeExtract %3 %190 0 %196 = OpCompositeExtract %3 %192 0 @@ -283,27 +283,27 @@ OpStore %56 %188 %200 = OpCompositeExtract %3 %192 1 %201 = OpIMul %3 %199 %200 %193 = OpIAdd %3 %198 %201 -%202 = OpLoad %3 %56 +%202 = OpLoad %3 %57 %203 = OpIAdd %3 %202 %193 -OpStore %56 %203 -%204 = OpLoad %3 %56 -%205 = OpLoad %3 %56 +OpStore %57 %203 +%204 = OpLoad %3 %57 +%205 = OpLoad %3 %57 %206 = OpExtInst %3 %1 SMax %204 %205 -%207 = OpLoad %3 %56 +%207 = OpLoad %3 %57 %208 = OpIAdd %3 %207 %206 -OpStore %56 %208 -%209 = OpLoad %3 %56 -%210 = OpLoad %3 %56 +OpStore %57 %208 +%209 = OpLoad %3 %57 +%210 = OpLoad %3 %57 %211 = OpExtInst %3 %1 SMin %209 %210 -%212 = OpLoad %3 %56 +%212 = OpLoad %3 %57 %213 = OpIAdd %3 %212 %211 -OpStore %56 %213 -%214 = OpLoad %3 %56 +OpStore %57 %213 +%214 = OpLoad %3 %57 %215 = OpExtInst %3 %1 SSign %214 -%216 = OpLoad %3 %56 +%216 = OpLoad %3 %57 %217 = OpIAdd %3 %216 %215 -OpStore %56 %217 -%218 = OpLoad %3 %56 +OpStore %57 %217 +%218 = OpLoad %3 %57 OpReturnValue %218 OpFunctionEnd %221 = OpFunction %4 None %222 @@ -327,7 +327,7 @@ OpStore %231 %237 %240 = OpLoad %4 %231 %241 = OpIAdd %4 %240 %239 OpStore %231 %241 -%242 = OpAccessChain %67 %223 %43 +%242 = OpAccessChain %68 %223 %43 %243 = OpLoad %5 %242 %244 = OpLoad %4 %231 %245 = OpUConvert %5 %244 @@ -336,7 +336,7 @@ OpStore %231 %241 %248 = OpLoad %4 %231 %249 = OpIAdd %4 %248 %247 OpStore %231 %249 -%250 = OpAccessChain %76 %223 %77 +%250 = OpAccessChain %77 %223 %78 %251 = OpLoad %6 %250 %252 = OpLoad %4 %231 %253 = OpSConvert %6 %252 @@ -345,7 +345,7 @@ OpStore %231 %249 %256 = OpLoad %4 %231 %257 = OpIAdd %4 %256 %255 OpStore %231 %257 -%258 = OpAccessChain %86 %223 %16 +%258 = OpAccessChain %87 %223 %16 %259 = OpLoad %7 %258 %260 = OpLoad %4 %231 %261 = OpConvertUToF %7 %260 @@ -355,14 +355,14 @@ OpStore %231 %257 %267 = OpLoad %4 %231 %268 = OpIAdd %4 %267 %266 OpStore %231 %268 -%269 = OpAccessChain %106 %223 %107 +%269 = OpAccessChain %107 %223 %108 %270 = OpLoad %4 %269 %271 = OpCompositeConstruct %9 %270 %270 %270 %272 = OpCompositeExtract %4 %271 2 %273 = OpLoad %4 %231 %274 = OpIAdd %4 %273 %272 OpStore %231 %274 -%275 = OpAccessChain %98 %223 %99 +%275 = OpAccessChain %99 %223 %100 %276 = OpLoad %3 %275 %277 = OpBitcast %4 %276 %278 = OpLoad %4 %231 @@ -389,33 +389,33 @@ OpStore %231 %291 %296 = OpLoad %4 %231 %297 = OpIAdd %4 %296 %295 OpStore %231 %297 -%299 = OpAccessChain %106 %223 %107 +%299 = OpAccessChain %107 %223 %108 %300 = OpLoad %4 %299 -%301 = OpAccessChain %298 %224 %107 +%301 = OpAccessChain %298 %224 %108 %302 = OpLoad %4 %301 %303 = OpIAdd %4 %300 %302 -%304 = OpAccessChain %298 %226 %107 +%304 = OpAccessChain %298 %226 %108 OpStore %304 %303 -%306 = OpAccessChain %113 %223 %114 +%306 = OpAccessChain %114 %223 %115 %307 = OpLoad %8 %306 -%308 = OpAccessChain %305 %224 %114 +%308 = OpAccessChain %305 %224 %115 %309 = OpLoad %8 %308 %310 = OpIAdd %8 %307 %309 -%311 = OpAccessChain %305 %226 %114 +%311 = OpAccessChain %305 %226 %115 OpStore %311 %310 -%313 = OpAccessChain %121 %223 %122 +%313 = OpAccessChain %122 %223 %123 %314 = OpLoad %9 %313 -%315 = OpAccessChain %312 %224 %122 +%315 = OpAccessChain %312 %224 %123 %316 = OpLoad %9 %315 %317 = OpIAdd %9 %314 %316 -%318 = OpAccessChain %312 %226 %122 +%318 = OpAccessChain %312 %226 %123 OpStore %318 %317 -%320 = OpAccessChain %129 %223 %130 +%320 = OpAccessChain %130 %223 %131 %321 = OpLoad %10 %320 -%322 = OpAccessChain %319 %224 %130 +%322 = OpAccessChain %319 %224 %131 %323 = OpLoad %10 %322 %324 = OpIAdd %10 %321 %323 -%325 = OpAccessChain %319 %226 %130 +%325 = OpAccessChain %319 %226 %131 OpStore %325 %324 %327 = OpAccessChain %326 %225 %43 %328 = OpLoad %15 %327 diff --git a/naga/tests/out/wgsl/wgsl-conversion-float-to-int.wgsl b/naga/tests/out/wgsl/wgsl-conversion-float-to-int.wgsl index a185b81dea..8435f01bf5 100644 --- a/naga/tests/out/wgsl/wgsl-conversion-float-to-int.wgsl +++ b/naga/tests/out/wgsl/wgsl-conversion-float-to-int.wgsl @@ -20,11 +20,11 @@ fn test_const_eval() { var max_f32_to_i32_: i32 = 2147483520i; var min_f32_to_u32_: u32 = 0u; var max_f32_to_u32_: u32 = 4294967040u; - var min_f32_to_i64_: i64 = -9223372036854775807li - 1li; + var min_f32_to_i64_: i64 = i64(-9223372036854775807 - 1); var max_f32_to_i64_: i64 = 9223371487098961920li; var min_f32_to_u64_: u64 = 0lu; var max_f32_to_u64_: u64 = 18446742974197923840lu; - var min_f64_to_i64_: i64 = -9223372036854775807li - 1li; + var min_f64_to_i64_: i64 = i64(-9223372036854775807 - 1); var max_f64_to_i64_: i64 = 9223372036854774784li; var min_f64_to_u64_: u64 = 0lu; var max_f64_to_u64_: u64 = 18446744073709549568lu; @@ -32,7 +32,7 @@ fn test_const_eval() { var max_abstract_float_to_i32_: i32 = 2147483647i; var min_abstract_float_to_u32_: u32 = 0u; var max_abstract_float_to_u32_: u32 = 4294967295u; - var min_abstract_float_to_i64_: i64 = -9223372036854775807li - 1li; + var min_abstract_float_to_i64_: i64 = i64(-9223372036854775807 - 1); var max_abstract_float_to_i64_: i64 = 9223372036854774784li; var min_abstract_float_to_u64_: u64 = 0lu; var max_abstract_float_to_u64_: u64 = 18446744073709549568lu; diff --git a/naga/tests/out/wgsl/wgsl-int64.wgsl b/naga/tests/out/wgsl/wgsl-int64.wgsl index c339d8d4c3..b554457700 100644 --- a/naga/tests/out/wgsl/wgsl-int64.wgsl +++ b/naga/tests/out/wgsl/wgsl-int64.wgsl @@ -67,47 +67,47 @@ fn int64_function(x: i64) -> i64 { let _e71 = input_uniform.val_u64_4_; let _e74 = val; val = (_e74 + bitcast>(_e71).w); - let _e79 = val; - val = (_e79 + (-9223372036854775807li - 1li)); - let _e85 = input_uniform.val_i64_; - let _e88 = input_storage.val_i64_; - output.val_i64_ = (_e85 + _e88); - let _e94 = input_uniform.val_i64_2_; - let _e97 = input_storage.val_i64_2_; - output.val_i64_2_ = (_e94 + _e97); - let _e103 = input_uniform.val_i64_3_; - let _e106 = input_storage.val_i64_3_; - output.val_i64_3_ = (_e103 + _e106); - let _e112 = input_uniform.val_i64_4_; - let _e115 = input_storage.val_i64_4_; - output.val_i64_4_ = (_e112 + _e115); - let _e121 = input_arrays.val_i64_array_2_; - output_arrays.val_i64_array_2_ = _e121; + let _e77 = val; + val = (_e77 + i64(-9223372036854775807 - 1)); + let _e83 = input_uniform.val_i64_; + let _e86 = input_storage.val_i64_; + output.val_i64_ = (_e83 + _e86); + let _e92 = input_uniform.val_i64_2_; + let _e95 = input_storage.val_i64_2_; + output.val_i64_2_ = (_e92 + _e95); + let _e101 = input_uniform.val_i64_3_; + let _e104 = input_storage.val_i64_3_; + output.val_i64_3_ = (_e101 + _e104); + let _e110 = input_uniform.val_i64_4_; + let _e113 = input_storage.val_i64_4_; + output.val_i64_4_ = (_e110 + _e113); + let _e119 = input_arrays.val_i64_array_2_; + output_arrays.val_i64_array_2_ = _e119; + let _e120 = val; let _e122 = val; + val = (_e122 + abs(_e120)); let _e124 = val; - val = (_e124 + abs(_e122)); + let _e125 = val; let _e126 = val; - let _e127 = val; let _e128 = val; + val = (_e128 + clamp(_e124, _e125, _e126)); let _e130 = val; - val = (_e130 + clamp(_e126, _e127, _e128)); let _e132 = val; - let _e134 = val; + let _e135 = val; + val = (_e135 + dot(vec2(_e130), vec2(_e132))); let _e137 = val; - val = (_e137 + dot(vec2(_e132), vec2(_e134))); - let _e139 = val; + let _e138 = val; let _e140 = val; + val = (_e140 + max(_e137, _e138)); let _e142 = val; - val = (_e142 + max(_e139, _e140)); - let _e144 = val; + let _e143 = val; let _e145 = val; + val = (_e145 + min(_e142, _e143)); let _e147 = val; - val = (_e147 + min(_e144, _e145)); let _e149 = val; + val = (_e149 + sign(_e147)); let _e151 = val; - val = (_e151 + sign(_e149)); - let _e153 = val; - return _e153; + return _e151; } fn uint64_function(x_1: u64) -> u64 {