[naga] Move array-by-value snapshot tests to index-by-value.wgsl.

This commit is contained in:
Jim Blandy 2024-10-11 08:12:32 -07:00
parent ed3006ccc6
commit 1047fa57f0
17 changed files with 499 additions and 814 deletions

View File

@ -168,18 +168,6 @@ fn assign_through_ptr() {
assign_array_through_ptr_fn(&arr);
}
@vertex
fn foo(@builtin(vertex_index) vi: u32) -> @builtin(position) vec4<f32> {
let arr = array<i32, 5>(1, 2, 3, 4, 5);
let value = arr[vi];
return vec4<f32>(vec4<i32>(value));
}
fn array_by_value(a: array<i32, 5>, i: i32) -> i32 {
return a[i];
}
struct AssignToMember {
x: u32,
}

View File

@ -11,3 +11,10 @@ fn index_let_matrix(i: i32, j: i32) -> f32 {
let a = mat2x2<f32>(1, 2, 3, 4);
return a[i][j];
}
@vertex
fn index_let_array_1d(@builtin(vertex_index) vi: u32) -> @builtin(position) vec4<f32> {
let arr = array<i32, 5>(1, 2, 3, 4, 5);
let value = arr[vi];
return vec4<f32>(vec4<i32>(value));
}

View File

@ -2739,54 +2739,6 @@
sampling: [],
dual_source_blending: false,
),
(
flags: ("EXPRESSIONS | BLOCKS | CONTROL_FLOW_UNIFORMITY | STRUCT_LAYOUTS | CONSTANTS | BINDINGS"),
available_stages: ("VERTEX | FRAGMENT | COMPUTE"),
uniformity: (
non_uniform_result: Some(0),
requirements: (""),
),
may_kill: false,
sampling_set: [],
global_uses: [
(""),
(""),
(""),
(""),
(""),
],
expressions: [
(
uniformity: (
non_uniform_result: Some(0),
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Handle(25),
),
(
uniformity: (
non_uniform_result: Some(1),
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Handle(2),
),
(
uniformity: (
non_uniform_result: Some(0),
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Handle(2),
),
],
sampling: [],
dual_source_blending: false,
),
(
flags: ("EXPRESSIONS | BLOCKS | CONTROL_FLOW_UNIFORMITY | STRUCT_LAYOUTS | CONSTANTS | BINDINGS"),
available_stages: ("VERTEX | FRAGMENT | COMPUTE"),
@ -4243,144 +4195,6 @@
sampling: [],
dual_source_blending: false,
),
(
flags: ("EXPRESSIONS | BLOCKS | CONTROL_FLOW_UNIFORMITY | STRUCT_LAYOUTS | CONSTANTS | BINDINGS"),
available_stages: ("VERTEX | FRAGMENT | COMPUTE"),
uniformity: (
non_uniform_result: Some(0),
requirements: (""),
),
may_kill: false,
sampling_set: [],
global_uses: [
(""),
(""),
(""),
(""),
(""),
],
expressions: [
(
uniformity: (
non_uniform_result: Some(0),
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Handle(0),
),
(
uniformity: (
non_uniform_result: None,
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Value(Scalar((
kind: Sint,
width: 4,
))),
),
(
uniformity: (
non_uniform_result: None,
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Value(Scalar((
kind: Sint,
width: 4,
))),
),
(
uniformity: (
non_uniform_result: None,
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Value(Scalar((
kind: Sint,
width: 4,
))),
),
(
uniformity: (
non_uniform_result: None,
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Value(Scalar((
kind: Sint,
width: 4,
))),
),
(
uniformity: (
non_uniform_result: None,
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Value(Scalar((
kind: Sint,
width: 4,
))),
),
(
uniformity: (
non_uniform_result: None,
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Handle(25),
),
(
uniformity: (
non_uniform_result: Some(0),
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Handle(2),
),
(
uniformity: (
non_uniform_result: Some(0),
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Value(Vector(
size: Quad,
scalar: (
kind: Sint,
width: 4,
),
)),
),
(
uniformity: (
non_uniform_result: Some(0),
requirements: (""),
),
ref_count: 1,
assignable_global: None,
ty: Value(Vector(
size: Quad,
scalar: (
kind: Float,
width: 4,
),
)),
),
],
sampling: [],
dual_source_blending: false,
),
(
flags: ("EXPRESSIONS | BLOCKS | CONTROL_FLOW_UNIFORMITY | STRUCT_LAYOUTS | CONSTANTS | BINDINGS"),
available_stages: ("VERTEX | FRAGMENT | COMPUTE"),

View File

@ -23,8 +23,8 @@ struct AssignToMember {
uint x;
};
float read_from_private(inout float foo_2) {
float _e1 = foo_2;
float read_from_private(inout float foo_1) {
float _e1 = foo_1;
return _e1;
}
@ -37,15 +37,11 @@ void assign_through_ptr_fn(inout uint p) {
return;
}
void assign_array_through_ptr_fn(inout vec4 foo_3[2]) {
foo_3 = vec4[2](vec4(1.0), vec4(2.0));
void assign_array_through_ptr_fn(inout vec4 foo_2[2]) {
foo_2 = vec4[2](vec4(1.0), vec4(2.0));
return;
}
int array_by_value(int a_1[5], int i) {
return a_1[i];
}
uint fetch_arg_ptr_member(inout AssignToMember p_1) {
uint _e2 = p_1.x;
return _e2;

View File

@ -23,8 +23,8 @@ struct AssignToMember {
uint x;
};
float read_from_private(inout float foo_2) {
float _e1 = foo_2;
float read_from_private(inout float foo_1) {
float _e1 = foo_1;
return _e1;
}
@ -37,15 +37,11 @@ void assign_through_ptr_fn(inout uint p) {
return;
}
void assign_array_through_ptr_fn(inout vec4 foo_3[2]) {
foo_3 = vec4[2](vec4(1.0), vec4(2.0));
void assign_array_through_ptr_fn(inout vec4 foo_2[2]) {
foo_2 = vec4[2](vec4(1.0), vec4(2.0));
return;
}
int array_by_value(int a_1[5], int i) {
return a_1[i];
}
uint fetch_arg_ptr_member(inout AssignToMember p_1) {
uint _e2 = p_1.x;
return _e2;

View File

@ -33,8 +33,8 @@ layout(std430) buffer type_13_block_1Fragment { ivec2 _group_0_binding_2_fs; };
layout(location = 0) out vec4 _fs2p_location0;
float read_from_private(inout float foo_2) {
float _e1 = foo_2;
float read_from_private(inout float foo_1) {
float _e1 = foo_1;
return _e1;
}
@ -47,15 +47,11 @@ void assign_through_ptr_fn(inout uint p) {
return;
}
void assign_array_through_ptr_fn(inout vec4 foo_3[2]) {
foo_3 = vec4[2](vec4(1.0), vec4(2.0));
void assign_array_through_ptr_fn(inout vec4 foo_2[2]) {
foo_2 = vec4[2](vec4(1.0), vec4(2.0));
return;
}
int array_by_value(int a_1[5], int i) {
return a_1[i];
}
uint fetch_arg_ptr_member(inout AssignToMember p_1) {
uint _e2 = p_1.x;
return _e2;

View File

@ -106,8 +106,8 @@ void test_matrix_within_array_within_struct_accesses() {
return;
}
float read_from_private(inout float foo_2) {
float _e1 = foo_2;
float read_from_private(inout float foo_1) {
float _e1 = foo_1;
return _e1;
}
@ -120,15 +120,11 @@ void assign_through_ptr_fn(inout uint p) {
return;
}
void assign_array_through_ptr_fn(inout vec4 foo_3[2]) {
foo_3 = vec4[2](vec4(1.0), vec4(2.0));
void assign_array_through_ptr_fn(inout vec4 foo_2[2]) {
foo_2 = vec4[2](vec4(1.0), vec4(2.0));
return;
}
int array_by_value(int a_1[5], int i) {
return a_1[i];
}
uint fetch_arg_ptr_member(inout AssignToMember p_1) {
uint _e2 = p_1.x;
return _e2;
@ -160,10 +156,10 @@ void main() {
mat4x3 _matrix = _group_0_binding_0_vs._matrix;
uvec2 arr_1[2] = _group_0_binding_0_vs.arr;
float b = _group_0_binding_0_vs._matrix[3u][0];
int a_2 = _group_0_binding_0_vs.data[(uint(_group_0_binding_0_vs.data.length()) - 2u)].value;
int a_1 = _group_0_binding_0_vs.data[(uint(_group_0_binding_0_vs.data.length()) - 2u)].value;
ivec2 c = _group_0_binding_2_vs;
float _e33 = read_from_private(foo);
c2_ = int[5](a_2, int(b), 3, 4, 5);
c2_ = int[5](a_1, int(b), 3, 4, 5);
c2_[(vi + 1u)] = 42;
int value = c2_[vi];
float _e47 = test_arr_as_arg(float[5][10](float[10](0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0), float[10](0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0), float[10](0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0), float[10](0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0), float[10](0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0, 0.0)));

View File

@ -205,9 +205,9 @@ void test_matrix_within_array_within_struct_accesses()
return;
}
float read_from_private(inout float foo_2)
float read_from_private(inout float foo_1)
{
float _e1 = foo_2;
float _e1 = foo_1;
return _e1;
}
@ -228,17 +228,12 @@ ret_Constructarray2_float4_ Constructarray2_float4_(float4 arg0, float4 arg1) {
return ret;
}
void assign_array_through_ptr_fn(inout float4 foo_3[2])
void assign_array_through_ptr_fn(inout float4 foo_2[2])
{
foo_3 = Constructarray2_float4_((1.0).xxxx, (2.0).xxxx);
foo_2 = Constructarray2_float4_((1.0).xxxx, (2.0).xxxx);
return;
}
int array_by_value(int a_1[5], int i)
{
return a_1[i];
}
uint fetch_arg_ptr_member(inout AssignToMember p_1)
{
uint _e2 = p_1.x;
@ -299,10 +294,10 @@ float4 foo_vert(uint vi : SV_VertexID) : SV_Position
float4x3 _matrix = float4x3(asfloat(bar.Load3(0+0)), asfloat(bar.Load3(0+16)), asfloat(bar.Load3(0+32)), asfloat(bar.Load3(0+48)));
uint2 arr_1[2] = Constructarray2_uint2_(asuint(bar.Load2(144+0)), asuint(bar.Load2(144+8)));
float b = asfloat(bar.Load(0+3u*16+0));
int a_2 = asint(bar.Load(0+(((NagaBufferLengthRW(bar) - 160) / 8) - 2u)*8+160));
int a_1 = asint(bar.Load(0+(((NagaBufferLengthRW(bar) - 160) / 8) - 2u)*8+160));
int2 c = asint(qux.Load2(0));
const float _e33 = read_from_private(foo);
c2_ = Constructarray5_int_(a_2, int(b), 3, 4, 5);
c2_ = Constructarray5_int_(a_1, int(b), 3, 4, 5);
c2_[(vi + 1u)] = 42;
int value = c2_[vi];
const float _e47 = test_arr_as_arg(ZeroValuearray5_array10_float__());
@ -344,13 +339,6 @@ void assign_through_ptr()
return;
}
float4 foo_1(uint vi_1 : SV_VertexID) : SV_Position
{
int arr_2[5] = Constructarray5_int_(1, 2, 3, 4, 5);
int value_1 = arr_2[vi_1];
return float4((value_1).xxxx);
}
[numthreads(1, 1, 1)]
void assign_to_ptr_components()
{

View File

@ -4,10 +4,6 @@
entry_point:"foo_vert",
target_profile:"vs_5_1",
),
(
entry_point:"foo_1",
target_profile:"vs_5_1",
),
],
fragment:[
(

View File

@ -1691,47 +1691,6 @@
),
],
),
(
name: Some("array_by_value"),
arguments: [
(
name: Some("a"),
ty: 25,
binding: None,
),
(
name: Some("i"),
ty: 2,
binding: None,
),
],
result: Some((
ty: 2,
binding: None,
)),
local_variables: [],
expressions: [
FunctionArgument(0),
FunctionArgument(1),
Access(
base: 0,
index: 1,
),
],
named_expressions: {
0: "a",
1: "i",
},
body: [
Emit((
start: 2,
end: 3,
)),
Return(
value: Some(2),
),
],
),
(
name: Some("fetch_arg_ptr_member"),
arguments: [
@ -2453,82 +2412,6 @@
],
),
),
(
name: "foo",
stage: Vertex,
early_depth_test: None,
workgroup_size: (0, 0, 0),
function: (
name: Some("foo"),
arguments: [
(
name: Some("vi"),
ty: 0,
binding: Some(BuiltIn(VertexIndex)),
),
],
result: Some((
ty: 24,
binding: Some(BuiltIn(Position(
invariant: false,
))),
)),
local_variables: [],
expressions: [
FunctionArgument(0),
Literal(I32(1)),
Literal(I32(2)),
Literal(I32(3)),
Literal(I32(4)),
Literal(I32(5)),
Compose(
ty: 25,
components: [
1,
2,
3,
4,
5,
],
),
Access(
base: 6,
index: 0,
),
Splat(
size: Quad,
value: 7,
),
As(
expr: 8,
kind: Float,
convert: Some(4),
),
],
named_expressions: {
0: "vi",
6: "arr",
7: "value",
},
body: [
Emit((
start: 6,
end: 7,
)),
Emit((
start: 7,
end: 8,
)),
Emit((
start: 8,
end: 10,
)),
Return(
value: Some(9),
),
],
),
),
(
name: "assign_to_ptr_components",
stage: Compute,
@ -2552,35 +2435,35 @@
],
expressions: [
LocalVariable(0),
CallResult(7),
CallResult(6),
LocalVariable(1),
CallResult(9),
CallResult(8),
],
named_expressions: {},
body: [
Call(
function: 8,
function: 7,
arguments: [
0,
],
result: None,
),
Call(
function: 7,
function: 6,
arguments: [
0,
],
result: Some(1),
),
Call(
function: 10,
function: 9,
arguments: [
2,
],
result: None,
),
Call(
function: 9,
function: 8,
arguments: [
2,
],

View File

@ -1691,47 +1691,6 @@
),
],
),
(
name: Some("array_by_value"),
arguments: [
(
name: Some("a"),
ty: 25,
binding: None,
),
(
name: Some("i"),
ty: 2,
binding: None,
),
],
result: Some((
ty: 2,
binding: None,
)),
local_variables: [],
expressions: [
FunctionArgument(0),
FunctionArgument(1),
Access(
base: 0,
index: 1,
),
],
named_expressions: {
0: "a",
1: "i",
},
body: [
Emit((
start: 2,
end: 3,
)),
Return(
value: Some(2),
),
],
),
(
name: Some("fetch_arg_ptr_member"),
arguments: [
@ -2453,82 +2412,6 @@
],
),
),
(
name: "foo",
stage: Vertex,
early_depth_test: None,
workgroup_size: (0, 0, 0),
function: (
name: Some("foo"),
arguments: [
(
name: Some("vi"),
ty: 0,
binding: Some(BuiltIn(VertexIndex)),
),
],
result: Some((
ty: 24,
binding: Some(BuiltIn(Position(
invariant: false,
))),
)),
local_variables: [],
expressions: [
FunctionArgument(0),
Literal(I32(1)),
Literal(I32(2)),
Literal(I32(3)),
Literal(I32(4)),
Literal(I32(5)),
Compose(
ty: 25,
components: [
1,
2,
3,
4,
5,
],
),
Access(
base: 6,
index: 0,
),
Splat(
size: Quad,
value: 7,
),
As(
expr: 8,
kind: Float,
convert: Some(4),
),
],
named_expressions: {
0: "vi",
6: "arr",
7: "value",
},
body: [
Emit((
start: 6,
end: 7,
)),
Emit((
start: 7,
end: 8,
)),
Emit((
start: 8,
end: 10,
)),
Return(
value: Some(9),
),
],
),
),
(
name: "assign_to_ptr_components",
stage: Compute,
@ -2552,35 +2435,35 @@
],
expressions: [
LocalVariable(0),
CallResult(7),
CallResult(6),
LocalVariable(1),
CallResult(9),
CallResult(8),
],
named_expressions: {},
body: [
Call(
function: 8,
function: 7,
arguments: [
0,
],
result: None,
),
Call(
function: 7,
function: 6,
arguments: [
0,
],
result: Some(1),
),
Call(
function: 10,
function: 9,
arguments: [
2,
],
result: None,
),
Call(
function: 9,
function: 8,
arguments: [
2,
],

View File

@ -59,6 +59,23 @@
),
),
),
(
name: None,
inner: Scalar((
kind: Uint,
width: 4,
)),
),
(
name: None,
inner: Vector(
size: Quad,
scalar: (
kind: Float,
width: 4,
),
),
),
],
special_types: (
ray_desc: None,
@ -274,5 +291,82 @@
],
),
],
entry_points: [],
entry_points: [
(
name: "index_let_array_1d",
stage: Vertex,
early_depth_test: None,
workgroup_size: (0, 0, 0),
function: (
name: Some("index_let_array_1d"),
arguments: [
(
name: Some("vi"),
ty: 7,
binding: Some(BuiltIn(VertexIndex)),
),
],
result: Some((
ty: 8,
binding: Some(BuiltIn(Position(
invariant: false,
))),
)),
local_variables: [],
expressions: [
FunctionArgument(0),
Literal(I32(1)),
Literal(I32(2)),
Literal(I32(3)),
Literal(I32(4)),
Literal(I32(5)),
Compose(
ty: 1,
components: [
1,
2,
3,
4,
5,
],
),
Access(
base: 6,
index: 0,
),
Splat(
size: Quad,
value: 7,
),
As(
expr: 8,
kind: Float,
convert: Some(4),
),
],
named_expressions: {
0: "vi",
6: "arr",
7: "value",
},
body: [
Emit((
start: 6,
end: 7,
)),
Emit((
start: 7,
end: 8,
)),
Emit((
start: 8,
end: 10,
)),
Return(
value: Some(9),
),
],
),
),
],
)

View File

@ -59,6 +59,23 @@
),
),
),
(
name: None,
inner: Scalar((
kind: Uint,
width: 4,
)),
),
(
name: None,
inner: Vector(
size: Quad,
scalar: (
kind: Float,
width: 4,
),
),
),
],
special_types: (
ray_desc: None,
@ -274,5 +291,82 @@
],
),
],
entry_points: [],
entry_points: [
(
name: "index_let_array_1d",
stage: Vertex,
early_depth_test: None,
workgroup_size: (0, 0, 0),
function: (
name: Some("index_let_array_1d"),
arguments: [
(
name: Some("vi"),
ty: 7,
binding: Some(BuiltIn(VertexIndex)),
),
],
result: Some((
ty: 8,
binding: Some(BuiltIn(Position(
invariant: false,
))),
)),
local_variables: [],
expressions: [
FunctionArgument(0),
Literal(I32(1)),
Literal(I32(2)),
Literal(I32(3)),
Literal(I32(4)),
Literal(I32(5)),
Compose(
ty: 1,
components: [
1,
2,
3,
4,
5,
],
),
Access(
base: 6,
index: 0,
),
Splat(
size: Quad,
value: 7,
),
As(
expr: 8,
kind: Float,
convert: Some(4),
),
],
named_expressions: {
0: "vi",
6: "arr",
7: "value",
},
body: [
Emit((
start: 6,
end: 7,
)),
Emit((
start: 7,
end: 8,
)),
Emit((
start: 8,
end: 10,
)),
Return(
value: Some(9),
),
],
),
),
],
)

View File

@ -139,9 +139,9 @@ void test_matrix_within_array_within_struct_accesses(
}
float read_from_private(
thread float& foo_2
thread float& foo_1
) {
float _e1 = foo_2;
float _e1 = foo_1;
return _e1;
}
@ -159,19 +159,12 @@ void assign_through_ptr_fn(
}
void assign_array_through_ptr_fn(
thread type_22& foo_3
thread type_22& foo_2
) {
foo_3 = type_22 {metal::float4(1.0), metal::float4(2.0)};
foo_2 = type_22 {metal::float4(1.0), metal::float4(2.0)};
return;
}
int array_by_value(
type_20 a_1,
int i
) {
return a_1.inner[i];
}
uint fetch_arg_ptr_member(
thread AssignToMember& p_1
) {
@ -222,10 +215,10 @@ vertex foo_vertOutput foo_vert(
metal::float4x3 _matrix = bar._matrix;
type_10 arr_1 = bar.arr;
float b = bar._matrix[3u].x;
int a_2 = bar.data[(1 + (_buffer_sizes.size1 - 160 - 8) / 8) - 2u].value;
int a_1 = bar.data[(1 + (_buffer_sizes.size1 - 160 - 8) / 8) - 2u].value;
metal::int2 c = qux;
float _e33 = read_from_private(foo);
c2_ = type_20 {a_2, static_cast<int>(b), 3, 4, 5};
c2_ = type_20 {a_1, static_cast<int>(b), 3, 4, 5};
c2_.inner[vi + 1u] = 42;
int value = c2_.inner[vi];
float _e47 = test_arr_as_arg(type_18 {});
@ -260,20 +253,6 @@ kernel void assign_through_ptr(
}
struct foo_1Input {
};
struct foo_1Output {
metal::float4 member_3 [[position]];
};
vertex foo_1Output foo_1(
uint vi_1 [[vertex_id]]
) {
type_20 arr_2 = type_20 {1, 2, 3, 4, 5};
int value_1 = arr_2.inner[vi_1];
return foo_1Output { static_cast<metal::float4>(metal::int4(value_1)) };
}
kernel void assign_to_ptr_components(
) {
AssignToMember s1_ = {};

View File

@ -1,19 +1,18 @@
; SPIR-V
; Version: 1.1
; Generator: rspirv
; Bound: 364
; Bound: 342
OpCapability Shader
OpExtension "SPV_KHR_storage_buffer_storage_class"
%1 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint Vertex %260 "foo_vert" %255 %258
OpEntryPoint Fragment %312 "foo_frag" %311
OpEntryPoint GLCompute %330 "assign_through_ptr"
OpEntryPoint Vertex %344 "foo" %341 %343
OpEntryPoint GLCompute %354 "assign_to_ptr_components"
OpExecutionMode %312 OriginUpperLeft
OpExecutionMode %330 LocalSize 1 1 1
OpExecutionMode %354 LocalSize 1 1 1
OpEntryPoint Vertex %250 "foo_vert" %245 %248
OpEntryPoint Fragment %303 "foo_frag" %302
OpEntryPoint GLCompute %321 "assign_through_ptr"
OpEntryPoint GLCompute %332 "assign_to_ptr_components"
OpExecutionMode %303 OriginUpperLeft
OpExecutionMode %321 LocalSize 1 1 1
OpExecutionMode %332 LocalSize 1 1 1
OpMemberName %6 0 "a"
OpMemberName %6 1 "b"
OpMemberName %6 2 "c"
@ -52,30 +51,25 @@ OpName %205 "p"
OpName %206 "assign_through_ptr_fn"
OpName %211 "foo"
OpName %212 "assign_array_through_ptr_fn"
OpName %219 "a"
OpName %220 "i"
OpName %221 "array_by_value"
OpName %229 "p"
OpName %230 "fetch_arg_ptr_member"
OpName %236 "p"
OpName %237 "assign_to_arg_ptr_member"
OpName %242 "p"
OpName %243 "fetch_arg_ptr_array_element"
OpName %249 "p"
OpName %250 "assign_to_arg_ptr_array_element"
OpName %255 "vi"
OpName %260 "foo_vert"
OpName %272 "foo"
OpName %273 "c2"
OpName %312 "foo_frag"
OpName %330 "assign_through_ptr"
OpName %335 "val"
OpName %336 "arr"
OpName %341 "vi"
OpName %344 "foo"
OpName %354 "assign_to_ptr_components"
OpName %355 "s1"
OpName %357 "a1"
OpName %219 "p"
OpName %220 "fetch_arg_ptr_member"
OpName %226 "p"
OpName %227 "assign_to_arg_ptr_member"
OpName %232 "p"
OpName %233 "fetch_arg_ptr_array_element"
OpName %239 "p"
OpName %240 "assign_to_arg_ptr_array_element"
OpName %245 "vi"
OpName %250 "foo_vert"
OpName %262 "foo"
OpName %263 "c2"
OpName %303 "foo_frag"
OpName %321 "assign_through_ptr"
OpName %326 "val"
OpName %327 "arr"
OpName %332 "assign_to_ptr_components"
OpName %333 "s1"
OpName %335 "a1"
OpMemberDecorate %6 0 Offset 0
OpMemberDecorate %6 1 Offset 16
OpMemberDecorate %6 2 Offset 28
@ -122,11 +116,9 @@ OpDecorate %55 DescriptorSet 0
OpDecorate %55 Binding 3
OpDecorate %56 Block
OpMemberDecorate %56 0 Offset 0
OpDecorate %255 BuiltIn VertexIndex
OpDecorate %258 BuiltIn Position
OpDecorate %311 Location 0
OpDecorate %341 BuiltIn VertexIndex
OpDecorate %343 BuiltIn Position
OpDecorate %245 BuiltIn VertexIndex
OpDecorate %248 BuiltIn Position
OpDecorate %302 Location 0
%2 = OpTypeVoid
%3 = OpTypeInt 32 0
%4 = OpTypeVector %3 3
@ -239,54 +231,49 @@ OpDecorate %343 BuiltIn Position
%214 = OpConstantComposite %31 %64 %64 %64 %64
%215 = OpConstantComposite %31 %66 %66 %66 %66
%216 = OpConstantComposite %34 %214 %215
%222 = OpTypeFunction %5 %32 %5
%225 = OpTypePointer Function %32
%231 = OpTypeFunction %3 %37
%238 = OpTypeFunction %2 %37
%244 = OpTypeFunction %3 %40
%251 = OpTypeFunction %2 %40
%256 = OpTypePointer Input %3
%255 = OpVariable %256 Input
%259 = OpTypePointer Output %31
%258 = OpVariable %259 Output
%262 = OpTypePointer StorageBuffer %23
%265 = OpConstant %8 0.0
%266 = OpConstant %3 3
%267 = OpConstant %5 3
%268 = OpConstant %5 4
%269 = OpConstant %5 5
%270 = OpConstant %5 42
%271 = OpConstantNull %29
%274 = OpConstantNull %32
%279 = OpTypePointer StorageBuffer %9
%282 = OpTypePointer StorageBuffer %18
%285 = OpTypePointer StorageBuffer %10
%286 = OpTypePointer StorageBuffer %8
%289 = OpTypePointer StorageBuffer %19
%292 = OpTypePointer StorageBuffer %7
%293 = OpTypePointer StorageBuffer %5
%305 = OpTypeVector %5 4
%311 = OpVariable %259 Output
%314 = OpConstantComposite %10 %265 %265 %265
%315 = OpConstantComposite %10 %64 %64 %64
%316 = OpConstantComposite %10 %66 %66 %66
%317 = OpConstantComposite %10 %68 %68 %68
%318 = OpConstantComposite %9 %314 %315 %316 %317
%319 = OpConstantComposite %17 %41 %41
%320 = OpConstantComposite %17 %104 %104
%321 = OpConstantComposite %18 %319 %320
%322 = OpConstantNull %23
%323 = OpConstantComposite %31 %265 %265 %265 %265
%331 = OpConstant %3 33
%332 = OpConstantComposite %31 %72 %72 %72 %72
%333 = OpConstantComposite %31 %142 %142 %142 %142
%334 = OpConstantComposite %34 %332 %333
%341 = OpVariable %256 Input
%343 = OpVariable %259 Output
%345 = OpConstant %5 2
%346 = OpConstantComposite %32 %63 %345 %267 %268 %269
%356 = OpConstantNull %36
%358 = OpConstantNull %38
%221 = OpTypeFunction %3 %37
%228 = OpTypeFunction %2 %37
%234 = OpTypeFunction %3 %40
%241 = OpTypeFunction %2 %40
%246 = OpTypePointer Input %3
%245 = OpVariable %246 Input
%249 = OpTypePointer Output %31
%248 = OpVariable %249 Output
%252 = OpTypePointer StorageBuffer %23
%255 = OpConstant %8 0.0
%256 = OpConstant %3 3
%257 = OpConstant %5 3
%258 = OpConstant %5 4
%259 = OpConstant %5 5
%260 = OpConstant %5 42
%261 = OpConstantNull %29
%264 = OpTypePointer Function %32
%265 = OpConstantNull %32
%270 = OpTypePointer StorageBuffer %9
%273 = OpTypePointer StorageBuffer %18
%276 = OpTypePointer StorageBuffer %10
%277 = OpTypePointer StorageBuffer %8
%280 = OpTypePointer StorageBuffer %19
%283 = OpTypePointer StorageBuffer %7
%284 = OpTypePointer StorageBuffer %5
%296 = OpTypeVector %5 4
%302 = OpVariable %249 Output
%305 = OpConstantComposite %10 %255 %255 %255
%306 = OpConstantComposite %10 %64 %64 %64
%307 = OpConstantComposite %10 %66 %66 %66
%308 = OpConstantComposite %10 %68 %68 %68
%309 = OpConstantComposite %9 %305 %306 %307 %308
%310 = OpConstantComposite %17 %41 %41
%311 = OpConstantComposite %17 %104 %104
%312 = OpConstantComposite %18 %310 %311
%313 = OpConstantNull %23
%314 = OpConstantComposite %31 %255 %255 %255 %255
%322 = OpConstant %3 33
%323 = OpConstantComposite %31 %72 %72 %72 %72
%324 = OpConstantComposite %31 %142 %142 %142 %142
%325 = OpConstantComposite %34 %323 %324
%334 = OpConstantNull %36
%336 = OpConstantNull %38
%59 = OpFunction %2 None %60
%58 = OpLabel
%87 = OpVariable %88 Function %63
@ -430,146 +417,120 @@ OpBranch %217
OpStore %211 %216
OpReturn
OpFunctionEnd
%221 = OpFunction %5 None %222
%219 = OpFunctionParameter %32
%220 = OpFunctionParameter %5
%220 = OpFunction %3 None %221
%219 = OpFunctionParameter %37
%218 = OpLabel
%224 = OpVariable %225 Function
OpBranch %223
%223 = OpLabel
OpStore %224 %219
%226 = OpAccessChain %88 %224 %220
%227 = OpLoad %5 %226
OpReturnValue %227
OpBranch %222
%222 = OpLabel
%223 = OpAccessChain %33 %219 %41
%224 = OpLoad %3 %223
OpReturnValue %224
OpFunctionEnd
%230 = OpFunction %3 None %231
%229 = OpFunctionParameter %37
%228 = OpLabel
OpBranch %232
%232 = OpLabel
%233 = OpAccessChain %33 %229 %41
%234 = OpLoad %3 %233
OpReturnValue %234
%227 = OpFunction %2 None %228
%226 = OpFunctionParameter %37
%225 = OpLabel
OpBranch %229
%229 = OpLabel
%230 = OpAccessChain %33 %226 %41
OpStore %230 %16
OpReturn
OpFunctionEnd
%237 = OpFunction %2 None %238
%236 = OpFunctionParameter %37
%233 = OpFunction %3 None %234
%232 = OpFunctionParameter %40
%231 = OpLabel
OpBranch %235
%235 = OpLabel
OpBranch %239
%239 = OpLabel
%240 = OpAccessChain %33 %236 %41
OpStore %240 %16
%236 = OpAccessChain %33 %232 %104
%237 = OpLoad %3 %236
OpReturnValue %237
OpFunctionEnd
%240 = OpFunction %2 None %241
%239 = OpFunctionParameter %40
%238 = OpLabel
OpBranch %242
%242 = OpLabel
%243 = OpAccessChain %33 %239 %104
OpStore %243 %16
OpReturn
OpFunctionEnd
%243 = OpFunction %3 None %244
%242 = OpFunctionParameter %40
%241 = OpLabel
OpBranch %245
%245 = OpLabel
%246 = OpAccessChain %33 %242 %104
%247 = OpLoad %3 %246
OpReturnValue %247
OpFunctionEnd
%250 = OpFunction %2 None %251
%249 = OpFunctionParameter %40
%248 = OpLabel
OpBranch %252
%252 = OpLabel
%253 = OpAccessChain %33 %249 %104
OpStore %253 %16
%250 = OpFunction %2 None %60
%244 = OpLabel
%262 = OpVariable %27 Function %255
%263 = OpVariable %264 Function %265
%247 = OpLoad %3 %245
%251 = OpAccessChain %61 %49 %41
%253 = OpAccessChain %252 %52 %41
%254 = OpAccessChain %136 %55 %41
OpBranch %266
%266 = OpLabel
%267 = OpLoad %8 %262
OpStore %262 %64
%268 = OpFunctionCall %2 %59
%269 = OpFunctionCall %2 %135
%271 = OpAccessChain %270 %47 %41
%272 = OpLoad %9 %271
%274 = OpAccessChain %273 %47 %39
%275 = OpLoad %18 %274
%278 = OpAccessChain %277 %47 %41 %256 %41
%279 = OpLoad %8 %278
%281 = OpArrayLength %3 %47 5
%282 = OpISub %3 %281 %14
%285 = OpAccessChain %284 %47 %30 %282 %41
%286 = OpLoad %5 %285
%287 = OpLoad %23 %253
%288 = OpFunctionCall %8 %193 %262
%289 = OpConvertFToS %5 %279
%290 = OpCompositeConstruct %32 %286 %289 %257 %258 %259
OpStore %263 %290
%291 = OpIAdd %3 %247 %104
%292 = OpAccessChain %88 %263 %291
OpStore %292 %260
%293 = OpAccessChain %88 %263 %247
%294 = OpLoad %5 %293
%295 = OpFunctionCall %8 %199 %261
%297 = OpCompositeConstruct %296 %294 %294 %294 %294
%298 = OpConvertSToF %31 %297
%299 = OpMatrixTimesVector %10 %272 %298
%300 = OpCompositeConstruct %31 %299 %66
OpStore %248 %300
OpReturn
OpFunctionEnd
%260 = OpFunction %2 None %60
%254 = OpLabel
%272 = OpVariable %27 Function %265
%273 = OpVariable %225 Function %274
%257 = OpLoad %3 %255
%261 = OpAccessChain %61 %49 %41
%263 = OpAccessChain %262 %52 %41
%264 = OpAccessChain %136 %55 %41
OpBranch %275
%275 = OpLabel
%276 = OpLoad %8 %272
OpStore %272 %64
%277 = OpFunctionCall %2 %59
%278 = OpFunctionCall %2 %135
%280 = OpAccessChain %279 %47 %41
%281 = OpLoad %9 %280
%283 = OpAccessChain %282 %47 %39
%284 = OpLoad %18 %283
%287 = OpAccessChain %286 %47 %41 %266 %41
%288 = OpLoad %8 %287
%290 = OpArrayLength %3 %47 5
%291 = OpISub %3 %290 %14
%294 = OpAccessChain %293 %47 %30 %291 %41
%295 = OpLoad %5 %294
%296 = OpLoad %23 %263
%297 = OpFunctionCall %8 %193 %272
%298 = OpConvertFToS %5 %288
%299 = OpCompositeConstruct %32 %295 %298 %267 %268 %269
OpStore %273 %299
%300 = OpIAdd %3 %257 %104
%301 = OpAccessChain %88 %273 %300
OpStore %301 %270
%302 = OpAccessChain %88 %273 %257
%303 = OpLoad %5 %302
%304 = OpFunctionCall %8 %199 %271
%306 = OpCompositeConstruct %305 %303 %303 %303 %303
%307 = OpConvertSToF %31 %306
%308 = OpMatrixTimesVector %10 %281 %307
%309 = OpCompositeConstruct %31 %308 %66
OpStore %258 %309
%303 = OpFunction %2 None %60
%301 = OpLabel
%304 = OpAccessChain %252 %52 %41
OpBranch %315
%315 = OpLabel
%316 = OpAccessChain %277 %47 %41 %104 %14
OpStore %316 %64
%317 = OpAccessChain %270 %47 %41
OpStore %317 %309
%318 = OpAccessChain %273 %47 %39
OpStore %318 %312
%319 = OpAccessChain %284 %47 %30 %104 %41
OpStore %319 %63
OpStore %304 %313
OpStore %302 %314
OpReturn
OpFunctionEnd
%312 = OpFunction %2 None %60
%310 = OpLabel
%313 = OpAccessChain %262 %52 %41
OpBranch %324
%324 = OpLabel
%325 = OpAccessChain %286 %47 %41 %104 %14
OpStore %325 %64
%326 = OpAccessChain %279 %47 %41
OpStore %326 %318
%327 = OpAccessChain %282 %47 %39
OpStore %327 %321
%328 = OpAccessChain %293 %47 %30 %104 %41
OpStore %328 %63
OpStore %313 %322
OpStore %311 %323
%321 = OpFunction %2 None %60
%320 = OpLabel
%326 = OpVariable %33 Function %322
%327 = OpVariable %35 Function %325
OpBranch %328
%328 = OpLabel
%329 = OpFunctionCall %2 %206 %326
%330 = OpFunctionCall %2 %212 %327
OpReturn
OpFunctionEnd
%330 = OpFunction %2 None %60
%329 = OpLabel
%335 = OpVariable %33 Function %331
%336 = OpVariable %35 Function %334
%332 = OpFunction %2 None %60
%331 = OpLabel
%333 = OpVariable %37 Function %334
%335 = OpVariable %40 Function %336
OpBranch %337
%337 = OpLabel
%338 = OpFunctionCall %2 %206 %335
%339 = OpFunctionCall %2 %212 %336
OpReturn
OpFunctionEnd
%344 = OpFunction %2 None %60
%340 = OpLabel
%348 = OpVariable %225 Function
%342 = OpLoad %3 %341
OpBranch %347
%347 = OpLabel
OpStore %348 %346
%349 = OpAccessChain %88 %348 %342
%350 = OpLoad %5 %349
%351 = OpCompositeConstruct %305 %350 %350 %350 %350
%352 = OpConvertSToF %31 %351
OpStore %343 %352
OpReturn
OpFunctionEnd
%354 = OpFunction %2 None %60
%353 = OpLabel
%355 = OpVariable %37 Function %356
%357 = OpVariable %40 Function %358
OpBranch %359
%359 = OpLabel
%360 = OpFunctionCall %2 %237 %355
%361 = OpFunctionCall %3 %230 %355
%362 = OpFunctionCall %2 %250 %357
%363 = OpFunctionCall %3 %243 %357
%338 = OpFunctionCall %2 %227 %333
%339 = OpFunctionCall %3 %220 %333
%340 = OpFunctionCall %2 %240 %335
%341 = OpFunctionCall %3 %233 %335
OpReturn
OpFunctionEnd

View File

@ -1,14 +1,16 @@
; SPIR-V
; Version: 1.1
; Generator: rspirv
; Bound: 59
; Bound: 77
OpCapability Shader
OpCapability Linkage
%1 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint Vertex %66 "index_let_array_1d" %61 %64
OpDecorate %4 ArrayStride 4
OpDecorate %7 ArrayStride 4
OpDecorate %9 ArrayStride 8
OpDecorate %61 BuiltIn VertexIndex
OpDecorate %64 BuiltIn Position
%2 = OpTypeVoid
%3 = OpTypeInt 32 1
%6 = OpTypeInt 32 0
@ -20,61 +22,84 @@ OpDecorate %9 ArrayStride 8
%10 = OpTypeFloat 32
%12 = OpTypeVector %10 2
%11 = OpTypeMatrix %12 2
%17 = OpTypeFunction %3 %4 %3
%20 = OpTypePointer Function %4
%21 = OpTypePointer Function %3
%28 = OpTypeFunction %3 %3 %3
%29 = OpConstant %3 1
%30 = OpConstant %3 2
%31 = OpConstantComposite %7 %29 %30
%32 = OpConstant %3 3
%33 = OpConstant %3 4
%34 = OpConstantComposite %7 %32 %33
%35 = OpConstantComposite %9 %31 %34
%38 = OpTypePointer Function %9
%45 = OpTypeFunction %10 %3 %3
%46 = OpConstant %10 1.0
%47 = OpConstant %10 2.0
%48 = OpConstant %10 3.0
%49 = OpConstant %10 4.0
%50 = OpConstantComposite %12 %46 %47
%51 = OpConstantComposite %12 %48 %49
%52 = OpConstantComposite %11 %50 %51
%55 = OpTypePointer Function %11
%56 = OpTypePointer Function %10
%16 = OpFunction %3 None %17
%14 = OpFunctionParameter %4
%15 = OpFunctionParameter %3
%13 = OpLabel
%19 = OpVariable %20 Function
OpBranch %18
%18 = OpLabel
OpStore %19 %14
%22 = OpAccessChain %21 %19 %15
%23 = OpLoad %3 %22
OpReturnValue %23
%13 = OpTypeVector %10 4
%18 = OpTypeFunction %3 %4 %3
%21 = OpTypePointer Function %4
%22 = OpTypePointer Function %3
%29 = OpTypeFunction %3 %3 %3
%30 = OpConstant %3 1
%31 = OpConstant %3 2
%32 = OpConstantComposite %7 %30 %31
%33 = OpConstant %3 3
%34 = OpConstant %3 4
%35 = OpConstantComposite %7 %33 %34
%36 = OpConstantComposite %9 %32 %35
%39 = OpTypePointer Function %9
%46 = OpTypeFunction %10 %3 %3
%47 = OpConstant %10 1.0
%48 = OpConstant %10 2.0
%49 = OpConstant %10 3.0
%50 = OpConstant %10 4.0
%51 = OpConstantComposite %12 %47 %48
%52 = OpConstantComposite %12 %49 %50
%53 = OpConstantComposite %11 %51 %52
%56 = OpTypePointer Function %11
%57 = OpTypePointer Function %10
%62 = OpTypePointer Input %6
%61 = OpVariable %62 Input
%65 = OpTypePointer Output %13
%64 = OpVariable %65 Output
%67 = OpTypeFunction %2
%68 = OpConstant %3 5
%69 = OpConstantComposite %4 %30 %31 %33 %34 %68
%74 = OpTypeVector %3 4
%17 = OpFunction %3 None %18
%15 = OpFunctionParameter %4
%16 = OpFunctionParameter %3
%14 = OpLabel
%20 = OpVariable %21 Function
OpBranch %19
%19 = OpLabel
OpStore %20 %15
%23 = OpAccessChain %22 %20 %16
%24 = OpLoad %3 %23
OpReturnValue %24
OpFunctionEnd
%27 = OpFunction %3 None %28
%25 = OpFunctionParameter %3
%28 = OpFunction %3 None %29
%26 = OpFunctionParameter %3
%24 = OpLabel
%37 = OpVariable %38 Function
OpBranch %36
%36 = OpLabel
OpStore %37 %35
%39 = OpAccessChain %21 %37 %25 %26
%40 = OpLoad %3 %39
OpReturnValue %40
%27 = OpFunctionParameter %3
%25 = OpLabel
%38 = OpVariable %39 Function
OpBranch %37
%37 = OpLabel
OpStore %38 %36
%40 = OpAccessChain %22 %38 %26 %27
%41 = OpLoad %3 %40
OpReturnValue %41
OpFunctionEnd
%44 = OpFunction %10 None %45
%42 = OpFunctionParameter %3
%45 = OpFunction %10 None %46
%43 = OpFunctionParameter %3
%41 = OpLabel
%54 = OpVariable %55 Function
OpBranch %53
%53 = OpLabel
OpStore %54 %52
%57 = OpAccessChain %56 %54 %42 %43
%58 = OpLoad %10 %57
OpReturnValue %58
%44 = OpFunctionParameter %3
%42 = OpLabel
%55 = OpVariable %56 Function
OpBranch %54
%54 = OpLabel
OpStore %55 %53
%58 = OpAccessChain %57 %55 %43 %44
%59 = OpLoad %10 %58
OpReturnValue %59
OpFunctionEnd
%66 = OpFunction %2 None %67
%60 = OpLabel
%71 = OpVariable %21 Function
%63 = OpLoad %6 %61
OpBranch %70
%70 = OpLabel
OpStore %71 %69
%72 = OpAccessChain %22 %71 %63
%73 = OpLoad %3 %72
%75 = OpCompositeConstruct %74 %73 %73 %73 %73
%76 = OpConvertSToF %13 %75
OpStore %64 %76
OpReturn
OpFunctionEnd

View File

@ -111,8 +111,8 @@ fn test_matrix_within_array_within_struct_accesses() {
return;
}
fn read_from_private(foo_2: ptr<function, f32>) -> f32 {
let _e1 = (*foo_2);
fn read_from_private(foo_1: ptr<function, f32>) -> f32 {
let _e1 = (*foo_1);
return _e1;
}
@ -125,15 +125,11 @@ fn assign_through_ptr_fn(p: ptr<function, u32>) {
return;
}
fn assign_array_through_ptr_fn(foo_3: ptr<function, array<vec4<f32>, 2>>) {
(*foo_3) = array<vec4<f32>, 2>(vec4(1f), vec4(2f));
fn assign_array_through_ptr_fn(foo_2: ptr<function, array<vec4<f32>, 2>>) {
(*foo_2) = array<vec4<f32>, 2>(vec4(1f), vec4(2f));
return;
}
fn array_by_value(a_1: array<i32, 5>, i: i32) -> i32 {
return a_1[i];
}
fn fetch_arg_ptr_member(p_1: ptr<function, AssignToMember>) -> u32 {
let _e2 = (*p_1).x;
return _e2;
@ -166,11 +162,11 @@ fn foo_vert(@builtin(vertex_index) vi: u32) -> @builtin(position) vec4<f32> {
let _matrix = bar._matrix;
let arr_1 = bar.arr;
let b = bar._matrix[3u][0];
let a_2 = bar.data[(arrayLength((&bar.data)) - 2u)].value;
let a_1 = bar.data[(arrayLength((&bar.data)) - 2u)].value;
let c = qux;
let data_pointer = (&bar.data[0].value);
let _e33 = read_from_private((&foo));
c2_ = array<i32, 5>(a_2, i32(b), 3i, 4i, 5i);
c2_ = array<i32, 5>(a_1, i32(b), 3i, 4i, 5i);
c2_[(vi + 1u)] = 42i;
let value = c2_[vi];
let _e47 = test_arr_as_arg(array<array<f32, 10>, 5>());
@ -197,13 +193,6 @@ fn assign_through_ptr() {
return;
}
@vertex
fn foo_1(@builtin(vertex_index) vi_1: u32) -> @builtin(position) vec4<f32> {
const arr_2 = array<i32, 5>(1i, 2i, 3i, 4i, 5i);
let value_1 = arr_2[vi_1];
return vec4<f32>(vec4(value_1));
}
@compute @workgroup_size(1, 1, 1)
fn assign_to_ptr_components() {
var s1_: AssignToMember;