mirror of
https://github.com/gfx-rs/wgpu.git
synced 2025-04-13 04:26:34 +00:00
Allow 64-bit hex literals and const-evaluate unary ops (#6616)
* Allow 64-bit hex literals * Implement const evaluation for unary 64-bit ops * Test 64-bit hex literals * Add changelog entry
This commit is contained in:
Sludge
GitHub
parent
3016c56db6
commit
a9f14c7a90
@ -98,6 +98,7 @@ By @ErichDonGubler in [#6456](https://github.com/gfx-rs/wgpu/pull/6456), [#6148]
|
||||
|
||||
- Fix an issue where `naga` CLI would incorrectly skip the first positional argument when `--stdin-file-path` was specified. By @ErichDonGubler in [#6480](https://github.com/gfx-rs/wgpu/pull/6480).
|
||||
- Fix textureNumLevels in the GLSL backend. By @magcius in [#6483](https://github.com/gfx-rs/wgpu/pull/6483).
|
||||
- Support 64-bit hex literals and unary operations in constants [#6616](https://github.com/gfx-rs/wgpu/pull/6616).
|
||||
- Implement `quantizeToF16()` for WGSL frontend, and WGSL, SPIR-V, HLSL, MSL, and GLSL backends. By @jamienicol in [#6519](https://github.com/gfx-rs/wgpu/pull/6519).
|
||||
- Add support for GLSL `usampler*` and `isampler*`. By @DavidPeicho in [#6513](https://github.com/gfx-rs/wgpu/pull/6513).
|
||||
|
||||
@ -128,7 +129,7 @@ By @ErichDonGubler in [#6456](https://github.com/gfx-rs/wgpu/pull/6456), [#6148]
|
||||
|
||||
- Handle query set creation failure as an internal error that loses the `Device`, rather than panicking. By @ErichDonGubler in [#6505](https://github.com/gfx-rs/wgpu/pull/6505).
|
||||
- Ensure that `Features::TIMESTAMP_QUERY` is set when using timestamp writes in render and compute passes. By @ErichDonGubler in [#6497](https://github.com/gfx-rs/wgpu/pull/6497).
|
||||
- Check for device mismatches when beginning render and compute passes. By @ErichDonGubler in [#6497](https://github.com/gfx-rs/wgpu/pull/6497).
|
||||
- Check for device mismatches when beginning render and compute passes. By @ErichDonGubler in [#6497](https://github.com/gfx-rs/wgpu/pull/6497).
|
||||
- Lower `QUERY_SET_MAX_QUERIES` (and enforced limits) from 8192 to 4096 to match WebGPU spec. By @ErichDonGubler in [#6525](https://github.com/gfx-rs/wgpu/pull/6525).
|
||||
- Allow non-filterable float on texture bindings never used with samplers when using a derived bind group layout. By @ErichDonGubler in [#6531](https://github.com/gfx-rs/wgpu/pull/6531/).
|
||||
- Replace potentially unsound usage of `PreHashedMap` with `FastHashMap`. By @jamienicol in [#6541](https://github.com/gfx-rs/wgpu/pull/6541).
|
||||
|
||||
@ -216,7 +216,12 @@ fn parse(input: &str) -> (Result<Number, NumberError>, &str) {
|
||||
rest_to_str!(bytes),
|
||||
)
|
||||
} else {
|
||||
let kind = consume_map!(bytes, [b'i' => IntKind::I32, b'u' => IntKind::U32]);
|
||||
let kind = consume_map!(bytes, [
|
||||
b'i' => IntKind::I32,
|
||||
b'u' => IntKind::U32,
|
||||
b'l', b'i' => IntKind::I64,
|
||||
b'l', b'u' => IntKind::U64,
|
||||
]);
|
||||
|
||||
(parse_hex_int(digits, kind), rest_to_str!(bytes))
|
||||
}
|
||||
|
||||
@ -1742,6 +1742,7 @@ impl<'a> ConstantEvaluator<'a> {
|
||||
Expression::Literal(value) => Expression::Literal(match op {
|
||||
UnaryOperator::Negate => match value {
|
||||
Literal::I32(v) => Literal::I32(v.wrapping_neg()),
|
||||
Literal::I64(v) => Literal::I64(v.wrapping_neg()),
|
||||
Literal::F32(v) => Literal::F32(-v),
|
||||
Literal::AbstractInt(v) => Literal::AbstractInt(v.wrapping_neg()),
|
||||
Literal::AbstractFloat(v) => Literal::AbstractFloat(-v),
|
||||
@ -1753,7 +1754,9 @@ impl<'a> ConstantEvaluator<'a> {
|
||||
},
|
||||
UnaryOperator::BitwiseNot => match value {
|
||||
Literal::I32(v) => Literal::I32(!v),
|
||||
Literal::I64(v) => Literal::I64(!v),
|
||||
Literal::U32(v) => Literal::U32(!v),
|
||||
Literal::U64(v) => Literal::U64(!v),
|
||||
Literal::AbstractInt(v) => Literal::AbstractInt(!v),
|
||||
_ => return Err(ConstantEvaluatorError::InvalidUnaryOpArg),
|
||||
},
|
||||
|
||||
@ -41,7 +41,7 @@ var<storage, read_write> output_arrays: StorageCompatible;
|
||||
fn int64_function(x: i64) -> i64 {
|
||||
var val: i64 = i64(constant_variable);
|
||||
// A number too big for i32
|
||||
val += 31li - 1002003004005006li;
|
||||
val += 31li - 1002003004005006li + -0x7fffffffffffffffli;
|
||||
// Constructing an i64 from an AbstractInt
|
||||
val += val + i64(5);
|
||||
// Constructing a i64 from other types and other types from u64.
|
||||
@ -89,8 +89,8 @@ fn int64_function(x: i64) -> i64 {
|
||||
|
||||
fn uint64_function(x: u64) -> u64 {
|
||||
var val: u64 = u64(constant_variable);
|
||||
// A number too big for u32
|
||||
val += 31lu + 1002003004005006lu;
|
||||
// Numbers too big for u32 (u64::MAX)
|
||||
val += 31lu + 18446744073709551615lu - 0xfffffffffffffffflu;
|
||||
// Constructing a u64 from an AbstractInt
|
||||
val += val + u64(5);
|
||||
// Constructing a u64 from other types and other types from u64.
|
||||
|
||||
@ -63,81 +63,81 @@ int64_t int64_function(int64_t x)
|
||||
{
|
||||
int64_t val = 20L;
|
||||
|
||||
int64_t _e6 = val;
|
||||
val = (_e6 + (31L - 1002003004005006L));
|
||||
int64_t _e8 = val;
|
||||
int64_t _e11 = val;
|
||||
val = (_e11 + (_e8 + 5L));
|
||||
uint _e15 = input_uniform.val_u32_;
|
||||
int64_t _e16 = val;
|
||||
int64_t _e20 = val;
|
||||
val = (_e20 + int64_t((_e15 + uint(_e16))));
|
||||
int _e24 = input_uniform.val_i32_;
|
||||
int64_t _e25 = val;
|
||||
int64_t _e29 = val;
|
||||
val = (_e29 + int64_t((_e24 + int(_e25))));
|
||||
float _e33 = input_uniform.val_f32_;
|
||||
int64_t _e34 = val;
|
||||
int64_t _e38 = val;
|
||||
val = (_e38 + int64_t((_e33 + float(_e34))));
|
||||
int64_t _e42 = input_uniform.val_i64_;
|
||||
int64_t _e45 = val;
|
||||
val = (_e45 + (_e42).xxx.z);
|
||||
uint64_t _e49 = input_uniform.val_u64_;
|
||||
int64_t _e51 = val;
|
||||
val = (_e51 + _e49);
|
||||
uint64_t2 _e55 = input_uniform.val_u64_2_;
|
||||
int64_t _e58 = val;
|
||||
val = (_e58 + _e55.y);
|
||||
uint64_t3 _e62 = input_uniform.val_u64_3_;
|
||||
int64_t _e65 = val;
|
||||
val = (_e65 + _e62.z);
|
||||
uint64_t4 _e69 = input_uniform.val_u64_4_;
|
||||
int64_t _e72 = val;
|
||||
val = (_e72 + _e69.w);
|
||||
int64_t _e78 = input_uniform.val_i64_;
|
||||
int64_t _e81 = input_storage.Load<int64_t>(128);
|
||||
output.Store(128, (_e78 + _e81));
|
||||
int64_t2 _e87 = input_uniform.val_i64_2_;
|
||||
int64_t2 _e90 = input_storage.Load<int64_t2>(144);
|
||||
output.Store(144, (_e87 + _e90));
|
||||
int64_t3 _e96 = input_uniform.val_i64_3_;
|
||||
int64_t3 _e99 = input_storage.Load<int64_t3>(160);
|
||||
output.Store(160, (_e96 + _e99));
|
||||
int64_t4 _e105 = input_uniform.val_i64_4_;
|
||||
int64_t4 _e108 = input_storage.Load<int64_t4>(192);
|
||||
output.Store(192, (_e105 + _e108));
|
||||
int64_t _e114[2] = Constructarray2_int64_t_(input_arrays.Load<int64_t>(16+0), input_arrays.Load<int64_t>(16+8));
|
||||
val = (_e8 + ((31L - 1002003004005006L) + -9223372036854775807L));
|
||||
int64_t _e10 = val;
|
||||
int64_t _e13 = val;
|
||||
val = (_e13 + (_e10 + 5L));
|
||||
uint _e17 = input_uniform.val_u32_;
|
||||
int64_t _e18 = val;
|
||||
int64_t _e22 = val;
|
||||
val = (_e22 + int64_t((_e17 + uint(_e18))));
|
||||
int _e26 = input_uniform.val_i32_;
|
||||
int64_t _e27 = val;
|
||||
int64_t _e31 = val;
|
||||
val = (_e31 + int64_t((_e26 + int(_e27))));
|
||||
float _e35 = input_uniform.val_f32_;
|
||||
int64_t _e36 = val;
|
||||
int64_t _e40 = val;
|
||||
val = (_e40 + int64_t((_e35 + float(_e36))));
|
||||
int64_t _e44 = input_uniform.val_i64_;
|
||||
int64_t _e47 = val;
|
||||
val = (_e47 + (_e44).xxx.z);
|
||||
uint64_t _e51 = input_uniform.val_u64_;
|
||||
int64_t _e53 = val;
|
||||
val = (_e53 + _e51);
|
||||
uint64_t2 _e57 = input_uniform.val_u64_2_;
|
||||
int64_t _e60 = val;
|
||||
val = (_e60 + _e57.y);
|
||||
uint64_t3 _e64 = input_uniform.val_u64_3_;
|
||||
int64_t _e67 = val;
|
||||
val = (_e67 + _e64.z);
|
||||
uint64_t4 _e71 = input_uniform.val_u64_4_;
|
||||
int64_t _e74 = val;
|
||||
val = (_e74 + _e71.w);
|
||||
int64_t _e80 = input_uniform.val_i64_;
|
||||
int64_t _e83 = input_storage.Load<int64_t>(128);
|
||||
output.Store(128, (_e80 + _e83));
|
||||
int64_t2 _e89 = input_uniform.val_i64_2_;
|
||||
int64_t2 _e92 = input_storage.Load<int64_t2>(144);
|
||||
output.Store(144, (_e89 + _e92));
|
||||
int64_t3 _e98 = input_uniform.val_i64_3_;
|
||||
int64_t3 _e101 = input_storage.Load<int64_t3>(160);
|
||||
output.Store(160, (_e98 + _e101));
|
||||
int64_t4 _e107 = input_uniform.val_i64_4_;
|
||||
int64_t4 _e110 = input_storage.Load<int64_t4>(192);
|
||||
output.Store(192, (_e107 + _e110));
|
||||
int64_t _e116[2] = Constructarray2_int64_t_(input_arrays.Load<int64_t>(16+0), input_arrays.Load<int64_t>(16+8));
|
||||
{
|
||||
int64_t _value2[2] = _e114;
|
||||
int64_t _value2[2] = _e116;
|
||||
output_arrays.Store(16+0, _value2[0]);
|
||||
output_arrays.Store(16+8, _value2[1]);
|
||||
}
|
||||
int64_t _e115 = val;
|
||||
int64_t _e117 = val;
|
||||
val = (_e117 + abs(_e115));
|
||||
int64_t _e119 = val;
|
||||
int64_t _e120 = val;
|
||||
val = (_e119 + abs(_e117));
|
||||
int64_t _e121 = val;
|
||||
int64_t _e122 = val;
|
||||
int64_t _e123 = val;
|
||||
val = (_e123 + clamp(_e119, _e120, _e121));
|
||||
int64_t _e125 = val;
|
||||
val = (_e125 + clamp(_e121, _e122, _e123));
|
||||
int64_t _e127 = val;
|
||||
int64_t _e130 = val;
|
||||
val = (_e130 + dot((_e125).xx, (_e127).xx));
|
||||
int64_t _e129 = val;
|
||||
int64_t _e132 = val;
|
||||
int64_t _e133 = val;
|
||||
val = (_e132 + dot((_e127).xx, (_e129).xx));
|
||||
int64_t _e134 = val;
|
||||
int64_t _e135 = val;
|
||||
val = (_e135 + max(_e132, _e133));
|
||||
int64_t _e137 = val;
|
||||
int64_t _e138 = val;
|
||||
val = (_e137 + max(_e134, _e135));
|
||||
int64_t _e139 = val;
|
||||
int64_t _e140 = val;
|
||||
val = (_e140 + min(_e137, _e138));
|
||||
int64_t _e142 = val;
|
||||
val = (_e142 + min(_e139, _e140));
|
||||
int64_t _e144 = val;
|
||||
val = (_e144 + sign(_e142));
|
||||
int64_t _e146 = val;
|
||||
return _e146;
|
||||
val = (_e146 + sign(_e144));
|
||||
int64_t _e148 = val;
|
||||
return _e148;
|
||||
}
|
||||
|
||||
typedef uint64_t ret_Constructarray2_uint64_t_[2];
|
||||
@ -150,78 +150,78 @@ uint64_t uint64_function(uint64_t x_1)
|
||||
{
|
||||
uint64_t val_1 = 20uL;
|
||||
|
||||
uint64_t _e6 = val_1;
|
||||
val_1 = (_e6 + (31uL + 1002003004005006uL));
|
||||
uint64_t _e8 = val_1;
|
||||
uint64_t _e11 = val_1;
|
||||
val_1 = (_e11 + (_e8 + 5uL));
|
||||
uint _e15 = input_uniform.val_u32_;
|
||||
uint64_t _e16 = val_1;
|
||||
uint64_t _e20 = val_1;
|
||||
val_1 = (_e20 + uint64_t((_e15 + uint(_e16))));
|
||||
int _e24 = input_uniform.val_i32_;
|
||||
uint64_t _e25 = val_1;
|
||||
uint64_t _e29 = val_1;
|
||||
val_1 = (_e29 + uint64_t((_e24 + int(_e25))));
|
||||
float _e33 = input_uniform.val_f32_;
|
||||
uint64_t _e34 = val_1;
|
||||
uint64_t _e38 = val_1;
|
||||
val_1 = (_e38 + uint64_t((_e33 + float(_e34))));
|
||||
uint64_t _e42 = input_uniform.val_u64_;
|
||||
uint64_t _e45 = val_1;
|
||||
val_1 = (_e45 + (_e42).xxx.z);
|
||||
int64_t _e49 = input_uniform.val_i64_;
|
||||
uint64_t _e51 = val_1;
|
||||
val_1 = (_e51 + _e49);
|
||||
int64_t2 _e55 = input_uniform.val_i64_2_;
|
||||
uint64_t _e58 = val_1;
|
||||
val_1 = (_e58 + _e55.y);
|
||||
int64_t3 _e62 = input_uniform.val_i64_3_;
|
||||
uint64_t _e65 = val_1;
|
||||
val_1 = (_e65 + _e62.z);
|
||||
int64_t4 _e69 = input_uniform.val_i64_4_;
|
||||
uint64_t _e72 = val_1;
|
||||
val_1 = (_e72 + _e69.w);
|
||||
uint64_t _e78 = input_uniform.val_u64_;
|
||||
uint64_t _e81 = input_storage.Load<uint64_t>(16);
|
||||
output.Store(16, (_e78 + _e81));
|
||||
uint64_t2 _e87 = input_uniform.val_u64_2_;
|
||||
uint64_t2 _e90 = input_storage.Load<uint64_t2>(32);
|
||||
output.Store(32, (_e87 + _e90));
|
||||
uint64_t3 _e96 = input_uniform.val_u64_3_;
|
||||
uint64_t3 _e99 = input_storage.Load<uint64_t3>(64);
|
||||
output.Store(64, (_e96 + _e99));
|
||||
uint64_t4 _e105 = input_uniform.val_u64_4_;
|
||||
uint64_t4 _e108 = input_storage.Load<uint64_t4>(96);
|
||||
output.Store(96, (_e105 + _e108));
|
||||
uint64_t _e114[2] = Constructarray2_uint64_t_(input_arrays.Load<uint64_t>(0+0), input_arrays.Load<uint64_t>(0+8));
|
||||
val_1 = (_e8 + ((31uL + 18446744073709551615uL) - 18446744073709551615uL));
|
||||
uint64_t _e10 = val_1;
|
||||
uint64_t _e13 = val_1;
|
||||
val_1 = (_e13 + (_e10 + 5uL));
|
||||
uint _e17 = input_uniform.val_u32_;
|
||||
uint64_t _e18 = val_1;
|
||||
uint64_t _e22 = val_1;
|
||||
val_1 = (_e22 + uint64_t((_e17 + uint(_e18))));
|
||||
int _e26 = input_uniform.val_i32_;
|
||||
uint64_t _e27 = val_1;
|
||||
uint64_t _e31 = val_1;
|
||||
val_1 = (_e31 + uint64_t((_e26 + int(_e27))));
|
||||
float _e35 = input_uniform.val_f32_;
|
||||
uint64_t _e36 = val_1;
|
||||
uint64_t _e40 = val_1;
|
||||
val_1 = (_e40 + uint64_t((_e35 + float(_e36))));
|
||||
uint64_t _e44 = input_uniform.val_u64_;
|
||||
uint64_t _e47 = val_1;
|
||||
val_1 = (_e47 + (_e44).xxx.z);
|
||||
int64_t _e51 = input_uniform.val_i64_;
|
||||
uint64_t _e53 = val_1;
|
||||
val_1 = (_e53 + _e51);
|
||||
int64_t2 _e57 = input_uniform.val_i64_2_;
|
||||
uint64_t _e60 = val_1;
|
||||
val_1 = (_e60 + _e57.y);
|
||||
int64_t3 _e64 = input_uniform.val_i64_3_;
|
||||
uint64_t _e67 = val_1;
|
||||
val_1 = (_e67 + _e64.z);
|
||||
int64_t4 _e71 = input_uniform.val_i64_4_;
|
||||
uint64_t _e74 = val_1;
|
||||
val_1 = (_e74 + _e71.w);
|
||||
uint64_t _e80 = input_uniform.val_u64_;
|
||||
uint64_t _e83 = input_storage.Load<uint64_t>(16);
|
||||
output.Store(16, (_e80 + _e83));
|
||||
uint64_t2 _e89 = input_uniform.val_u64_2_;
|
||||
uint64_t2 _e92 = input_storage.Load<uint64_t2>(32);
|
||||
output.Store(32, (_e89 + _e92));
|
||||
uint64_t3 _e98 = input_uniform.val_u64_3_;
|
||||
uint64_t3 _e101 = input_storage.Load<uint64_t3>(64);
|
||||
output.Store(64, (_e98 + _e101));
|
||||
uint64_t4 _e107 = input_uniform.val_u64_4_;
|
||||
uint64_t4 _e110 = input_storage.Load<uint64_t4>(96);
|
||||
output.Store(96, (_e107 + _e110));
|
||||
uint64_t _e116[2] = Constructarray2_uint64_t_(input_arrays.Load<uint64_t>(0+0), input_arrays.Load<uint64_t>(0+8));
|
||||
{
|
||||
uint64_t _value2[2] = _e114;
|
||||
uint64_t _value2[2] = _e116;
|
||||
output_arrays.Store(0+0, _value2[0]);
|
||||
output_arrays.Store(0+8, _value2[1]);
|
||||
}
|
||||
uint64_t _e115 = val_1;
|
||||
uint64_t _e117 = val_1;
|
||||
val_1 = (_e117 + abs(_e115));
|
||||
uint64_t _e119 = val_1;
|
||||
uint64_t _e120 = val_1;
|
||||
val_1 = (_e119 + abs(_e117));
|
||||
uint64_t _e121 = val_1;
|
||||
uint64_t _e122 = val_1;
|
||||
uint64_t _e123 = val_1;
|
||||
val_1 = (_e123 + clamp(_e119, _e120, _e121));
|
||||
uint64_t _e125 = val_1;
|
||||
val_1 = (_e125 + clamp(_e121, _e122, _e123));
|
||||
uint64_t _e127 = val_1;
|
||||
uint64_t _e130 = val_1;
|
||||
val_1 = (_e130 + dot((_e125).xx, (_e127).xx));
|
||||
uint64_t _e129 = val_1;
|
||||
uint64_t _e132 = val_1;
|
||||
uint64_t _e133 = val_1;
|
||||
val_1 = (_e132 + dot((_e127).xx, (_e129).xx));
|
||||
uint64_t _e134 = val_1;
|
||||
uint64_t _e135 = val_1;
|
||||
val_1 = (_e135 + max(_e132, _e133));
|
||||
uint64_t _e137 = val_1;
|
||||
uint64_t _e138 = val_1;
|
||||
val_1 = (_e137 + max(_e134, _e135));
|
||||
uint64_t _e139 = val_1;
|
||||
uint64_t _e140 = val_1;
|
||||
val_1 = (_e140 + min(_e137, _e138));
|
||||
uint64_t _e142 = val_1;
|
||||
return _e142;
|
||||
val_1 = (_e142 + min(_e139, _e140));
|
||||
uint64_t _e144 = val_1;
|
||||
return _e144;
|
||||
}
|
||||
|
||||
[numthreads(1, 1, 1)]
|
||||
|
||||
@ -43,79 +43,79 @@ long int64_function(
|
||||
device StorageCompatible& output_arrays
|
||||
) {
|
||||
long val = 20L;
|
||||
long _e6 = val;
|
||||
val = _e6 + (31L - 1002003004005006L);
|
||||
long _e8 = val;
|
||||
long _e11 = val;
|
||||
val = _e11 + (_e8 + 5L);
|
||||
uint _e15 = input_uniform.val_u32_;
|
||||
long _e16 = val;
|
||||
long _e20 = val;
|
||||
val = _e20 + static_cast<long>(_e15 + static_cast<uint>(_e16));
|
||||
int _e24 = input_uniform.val_i32_;
|
||||
long _e25 = val;
|
||||
long _e29 = val;
|
||||
val = _e29 + static_cast<long>(_e24 + static_cast<int>(_e25));
|
||||
float _e33 = input_uniform.val_f32_;
|
||||
long _e34 = val;
|
||||
long _e38 = val;
|
||||
val = _e38 + static_cast<long>(_e33 + static_cast<float>(_e34));
|
||||
long _e42 = input_uniform.val_i64_;
|
||||
long _e45 = val;
|
||||
val = _e45 + metal::long3(_e42).z;
|
||||
ulong _e49 = input_uniform.val_u64_;
|
||||
long _e51 = val;
|
||||
val = _e51 + as_type<long>(_e49);
|
||||
metal::ulong2 _e55 = input_uniform.val_u64_2_;
|
||||
long _e58 = val;
|
||||
val = _e58 + as_type<metal::long2>(_e55).y;
|
||||
metal::ulong3 _e62 = input_uniform.val_u64_3_;
|
||||
long _e65 = val;
|
||||
val = _e65 + as_type<metal::long3>(_e62).z;
|
||||
metal::ulong4 _e69 = input_uniform.val_u64_4_;
|
||||
long _e72 = val;
|
||||
val = _e72 + as_type<metal::long4>(_e69).w;
|
||||
long _e78 = input_uniform.val_i64_;
|
||||
long _e81 = input_storage.val_i64_;
|
||||
output.val_i64_ = _e78 + _e81;
|
||||
metal::long2 _e87 = input_uniform.val_i64_2_;
|
||||
metal::long2 _e90 = input_storage.val_i64_2_;
|
||||
output.val_i64_2_ = _e87 + _e90;
|
||||
metal::long3 _e96 = input_uniform.val_i64_3_;
|
||||
metal::long3 _e99 = input_storage.val_i64_3_;
|
||||
output.val_i64_3_ = _e96 + _e99;
|
||||
metal::long4 _e105 = input_uniform.val_i64_4_;
|
||||
metal::long4 _e108 = input_storage.val_i64_4_;
|
||||
output.val_i64_4_ = _e105 + _e108;
|
||||
type_12 _e114 = input_arrays.val_i64_array_2_;
|
||||
output_arrays.val_i64_array_2_ = _e114;
|
||||
long _e115 = val;
|
||||
val = _e8 + ((31L - 1002003004005006L) + -9223372036854775807L);
|
||||
long _e10 = val;
|
||||
long _e13 = val;
|
||||
val = _e13 + (_e10 + 5L);
|
||||
uint _e17 = input_uniform.val_u32_;
|
||||
long _e18 = val;
|
||||
long _e22 = val;
|
||||
val = _e22 + static_cast<long>(_e17 + static_cast<uint>(_e18));
|
||||
int _e26 = input_uniform.val_i32_;
|
||||
long _e27 = val;
|
||||
long _e31 = val;
|
||||
val = _e31 + static_cast<long>(_e26 + static_cast<int>(_e27));
|
||||
float _e35 = input_uniform.val_f32_;
|
||||
long _e36 = val;
|
||||
long _e40 = val;
|
||||
val = _e40 + static_cast<long>(_e35 + static_cast<float>(_e36));
|
||||
long _e44 = input_uniform.val_i64_;
|
||||
long _e47 = val;
|
||||
val = _e47 + metal::long3(_e44).z;
|
||||
ulong _e51 = input_uniform.val_u64_;
|
||||
long _e53 = val;
|
||||
val = _e53 + as_type<long>(_e51);
|
||||
metal::ulong2 _e57 = input_uniform.val_u64_2_;
|
||||
long _e60 = val;
|
||||
val = _e60 + as_type<metal::long2>(_e57).y;
|
||||
metal::ulong3 _e64 = input_uniform.val_u64_3_;
|
||||
long _e67 = val;
|
||||
val = _e67 + as_type<metal::long3>(_e64).z;
|
||||
metal::ulong4 _e71 = input_uniform.val_u64_4_;
|
||||
long _e74 = val;
|
||||
val = _e74 + as_type<metal::long4>(_e71).w;
|
||||
long _e80 = input_uniform.val_i64_;
|
||||
long _e83 = input_storage.val_i64_;
|
||||
output.val_i64_ = _e80 + _e83;
|
||||
metal::long2 _e89 = input_uniform.val_i64_2_;
|
||||
metal::long2 _e92 = input_storage.val_i64_2_;
|
||||
output.val_i64_2_ = _e89 + _e92;
|
||||
metal::long3 _e98 = input_uniform.val_i64_3_;
|
||||
metal::long3 _e101 = input_storage.val_i64_3_;
|
||||
output.val_i64_3_ = _e98 + _e101;
|
||||
metal::long4 _e107 = input_uniform.val_i64_4_;
|
||||
metal::long4 _e110 = input_storage.val_i64_4_;
|
||||
output.val_i64_4_ = _e107 + _e110;
|
||||
type_12 _e116 = input_arrays.val_i64_array_2_;
|
||||
output_arrays.val_i64_array_2_ = _e116;
|
||||
long _e117 = val;
|
||||
val = _e117 + metal::abs(_e115);
|
||||
long _e119 = val;
|
||||
long _e120 = val;
|
||||
val = _e119 + metal::abs(_e117);
|
||||
long _e121 = val;
|
||||
long _e122 = val;
|
||||
long _e123 = val;
|
||||
val = _e123 + metal::clamp(_e119, _e120, _e121);
|
||||
long _e125 = val;
|
||||
metal::long2 _e126 = metal::long2(_e125);
|
||||
val = _e125 + metal::clamp(_e121, _e122, _e123);
|
||||
long _e127 = val;
|
||||
metal::long2 _e128 = metal::long2(_e127);
|
||||
long _e130 = val;
|
||||
val = _e130 + ( + _e126.x * _e128.x + _e126.y * _e128.y);
|
||||
long _e129 = val;
|
||||
metal::long2 _e130 = metal::long2(_e129);
|
||||
long _e132 = val;
|
||||
long _e133 = val;
|
||||
val = _e132 + ( + _e128.x * _e130.x + _e128.y * _e130.y);
|
||||
long _e134 = val;
|
||||
long _e135 = val;
|
||||
val = _e135 + metal::max(_e132, _e133);
|
||||
long _e137 = val;
|
||||
long _e138 = val;
|
||||
val = _e137 + metal::max(_e134, _e135);
|
||||
long _e139 = val;
|
||||
long _e140 = val;
|
||||
val = _e140 + metal::min(_e137, _e138);
|
||||
long _e142 = val;
|
||||
val = _e142 + metal::min(_e139, _e140);
|
||||
long _e144 = val;
|
||||
val = _e144 + metal::select(metal::select(-1, 1, (_e142 > 0)), 0, (_e142 == 0));
|
||||
long _e146 = val;
|
||||
return _e146;
|
||||
val = _e146 + metal::select(metal::select(-1, 1, (_e144 > 0)), 0, (_e144 == 0));
|
||||
long _e148 = val;
|
||||
return _e148;
|
||||
}
|
||||
|
||||
ulong uint64_function(
|
||||
@ -127,76 +127,76 @@ ulong uint64_function(
|
||||
device StorageCompatible& output_arrays
|
||||
) {
|
||||
ulong val_1 = 20uL;
|
||||
ulong _e6 = val_1;
|
||||
val_1 = _e6 + (31uL + 1002003004005006uL);
|
||||
ulong _e8 = val_1;
|
||||
ulong _e11 = val_1;
|
||||
val_1 = _e11 + (_e8 + 5uL);
|
||||
uint _e15 = input_uniform.val_u32_;
|
||||
ulong _e16 = val_1;
|
||||
ulong _e20 = val_1;
|
||||
val_1 = _e20 + static_cast<ulong>(_e15 + static_cast<uint>(_e16));
|
||||
int _e24 = input_uniform.val_i32_;
|
||||
ulong _e25 = val_1;
|
||||
ulong _e29 = val_1;
|
||||
val_1 = _e29 + static_cast<ulong>(_e24 + static_cast<int>(_e25));
|
||||
float _e33 = input_uniform.val_f32_;
|
||||
ulong _e34 = val_1;
|
||||
ulong _e38 = val_1;
|
||||
val_1 = _e38 + static_cast<ulong>(_e33 + static_cast<float>(_e34));
|
||||
ulong _e42 = input_uniform.val_u64_;
|
||||
ulong _e45 = val_1;
|
||||
val_1 = _e45 + metal::ulong3(_e42).z;
|
||||
long _e49 = input_uniform.val_i64_;
|
||||
ulong _e51 = val_1;
|
||||
val_1 = _e51 + as_type<ulong>(_e49);
|
||||
metal::long2 _e55 = input_uniform.val_i64_2_;
|
||||
ulong _e58 = val_1;
|
||||
val_1 = _e58 + as_type<metal::ulong2>(_e55).y;
|
||||
metal::long3 _e62 = input_uniform.val_i64_3_;
|
||||
ulong _e65 = val_1;
|
||||
val_1 = _e65 + as_type<metal::ulong3>(_e62).z;
|
||||
metal::long4 _e69 = input_uniform.val_i64_4_;
|
||||
ulong _e72 = val_1;
|
||||
val_1 = _e72 + as_type<metal::ulong4>(_e69).w;
|
||||
ulong _e78 = input_uniform.val_u64_;
|
||||
ulong _e81 = input_storage.val_u64_;
|
||||
output.val_u64_ = _e78 + _e81;
|
||||
metal::ulong2 _e87 = input_uniform.val_u64_2_;
|
||||
metal::ulong2 _e90 = input_storage.val_u64_2_;
|
||||
output.val_u64_2_ = _e87 + _e90;
|
||||
metal::ulong3 _e96 = input_uniform.val_u64_3_;
|
||||
metal::ulong3 _e99 = input_storage.val_u64_3_;
|
||||
output.val_u64_3_ = _e96 + _e99;
|
||||
metal::ulong4 _e105 = input_uniform.val_u64_4_;
|
||||
metal::ulong4 _e108 = input_storage.val_u64_4_;
|
||||
output.val_u64_4_ = _e105 + _e108;
|
||||
type_11 _e114 = input_arrays.val_u64_array_2_;
|
||||
output_arrays.val_u64_array_2_ = _e114;
|
||||
ulong _e115 = val_1;
|
||||
val_1 = _e8 + ((31uL + 18446744073709551615uL) - 18446744073709551615uL);
|
||||
ulong _e10 = val_1;
|
||||
ulong _e13 = val_1;
|
||||
val_1 = _e13 + (_e10 + 5uL);
|
||||
uint _e17 = input_uniform.val_u32_;
|
||||
ulong _e18 = val_1;
|
||||
ulong _e22 = val_1;
|
||||
val_1 = _e22 + static_cast<ulong>(_e17 + static_cast<uint>(_e18));
|
||||
int _e26 = input_uniform.val_i32_;
|
||||
ulong _e27 = val_1;
|
||||
ulong _e31 = val_1;
|
||||
val_1 = _e31 + static_cast<ulong>(_e26 + static_cast<int>(_e27));
|
||||
float _e35 = input_uniform.val_f32_;
|
||||
ulong _e36 = val_1;
|
||||
ulong _e40 = val_1;
|
||||
val_1 = _e40 + static_cast<ulong>(_e35 + static_cast<float>(_e36));
|
||||
ulong _e44 = input_uniform.val_u64_;
|
||||
ulong _e47 = val_1;
|
||||
val_1 = _e47 + metal::ulong3(_e44).z;
|
||||
long _e51 = input_uniform.val_i64_;
|
||||
ulong _e53 = val_1;
|
||||
val_1 = _e53 + as_type<ulong>(_e51);
|
||||
metal::long2 _e57 = input_uniform.val_i64_2_;
|
||||
ulong _e60 = val_1;
|
||||
val_1 = _e60 + as_type<metal::ulong2>(_e57).y;
|
||||
metal::long3 _e64 = input_uniform.val_i64_3_;
|
||||
ulong _e67 = val_1;
|
||||
val_1 = _e67 + as_type<metal::ulong3>(_e64).z;
|
||||
metal::long4 _e71 = input_uniform.val_i64_4_;
|
||||
ulong _e74 = val_1;
|
||||
val_1 = _e74 + as_type<metal::ulong4>(_e71).w;
|
||||
ulong _e80 = input_uniform.val_u64_;
|
||||
ulong _e83 = input_storage.val_u64_;
|
||||
output.val_u64_ = _e80 + _e83;
|
||||
metal::ulong2 _e89 = input_uniform.val_u64_2_;
|
||||
metal::ulong2 _e92 = input_storage.val_u64_2_;
|
||||
output.val_u64_2_ = _e89 + _e92;
|
||||
metal::ulong3 _e98 = input_uniform.val_u64_3_;
|
||||
metal::ulong3 _e101 = input_storage.val_u64_3_;
|
||||
output.val_u64_3_ = _e98 + _e101;
|
||||
metal::ulong4 _e107 = input_uniform.val_u64_4_;
|
||||
metal::ulong4 _e110 = input_storage.val_u64_4_;
|
||||
output.val_u64_4_ = _e107 + _e110;
|
||||
type_11 _e116 = input_arrays.val_u64_array_2_;
|
||||
output_arrays.val_u64_array_2_ = _e116;
|
||||
ulong _e117 = val_1;
|
||||
val_1 = _e117 + metal::abs(_e115);
|
||||
ulong _e119 = val_1;
|
||||
ulong _e120 = val_1;
|
||||
val_1 = _e119 + metal::abs(_e117);
|
||||
ulong _e121 = val_1;
|
||||
ulong _e122 = val_1;
|
||||
ulong _e123 = val_1;
|
||||
val_1 = _e123 + metal::clamp(_e119, _e120, _e121);
|
||||
ulong _e125 = val_1;
|
||||
metal::ulong2 _e126 = metal::ulong2(_e125);
|
||||
val_1 = _e125 + metal::clamp(_e121, _e122, _e123);
|
||||
ulong _e127 = val_1;
|
||||
metal::ulong2 _e128 = metal::ulong2(_e127);
|
||||
ulong _e130 = val_1;
|
||||
val_1 = _e130 + ( + _e126.x * _e128.x + _e126.y * _e128.y);
|
||||
ulong _e129 = val_1;
|
||||
metal::ulong2 _e130 = metal::ulong2(_e129);
|
||||
ulong _e132 = val_1;
|
||||
ulong _e133 = val_1;
|
||||
val_1 = _e132 + ( + _e128.x * _e130.x + _e128.y * _e130.y);
|
||||
ulong _e134 = val_1;
|
||||
ulong _e135 = val_1;
|
||||
val_1 = _e135 + metal::max(_e132, _e133);
|
||||
ulong _e137 = val_1;
|
||||
ulong _e138 = val_1;
|
||||
val_1 = _e137 + metal::max(_e134, _e135);
|
||||
ulong _e139 = val_1;
|
||||
ulong _e140 = val_1;
|
||||
val_1 = _e140 + metal::min(_e137, _e138);
|
||||
ulong _e142 = val_1;
|
||||
return _e142;
|
||||
val_1 = _e142 + metal::min(_e139, _e140);
|
||||
ulong _e144 = val_1;
|
||||
return _e144;
|
||||
}
|
||||
|
||||
kernel void main_(
|
||||
|
||||
@ -1,14 +1,14 @@
|
||||
; SPIR-V
|
||||
; Version: 1.0
|
||||
; Generator: rspirv
|
||||
; Bound: 372
|
||||
; Bound: 375
|
||||
OpCapability Shader
|
||||
OpCapability Int64
|
||||
OpExtension "SPV_KHR_storage_buffer_storage_class"
|
||||
%1 = OpExtInstImport "GLSL.std.450"
|
||||
OpMemoryModel Logical GLSL450
|
||||
OpEntryPoint GLCompute %356 "main"
|
||||
OpExecutionMode %356 LocalSize 1 1 1
|
||||
OpEntryPoint GLCompute %359 "main"
|
||||
OpExecutionMode %359 LocalSize 1 1 1
|
||||
OpMemberDecorate %14 0 Offset 0
|
||||
OpMemberDecorate %14 1 Offset 4
|
||||
OpMemberDecorate %14 2 Offset 8
|
||||
@ -91,380 +91,383 @@ OpMemberDecorate %36 0 Offset 0
|
||||
%51 = OpConstant %3 20
|
||||
%52 = OpConstant %3 31
|
||||
%53 = OpConstant %3 1002003004005006
|
||||
%54 = OpConstant %3 5
|
||||
%56 = OpTypePointer Function %3
|
||||
%65 = OpTypePointer Uniform %5
|
||||
%74 = OpTypePointer Uniform %6
|
||||
%75 = OpConstant %5 1
|
||||
%84 = OpTypePointer Uniform %7
|
||||
%93 = OpTypePointer Uniform %3
|
||||
%94 = OpConstant %5 7
|
||||
%101 = OpTypePointer Uniform %4
|
||||
%102 = OpConstant %5 3
|
||||
%108 = OpTypePointer Uniform %8
|
||||
%109 = OpConstant %5 4
|
||||
%116 = OpTypePointer Uniform %9
|
||||
%117 = OpConstant %5 5
|
||||
%124 = OpTypePointer Uniform %10
|
||||
%125 = OpConstant %5 6
|
||||
%132 = OpTypePointer StorageBuffer %3
|
||||
%139 = OpTypePointer StorageBuffer %11
|
||||
%140 = OpTypePointer Uniform %11
|
||||
%141 = OpConstant %5 8
|
||||
%148 = OpTypePointer StorageBuffer %12
|
||||
%149 = OpTypePointer Uniform %12
|
||||
%150 = OpConstant %5 9
|
||||
%157 = OpTypePointer StorageBuffer %13
|
||||
%158 = OpTypePointer Uniform %13
|
||||
%159 = OpConstant %5 10
|
||||
%166 = OpTypePointer StorageBuffer %17
|
||||
%186 = OpConstantNull %3
|
||||
%214 = OpTypeFunction %4 %4
|
||||
%220 = OpConstant %4 31
|
||||
%221 = OpConstant %4 1002003004005006
|
||||
%222 = OpConstant %4 5
|
||||
%224 = OpTypePointer Function %4
|
||||
%286 = OpTypePointer StorageBuffer %4
|
||||
%293 = OpTypePointer StorageBuffer %8
|
||||
%300 = OpTypePointer StorageBuffer %9
|
||||
%307 = OpTypePointer StorageBuffer %10
|
||||
%314 = OpTypePointer StorageBuffer %15
|
||||
%334 = OpConstantNull %4
|
||||
%357 = OpTypeFunction %2
|
||||
%363 = OpConstant %4 67
|
||||
%364 = OpConstant %3 60
|
||||
%370 = OpConstant %5 11
|
||||
%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
|
||||
%95 = OpTypePointer Uniform %3
|
||||
%96 = OpConstant %5 7
|
||||
%103 = OpTypePointer Uniform %4
|
||||
%104 = OpConstant %5 3
|
||||
%110 = OpTypePointer Uniform %8
|
||||
%111 = OpConstant %5 4
|
||||
%118 = OpTypePointer Uniform %9
|
||||
%119 = OpConstant %5 5
|
||||
%126 = OpTypePointer Uniform %10
|
||||
%127 = OpConstant %5 6
|
||||
%134 = OpTypePointer StorageBuffer %3
|
||||
%141 = OpTypePointer StorageBuffer %11
|
||||
%142 = OpTypePointer Uniform %11
|
||||
%143 = OpConstant %5 8
|
||||
%150 = OpTypePointer StorageBuffer %12
|
||||
%151 = OpTypePointer Uniform %12
|
||||
%152 = OpConstant %5 9
|
||||
%159 = OpTypePointer StorageBuffer %13
|
||||
%160 = OpTypePointer Uniform %13
|
||||
%161 = OpConstant %5 10
|
||||
%168 = OpTypePointer StorageBuffer %17
|
||||
%188 = OpConstantNull %3
|
||||
%216 = OpTypeFunction %4 %4
|
||||
%222 = OpConstant %4 31
|
||||
%223 = OpConstant %4 18446744073709551615
|
||||
%224 = OpConstant %4 5
|
||||
%226 = OpTypePointer Function %4
|
||||
%289 = OpTypePointer StorageBuffer %4
|
||||
%296 = OpTypePointer StorageBuffer %8
|
||||
%303 = OpTypePointer StorageBuffer %9
|
||||
%310 = OpTypePointer StorageBuffer %10
|
||||
%317 = OpTypePointer StorageBuffer %15
|
||||
%337 = OpConstantNull %4
|
||||
%360 = OpTypeFunction %2
|
||||
%366 = OpConstant %4 67
|
||||
%367 = OpConstant %3 60
|
||||
%373 = OpConstant %5 11
|
||||
%40 = OpFunction %3 None %41
|
||||
%39 = OpFunctionParameter %3
|
||||
%38 = OpLabel
|
||||
%55 = OpVariable %56 Function %51
|
||||
%56 = OpVariable %57 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 %57
|
||||
%57 = OpLabel
|
||||
%58 = OpISub %3 %52 %53
|
||||
%59 = OpLoad %3 %55
|
||||
%60 = OpIAdd %3 %59 %58
|
||||
OpStore %55 %60
|
||||
%61 = OpLoad %3 %55
|
||||
%62 = OpIAdd %3 %61 %54
|
||||
%63 = OpLoad %3 %55
|
||||
%64 = OpIAdd %3 %63 %62
|
||||
OpStore %55 %64
|
||||
%66 = OpAccessChain %65 %44 %43
|
||||
%67 = OpLoad %5 %66
|
||||
%68 = OpLoad %3 %55
|
||||
%69 = OpUConvert %5 %68
|
||||
%70 = OpIAdd %5 %67 %69
|
||||
%71 = OpSConvert %3 %70
|
||||
%72 = OpLoad %3 %55
|
||||
%73 = OpIAdd %3 %72 %71
|
||||
OpStore %55 %73
|
||||
%76 = OpAccessChain %74 %44 %75
|
||||
%77 = OpLoad %6 %76
|
||||
%78 = OpLoad %3 %55
|
||||
%79 = OpSConvert %6 %78
|
||||
%80 = OpIAdd %6 %77 %79
|
||||
%81 = OpSConvert %3 %80
|
||||
%82 = OpLoad %3 %55
|
||||
%83 = OpIAdd %3 %82 %81
|
||||
OpStore %55 %83
|
||||
%85 = OpAccessChain %84 %44 %16
|
||||
%86 = OpLoad %7 %85
|
||||
%87 = OpLoad %3 %55
|
||||
%88 = OpConvertSToF %7 %87
|
||||
%89 = OpFAdd %7 %86 %88
|
||||
%90 = OpConvertFToS %3 %89
|
||||
%91 = OpLoad %3 %55
|
||||
%92 = OpIAdd %3 %91 %90
|
||||
OpStore %55 %92
|
||||
%95 = OpAccessChain %93 %44 %94
|
||||
%96 = OpLoad %3 %95
|
||||
%97 = OpCompositeConstruct %12 %96 %96 %96
|
||||
%98 = OpCompositeExtract %3 %97 2
|
||||
%99 = OpLoad %3 %55
|
||||
%100 = OpIAdd %3 %99 %98
|
||||
OpStore %55 %100
|
||||
%103 = OpAccessChain %101 %44 %102
|
||||
%104 = OpLoad %4 %103
|
||||
%105 = OpBitcast %3 %104
|
||||
%106 = OpLoad %3 %55
|
||||
%107 = OpIAdd %3 %106 %105
|
||||
OpStore %55 %107
|
||||
%110 = OpAccessChain %108 %44 %109
|
||||
%111 = OpLoad %8 %110
|
||||
%112 = OpBitcast %11 %111
|
||||
%113 = OpCompositeExtract %3 %112 1
|
||||
%114 = OpLoad %3 %55
|
||||
%115 = OpIAdd %3 %114 %113
|
||||
OpStore %55 %115
|
||||
%118 = OpAccessChain %116 %44 %117
|
||||
%119 = OpLoad %9 %118
|
||||
%120 = OpBitcast %12 %119
|
||||
%121 = OpCompositeExtract %3 %120 2
|
||||
%122 = OpLoad %3 %55
|
||||
%123 = OpIAdd %3 %122 %121
|
||||
OpStore %55 %123
|
||||
%126 = OpAccessChain %124 %44 %125
|
||||
%127 = OpLoad %10 %126
|
||||
%128 = OpBitcast %13 %127
|
||||
%129 = OpCompositeExtract %3 %128 3
|
||||
%130 = OpLoad %3 %55
|
||||
%131 = OpIAdd %3 %130 %129
|
||||
OpStore %55 %131
|
||||
%133 = OpAccessChain %93 %44 %94
|
||||
%134 = OpLoad %3 %133
|
||||
%135 = OpAccessChain %132 %46 %94
|
||||
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
|
||||
%92 = OpConvertFToS %3 %91
|
||||
%93 = OpLoad %3 %56
|
||||
%94 = OpIAdd %3 %93 %92
|
||||
OpStore %56 %94
|
||||
%97 = OpAccessChain %95 %44 %96
|
||||
%98 = OpLoad %3 %97
|
||||
%99 = OpCompositeConstruct %12 %98 %98 %98
|
||||
%100 = OpCompositeExtract %3 %99 2
|
||||
%101 = OpLoad %3 %56
|
||||
%102 = OpIAdd %3 %101 %100
|
||||
OpStore %56 %102
|
||||
%105 = OpAccessChain %103 %44 %104
|
||||
%106 = OpLoad %4 %105
|
||||
%107 = OpBitcast %3 %106
|
||||
%108 = OpLoad %3 %56
|
||||
%109 = OpIAdd %3 %108 %107
|
||||
OpStore %56 %109
|
||||
%112 = OpAccessChain %110 %44 %111
|
||||
%113 = OpLoad %8 %112
|
||||
%114 = OpBitcast %11 %113
|
||||
%115 = OpCompositeExtract %3 %114 1
|
||||
%116 = OpLoad %3 %56
|
||||
%117 = OpIAdd %3 %116 %115
|
||||
OpStore %56 %117
|
||||
%120 = OpAccessChain %118 %44 %119
|
||||
%121 = OpLoad %9 %120
|
||||
%122 = OpBitcast %12 %121
|
||||
%123 = OpCompositeExtract %3 %122 2
|
||||
%124 = OpLoad %3 %56
|
||||
%125 = OpIAdd %3 %124 %123
|
||||
OpStore %56 %125
|
||||
%128 = OpAccessChain %126 %44 %127
|
||||
%129 = OpLoad %10 %128
|
||||
%130 = OpBitcast %13 %129
|
||||
%131 = OpCompositeExtract %3 %130 3
|
||||
%132 = OpLoad %3 %56
|
||||
%133 = OpIAdd %3 %132 %131
|
||||
OpStore %56 %133
|
||||
%135 = OpAccessChain %95 %44 %96
|
||||
%136 = OpLoad %3 %135
|
||||
%137 = OpIAdd %3 %134 %136
|
||||
%138 = OpAccessChain %132 %49 %94
|
||||
OpStore %138 %137
|
||||
%142 = OpAccessChain %140 %44 %141
|
||||
%143 = OpLoad %11 %142
|
||||
%144 = OpAccessChain %139 %46 %141
|
||||
%137 = OpAccessChain %134 %46 %96
|
||||
%138 = OpLoad %3 %137
|
||||
%139 = OpIAdd %3 %136 %138
|
||||
%140 = OpAccessChain %134 %49 %96
|
||||
OpStore %140 %139
|
||||
%144 = OpAccessChain %142 %44 %143
|
||||
%145 = OpLoad %11 %144
|
||||
%146 = OpIAdd %11 %143 %145
|
||||
%147 = OpAccessChain %139 %49 %141
|
||||
OpStore %147 %146
|
||||
%151 = OpAccessChain %149 %44 %150
|
||||
%152 = OpLoad %12 %151
|
||||
%153 = OpAccessChain %148 %46 %150
|
||||
%146 = OpAccessChain %141 %46 %143
|
||||
%147 = OpLoad %11 %146
|
||||
%148 = OpIAdd %11 %145 %147
|
||||
%149 = OpAccessChain %141 %49 %143
|
||||
OpStore %149 %148
|
||||
%153 = OpAccessChain %151 %44 %152
|
||||
%154 = OpLoad %12 %153
|
||||
%155 = OpIAdd %12 %152 %154
|
||||
%156 = OpAccessChain %148 %49 %150
|
||||
OpStore %156 %155
|
||||
%160 = OpAccessChain %158 %44 %159
|
||||
%161 = OpLoad %13 %160
|
||||
%162 = OpAccessChain %157 %46 %159
|
||||
%155 = OpAccessChain %150 %46 %152
|
||||
%156 = OpLoad %12 %155
|
||||
%157 = OpIAdd %12 %154 %156
|
||||
%158 = OpAccessChain %150 %49 %152
|
||||
OpStore %158 %157
|
||||
%162 = OpAccessChain %160 %44 %161
|
||||
%163 = OpLoad %13 %162
|
||||
%164 = OpIAdd %13 %161 %163
|
||||
%165 = OpAccessChain %157 %49 %159
|
||||
OpStore %165 %164
|
||||
%167 = OpAccessChain %166 %48 %75
|
||||
%168 = OpLoad %17 %167
|
||||
%169 = OpAccessChain %166 %50 %75
|
||||
OpStore %169 %168
|
||||
%170 = OpLoad %3 %55
|
||||
%171 = OpExtInst %3 %1 SAbs %170
|
||||
%172 = OpLoad %3 %55
|
||||
%173 = OpIAdd %3 %172 %171
|
||||
OpStore %55 %173
|
||||
%174 = OpLoad %3 %55
|
||||
%175 = OpLoad %3 %55
|
||||
%176 = OpLoad %3 %55
|
||||
%178 = OpExtInst %3 %1 SMax %174 %175
|
||||
%177 = OpExtInst %3 %1 SMin %178 %176
|
||||
%179 = OpLoad %3 %55
|
||||
%180 = OpIAdd %3 %179 %177
|
||||
OpStore %55 %180
|
||||
%181 = OpLoad %3 %55
|
||||
%182 = OpCompositeConstruct %11 %181 %181
|
||||
%183 = OpLoad %3 %55
|
||||
%164 = OpAccessChain %159 %46 %161
|
||||
%165 = OpLoad %13 %164
|
||||
%166 = OpIAdd %13 %163 %165
|
||||
%167 = OpAccessChain %159 %49 %161
|
||||
OpStore %167 %166
|
||||
%169 = OpAccessChain %168 %48 %77
|
||||
%170 = OpLoad %17 %169
|
||||
%171 = OpAccessChain %168 %50 %77
|
||||
OpStore %171 %170
|
||||
%172 = OpLoad %3 %56
|
||||
%173 = OpExtInst %3 %1 SAbs %172
|
||||
%174 = OpLoad %3 %56
|
||||
%175 = OpIAdd %3 %174 %173
|
||||
OpStore %56 %175
|
||||
%176 = OpLoad %3 %56
|
||||
%177 = OpLoad %3 %56
|
||||
%178 = OpLoad %3 %56
|
||||
%180 = OpExtInst %3 %1 SMax %176 %177
|
||||
%179 = OpExtInst %3 %1 SMin %180 %178
|
||||
%181 = OpLoad %3 %56
|
||||
%182 = OpIAdd %3 %181 %179
|
||||
OpStore %56 %182
|
||||
%183 = OpLoad %3 %56
|
||||
%184 = OpCompositeConstruct %11 %183 %183
|
||||
%187 = OpCompositeExtract %3 %182 0
|
||||
%188 = OpCompositeExtract %3 %184 0
|
||||
%189 = OpIMul %3 %187 %188
|
||||
%190 = OpIAdd %3 %186 %189
|
||||
%191 = OpCompositeExtract %3 %182 1
|
||||
%192 = OpCompositeExtract %3 %184 1
|
||||
%193 = OpIMul %3 %191 %192
|
||||
%185 = OpIAdd %3 %190 %193
|
||||
%194 = OpLoad %3 %55
|
||||
%195 = OpIAdd %3 %194 %185
|
||||
OpStore %55 %195
|
||||
%196 = OpLoad %3 %55
|
||||
%197 = OpLoad %3 %55
|
||||
%198 = OpExtInst %3 %1 SMax %196 %197
|
||||
%199 = OpLoad %3 %55
|
||||
%200 = OpIAdd %3 %199 %198
|
||||
OpStore %55 %200
|
||||
%201 = OpLoad %3 %55
|
||||
%202 = OpLoad %3 %55
|
||||
%203 = OpExtInst %3 %1 SMin %201 %202
|
||||
%204 = OpLoad %3 %55
|
||||
%205 = OpIAdd %3 %204 %203
|
||||
OpStore %55 %205
|
||||
%206 = OpLoad %3 %55
|
||||
%207 = OpExtInst %3 %1 SSign %206
|
||||
%208 = OpLoad %3 %55
|
||||
%209 = OpIAdd %3 %208 %207
|
||||
OpStore %55 %209
|
||||
%210 = OpLoad %3 %55
|
||||
OpReturnValue %210
|
||||
%185 = OpLoad %3 %56
|
||||
%186 = OpCompositeConstruct %11 %185 %185
|
||||
%189 = OpCompositeExtract %3 %184 0
|
||||
%190 = OpCompositeExtract %3 %186 0
|
||||
%191 = OpIMul %3 %189 %190
|
||||
%192 = OpIAdd %3 %188 %191
|
||||
%193 = OpCompositeExtract %3 %184 1
|
||||
%194 = OpCompositeExtract %3 %186 1
|
||||
%195 = OpIMul %3 %193 %194
|
||||
%187 = OpIAdd %3 %192 %195
|
||||
%196 = OpLoad %3 %56
|
||||
%197 = OpIAdd %3 %196 %187
|
||||
OpStore %56 %197
|
||||
%198 = OpLoad %3 %56
|
||||
%199 = OpLoad %3 %56
|
||||
%200 = OpExtInst %3 %1 SMax %198 %199
|
||||
%201 = OpLoad %3 %56
|
||||
%202 = OpIAdd %3 %201 %200
|
||||
OpStore %56 %202
|
||||
%203 = OpLoad %3 %56
|
||||
%204 = OpLoad %3 %56
|
||||
%205 = OpExtInst %3 %1 SMin %203 %204
|
||||
%206 = OpLoad %3 %56
|
||||
%207 = OpIAdd %3 %206 %205
|
||||
OpStore %56 %207
|
||||
%208 = OpLoad %3 %56
|
||||
%209 = OpExtInst %3 %1 SSign %208
|
||||
%210 = OpLoad %3 %56
|
||||
%211 = OpIAdd %3 %210 %209
|
||||
OpStore %56 %211
|
||||
%212 = OpLoad %3 %56
|
||||
OpReturnValue %212
|
||||
OpFunctionEnd
|
||||
%213 = OpFunction %4 None %214
|
||||
%212 = OpFunctionParameter %4
|
||||
%211 = OpLabel
|
||||
%223 = OpVariable %224 Function %20
|
||||
%215 = OpAccessChain %42 %23 %43
|
||||
%216 = OpAccessChain %45 %26 %43
|
||||
%217 = OpAccessChain %47 %29 %43
|
||||
%218 = OpAccessChain %45 %32 %43
|
||||
%219 = OpAccessChain %47 %35 %43
|
||||
OpBranch %225
|
||||
%225 = OpLabel
|
||||
%226 = OpIAdd %4 %220 %221
|
||||
%227 = OpLoad %4 %223
|
||||
%228 = OpIAdd %4 %227 %226
|
||||
OpStore %223 %228
|
||||
%229 = OpLoad %4 %223
|
||||
%230 = OpIAdd %4 %229 %222
|
||||
%231 = OpLoad %4 %223
|
||||
%232 = OpIAdd %4 %231 %230
|
||||
OpStore %223 %232
|
||||
%233 = OpAccessChain %65 %215 %43
|
||||
%234 = OpLoad %5 %233
|
||||
%235 = OpLoad %4 %223
|
||||
%236 = OpUConvert %5 %235
|
||||
%237 = OpIAdd %5 %234 %236
|
||||
%238 = OpUConvert %4 %237
|
||||
%239 = OpLoad %4 %223
|
||||
%240 = OpIAdd %4 %239 %238
|
||||
OpStore %223 %240
|
||||
%241 = OpAccessChain %74 %215 %75
|
||||
%242 = OpLoad %6 %241
|
||||
%243 = OpLoad %4 %223
|
||||
%244 = OpSConvert %6 %243
|
||||
%245 = OpIAdd %6 %242 %244
|
||||
%246 = OpUConvert %4 %245
|
||||
%247 = OpLoad %4 %223
|
||||
%248 = OpIAdd %4 %247 %246
|
||||
OpStore %223 %248
|
||||
%249 = OpAccessChain %84 %215 %16
|
||||
%250 = OpLoad %7 %249
|
||||
%251 = OpLoad %4 %223
|
||||
%252 = OpConvertUToF %7 %251
|
||||
%253 = OpFAdd %7 %250 %252
|
||||
%254 = OpConvertFToU %4 %253
|
||||
%255 = OpLoad %4 %223
|
||||
%256 = OpIAdd %4 %255 %254
|
||||
OpStore %223 %256
|
||||
%257 = OpAccessChain %101 %215 %102
|
||||
%258 = OpLoad %4 %257
|
||||
%259 = OpCompositeConstruct %9 %258 %258 %258
|
||||
%260 = OpCompositeExtract %4 %259 2
|
||||
%261 = OpLoad %4 %223
|
||||
%262 = OpIAdd %4 %261 %260
|
||||
OpStore %223 %262
|
||||
%263 = OpAccessChain %93 %215 %94
|
||||
%264 = OpLoad %3 %263
|
||||
%265 = OpBitcast %4 %264
|
||||
%266 = OpLoad %4 %223
|
||||
%267 = OpIAdd %4 %266 %265
|
||||
OpStore %223 %267
|
||||
%268 = OpAccessChain %140 %215 %141
|
||||
%269 = OpLoad %11 %268
|
||||
%270 = OpBitcast %8 %269
|
||||
%271 = OpCompositeExtract %4 %270 1
|
||||
%272 = OpLoad %4 %223
|
||||
%273 = OpIAdd %4 %272 %271
|
||||
OpStore %223 %273
|
||||
%274 = OpAccessChain %149 %215 %150
|
||||
%275 = OpLoad %12 %274
|
||||
%276 = OpBitcast %9 %275
|
||||
%277 = OpCompositeExtract %4 %276 2
|
||||
%278 = OpLoad %4 %223
|
||||
%279 = OpIAdd %4 %278 %277
|
||||
OpStore %223 %279
|
||||
%280 = OpAccessChain %158 %215 %159
|
||||
%281 = OpLoad %13 %280
|
||||
%282 = OpBitcast %10 %281
|
||||
%283 = OpCompositeExtract %4 %282 3
|
||||
%284 = OpLoad %4 %223
|
||||
%285 = OpIAdd %4 %284 %283
|
||||
OpStore %223 %285
|
||||
%287 = OpAccessChain %101 %215 %102
|
||||
%288 = OpLoad %4 %287
|
||||
%289 = OpAccessChain %286 %216 %102
|
||||
%290 = OpLoad %4 %289
|
||||
%291 = OpIAdd %4 %288 %290
|
||||
%292 = OpAccessChain %286 %218 %102
|
||||
OpStore %292 %291
|
||||
%294 = OpAccessChain %108 %215 %109
|
||||
%295 = OpLoad %8 %294
|
||||
%296 = OpAccessChain %293 %216 %109
|
||||
%297 = OpLoad %8 %296
|
||||
%298 = OpIAdd %8 %295 %297
|
||||
%299 = OpAccessChain %293 %218 %109
|
||||
OpStore %299 %298
|
||||
%301 = OpAccessChain %116 %215 %117
|
||||
%302 = OpLoad %9 %301
|
||||
%303 = OpAccessChain %300 %216 %117
|
||||
%304 = OpLoad %9 %303
|
||||
%305 = OpIAdd %9 %302 %304
|
||||
%306 = OpAccessChain %300 %218 %117
|
||||
OpStore %306 %305
|
||||
%308 = OpAccessChain %124 %215 %125
|
||||
%309 = OpLoad %10 %308
|
||||
%310 = OpAccessChain %307 %216 %125
|
||||
%311 = OpLoad %10 %310
|
||||
%312 = OpIAdd %10 %309 %311
|
||||
%313 = OpAccessChain %307 %218 %125
|
||||
OpStore %313 %312
|
||||
%315 = OpAccessChain %314 %217 %43
|
||||
%316 = OpLoad %15 %315
|
||||
%317 = OpAccessChain %314 %219 %43
|
||||
OpStore %317 %316
|
||||
%318 = OpLoad %4 %223
|
||||
%319 = OpCopyObject %4 %318
|
||||
%320 = OpLoad %4 %223
|
||||
%321 = OpIAdd %4 %320 %319
|
||||
OpStore %223 %321
|
||||
%322 = OpLoad %4 %223
|
||||
%323 = OpLoad %4 %223
|
||||
%324 = OpLoad %4 %223
|
||||
%326 = OpExtInst %4 %1 UMax %322 %323
|
||||
%325 = OpExtInst %4 %1 UMin %326 %324
|
||||
%327 = OpLoad %4 %223
|
||||
%328 = OpIAdd %4 %327 %325
|
||||
OpStore %223 %328
|
||||
%329 = OpLoad %4 %223
|
||||
%330 = OpCompositeConstruct %8 %329 %329
|
||||
%331 = OpLoad %4 %223
|
||||
%332 = OpCompositeConstruct %8 %331 %331
|
||||
%335 = OpCompositeExtract %4 %330 0
|
||||
%336 = OpCompositeExtract %4 %332 0
|
||||
%337 = OpIMul %4 %335 %336
|
||||
%338 = OpIAdd %4 %334 %337
|
||||
%339 = OpCompositeExtract %4 %330 1
|
||||
%340 = OpCompositeExtract %4 %332 1
|
||||
%341 = OpIMul %4 %339 %340
|
||||
%333 = OpIAdd %4 %338 %341
|
||||
%342 = OpLoad %4 %223
|
||||
%343 = OpIAdd %4 %342 %333
|
||||
OpStore %223 %343
|
||||
%344 = OpLoad %4 %223
|
||||
%345 = OpLoad %4 %223
|
||||
%346 = OpExtInst %4 %1 UMax %344 %345
|
||||
%347 = OpLoad %4 %223
|
||||
%348 = OpIAdd %4 %347 %346
|
||||
OpStore %223 %348
|
||||
%349 = OpLoad %4 %223
|
||||
%350 = OpLoad %4 %223
|
||||
%351 = OpExtInst %4 %1 UMin %349 %350
|
||||
%352 = OpLoad %4 %223
|
||||
%353 = OpIAdd %4 %352 %351
|
||||
OpStore %223 %353
|
||||
%354 = OpLoad %4 %223
|
||||
OpReturnValue %354
|
||||
%215 = OpFunction %4 None %216
|
||||
%214 = OpFunctionParameter %4
|
||||
%213 = OpLabel
|
||||
%225 = OpVariable %226 Function %20
|
||||
%217 = OpAccessChain %42 %23 %43
|
||||
%218 = OpAccessChain %45 %26 %43
|
||||
%219 = OpAccessChain %47 %29 %43
|
||||
%220 = OpAccessChain %45 %32 %43
|
||||
%221 = OpAccessChain %47 %35 %43
|
||||
OpBranch %227
|
||||
%227 = OpLabel
|
||||
%228 = OpIAdd %4 %222 %223
|
||||
%229 = OpISub %4 %228 %223
|
||||
%230 = OpLoad %4 %225
|
||||
%231 = OpIAdd %4 %230 %229
|
||||
OpStore %225 %231
|
||||
%232 = OpLoad %4 %225
|
||||
%233 = OpIAdd %4 %232 %224
|
||||
%234 = OpLoad %4 %225
|
||||
%235 = OpIAdd %4 %234 %233
|
||||
OpStore %225 %235
|
||||
%236 = OpAccessChain %67 %217 %43
|
||||
%237 = OpLoad %5 %236
|
||||
%238 = OpLoad %4 %225
|
||||
%239 = OpUConvert %5 %238
|
||||
%240 = OpIAdd %5 %237 %239
|
||||
%241 = OpUConvert %4 %240
|
||||
%242 = OpLoad %4 %225
|
||||
%243 = OpIAdd %4 %242 %241
|
||||
OpStore %225 %243
|
||||
%244 = OpAccessChain %76 %217 %77
|
||||
%245 = OpLoad %6 %244
|
||||
%246 = OpLoad %4 %225
|
||||
%247 = OpSConvert %6 %246
|
||||
%248 = OpIAdd %6 %245 %247
|
||||
%249 = OpUConvert %4 %248
|
||||
%250 = OpLoad %4 %225
|
||||
%251 = OpIAdd %4 %250 %249
|
||||
OpStore %225 %251
|
||||
%252 = OpAccessChain %86 %217 %16
|
||||
%253 = OpLoad %7 %252
|
||||
%254 = OpLoad %4 %225
|
||||
%255 = OpConvertUToF %7 %254
|
||||
%256 = OpFAdd %7 %253 %255
|
||||
%257 = OpConvertFToU %4 %256
|
||||
%258 = OpLoad %4 %225
|
||||
%259 = OpIAdd %4 %258 %257
|
||||
OpStore %225 %259
|
||||
%260 = OpAccessChain %103 %217 %104
|
||||
%261 = OpLoad %4 %260
|
||||
%262 = OpCompositeConstruct %9 %261 %261 %261
|
||||
%263 = OpCompositeExtract %4 %262 2
|
||||
%264 = OpLoad %4 %225
|
||||
%265 = OpIAdd %4 %264 %263
|
||||
OpStore %225 %265
|
||||
%266 = OpAccessChain %95 %217 %96
|
||||
%267 = OpLoad %3 %266
|
||||
%268 = OpBitcast %4 %267
|
||||
%269 = OpLoad %4 %225
|
||||
%270 = OpIAdd %4 %269 %268
|
||||
OpStore %225 %270
|
||||
%271 = OpAccessChain %142 %217 %143
|
||||
%272 = OpLoad %11 %271
|
||||
%273 = OpBitcast %8 %272
|
||||
%274 = OpCompositeExtract %4 %273 1
|
||||
%275 = OpLoad %4 %225
|
||||
%276 = OpIAdd %4 %275 %274
|
||||
OpStore %225 %276
|
||||
%277 = OpAccessChain %151 %217 %152
|
||||
%278 = OpLoad %12 %277
|
||||
%279 = OpBitcast %9 %278
|
||||
%280 = OpCompositeExtract %4 %279 2
|
||||
%281 = OpLoad %4 %225
|
||||
%282 = OpIAdd %4 %281 %280
|
||||
OpStore %225 %282
|
||||
%283 = OpAccessChain %160 %217 %161
|
||||
%284 = OpLoad %13 %283
|
||||
%285 = OpBitcast %10 %284
|
||||
%286 = OpCompositeExtract %4 %285 3
|
||||
%287 = OpLoad %4 %225
|
||||
%288 = OpIAdd %4 %287 %286
|
||||
OpStore %225 %288
|
||||
%290 = OpAccessChain %103 %217 %104
|
||||
%291 = OpLoad %4 %290
|
||||
%292 = OpAccessChain %289 %218 %104
|
||||
%293 = OpLoad %4 %292
|
||||
%294 = OpIAdd %4 %291 %293
|
||||
%295 = OpAccessChain %289 %220 %104
|
||||
OpStore %295 %294
|
||||
%297 = OpAccessChain %110 %217 %111
|
||||
%298 = OpLoad %8 %297
|
||||
%299 = OpAccessChain %296 %218 %111
|
||||
%300 = OpLoad %8 %299
|
||||
%301 = OpIAdd %8 %298 %300
|
||||
%302 = OpAccessChain %296 %220 %111
|
||||
OpStore %302 %301
|
||||
%304 = OpAccessChain %118 %217 %119
|
||||
%305 = OpLoad %9 %304
|
||||
%306 = OpAccessChain %303 %218 %119
|
||||
%307 = OpLoad %9 %306
|
||||
%308 = OpIAdd %9 %305 %307
|
||||
%309 = OpAccessChain %303 %220 %119
|
||||
OpStore %309 %308
|
||||
%311 = OpAccessChain %126 %217 %127
|
||||
%312 = OpLoad %10 %311
|
||||
%313 = OpAccessChain %310 %218 %127
|
||||
%314 = OpLoad %10 %313
|
||||
%315 = OpIAdd %10 %312 %314
|
||||
%316 = OpAccessChain %310 %220 %127
|
||||
OpStore %316 %315
|
||||
%318 = OpAccessChain %317 %219 %43
|
||||
%319 = OpLoad %15 %318
|
||||
%320 = OpAccessChain %317 %221 %43
|
||||
OpStore %320 %319
|
||||
%321 = OpLoad %4 %225
|
||||
%322 = OpCopyObject %4 %321
|
||||
%323 = OpLoad %4 %225
|
||||
%324 = OpIAdd %4 %323 %322
|
||||
OpStore %225 %324
|
||||
%325 = OpLoad %4 %225
|
||||
%326 = OpLoad %4 %225
|
||||
%327 = OpLoad %4 %225
|
||||
%329 = OpExtInst %4 %1 UMax %325 %326
|
||||
%328 = OpExtInst %4 %1 UMin %329 %327
|
||||
%330 = OpLoad %4 %225
|
||||
%331 = OpIAdd %4 %330 %328
|
||||
OpStore %225 %331
|
||||
%332 = OpLoad %4 %225
|
||||
%333 = OpCompositeConstruct %8 %332 %332
|
||||
%334 = OpLoad %4 %225
|
||||
%335 = OpCompositeConstruct %8 %334 %334
|
||||
%338 = OpCompositeExtract %4 %333 0
|
||||
%339 = OpCompositeExtract %4 %335 0
|
||||
%340 = OpIMul %4 %338 %339
|
||||
%341 = OpIAdd %4 %337 %340
|
||||
%342 = OpCompositeExtract %4 %333 1
|
||||
%343 = OpCompositeExtract %4 %335 1
|
||||
%344 = OpIMul %4 %342 %343
|
||||
%336 = OpIAdd %4 %341 %344
|
||||
%345 = OpLoad %4 %225
|
||||
%346 = OpIAdd %4 %345 %336
|
||||
OpStore %225 %346
|
||||
%347 = OpLoad %4 %225
|
||||
%348 = OpLoad %4 %225
|
||||
%349 = OpExtInst %4 %1 UMax %347 %348
|
||||
%350 = OpLoad %4 %225
|
||||
%351 = OpIAdd %4 %350 %349
|
||||
OpStore %225 %351
|
||||
%352 = OpLoad %4 %225
|
||||
%353 = OpLoad %4 %225
|
||||
%354 = OpExtInst %4 %1 UMin %352 %353
|
||||
%355 = OpLoad %4 %225
|
||||
%356 = OpIAdd %4 %355 %354
|
||||
OpStore %225 %356
|
||||
%357 = OpLoad %4 %225
|
||||
OpReturnValue %357
|
||||
OpFunctionEnd
|
||||
%356 = OpFunction %2 None %357
|
||||
%355 = OpLabel
|
||||
%358 = OpAccessChain %42 %23 %43
|
||||
%359 = OpAccessChain %45 %26 %43
|
||||
%360 = OpAccessChain %47 %29 %43
|
||||
%361 = OpAccessChain %45 %32 %43
|
||||
%362 = OpAccessChain %47 %35 %43
|
||||
OpBranch %365
|
||||
%365 = OpLabel
|
||||
%366 = OpFunctionCall %4 %213 %363
|
||||
%367 = OpFunctionCall %3 %40 %364
|
||||
%368 = OpBitcast %4 %367
|
||||
%369 = OpIAdd %4 %366 %368
|
||||
%371 = OpAccessChain %286 %361 %370
|
||||
OpStore %371 %369
|
||||
%359 = OpFunction %2 None %360
|
||||
%358 = OpLabel
|
||||
%361 = OpAccessChain %42 %23 %43
|
||||
%362 = OpAccessChain %45 %26 %43
|
||||
%363 = OpAccessChain %47 %29 %43
|
||||
%364 = OpAccessChain %45 %32 %43
|
||||
%365 = OpAccessChain %47 %35 %43
|
||||
OpBranch %368
|
||||
%368 = OpLabel
|
||||
%369 = OpFunctionCall %4 %215 %366
|
||||
%370 = OpFunctionCall %3 %40 %367
|
||||
%371 = OpBitcast %4 %370
|
||||
%372 = OpIAdd %4 %369 %371
|
||||
%374 = OpAccessChain %289 %364 %373
|
||||
OpStore %374 %372
|
||||
OpReturn
|
||||
OpFunctionEnd
|
||||
@ -35,150 +35,150 @@ var<storage, read_write> output_arrays: StorageCompatible;
|
||||
fn int64_function(x: i64) -> i64 {
|
||||
var val: i64 = 20li;
|
||||
|
||||
let _e6 = val;
|
||||
val = (_e6 + (31li - 1002003004005006li));
|
||||
let _e8 = val;
|
||||
let _e11 = val;
|
||||
val = (_e11 + (_e8 + 5li));
|
||||
let _e15 = input_uniform.val_u32_;
|
||||
let _e16 = val;
|
||||
let _e20 = val;
|
||||
val = (_e20 + i64((_e15 + u32(_e16))));
|
||||
let _e24 = input_uniform.val_i32_;
|
||||
let _e25 = val;
|
||||
let _e29 = val;
|
||||
val = (_e29 + i64((_e24 + i32(_e25))));
|
||||
let _e33 = input_uniform.val_f32_;
|
||||
let _e34 = val;
|
||||
let _e38 = val;
|
||||
val = (_e38 + i64((_e33 + f32(_e34))));
|
||||
let _e42 = input_uniform.val_i64_;
|
||||
let _e45 = val;
|
||||
val = (_e45 + vec3(_e42).z);
|
||||
let _e49 = input_uniform.val_u64_;
|
||||
let _e51 = val;
|
||||
val = (_e51 + bitcast<i64>(_e49));
|
||||
let _e55 = input_uniform.val_u64_2_;
|
||||
let _e58 = val;
|
||||
val = (_e58 + bitcast<vec2<i64>>(_e55).y);
|
||||
let _e62 = input_uniform.val_u64_3_;
|
||||
let _e65 = val;
|
||||
val = (_e65 + bitcast<vec3<i64>>(_e62).z);
|
||||
let _e69 = input_uniform.val_u64_4_;
|
||||
let _e72 = val;
|
||||
val = (_e72 + bitcast<vec4<i64>>(_e69).w);
|
||||
let _e78 = input_uniform.val_i64_;
|
||||
let _e81 = input_storage.val_i64_;
|
||||
output.val_i64_ = (_e78 + _e81);
|
||||
let _e87 = input_uniform.val_i64_2_;
|
||||
let _e90 = input_storage.val_i64_2_;
|
||||
output.val_i64_2_ = (_e87 + _e90);
|
||||
let _e96 = input_uniform.val_i64_3_;
|
||||
let _e99 = input_storage.val_i64_3_;
|
||||
output.val_i64_3_ = (_e96 + _e99);
|
||||
let _e105 = input_uniform.val_i64_4_;
|
||||
let _e108 = input_storage.val_i64_4_;
|
||||
output.val_i64_4_ = (_e105 + _e108);
|
||||
let _e114 = input_arrays.val_i64_array_2_;
|
||||
output_arrays.val_i64_array_2_ = _e114;
|
||||
let _e115 = val;
|
||||
val = (_e8 + ((31li - 1002003004005006li) + -9223372036854775807li));
|
||||
let _e10 = val;
|
||||
let _e13 = val;
|
||||
val = (_e13 + (_e10 + 5li));
|
||||
let _e17 = input_uniform.val_u32_;
|
||||
let _e18 = val;
|
||||
let _e22 = val;
|
||||
val = (_e22 + i64((_e17 + u32(_e18))));
|
||||
let _e26 = input_uniform.val_i32_;
|
||||
let _e27 = val;
|
||||
let _e31 = val;
|
||||
val = (_e31 + i64((_e26 + i32(_e27))));
|
||||
let _e35 = input_uniform.val_f32_;
|
||||
let _e36 = val;
|
||||
let _e40 = val;
|
||||
val = (_e40 + i64((_e35 + f32(_e36))));
|
||||
let _e44 = input_uniform.val_i64_;
|
||||
let _e47 = val;
|
||||
val = (_e47 + vec3(_e44).z);
|
||||
let _e51 = input_uniform.val_u64_;
|
||||
let _e53 = val;
|
||||
val = (_e53 + bitcast<i64>(_e51));
|
||||
let _e57 = input_uniform.val_u64_2_;
|
||||
let _e60 = val;
|
||||
val = (_e60 + bitcast<vec2<i64>>(_e57).y);
|
||||
let _e64 = input_uniform.val_u64_3_;
|
||||
let _e67 = val;
|
||||
val = (_e67 + bitcast<vec3<i64>>(_e64).z);
|
||||
let _e71 = input_uniform.val_u64_4_;
|
||||
let _e74 = val;
|
||||
val = (_e74 + bitcast<vec4<i64>>(_e71).w);
|
||||
let _e80 = input_uniform.val_i64_;
|
||||
let _e83 = input_storage.val_i64_;
|
||||
output.val_i64_ = (_e80 + _e83);
|
||||
let _e89 = input_uniform.val_i64_2_;
|
||||
let _e92 = input_storage.val_i64_2_;
|
||||
output.val_i64_2_ = (_e89 + _e92);
|
||||
let _e98 = input_uniform.val_i64_3_;
|
||||
let _e101 = input_storage.val_i64_3_;
|
||||
output.val_i64_3_ = (_e98 + _e101);
|
||||
let _e107 = input_uniform.val_i64_4_;
|
||||
let _e110 = input_storage.val_i64_4_;
|
||||
output.val_i64_4_ = (_e107 + _e110);
|
||||
let _e116 = input_arrays.val_i64_array_2_;
|
||||
output_arrays.val_i64_array_2_ = _e116;
|
||||
let _e117 = val;
|
||||
val = (_e117 + abs(_e115));
|
||||
let _e119 = val;
|
||||
let _e120 = val;
|
||||
val = (_e119 + abs(_e117));
|
||||
let _e121 = val;
|
||||
let _e122 = val;
|
||||
let _e123 = val;
|
||||
val = (_e123 + clamp(_e119, _e120, _e121));
|
||||
let _e125 = val;
|
||||
val = (_e125 + clamp(_e121, _e122, _e123));
|
||||
let _e127 = val;
|
||||
let _e130 = val;
|
||||
val = (_e130 + dot(vec2(_e125), vec2(_e127)));
|
||||
let _e129 = val;
|
||||
let _e132 = val;
|
||||
let _e133 = val;
|
||||
val = (_e132 + dot(vec2(_e127), vec2(_e129)));
|
||||
let _e134 = val;
|
||||
let _e135 = val;
|
||||
val = (_e135 + max(_e132, _e133));
|
||||
let _e137 = val;
|
||||
let _e138 = val;
|
||||
val = (_e137 + max(_e134, _e135));
|
||||
let _e139 = val;
|
||||
let _e140 = val;
|
||||
val = (_e140 + min(_e137, _e138));
|
||||
let _e142 = val;
|
||||
val = (_e142 + min(_e139, _e140));
|
||||
let _e144 = val;
|
||||
val = (_e144 + sign(_e142));
|
||||
let _e146 = val;
|
||||
return _e146;
|
||||
val = (_e146 + sign(_e144));
|
||||
let _e148 = val;
|
||||
return _e148;
|
||||
}
|
||||
|
||||
fn uint64_function(x_1: u64) -> u64 {
|
||||
var val_1: u64 = 20lu;
|
||||
|
||||
let _e6 = val_1;
|
||||
val_1 = (_e6 + (31lu + 1002003004005006lu));
|
||||
let _e8 = val_1;
|
||||
let _e11 = val_1;
|
||||
val_1 = (_e11 + (_e8 + 5lu));
|
||||
let _e15 = input_uniform.val_u32_;
|
||||
let _e16 = val_1;
|
||||
let _e20 = val_1;
|
||||
val_1 = (_e20 + u64((_e15 + u32(_e16))));
|
||||
let _e24 = input_uniform.val_i32_;
|
||||
let _e25 = val_1;
|
||||
let _e29 = val_1;
|
||||
val_1 = (_e29 + u64((_e24 + i32(_e25))));
|
||||
let _e33 = input_uniform.val_f32_;
|
||||
let _e34 = val_1;
|
||||
let _e38 = val_1;
|
||||
val_1 = (_e38 + u64((_e33 + f32(_e34))));
|
||||
let _e42 = input_uniform.val_u64_;
|
||||
let _e45 = val_1;
|
||||
val_1 = (_e45 + vec3(_e42).z);
|
||||
let _e49 = input_uniform.val_i64_;
|
||||
let _e51 = val_1;
|
||||
val_1 = (_e51 + bitcast<u64>(_e49));
|
||||
let _e55 = input_uniform.val_i64_2_;
|
||||
let _e58 = val_1;
|
||||
val_1 = (_e58 + bitcast<vec2<u64>>(_e55).y);
|
||||
let _e62 = input_uniform.val_i64_3_;
|
||||
let _e65 = val_1;
|
||||
val_1 = (_e65 + bitcast<vec3<u64>>(_e62).z);
|
||||
let _e69 = input_uniform.val_i64_4_;
|
||||
let _e72 = val_1;
|
||||
val_1 = (_e72 + bitcast<vec4<u64>>(_e69).w);
|
||||
let _e78 = input_uniform.val_u64_;
|
||||
let _e81 = input_storage.val_u64_;
|
||||
output.val_u64_ = (_e78 + _e81);
|
||||
let _e87 = input_uniform.val_u64_2_;
|
||||
let _e90 = input_storage.val_u64_2_;
|
||||
output.val_u64_2_ = (_e87 + _e90);
|
||||
let _e96 = input_uniform.val_u64_3_;
|
||||
let _e99 = input_storage.val_u64_3_;
|
||||
output.val_u64_3_ = (_e96 + _e99);
|
||||
let _e105 = input_uniform.val_u64_4_;
|
||||
let _e108 = input_storage.val_u64_4_;
|
||||
output.val_u64_4_ = (_e105 + _e108);
|
||||
let _e114 = input_arrays.val_u64_array_2_;
|
||||
output_arrays.val_u64_array_2_ = _e114;
|
||||
let _e115 = val_1;
|
||||
val_1 = (_e8 + ((31lu + 18446744073709551615lu) - 18446744073709551615lu));
|
||||
let _e10 = val_1;
|
||||
let _e13 = val_1;
|
||||
val_1 = (_e13 + (_e10 + 5lu));
|
||||
let _e17 = input_uniform.val_u32_;
|
||||
let _e18 = val_1;
|
||||
let _e22 = val_1;
|
||||
val_1 = (_e22 + u64((_e17 + u32(_e18))));
|
||||
let _e26 = input_uniform.val_i32_;
|
||||
let _e27 = val_1;
|
||||
let _e31 = val_1;
|
||||
val_1 = (_e31 + u64((_e26 + i32(_e27))));
|
||||
let _e35 = input_uniform.val_f32_;
|
||||
let _e36 = val_1;
|
||||
let _e40 = val_1;
|
||||
val_1 = (_e40 + u64((_e35 + f32(_e36))));
|
||||
let _e44 = input_uniform.val_u64_;
|
||||
let _e47 = val_1;
|
||||
val_1 = (_e47 + vec3(_e44).z);
|
||||
let _e51 = input_uniform.val_i64_;
|
||||
let _e53 = val_1;
|
||||
val_1 = (_e53 + bitcast<u64>(_e51));
|
||||
let _e57 = input_uniform.val_i64_2_;
|
||||
let _e60 = val_1;
|
||||
val_1 = (_e60 + bitcast<vec2<u64>>(_e57).y);
|
||||
let _e64 = input_uniform.val_i64_3_;
|
||||
let _e67 = val_1;
|
||||
val_1 = (_e67 + bitcast<vec3<u64>>(_e64).z);
|
||||
let _e71 = input_uniform.val_i64_4_;
|
||||
let _e74 = val_1;
|
||||
val_1 = (_e74 + bitcast<vec4<u64>>(_e71).w);
|
||||
let _e80 = input_uniform.val_u64_;
|
||||
let _e83 = input_storage.val_u64_;
|
||||
output.val_u64_ = (_e80 + _e83);
|
||||
let _e89 = input_uniform.val_u64_2_;
|
||||
let _e92 = input_storage.val_u64_2_;
|
||||
output.val_u64_2_ = (_e89 + _e92);
|
||||
let _e98 = input_uniform.val_u64_3_;
|
||||
let _e101 = input_storage.val_u64_3_;
|
||||
output.val_u64_3_ = (_e98 + _e101);
|
||||
let _e107 = input_uniform.val_u64_4_;
|
||||
let _e110 = input_storage.val_u64_4_;
|
||||
output.val_u64_4_ = (_e107 + _e110);
|
||||
let _e116 = input_arrays.val_u64_array_2_;
|
||||
output_arrays.val_u64_array_2_ = _e116;
|
||||
let _e117 = val_1;
|
||||
val_1 = (_e117 + abs(_e115));
|
||||
let _e119 = val_1;
|
||||
let _e120 = val_1;
|
||||
val_1 = (_e119 + abs(_e117));
|
||||
let _e121 = val_1;
|
||||
let _e122 = val_1;
|
||||
let _e123 = val_1;
|
||||
val_1 = (_e123 + clamp(_e119, _e120, _e121));
|
||||
let _e125 = val_1;
|
||||
val_1 = (_e125 + clamp(_e121, _e122, _e123));
|
||||
let _e127 = val_1;
|
||||
let _e130 = val_1;
|
||||
val_1 = (_e130 + dot(vec2(_e125), vec2(_e127)));
|
||||
let _e129 = val_1;
|
||||
let _e132 = val_1;
|
||||
let _e133 = val_1;
|
||||
val_1 = (_e132 + dot(vec2(_e127), vec2(_e129)));
|
||||
let _e134 = val_1;
|
||||
let _e135 = val_1;
|
||||
val_1 = (_e135 + max(_e132, _e133));
|
||||
let _e137 = val_1;
|
||||
let _e138 = val_1;
|
||||
val_1 = (_e137 + max(_e134, _e135));
|
||||
let _e139 = val_1;
|
||||
let _e140 = val_1;
|
||||
val_1 = (_e140 + min(_e137, _e138));
|
||||
let _e142 = val_1;
|
||||
return _e142;
|
||||
val_1 = (_e142 + min(_e139, _e140));
|
||||
let _e144 = val_1;
|
||||
return _e144;
|
||||
}
|
||||
|
||||
@compute @workgroup_size(1, 1, 1)
|
||||
|
||||
Loading…
Reference in New Issue
Block a user