diff --git a/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm b/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm new file mode 100644 index 0000000000..99a3deacf8 --- /dev/null +++ b/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm @@ -0,0 +1,99 @@ +; Test: ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm +; SPIR-V +; Version: 1.0 +; Generator: Khronos SPIR-V Tools Assembler; 0 +; Bound: 100 +; Schema: 0 +OpCapability Shader +OpCapability Sampled1D +OpCapability Image1D +OpCapability StorageImageExtendedFormats +OpCapability ImageQuery +OpMemoryModel Logical Simple +OpEntryPoint Fragment %1 "main" +OpExecutionMode %1 OriginUpperLeft +OpName %f1 "f1" +OpName %vf12 "vf12" +OpName %vf123 "vf123" +OpName %vf1234 "vf1234" +OpName %u1 "u1" +OpName %vu12 "vu12" +OpName %vu123 "vu123" +OpName %vu1234 "vu1234" +OpName %i1 "i1" +OpName %vi12 "vi12" +OpName %vi123 "vi123" +OpName %vi1234 "vi1234" +OpName %offsets2d "offsets2d" +OpDecorate %20 DescriptorSet 2 +OpDecorate %20 Binding 1 +%void = OpTypeVoid +%16 = OpTypeFunction %void +%float = OpTypeFloat 32 +%uint = OpTypeInt 32 0 +%int = OpTypeInt 32 1 +%int_0 = OpConstant %int 0 +%int_1 = OpConstant %int 1 +%int_2 = OpConstant %int 2 +%int_3 = OpConstant %int 3 +%int_4 = OpConstant %int 4 +%uint_0 = OpConstant %uint 0 +%uint_1 = OpConstant %uint 1 +%uint_2 = OpConstant %uint 2 +%uint_3 = OpConstant %uint 3 +%uint_4 = OpConstant %uint 4 +%uint_100 = OpConstant %uint 100 +%v2int = OpTypeVector %int 2 +%v3int = OpTypeVector %int 3 +%v4int = OpTypeVector %int 4 +%v2uint = OpTypeVector %uint 2 +%v3uint = OpTypeVector %uint 3 +%v4uint = OpTypeVector %uint 4 +%v2float = OpTypeVector %float 2 +%v3float = OpTypeVector %float 3 +%v4float = OpTypeVector %float 4 +%41 = OpConstantNull %float +%float_0 = OpConstant %float 0 +%float_1 = OpConstant %float 1 +%float_2 = OpConstant %float 2 +%float_3 = OpConstant %float 3 +%float_4 = OpConstant %float 4 +%float_7 = OpConstant %float 7 +%48 = OpConstantNull %v2float +%49 = OpConstantNull %v3float +%50 = OpConstantNull %v4float +%51 = OpConstantComposite %v2int %int_1 %int_2 +%52 = OpConstantComposite %v3int %int_1 %int_2 %int_3 +%53 = OpConstantComposite %v4int %int_1 %int_2 %int_3 %int_4 +%54 = OpConstantComposite %v2uint %uint_1 %uint_2 +%55 = OpConstantComposite %v3uint %uint_1 %uint_2 %uint_3 +%56 = OpConstantComposite %v4uint %uint_1 %uint_2 %uint_3 %uint_4 +%57 = OpConstantComposite %v2float %float_1 %float_2 +%58 = OpConstantComposite %v2float %float_2 %float_1 +%59 = OpConstantComposite %v3float %float_1 %float_2 %float_3 +%60 = OpConstantComposite %v4float %float_1 %float_2 %float_3 %float_4 +%float_0_200000003 = OpConstant %float 0.200000003 +%62 = OpTypeImage %float 2D 1 0 1 1 Unknown +%_ptr_UniformConstant_62 = OpTypePointer UniformConstant %62 +%20 = OpVariable %_ptr_UniformConstant_62 UniformConstant +%1 = OpFunction %void None %16 +%64 = OpLabel +%f1 = OpCopyObject %float %float_1 +%vf12 = OpCopyObject %v2float %57 +%vf123 = OpCopyObject %v3float %59 +%vf1234 = OpCopyObject %v4float %60 +%i1 = OpCopyObject %int %int_1 +%vi12 = OpCopyObject %v2int %51 +%vi123 = OpCopyObject %v3int %52 +%vi1234 = OpCopyObject %v4int %53 +%u1 = OpCopyObject %uint %uint_1 +%vu12 = OpCopyObject %v2uint %54 +%vu123 = OpCopyObject %v3uint %55 +%vu1234 = OpCopyObject %v4uint %56 +%65 = OpCompositeConstruct %v2int %int_3 %int_4 +%offsets2d = OpCopyObject %v2int %65 +%66 = OpLoad %62 %20 +%99 = OpImageFetch %v4float %66 %vi12 Sample %i1 +OpReturn +OpFunctionEnd + diff --git a/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.hlsl b/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.hlsl new file mode 100644 index 0000000000..94c1196ef5 --- /dev/null +++ b/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.hlsl @@ -0,0 +1,24 @@ +Texture2DMS x_20 : register(t1, space2); + +void main_1() { + const float f1 = 1.0f; + const float2 vf12 = float2(1.0f, 2.0f); + const float3 vf123 = float3(1.0f, 2.0f, 3.0f); + const float4 vf1234 = float4(1.0f, 2.0f, 3.0f, 4.0f); + const int i1 = 1; + const int2 vi12 = int2(1, 2); + const int3 vi123 = int3(1, 2, 3); + const int4 vi1234 = int4(1, 2, 3, 4); + const uint u1 = 1u; + const uint2 vu12 = uint2(1u, 2u); + const uint3 vu123 = uint3(1u, 2u, 3u); + const uint4 vu1234 = uint4(1u, 2u, 3u, 4u); + const int2 offsets2d = int2(3, 4); + const float4 x_99 = float4(x_20.Load(int3(vi12, 0), i1).x, 0.0f, 0.0f, 0.0f); + return; +} + +void main() { + main_1(); + return; +} diff --git a/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.msl b/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.msl new file mode 100644 index 0000000000..a41a1760ab --- /dev/null +++ b/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.msl @@ -0,0 +1,26 @@ +#include + +using namespace metal; +void main_1(depth2d_ms tint_symbol_1) { + float const f1 = 1.0f; + float2 const vf12 = float2(1.0f, 2.0f); + float3 const vf123 = float3(1.0f, 2.0f, 3.0f); + float4 const vf1234 = float4(1.0f, 2.0f, 3.0f, 4.0f); + int const i1 = 1; + int2 const vi12 = int2(1, 2); + int3 const vi123 = int3(1, 2, 3); + int4 const vi1234 = int4(1, 2, 3, 4); + uint const u1 = 1u; + uint2 const vu12 = uint2(1u, 2u); + uint3 const vu123 = uint3(1u, 2u, 3u); + uint4 const vu1234 = uint4(1u, 2u, 3u, 4u); + int2 const offsets2d = int2(3, 4); + float4 const x_99 = float4(tint_symbol_1.read(uint2(vi12), i1), 0.0f, 0.0f, 0.0f); + return; +} + +fragment void tint_symbol(depth2d_ms tint_symbol_2 [[texture(1)]]) { + main_1(tint_symbol_2); + return; +} + diff --git a/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.spvasm b/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.spvasm new file mode 100644 index 0000000000..a49c62c887 --- /dev/null +++ b/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.spvasm @@ -0,0 +1,67 @@ +; SPIR-V +; Version: 1.3 +; Generator: Google Tint Compiler; 0 +; Bound: 50 +; Schema: 0 + OpCapability Shader + OpMemoryModel Logical GLSL450 + OpEntryPoint Fragment %main "main" + OpExecutionMode %main OriginUpperLeft + OpName %x_20 "x_20" + OpName %main_1 "main_1" + OpName %main "main" + OpDecorate %x_20 DescriptorSet 2 + OpDecorate %x_20 Binding 1 + %float = OpTypeFloat 32 + %3 = OpTypeImage %float 2D 1 0 1 1 Unknown +%_ptr_UniformConstant_3 = OpTypePointer UniformConstant %3 + %x_20 = OpVariable %_ptr_UniformConstant_3 UniformConstant + %void = OpTypeVoid + %5 = OpTypeFunction %void + %float_1 = OpConstant %float 1 + %v2float = OpTypeVector %float 2 + %float_2 = OpConstant %float 2 + %12 = OpConstantComposite %v2float %float_1 %float_2 + %v3float = OpTypeVector %float 3 + %float_3 = OpConstant %float 3 + %15 = OpConstantComposite %v3float %float_1 %float_2 %float_3 + %v4float = OpTypeVector %float 4 + %float_4 = OpConstant %float 4 + %18 = OpConstantComposite %v4float %float_1 %float_2 %float_3 %float_4 + %int = OpTypeInt 32 1 + %int_1 = OpConstant %int 1 + %v2int = OpTypeVector %int 2 + %int_2 = OpConstant %int 2 + %23 = OpConstantComposite %v2int %int_1 %int_2 + %v3int = OpTypeVector %int 3 + %int_3 = OpConstant %int 3 + %26 = OpConstantComposite %v3int %int_1 %int_2 %int_3 + %v4int = OpTypeVector %int 4 + %int_4 = OpConstant %int 4 + %29 = OpConstantComposite %v4int %int_1 %int_2 %int_3 %int_4 + %uint = OpTypeInt 32 0 + %uint_1 = OpConstant %uint 1 + %v2uint = OpTypeVector %uint 2 + %uint_2 = OpConstant %uint 2 + %34 = OpConstantComposite %v2uint %uint_1 %uint_2 + %v3uint = OpTypeVector %uint 3 + %uint_3 = OpConstant %uint 3 + %37 = OpConstantComposite %v3uint %uint_1 %uint_2 %uint_3 + %v4uint = OpTypeVector %uint 4 + %uint_4 = OpConstant %uint 4 + %40 = OpConstantComposite %v4uint %uint_1 %uint_2 %uint_3 %uint_4 + %41 = OpConstantComposite %v2int %int_3 %int_4 + %float_0 = OpConstant %float 0 + %main_1 = OpFunction %void None %5 + %8 = OpLabel + %44 = OpLoad %3 %x_20 + %43 = OpImageFetch %v4float %44 %23 Sample %int_1 + %42 = OpCompositeExtract %float %43 0 + %46 = OpCompositeConstruct %v4float %42 %float_0 %float_0 %float_0 + OpReturn + OpFunctionEnd + %main = OpFunction %void None %5 + %48 = OpLabel + %49 = OpFunctionCall %void %main_1 + OpReturn + OpFunctionEnd diff --git a/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.wgsl b/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.wgsl new file mode 100644 index 0000000000..1f17b47a8a --- /dev/null +++ b/test/unittest/reader/spirv/ImageFetch_DepthMultisampled_SpvParserHandleTest_ImageAccessTest_Variable_0.spvasm.expected.wgsl @@ -0,0 +1,24 @@ +[[group(2), binding(1)]] var x_20 : texture_depth_multisampled_2d; + +fn main_1() { + let f1 : f32 = 1.0; + let vf12 : vec2 = vec2(1.0, 2.0); + let vf123 : vec3 = vec3(1.0, 2.0, 3.0); + let vf1234 : vec4 = vec4(1.0, 2.0, 3.0, 4.0); + let i1 : i32 = 1; + let vi12 : vec2 = vec2(1, 2); + let vi123 : vec3 = vec3(1, 2, 3); + let vi1234 : vec4 = vec4(1, 2, 3, 4); + let u1 : u32 = 1u; + let vu12 : vec2 = vec2(1u, 2u); + let vu123 : vec3 = vec3(1u, 2u, 3u); + let vu1234 : vec4 = vec4(1u, 2u, 3u, 4u); + let offsets2d : vec2 = vec2(3, 4); + let x_99 : vec4 = vec4(textureLoad(x_20, vi12, i1), 0.0, 0.0, 0.0); + return; +} + +[[stage(fragment)]] +fn main() { + main_1(); +} diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm index 55e55eebb4..bae6f5957e 100644 --- a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm @@ -12,7 +12,7 @@ OpExecutionMode %5 OriginUpperLeft OpDecorate %1 Location 0 OpDecorate %2 Location 0 OpDecorate %3 Location 30 -OpDecorate %4 Location 40 +OpDecorate %4 Location 6 %void = OpTypeVoid %7 = OpTypeFunction %void %bool = OpTypeBool diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.hlsl b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.hlsl index 7017765db0..d2eadc7035 100644 --- a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.hlsl +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.hlsl @@ -1,5 +1,3 @@ -SKIP: FAILED - static uint x_1 = 0u; static uint x_2 = 0u; static uint x_3 = 0u; @@ -19,7 +17,7 @@ struct tint_symbol_1 { }; struct tint_symbol_2 { uint x_2_1 : SV_Target0; - uint x_4_1 : SV_Target40; + uint x_4_1 : SV_Target6; }; tint_symbol_2 main(tint_symbol_1 tint_symbol) { @@ -32,11 +30,3 @@ tint_symbol_2 main(tint_symbol_1 tint_symbol) { const tint_symbol_2 tint_symbol_4 = {tint_symbol_3.x_2_1, tint_symbol_3.x_4_1}; return tint_symbol_4; } -warning: DXIL.dll not found. Resulting DXIL will not be signed for use in release environments. - -error: validation errors -error: SV_Target semantic index exceeds maximum (7). -Validation failed. - - - diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.spvasm b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.spvasm index 07d96dd40c..85a1b1bca8 100644 --- a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.spvasm +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.spvasm @@ -27,7 +27,7 @@ OpDecorate %tint_symbol_1 Location 30 OpDecorate %tint_symbol_1 Flat OpDecorate %tint_symbol_3 Location 0 - OpDecorate %tint_symbol_4 Location 40 + OpDecorate %tint_symbol_4 Location 6 OpMemberDecorate %main_out 0 Offset 0 OpMemberDecorate %main_out 1 Offset 4 %uint = OpTypeInt 32 0 diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.wgsl b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.wgsl index 542ae52e13..59ed023dc6 100644 --- a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.wgsl +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_EntryPointWrapping_IOLocations.spvasm.expected.wgsl @@ -13,7 +13,7 @@ fn main_1() { struct main_out { [[location(0)]] x_2_1 : u32; - [[location(40)]] + [[location(6)]] x_4_1 : u32; }; diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.hlsl b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.hlsl new file mode 100644 index 0000000000..ef0855d437 --- /dev/null +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.hlsl @@ -0,0 +1,10 @@ +RWByteAddressBuffer myvar : register(u0, space0); + +void main_1() { + return; +} + +void main() { + main_1(); + return; +} diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.msl b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.msl new file mode 100644 index 0000000000..b2a51ad5dc --- /dev/null +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.msl @@ -0,0 +1,23 @@ +#include + +using namespace metal; +struct tint_padded_array_element { + /* 0x0000 */ packed_float2 el; + /* 0x0008 */ int8_t tint_pad[56]; +}; +struct tint_array_wrapper { + /* 0x0000 */ tint_padded_array_element arr[3]; +}; +struct S { + /* 0x0000 */ tint_array_wrapper field0; +}; + +void main_1() { + return; +} + +fragment void tint_symbol() { + main_1(); + return; +} + diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.spvasm b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.spvasm new file mode 100644 index 0000000000..33921e1e34 --- /dev/null +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.spvasm @@ -0,0 +1,38 @@ +; SPIR-V +; Version: 1.3 +; Generator: Google Tint Compiler; 0 +; Bound: 16 +; Schema: 0 + OpCapability Shader + OpMemoryModel Logical GLSL450 + OpEntryPoint Fragment %main "main" + OpExecutionMode %main OriginUpperLeft + OpName %S "S" + OpMemberName %S 0 "field0" + OpName %myvar "myvar" + OpName %main_1 "main_1" + OpName %main "main" + OpDecorate %S Block + OpMemberDecorate %S 0 Offset 0 + OpDecorate %_arr_v2float_uint_3 ArrayStride 64 + OpDecorate %myvar DescriptorSet 0 + OpDecorate %myvar Binding 0 + %float = OpTypeFloat 32 + %v2float = OpTypeVector %float 2 + %uint = OpTypeInt 32 0 + %uint_3 = OpConstant %uint 3 +%_arr_v2float_uint_3 = OpTypeArray %v2float %uint_3 + %S = OpTypeStruct %_arr_v2float_uint_3 +%_ptr_StorageBuffer_S = OpTypePointer StorageBuffer %S + %myvar = OpVariable %_ptr_StorageBuffer_S StorageBuffer + %void = OpTypeVoid + %9 = OpTypeFunction %void + %main_1 = OpFunction %void None %9 + %12 = OpLabel + OpReturn + OpFunctionEnd + %main = OpFunction %void None %9 + %14 = OpLabel + %15 = OpFunctionCall %void %main_1 + OpReturn + OpFunctionEnd diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.wgsl b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.wgsl new file mode 100644 index 0000000000..996fa56a4f --- /dev/null +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration.spvasm.expected.wgsl @@ -0,0 +1,15 @@ +[[block]] +struct S { + field0 : [[stride(64)]] array, 3>; +}; + +[[group(0), binding(0)]] var myvar : S; + +fn main_1() { + return; +} + +[[stage(fragment)]] +fn main() { + main_1(); +} diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.hlsl b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.hlsl new file mode 100644 index 0000000000..ef0855d437 --- /dev/null +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.hlsl @@ -0,0 +1,10 @@ +RWByteAddressBuffer myvar : register(u0, space0); + +void main_1() { + return; +} + +void main() { + main_1(); + return; +} diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.msl b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.msl new file mode 100644 index 0000000000..ecdf927bf2 --- /dev/null +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.msl @@ -0,0 +1,16 @@ +#include + +using namespace metal; +struct S { + /* 0x0000 */ float3x2 field0; +}; + +void main_1() { + return; +} + +fragment void tint_symbol() { + main_1(); + return; +} + diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.spvasm b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.spvasm new file mode 100644 index 0000000000..64446e5b5a --- /dev/null +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.spvasm @@ -0,0 +1,37 @@ +; SPIR-V +; Version: 1.3 +; Generator: Google Tint Compiler; 0 +; Bound: 14 +; Schema: 0 + OpCapability Shader + OpMemoryModel Logical GLSL450 + OpEntryPoint Fragment %main "main" + OpExecutionMode %main OriginUpperLeft + OpName %S "S" + OpMemberName %S 0 "field0" + OpName %myvar "myvar" + OpName %main_1 "main_1" + OpName %main "main" + OpDecorate %S Block + OpMemberDecorate %S 0 Offset 0 + OpMemberDecorate %S 0 ColMajor + OpMemberDecorate %S 0 MatrixStride 8 + OpDecorate %myvar DescriptorSet 0 + OpDecorate %myvar Binding 0 + %float = OpTypeFloat 32 + %v2float = OpTypeVector %float 2 +%mat3v2float = OpTypeMatrix %v2float 3 + %S = OpTypeStruct %mat3v2float +%_ptr_StorageBuffer_S = OpTypePointer StorageBuffer %S + %myvar = OpVariable %_ptr_StorageBuffer_S StorageBuffer + %void = OpTypeVoid + %7 = OpTypeFunction %void + %main_1 = OpFunction %void None %7 + %10 = OpLabel + OpReturn + OpFunctionEnd + %main = OpFunction %void None %7 + %12 = OpLabel + %13 = OpFunctionCall %void %main_1 + OpReturn + OpFunctionEnd diff --git a/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.wgsl b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.wgsl new file mode 100644 index 0000000000..79b8f4855c --- /dev/null +++ b/test/unittest/reader/spirv/SpvModuleScopeVarParserTest_MatrixStrideDecoration_Natural_Dropped.spvasm.expected.wgsl @@ -0,0 +1,15 @@ +[[block]] +struct S { + field0 : mat3x2; +}; + +[[group(0), binding(0)]] var myvar : S; + +fn main_1() { + return; +} + +[[stage(fragment)]] +fn main() { + main_1(); +} diff --git a/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.hlsl b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.hlsl new file mode 100644 index 0000000000..a889e7888b --- /dev/null +++ b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.hlsl @@ -0,0 +1,32 @@ +void main_1() { + const uint u1 = 10u; + const uint u2 = 15u; + const uint u3 = 20u; + const int i1 = 30; + const int i2 = 35; + const int i3 = 40; + const float f1 = 50.0f; + const float f2 = 60.0f; + const float f3 = 70.0f; + const uint2 v2u1 = uint2(10u, 20u); + const uint2 v2u2 = uint2(20u, 10u); + const uint2 v2u3 = uint2(15u, 15u); + const int2 v2i1 = int2(30, 40); + const int2 v2i2 = int2(40, 30); + const int2 v2i3 = int2(35, 35); + const float2 v2f1 = float2(50.0f, 60.0f); + const float2 v2f2 = float2(60.0f, 50.0f); + const float2 v2f3 = float2(70.0f, 70.0f); + const float3 v3f1 = float3(50.0f, 60.0f, 70.0f); + const float3 v3f2 = float3(60.0f, 70.0f, 50.0f); + const float4 v4f1 = float4(50.0f, 50.0f, 50.0f, 50.0f); + const float4 v4f2 = v4f1; + const float x_1 = ldexp(f1, int(u1)); + return; +} + +[numthreads(1, 1, 1)] +void main() { + main_1(); + return; +} diff --git a/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.msl b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.msl new file mode 100644 index 0000000000..9484ddac44 --- /dev/null +++ b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.msl @@ -0,0 +1,35 @@ +#include + +using namespace metal; +void main_1() { + uint const u1 = 10u; + uint const u2 = 15u; + uint const u3 = 20u; + int const i1 = 30; + int const i2 = 35; + int const i3 = 40; + float const f1 = 50.0f; + float const f2 = 60.0f; + float const f3 = 70.0f; + uint2 const v2u1 = uint2(10u, 20u); + uint2 const v2u2 = uint2(20u, 10u); + uint2 const v2u3 = uint2(15u, 15u); + int2 const v2i1 = int2(30, 40); + int2 const v2i2 = int2(40, 30); + int2 const v2i3 = int2(35, 35); + float2 const v2f1 = float2(50.0f, 60.0f); + float2 const v2f2 = float2(60.0f, 50.0f); + float2 const v2f3 = float2(70.0f, 70.0f); + float3 const v3f1 = float3(50.0f, 60.0f, 70.0f); + float3 const v3f2 = float3(60.0f, 70.0f, 50.0f); + float4 const v4f1 = float4(50.0f, 50.0f, 50.0f, 50.0f); + float4 const v4f2 = v4f1; + float const x_1 = ldexp(f1, int(u1)); + return; +} + +kernel void tint_symbol() { + main_1(); + return; +} + diff --git a/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.spvasm b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.spvasm new file mode 100644 index 0000000000..2eab6d45b1 --- /dev/null +++ b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.spvasm @@ -0,0 +1,54 @@ +; SPIR-V +; Version: 1.3 +; Generator: Google Tint Compiler; 0 +; Bound: 40 +; Schema: 0 + OpCapability Shader + %35 = OpExtInstImport "GLSL.std.450" + OpMemoryModel Logical GLSL450 + OpEntryPoint GLCompute %main "main" + OpExecutionMode %main LocalSize 1 1 1 + OpName %main_1 "main_1" + OpName %main "main" + %void = OpTypeVoid + %1 = OpTypeFunction %void + %uint = OpTypeInt 32 0 + %uint_10 = OpConstant %uint 10 + %uint_15 = OpConstant %uint 15 + %uint_20 = OpConstant %uint 20 + %int = OpTypeInt 32 1 + %int_30 = OpConstant %int 30 + %int_35 = OpConstant %int 35 + %int_40 = OpConstant %int 40 + %float = OpTypeFloat 32 + %float_50 = OpConstant %float 50 + %float_60 = OpConstant %float 60 + %float_70 = OpConstant %float 70 + %v2uint = OpTypeVector %uint 2 + %18 = OpConstantComposite %v2uint %uint_10 %uint_20 + %19 = OpConstantComposite %v2uint %uint_20 %uint_10 + %20 = OpConstantComposite %v2uint %uint_15 %uint_15 + %v2int = OpTypeVector %int 2 + %22 = OpConstantComposite %v2int %int_30 %int_40 + %23 = OpConstantComposite %v2int %int_40 %int_30 + %24 = OpConstantComposite %v2int %int_35 %int_35 + %v2float = OpTypeVector %float 2 + %26 = OpConstantComposite %v2float %float_50 %float_60 + %27 = OpConstantComposite %v2float %float_60 %float_50 + %28 = OpConstantComposite %v2float %float_70 %float_70 + %v3float = OpTypeVector %float 3 + %30 = OpConstantComposite %v3float %float_50 %float_60 %float_70 + %31 = OpConstantComposite %v3float %float_60 %float_70 %float_50 + %v4float = OpTypeVector %float 4 + %33 = OpConstantComposite %v4float %float_50 %float_50 %float_50 %float_50 + %main_1 = OpFunction %void None %1 + %4 = OpLabel + %36 = OpBitcast %int %uint_10 + %34 = OpExtInst %float %35 Ldexp %float_50 %36 + OpReturn + OpFunctionEnd + %main = OpFunction %void None %1 + %38 = OpLabel + %39 = OpFunctionCall %void %main_1 + OpReturn + OpFunctionEnd diff --git a/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.wgsl b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.wgsl new file mode 100644 index 0000000000..45e078423c --- /dev/null +++ b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Scalar_Float_Uint.spvasm.expected.wgsl @@ -0,0 +1,31 @@ +fn main_1() { + let u1 : u32 = 10u; + let u2 : u32 = 15u; + let u3 : u32 = 20u; + let i1 : i32 = 30; + let i2 : i32 = 35; + let i3 : i32 = 40; + let f1 : f32 = 50.0; + let f2 : f32 = 60.0; + let f3 : f32 = 70.0; + let v2u1 : vec2 = vec2(10u, 20u); + let v2u2 : vec2 = vec2(20u, 10u); + let v2u3 : vec2 = vec2(15u, 15u); + let v2i1 : vec2 = vec2(30, 40); + let v2i2 : vec2 = vec2(40, 30); + let v2i3 : vec2 = vec2(35, 35); + let v2f1 : vec2 = vec2(50.0, 60.0); + let v2f2 : vec2 = vec2(60.0, 50.0); + let v2f3 : vec2 = vec2(70.0, 70.0); + let v3f1 : vec3 = vec3(50.0, 60.0, 70.0); + let v3f2 : vec3 = vec3(60.0, 70.0, 50.0); + let v4f1 : vec4 = vec4(50.0, 50.0, 50.0, 50.0); + let v4f2 : vec4 = v4f1; + let x_1 : f32 = ldexp(f1, i32(u1)); + return; +} + +[[stage(compute), workgroup_size(1, 1, 1)]] +fn main() { + main_1(); +} diff --git a/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.hlsl b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.hlsl new file mode 100644 index 0000000000..896884350a --- /dev/null +++ b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.hlsl @@ -0,0 +1,32 @@ +void main_1() { + const uint u1 = 10u; + const uint u2 = 15u; + const uint u3 = 20u; + const int i1 = 30; + const int i2 = 35; + const int i3 = 40; + const float f1 = 50.0f; + const float f2 = 60.0f; + const float f3 = 70.0f; + const uint2 v2u1 = uint2(10u, 20u); + const uint2 v2u2 = uint2(20u, 10u); + const uint2 v2u3 = uint2(15u, 15u); + const int2 v2i1 = int2(30, 40); + const int2 v2i2 = int2(40, 30); + const int2 v2i3 = int2(35, 35); + const float2 v2f1 = float2(50.0f, 60.0f); + const float2 v2f2 = float2(60.0f, 50.0f); + const float2 v2f3 = float2(70.0f, 70.0f); + const float3 v3f1 = float3(50.0f, 60.0f, 70.0f); + const float3 v3f2 = float3(60.0f, 70.0f, 50.0f); + const float4 v4f1 = float4(50.0f, 50.0f, 50.0f, 50.0f); + const float4 v4f2 = v4f1; + const float2 x_1 = ldexp(v2f1, int2(v2u1)); + return; +} + +[numthreads(1, 1, 1)] +void main() { + main_1(); + return; +} diff --git a/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.msl b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.msl new file mode 100644 index 0000000000..937bf3d103 --- /dev/null +++ b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.msl @@ -0,0 +1,35 @@ +#include + +using namespace metal; +void main_1() { + uint const u1 = 10u; + uint const u2 = 15u; + uint const u3 = 20u; + int const i1 = 30; + int const i2 = 35; + int const i3 = 40; + float const f1 = 50.0f; + float const f2 = 60.0f; + float const f3 = 70.0f; + uint2 const v2u1 = uint2(10u, 20u); + uint2 const v2u2 = uint2(20u, 10u); + uint2 const v2u3 = uint2(15u, 15u); + int2 const v2i1 = int2(30, 40); + int2 const v2i2 = int2(40, 30); + int2 const v2i3 = int2(35, 35); + float2 const v2f1 = float2(50.0f, 60.0f); + float2 const v2f2 = float2(60.0f, 50.0f); + float2 const v2f3 = float2(70.0f, 70.0f); + float3 const v3f1 = float3(50.0f, 60.0f, 70.0f); + float3 const v3f2 = float3(60.0f, 70.0f, 50.0f); + float4 const v4f1 = float4(50.0f, 50.0f, 50.0f, 50.0f); + float4 const v4f2 = v4f1; + float2 const x_1 = ldexp(v2f1, int2(v2u1)); + return; +} + +kernel void tint_symbol() { + main_1(); + return; +} + diff --git a/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.spvasm b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.spvasm new file mode 100644 index 0000000000..69e8594ecf --- /dev/null +++ b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.spvasm @@ -0,0 +1,54 @@ +; SPIR-V +; Version: 1.3 +; Generator: Google Tint Compiler; 0 +; Bound: 40 +; Schema: 0 + OpCapability Shader + %35 = OpExtInstImport "GLSL.std.450" + OpMemoryModel Logical GLSL450 + OpEntryPoint GLCompute %main "main" + OpExecutionMode %main LocalSize 1 1 1 + OpName %main_1 "main_1" + OpName %main "main" + %void = OpTypeVoid + %1 = OpTypeFunction %void + %uint = OpTypeInt 32 0 + %uint_10 = OpConstant %uint 10 + %uint_15 = OpConstant %uint 15 + %uint_20 = OpConstant %uint 20 + %int = OpTypeInt 32 1 + %int_30 = OpConstant %int 30 + %int_35 = OpConstant %int 35 + %int_40 = OpConstant %int 40 + %float = OpTypeFloat 32 + %float_50 = OpConstant %float 50 + %float_60 = OpConstant %float 60 + %float_70 = OpConstant %float 70 + %v2uint = OpTypeVector %uint 2 + %18 = OpConstantComposite %v2uint %uint_10 %uint_20 + %19 = OpConstantComposite %v2uint %uint_20 %uint_10 + %20 = OpConstantComposite %v2uint %uint_15 %uint_15 + %v2int = OpTypeVector %int 2 + %22 = OpConstantComposite %v2int %int_30 %int_40 + %23 = OpConstantComposite %v2int %int_40 %int_30 + %24 = OpConstantComposite %v2int %int_35 %int_35 + %v2float = OpTypeVector %float 2 + %26 = OpConstantComposite %v2float %float_50 %float_60 + %27 = OpConstantComposite %v2float %float_60 %float_50 + %28 = OpConstantComposite %v2float %float_70 %float_70 + %v3float = OpTypeVector %float 3 + %30 = OpConstantComposite %v3float %float_50 %float_60 %float_70 + %31 = OpConstantComposite %v3float %float_60 %float_70 %float_50 + %v4float = OpTypeVector %float 4 + %33 = OpConstantComposite %v4float %float_50 %float_50 %float_50 %float_50 + %main_1 = OpFunction %void None %1 + %4 = OpLabel + %36 = OpBitcast %v2int %18 + %34 = OpExtInst %v2float %35 Ldexp %26 %36 + OpReturn + OpFunctionEnd + %main = OpFunction %void None %1 + %38 = OpLabel + %39 = OpFunctionCall %void %main_1 + OpReturn + OpFunctionEnd diff --git a/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.wgsl b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.wgsl new file mode 100644 index 0000000000..100a2a971e --- /dev/null +++ b/test/unittest/reader/spirv/SpvParserTest_GlslStd450_Ldexp_Vector_Floatvec_Uintvec.spvasm.expected.wgsl @@ -0,0 +1,31 @@ +fn main_1() { + let u1 : u32 = 10u; + let u2 : u32 = 15u; + let u3 : u32 = 20u; + let i1 : i32 = 30; + let i2 : i32 = 35; + let i3 : i32 = 40; + let f1 : f32 = 50.0; + let f2 : f32 = 60.0; + let f3 : f32 = 70.0; + let v2u1 : vec2 = vec2(10u, 20u); + let v2u2 : vec2 = vec2(20u, 10u); + let v2u3 : vec2 = vec2(15u, 15u); + let v2i1 : vec2 = vec2(30, 40); + let v2i2 : vec2 = vec2(40, 30); + let v2i3 : vec2 = vec2(35, 35); + let v2f1 : vec2 = vec2(50.0, 60.0); + let v2f2 : vec2 = vec2(60.0, 50.0); + let v2f3 : vec2 = vec2(70.0, 70.0); + let v3f1 : vec3 = vec3(50.0, 60.0, 70.0); + let v3f2 : vec3 = vec3(60.0, 70.0, 50.0); + let v4f1 : vec4 = vec4(50.0, 50.0, 50.0, 50.0); + let v4f2 : vec4 = v4f1; + let x_1 : vec2 = ldexp(v2f1, vec2(v2u1)); + return; +} + +[[stage(compute), workgroup_size(1, 1, 1)]] +fn main() { + main_1(); +}