diff --git a/reference/shaders-msl-no-opt/packing/matrix-2x2-std430.comp b/reference/shaders-msl-no-opt/packing/matrix-2x2-std430.comp new file mode 100644 index 00000000..a0fdf9d9 --- /dev/null +++ b/reference/shaders-msl-no-opt/packing/matrix-2x2-std430.comp @@ -0,0 +1,76 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" + +#include +#include + +using namespace metal; + +struct SSBOCol +{ + float2x2 col_major0; + float2x2 col_major1; +}; + +struct SSBORow +{ + float2x2 row_major0; + float2x2 row_major1; +}; + +void load_store_to_variable_col_major(device SSBOCol& v_29) +{ + float2x2 loaded = v_29.col_major0; + v_29.col_major1 = loaded; +} + +void load_store_to_variable_row_major(device SSBORow& v_41) +{ + float2x2 loaded = transpose(v_41.row_major0); + v_41.row_major0 = transpose(loaded); +} + +void copy_col_major_to_col_major(device SSBOCol& v_29) +{ + v_29.col_major0 = v_29.col_major1; +} + +void copy_col_major_to_row_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_41.row_major0 = transpose(v_29.col_major0); +} + +void copy_row_major_to_col_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0 = transpose(v_41.row_major0); +} + +void copy_row_major_to_row_major(device SSBORow& v_41) +{ + v_41.row_major0 = v_41.row_major1; +} + +void copy_columns(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[1] = float2(v_41.row_major0[0][1], v_41.row_major0[1][1]); + v_41.row_major0[0][1] = v_29.col_major0[1].x; + v_41.row_major0[1][1] = v_29.col_major0[1].y; +} + +void copy_elements(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[0].y = v_41.row_major0[1u][0]; + v_41.row_major0[1u][0] = v_29.col_major0[0].y; +} + +kernel void main0(device SSBOCol& v_29 [[buffer(0)]], device SSBORow& v_41 [[buffer(1)]]) +{ + load_store_to_variable_col_major(v_29); + load_store_to_variable_row_major(v_41); + copy_col_major_to_col_major(v_29); + copy_col_major_to_row_major(v_29, v_41); + copy_row_major_to_col_major(v_29, v_41); + copy_row_major_to_row_major(v_41); + copy_columns(v_29, v_41); + copy_elements(v_29, v_41); +} + diff --git a/reference/shaders-msl-no-opt/packing/matrix-2x3-std430.comp b/reference/shaders-msl-no-opt/packing/matrix-2x3-std430.comp new file mode 100644 index 00000000..b86de347 --- /dev/null +++ b/reference/shaders-msl-no-opt/packing/matrix-2x3-std430.comp @@ -0,0 +1,77 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" + +#include +#include + +using namespace metal; + +struct SSBOCol +{ + float2x3 col_major0; + float2x3 col_major1; +}; + +struct SSBORow +{ + float3x2 row_major0; + float3x2 row_major1; +}; + +void load_store_to_variable_col_major(device SSBOCol& v_29) +{ + float2x3 loaded = v_29.col_major0; + v_29.col_major1 = loaded; +} + +void load_store_to_variable_row_major(device SSBORow& v_41) +{ + float2x3 loaded = transpose(v_41.row_major0); + v_41.row_major0 = transpose(loaded); +} + +void copy_col_major_to_col_major(device SSBOCol& v_29) +{ + v_29.col_major0 = v_29.col_major1; +} + +void copy_col_major_to_row_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_41.row_major0 = transpose(v_29.col_major0); +} + +void copy_row_major_to_col_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0 = transpose(v_41.row_major0); +} + +void copy_row_major_to_row_major(device SSBORow& v_41) +{ + v_41.row_major0 = v_41.row_major1; +} + +void copy_columns(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[1] = float3(v_41.row_major0[0][1], v_41.row_major0[1][1], v_41.row_major0[2][1]); + v_41.row_major0[0][1] = v_29.col_major0[1].x; + v_41.row_major0[1][1] = v_29.col_major0[1].y; + v_41.row_major0[2][1] = v_29.col_major0[1].z; +} + +void copy_elements(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[0].y = v_41.row_major0[1u][0]; + v_41.row_major0[1u][0] = v_29.col_major0[0].y; +} + +kernel void main0(device SSBOCol& v_29 [[buffer(0)]], device SSBORow& v_41 [[buffer(1)]]) +{ + load_store_to_variable_col_major(v_29); + load_store_to_variable_row_major(v_41); + copy_col_major_to_col_major(v_29); + copy_col_major_to_row_major(v_29, v_41); + copy_row_major_to_col_major(v_29, v_41); + copy_row_major_to_row_major(v_41); + copy_columns(v_29, v_41); + copy_elements(v_29, v_41); +} + diff --git a/reference/shaders-msl-no-opt/packing/matrix-2x4-std430.comp b/reference/shaders-msl-no-opt/packing/matrix-2x4-std430.comp new file mode 100644 index 00000000..2498d6b2 --- /dev/null +++ b/reference/shaders-msl-no-opt/packing/matrix-2x4-std430.comp @@ -0,0 +1,78 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" + +#include +#include + +using namespace metal; + +struct SSBOCol +{ + float2x4 col_major0; + float2x4 col_major1; +}; + +struct SSBORow +{ + float4x2 row_major0; + float4x2 row_major1; +}; + +void load_store_to_variable_col_major(device SSBOCol& v_29) +{ + float2x4 loaded = v_29.col_major0; + v_29.col_major1 = loaded; +} + +void load_store_to_variable_row_major(device SSBORow& v_41) +{ + float2x4 loaded = transpose(v_41.row_major0); + v_41.row_major0 = transpose(loaded); +} + +void copy_col_major_to_col_major(device SSBOCol& v_29) +{ + v_29.col_major0 = v_29.col_major1; +} + +void copy_col_major_to_row_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_41.row_major0 = transpose(v_29.col_major0); +} + +void copy_row_major_to_col_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0 = transpose(v_41.row_major0); +} + +void copy_row_major_to_row_major(device SSBORow& v_41) +{ + v_41.row_major0 = v_41.row_major1; +} + +void copy_columns(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[1] = float4(v_41.row_major0[0][1], v_41.row_major0[1][1], v_41.row_major0[2][1], v_41.row_major0[3][1]); + v_41.row_major0[0][1] = v_29.col_major0[1].x; + v_41.row_major0[1][1] = v_29.col_major0[1].y; + v_41.row_major0[2][1] = v_29.col_major0[1].z; + v_41.row_major0[3][1] = v_29.col_major0[1].w; +} + +void copy_elements(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[0].y = v_41.row_major0[1u][0]; + v_41.row_major0[1u][0] = v_29.col_major0[0].y; +} + +kernel void main0(device SSBOCol& v_29 [[buffer(0)]], device SSBORow& v_41 [[buffer(1)]]) +{ + load_store_to_variable_col_major(v_29); + load_store_to_variable_row_major(v_41); + copy_col_major_to_col_major(v_29); + copy_col_major_to_row_major(v_29, v_41); + copy_row_major_to_col_major(v_29, v_41); + copy_row_major_to_row_major(v_41); + copy_columns(v_29, v_41); + copy_elements(v_29, v_41); +} + diff --git a/reference/shaders-msl-no-opt/packing/matrix-3x2-std430.comp b/reference/shaders-msl-no-opt/packing/matrix-3x2-std430.comp new file mode 100644 index 00000000..04bd790b --- /dev/null +++ b/reference/shaders-msl-no-opt/packing/matrix-3x2-std430.comp @@ -0,0 +1,76 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" + +#include +#include + +using namespace metal; + +struct SSBOCol +{ + float3x2 col_major0; + float3x2 col_major1; +}; + +struct SSBORow +{ + float2x3 row_major0; + float2x3 row_major1; +}; + +void load_store_to_variable_col_major(device SSBOCol& v_29) +{ + float3x2 loaded = v_29.col_major0; + v_29.col_major1 = loaded; +} + +void load_store_to_variable_row_major(device SSBORow& v_41) +{ + float3x2 loaded = transpose(v_41.row_major0); + v_41.row_major0 = transpose(loaded); +} + +void copy_col_major_to_col_major(device SSBOCol& v_29) +{ + v_29.col_major0 = v_29.col_major1; +} + +void copy_col_major_to_row_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_41.row_major0 = transpose(v_29.col_major0); +} + +void copy_row_major_to_col_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0 = transpose(v_41.row_major0); +} + +void copy_row_major_to_row_major(device SSBORow& v_41) +{ + v_41.row_major0 = v_41.row_major1; +} + +void copy_columns(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[1] = float2(v_41.row_major0[0][1], v_41.row_major0[1][1]); + v_41.row_major0[0][1] = v_29.col_major0[1].x; + v_41.row_major0[1][1] = v_29.col_major0[1].y; +} + +void copy_elements(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[0].y = v_41.row_major0[1u][0]; + v_41.row_major0[1u][0] = v_29.col_major0[0].y; +} + +kernel void main0(device SSBOCol& v_29 [[buffer(0)]], device SSBORow& v_41 [[buffer(1)]]) +{ + load_store_to_variable_col_major(v_29); + load_store_to_variable_row_major(v_41); + copy_col_major_to_col_major(v_29); + copy_col_major_to_row_major(v_29, v_41); + copy_row_major_to_col_major(v_29, v_41); + copy_row_major_to_row_major(v_41); + copy_columns(v_29, v_41); + copy_elements(v_29, v_41); +} + diff --git a/reference/shaders-msl-no-opt/packing/matrix-3x3-std430.comp b/reference/shaders-msl-no-opt/packing/matrix-3x3-std430.comp new file mode 100644 index 00000000..ceff01e4 --- /dev/null +++ b/reference/shaders-msl-no-opt/packing/matrix-3x3-std430.comp @@ -0,0 +1,77 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" + +#include +#include + +using namespace metal; + +struct SSBOCol +{ + float3x3 col_major0; + float3x3 col_major1; +}; + +struct SSBORow +{ + float3x3 row_major0; + float3x3 row_major1; +}; + +void load_store_to_variable_col_major(device SSBOCol& v_29) +{ + float3x3 loaded = v_29.col_major0; + v_29.col_major1 = loaded; +} + +void load_store_to_variable_row_major(device SSBORow& v_41) +{ + float3x3 loaded = transpose(v_41.row_major0); + v_41.row_major0 = transpose(loaded); +} + +void copy_col_major_to_col_major(device SSBOCol& v_29) +{ + v_29.col_major0 = v_29.col_major1; +} + +void copy_col_major_to_row_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_41.row_major0 = transpose(v_29.col_major0); +} + +void copy_row_major_to_col_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0 = transpose(v_41.row_major0); +} + +void copy_row_major_to_row_major(device SSBORow& v_41) +{ + v_41.row_major0 = v_41.row_major1; +} + +void copy_columns(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[1] = float3(v_41.row_major0[0][1], v_41.row_major0[1][1], v_41.row_major0[2][1]); + v_41.row_major0[0][1] = v_29.col_major0[1].x; + v_41.row_major0[1][1] = v_29.col_major0[1].y; + v_41.row_major0[2][1] = v_29.col_major0[1].z; +} + +void copy_elements(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[0].y = v_41.row_major0[1u][0]; + v_41.row_major0[1u][0] = v_29.col_major0[0].y; +} + +kernel void main0(device SSBOCol& v_29 [[buffer(0)]], device SSBORow& v_41 [[buffer(1)]]) +{ + load_store_to_variable_col_major(v_29); + load_store_to_variable_row_major(v_41); + copy_col_major_to_col_major(v_29); + copy_col_major_to_row_major(v_29, v_41); + copy_row_major_to_col_major(v_29, v_41); + copy_row_major_to_row_major(v_41); + copy_columns(v_29, v_41); + copy_elements(v_29, v_41); +} + diff --git a/reference/shaders-msl-no-opt/packing/matrix-3x4-std430.comp b/reference/shaders-msl-no-opt/packing/matrix-3x4-std430.comp new file mode 100644 index 00000000..9f6a20ed --- /dev/null +++ b/reference/shaders-msl-no-opt/packing/matrix-3x4-std430.comp @@ -0,0 +1,78 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" + +#include +#include + +using namespace metal; + +struct SSBOCol +{ + float3x4 col_major0; + float3x4 col_major1; +}; + +struct SSBORow +{ + float4x3 row_major0; + float4x3 row_major1; +}; + +void load_store_to_variable_col_major(device SSBOCol& v_29) +{ + float3x4 loaded = v_29.col_major0; + v_29.col_major1 = loaded; +} + +void load_store_to_variable_row_major(device SSBORow& v_41) +{ + float3x4 loaded = transpose(v_41.row_major0); + v_41.row_major0 = transpose(loaded); +} + +void copy_col_major_to_col_major(device SSBOCol& v_29) +{ + v_29.col_major0 = v_29.col_major1; +} + +void copy_col_major_to_row_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_41.row_major0 = transpose(v_29.col_major0); +} + +void copy_row_major_to_col_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0 = transpose(v_41.row_major0); +} + +void copy_row_major_to_row_major(device SSBORow& v_41) +{ + v_41.row_major0 = v_41.row_major1; +} + +void copy_columns(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[1] = float4(v_41.row_major0[0][1], v_41.row_major0[1][1], v_41.row_major0[2][1], v_41.row_major0[3][1]); + v_41.row_major0[0][1] = v_29.col_major0[1].x; + v_41.row_major0[1][1] = v_29.col_major0[1].y; + v_41.row_major0[2][1] = v_29.col_major0[1].z; + v_41.row_major0[3][1] = v_29.col_major0[1].w; +} + +void copy_elements(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[0].y = v_41.row_major0[1u][0]; + v_41.row_major0[1u][0] = v_29.col_major0[0].y; +} + +kernel void main0(device SSBOCol& v_29 [[buffer(0)]], device SSBORow& v_41 [[buffer(1)]]) +{ + load_store_to_variable_col_major(v_29); + load_store_to_variable_row_major(v_41); + copy_col_major_to_col_major(v_29); + copy_col_major_to_row_major(v_29, v_41); + copy_row_major_to_col_major(v_29, v_41); + copy_row_major_to_row_major(v_41); + copy_columns(v_29, v_41); + copy_elements(v_29, v_41); +} + diff --git a/reference/shaders-msl-no-opt/packing/matrix-4x2-std430.comp b/reference/shaders-msl-no-opt/packing/matrix-4x2-std430.comp new file mode 100644 index 00000000..f863373e --- /dev/null +++ b/reference/shaders-msl-no-opt/packing/matrix-4x2-std430.comp @@ -0,0 +1,76 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" + +#include +#include + +using namespace metal; + +struct SSBOCol +{ + float4x2 col_major0; + float4x2 col_major1; +}; + +struct SSBORow +{ + float2x4 row_major0; + float2x4 row_major1; +}; + +void load_store_to_variable_col_major(device SSBOCol& v_29) +{ + float4x2 loaded = v_29.col_major0; + v_29.col_major1 = loaded; +} + +void load_store_to_variable_row_major(device SSBORow& v_41) +{ + float4x2 loaded = transpose(v_41.row_major0); + v_41.row_major0 = transpose(loaded); +} + +void copy_col_major_to_col_major(device SSBOCol& v_29) +{ + v_29.col_major0 = v_29.col_major1; +} + +void copy_col_major_to_row_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_41.row_major0 = transpose(v_29.col_major0); +} + +void copy_row_major_to_col_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0 = transpose(v_41.row_major0); +} + +void copy_row_major_to_row_major(device SSBORow& v_41) +{ + v_41.row_major0 = v_41.row_major1; +} + +void copy_columns(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[1] = float2(v_41.row_major0[0][1], v_41.row_major0[1][1]); + v_41.row_major0[0][1] = v_29.col_major0[1].x; + v_41.row_major0[1][1] = v_29.col_major0[1].y; +} + +void copy_elements(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[0].y = v_41.row_major0[1u][0]; + v_41.row_major0[1u][0] = v_29.col_major0[0].y; +} + +kernel void main0(device SSBOCol& v_29 [[buffer(0)]], device SSBORow& v_41 [[buffer(1)]]) +{ + load_store_to_variable_col_major(v_29); + load_store_to_variable_row_major(v_41); + copy_col_major_to_col_major(v_29); + copy_col_major_to_row_major(v_29, v_41); + copy_row_major_to_col_major(v_29, v_41); + copy_row_major_to_row_major(v_41); + copy_columns(v_29, v_41); + copy_elements(v_29, v_41); +} + diff --git a/reference/shaders-msl-no-opt/packing/matrix-4x3-std430.comp b/reference/shaders-msl-no-opt/packing/matrix-4x3-std430.comp new file mode 100644 index 00000000..574f02e8 --- /dev/null +++ b/reference/shaders-msl-no-opt/packing/matrix-4x3-std430.comp @@ -0,0 +1,77 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" + +#include +#include + +using namespace metal; + +struct SSBOCol +{ + float4x3 col_major0; + float4x3 col_major1; +}; + +struct SSBORow +{ + float3x4 row_major0; + float3x4 row_major1; +}; + +void load_store_to_variable_col_major(device SSBOCol& v_29) +{ + float4x3 loaded = v_29.col_major0; + v_29.col_major1 = loaded; +} + +void load_store_to_variable_row_major(device SSBORow& v_41) +{ + float4x3 loaded = transpose(v_41.row_major0); + v_41.row_major0 = transpose(loaded); +} + +void copy_col_major_to_col_major(device SSBOCol& v_29) +{ + v_29.col_major0 = v_29.col_major1; +} + +void copy_col_major_to_row_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_41.row_major0 = transpose(v_29.col_major0); +} + +void copy_row_major_to_col_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0 = transpose(v_41.row_major0); +} + +void copy_row_major_to_row_major(device SSBORow& v_41) +{ + v_41.row_major0 = v_41.row_major1; +} + +void copy_columns(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[1] = float3(v_41.row_major0[0][1], v_41.row_major0[1][1], v_41.row_major0[2][1]); + v_41.row_major0[0][1] = v_29.col_major0[1].x; + v_41.row_major0[1][1] = v_29.col_major0[1].y; + v_41.row_major0[2][1] = v_29.col_major0[1].z; +} + +void copy_elements(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[0].y = v_41.row_major0[1u][0]; + v_41.row_major0[1u][0] = v_29.col_major0[0].y; +} + +kernel void main0(device SSBOCol& v_29 [[buffer(0)]], device SSBORow& v_41 [[buffer(1)]]) +{ + load_store_to_variable_col_major(v_29); + load_store_to_variable_row_major(v_41); + copy_col_major_to_col_major(v_29); + copy_col_major_to_row_major(v_29, v_41); + copy_row_major_to_col_major(v_29, v_41); + copy_row_major_to_row_major(v_41); + copy_columns(v_29, v_41); + copy_elements(v_29, v_41); +} + diff --git a/reference/shaders-msl-no-opt/packing/matrix-4x4-std430.comp b/reference/shaders-msl-no-opt/packing/matrix-4x4-std430.comp new file mode 100644 index 00000000..45bf0d87 --- /dev/null +++ b/reference/shaders-msl-no-opt/packing/matrix-4x4-std430.comp @@ -0,0 +1,78 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" + +#include +#include + +using namespace metal; + +struct SSBOCol +{ + float4x4 col_major0; + float4x4 col_major1; +}; + +struct SSBORow +{ + float4x4 row_major0; + float4x4 row_major1; +}; + +void load_store_to_variable_col_major(device SSBOCol& v_29) +{ + float4x4 loaded = v_29.col_major0; + v_29.col_major1 = loaded; +} + +void load_store_to_variable_row_major(device SSBORow& v_41) +{ + float4x4 loaded = transpose(v_41.row_major0); + v_41.row_major0 = transpose(loaded); +} + +void copy_col_major_to_col_major(device SSBOCol& v_29) +{ + v_29.col_major0 = v_29.col_major1; +} + +void copy_col_major_to_row_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_41.row_major0 = transpose(v_29.col_major0); +} + +void copy_row_major_to_col_major(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0 = transpose(v_41.row_major0); +} + +void copy_row_major_to_row_major(device SSBORow& v_41) +{ + v_41.row_major0 = v_41.row_major1; +} + +void copy_columns(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[1] = float4(v_41.row_major0[0][1], v_41.row_major0[1][1], v_41.row_major0[2][1], v_41.row_major0[3][1]); + v_41.row_major0[0][1] = v_29.col_major0[1].x; + v_41.row_major0[1][1] = v_29.col_major0[1].y; + v_41.row_major0[2][1] = v_29.col_major0[1].z; + v_41.row_major0[3][1] = v_29.col_major0[1].w; +} + +void copy_elements(device SSBOCol& v_29, device SSBORow& v_41) +{ + v_29.col_major0[0].y = v_41.row_major0[1u][0]; + v_41.row_major0[1u][0] = v_29.col_major0[0].y; +} + +kernel void main0(device SSBOCol& v_29 [[buffer(0)]], device SSBORow& v_41 [[buffer(1)]]) +{ + load_store_to_variable_col_major(v_29); + load_store_to_variable_row_major(v_41); + copy_col_major_to_col_major(v_29); + copy_col_major_to_row_major(v_29, v_41); + copy_row_major_to_col_major(v_29, v_41); + copy_row_major_to_row_major(v_41); + copy_columns(v_29, v_41); + copy_elements(v_29, v_41); +} + diff --git a/shaders-msl-no-opt/packing/matrix-2x2-std430.comp b/shaders-msl-no-opt/packing/matrix-2x2-std430.comp new file mode 100644 index 00000000..342c3989 --- /dev/null +++ b/shaders-msl-no-opt/packing/matrix-2x2-std430.comp @@ -0,0 +1,85 @@ +#version 450 +layout(local_size_x = 1) in; + +#define T mat2 +#define PACKING std430 + +layout(set = 0, binding = 0, PACKING) buffer SSBOCol +{ + layout(column_major) T col_major0; + layout(column_major) T col_major1; +}; + +layout(set = 0, binding = 1, PACKING) buffer SSBORow +{ + layout(row_major) T row_major0; + layout(row_major) T row_major1; +}; + +void load_store_to_variable_col_major() +{ + // Load to variable. + T loaded = col_major0; + + // Store from variable. + col_major1 = loaded; +} + +void load_store_to_variable_row_major() +{ + // Load to variable. + T loaded = row_major0; + + // Store to variable. + row_major0 = loaded; +} + +void copy_col_major_to_col_major() +{ + // Copy col -> col + col_major0 = col_major1; +} + +void copy_row_major_to_col_major() +{ + // Copy row -> col + col_major0 = row_major0; +} + +void copy_col_major_to_row_major() +{ + // Copy col -> row + row_major0 = col_major0; +} + +void copy_row_major_to_row_major() +{ + // Copy row -> row + row_major0 = row_major1; +} + +void copy_columns() +{ + // Copy columns/rows. + col_major0[1] = row_major0[1]; + row_major0[1] = col_major0[1]; +} + +void copy_elements() +{ + // Copy individual elements. + col_major0[0][1] = row_major0[0][1]; + row_major0[0][1] = col_major0[0][1]; +} + +void main() +{ + load_store_to_variable_col_major(); + load_store_to_variable_row_major(); + copy_col_major_to_col_major(); + copy_col_major_to_row_major(); + copy_row_major_to_col_major(); + copy_row_major_to_row_major(); + copy_columns(); + copy_elements(); +} diff --git a/shaders-msl-no-opt/packing/matrix-2x3-std430.comp b/shaders-msl-no-opt/packing/matrix-2x3-std430.comp new file mode 100644 index 00000000..36a6bab1 --- /dev/null +++ b/shaders-msl-no-opt/packing/matrix-2x3-std430.comp @@ -0,0 +1,85 @@ +#version 450 +layout(local_size_x = 1) in; + +#define T mat2x3 +#define PACKING std430 + +layout(set = 0, binding = 0, PACKING) buffer SSBOCol +{ + layout(column_major) T col_major0; + layout(column_major) T col_major1; +}; + +layout(set = 0, binding = 1, PACKING) buffer SSBORow +{ + layout(row_major) T row_major0; + layout(row_major) T row_major1; +}; + +void load_store_to_variable_col_major() +{ + // Load to variable. + T loaded = col_major0; + + // Store from variable. + col_major1 = loaded; +} + +void load_store_to_variable_row_major() +{ + // Load to variable. + T loaded = row_major0; + + // Store to variable. + row_major0 = loaded; +} + +void copy_col_major_to_col_major() +{ + // Copy col -> col + col_major0 = col_major1; +} + +void copy_row_major_to_col_major() +{ + // Copy row -> col + col_major0 = row_major0; +} + +void copy_col_major_to_row_major() +{ + // Copy col -> row + row_major0 = col_major0; +} + +void copy_row_major_to_row_major() +{ + // Copy row -> row + row_major0 = row_major1; +} + +void copy_columns() +{ + // Copy columns/rows. + col_major0[1] = row_major0[1]; + row_major0[1] = col_major0[1]; +} + +void copy_elements() +{ + // Copy individual elements. + col_major0[0][1] = row_major0[0][1]; + row_major0[0][1] = col_major0[0][1]; +} + +void main() +{ + load_store_to_variable_col_major(); + load_store_to_variable_row_major(); + copy_col_major_to_col_major(); + copy_col_major_to_row_major(); + copy_row_major_to_col_major(); + copy_row_major_to_row_major(); + copy_columns(); + copy_elements(); +} diff --git a/shaders-msl-no-opt/packing/matrix-2x4-std430.comp b/shaders-msl-no-opt/packing/matrix-2x4-std430.comp new file mode 100644 index 00000000..177b9669 --- /dev/null +++ b/shaders-msl-no-opt/packing/matrix-2x4-std430.comp @@ -0,0 +1,85 @@ +#version 450 +layout(local_size_x = 1) in; + +#define T mat2x4 +#define PACKING std430 + +layout(set = 0, binding = 0, PACKING) buffer SSBOCol +{ + layout(column_major) T col_major0; + layout(column_major) T col_major1; +}; + +layout(set = 0, binding = 1, PACKING) buffer SSBORow +{ + layout(row_major) T row_major0; + layout(row_major) T row_major1; +}; + +void load_store_to_variable_col_major() +{ + // Load to variable. + T loaded = col_major0; + + // Store from variable. + col_major1 = loaded; +} + +void load_store_to_variable_row_major() +{ + // Load to variable. + T loaded = row_major0; + + // Store to variable. + row_major0 = loaded; +} + +void copy_col_major_to_col_major() +{ + // Copy col -> col + col_major0 = col_major1; +} + +void copy_row_major_to_col_major() +{ + // Copy row -> col + col_major0 = row_major0; +} + +void copy_col_major_to_row_major() +{ + // Copy col -> row + row_major0 = col_major0; +} + +void copy_row_major_to_row_major() +{ + // Copy row -> row + row_major0 = row_major1; +} + +void copy_columns() +{ + // Copy columns/rows. + col_major0[1] = row_major0[1]; + row_major0[1] = col_major0[1]; +} + +void copy_elements() +{ + // Copy individual elements. + col_major0[0][1] = row_major0[0][1]; + row_major0[0][1] = col_major0[0][1]; +} + +void main() +{ + load_store_to_variable_col_major(); + load_store_to_variable_row_major(); + copy_col_major_to_col_major(); + copy_col_major_to_row_major(); + copy_row_major_to_col_major(); + copy_row_major_to_row_major(); + copy_columns(); + copy_elements(); +} diff --git a/shaders-msl-no-opt/packing/matrix-3x2-std430.comp b/shaders-msl-no-opt/packing/matrix-3x2-std430.comp new file mode 100644 index 00000000..fe82993d --- /dev/null +++ b/shaders-msl-no-opt/packing/matrix-3x2-std430.comp @@ -0,0 +1,85 @@ +#version 450 +layout(local_size_x = 1) in; + +#define T mat3x2 +#define PACKING std430 + +layout(set = 0, binding = 0, PACKING) buffer SSBOCol +{ + layout(column_major) T col_major0; + layout(column_major) T col_major1; +}; + +layout(set = 0, binding = 1, PACKING) buffer SSBORow +{ + layout(row_major) T row_major0; + layout(row_major) T row_major1; +}; + +void load_store_to_variable_col_major() +{ + // Load to variable. + T loaded = col_major0; + + // Store from variable. + col_major1 = loaded; +} + +void load_store_to_variable_row_major() +{ + // Load to variable. + T loaded = row_major0; + + // Store to variable. + row_major0 = loaded; +} + +void copy_col_major_to_col_major() +{ + // Copy col -> col + col_major0 = col_major1; +} + +void copy_row_major_to_col_major() +{ + // Copy row -> col + col_major0 = row_major0; +} + +void copy_col_major_to_row_major() +{ + // Copy col -> row + row_major0 = col_major0; +} + +void copy_row_major_to_row_major() +{ + // Copy row -> row + row_major0 = row_major1; +} + +void copy_columns() +{ + // Copy columns/rows. + col_major0[1] = row_major0[1]; + row_major0[1] = col_major0[1]; +} + +void copy_elements() +{ + // Copy individual elements. + col_major0[0][1] = row_major0[0][1]; + row_major0[0][1] = col_major0[0][1]; +} + +void main() +{ + load_store_to_variable_col_major(); + load_store_to_variable_row_major(); + copy_col_major_to_col_major(); + copy_col_major_to_row_major(); + copy_row_major_to_col_major(); + copy_row_major_to_row_major(); + copy_columns(); + copy_elements(); +} diff --git a/shaders-msl-no-opt/packing/matrix-3x3-std430.comp b/shaders-msl-no-opt/packing/matrix-3x3-std430.comp new file mode 100644 index 00000000..8e48109e --- /dev/null +++ b/shaders-msl-no-opt/packing/matrix-3x3-std430.comp @@ -0,0 +1,85 @@ +#version 450 +layout(local_size_x = 1) in; + +#define T mat3 +#define PACKING std430 + +layout(set = 0, binding = 0, PACKING) buffer SSBOCol +{ + layout(column_major) T col_major0; + layout(column_major) T col_major1; +}; + +layout(set = 0, binding = 1, PACKING) buffer SSBORow +{ + layout(row_major) T row_major0; + layout(row_major) T row_major1; +}; + +void load_store_to_variable_col_major() +{ + // Load to variable. + T loaded = col_major0; + + // Store from variable. + col_major1 = loaded; +} + +void load_store_to_variable_row_major() +{ + // Load to variable. + T loaded = row_major0; + + // Store to variable. + row_major0 = loaded; +} + +void copy_col_major_to_col_major() +{ + // Copy col -> col + col_major0 = col_major1; +} + +void copy_row_major_to_col_major() +{ + // Copy row -> col + col_major0 = row_major0; +} + +void copy_col_major_to_row_major() +{ + // Copy col -> row + row_major0 = col_major0; +} + +void copy_row_major_to_row_major() +{ + // Copy row -> row + row_major0 = row_major1; +} + +void copy_columns() +{ + // Copy columns/rows. + col_major0[1] = row_major0[1]; + row_major0[1] = col_major0[1]; +} + +void copy_elements() +{ + // Copy individual elements. + col_major0[0][1] = row_major0[0][1]; + row_major0[0][1] = col_major0[0][1]; +} + +void main() +{ + load_store_to_variable_col_major(); + load_store_to_variable_row_major(); + copy_col_major_to_col_major(); + copy_col_major_to_row_major(); + copy_row_major_to_col_major(); + copy_row_major_to_row_major(); + copy_columns(); + copy_elements(); +} diff --git a/shaders-msl-no-opt/packing/matrix-3x4-std430.comp b/shaders-msl-no-opt/packing/matrix-3x4-std430.comp new file mode 100644 index 00000000..78c577f2 --- /dev/null +++ b/shaders-msl-no-opt/packing/matrix-3x4-std430.comp @@ -0,0 +1,85 @@ +#version 450 +layout(local_size_x = 1) in; + +#define T mat3x4 +#define PACKING std430 + +layout(set = 0, binding = 0, PACKING) buffer SSBOCol +{ + layout(column_major) T col_major0; + layout(column_major) T col_major1; +}; + +layout(set = 0, binding = 1, PACKING) buffer SSBORow +{ + layout(row_major) T row_major0; + layout(row_major) T row_major1; +}; + +void load_store_to_variable_col_major() +{ + // Load to variable. + T loaded = col_major0; + + // Store from variable. + col_major1 = loaded; +} + +void load_store_to_variable_row_major() +{ + // Load to variable. + T loaded = row_major0; + + // Store to variable. + row_major0 = loaded; +} + +void copy_col_major_to_col_major() +{ + // Copy col -> col + col_major0 = col_major1; +} + +void copy_row_major_to_col_major() +{ + // Copy row -> col + col_major0 = row_major0; +} + +void copy_col_major_to_row_major() +{ + // Copy col -> row + row_major0 = col_major0; +} + +void copy_row_major_to_row_major() +{ + // Copy row -> row + row_major0 = row_major1; +} + +void copy_columns() +{ + // Copy columns/rows. + col_major0[1] = row_major0[1]; + row_major0[1] = col_major0[1]; +} + +void copy_elements() +{ + // Copy individual elements. + col_major0[0][1] = row_major0[0][1]; + row_major0[0][1] = col_major0[0][1]; +} + +void main() +{ + load_store_to_variable_col_major(); + load_store_to_variable_row_major(); + copy_col_major_to_col_major(); + copy_col_major_to_row_major(); + copy_row_major_to_col_major(); + copy_row_major_to_row_major(); + copy_columns(); + copy_elements(); +} diff --git a/shaders-msl-no-opt/packing/matrix-4x2-std430.comp b/shaders-msl-no-opt/packing/matrix-4x2-std430.comp new file mode 100644 index 00000000..76aa9ae4 --- /dev/null +++ b/shaders-msl-no-opt/packing/matrix-4x2-std430.comp @@ -0,0 +1,85 @@ +#version 450 +layout(local_size_x = 1) in; + +#define T mat4x2 +#define PACKING std430 + +layout(set = 0, binding = 0, PACKING) buffer SSBOCol +{ + layout(column_major) T col_major0; + layout(column_major) T col_major1; +}; + +layout(set = 0, binding = 1, PACKING) buffer SSBORow +{ + layout(row_major) T row_major0; + layout(row_major) T row_major1; +}; + +void load_store_to_variable_col_major() +{ + // Load to variable. + T loaded = col_major0; + + // Store from variable. + col_major1 = loaded; +} + +void load_store_to_variable_row_major() +{ + // Load to variable. + T loaded = row_major0; + + // Store to variable. + row_major0 = loaded; +} + +void copy_col_major_to_col_major() +{ + // Copy col -> col + col_major0 = col_major1; +} + +void copy_row_major_to_col_major() +{ + // Copy row -> col + col_major0 = row_major0; +} + +void copy_col_major_to_row_major() +{ + // Copy col -> row + row_major0 = col_major0; +} + +void copy_row_major_to_row_major() +{ + // Copy row -> row + row_major0 = row_major1; +} + +void copy_columns() +{ + // Copy columns/rows. + col_major0[1] = row_major0[1]; + row_major0[1] = col_major0[1]; +} + +void copy_elements() +{ + // Copy individual elements. + col_major0[0][1] = row_major0[0][1]; + row_major0[0][1] = col_major0[0][1]; +} + +void main() +{ + load_store_to_variable_col_major(); + load_store_to_variable_row_major(); + copy_col_major_to_col_major(); + copy_col_major_to_row_major(); + copy_row_major_to_col_major(); + copy_row_major_to_row_major(); + copy_columns(); + copy_elements(); +} diff --git a/shaders-msl-no-opt/packing/matrix-4x3-std430.comp b/shaders-msl-no-opt/packing/matrix-4x3-std430.comp new file mode 100644 index 00000000..aa4d685c --- /dev/null +++ b/shaders-msl-no-opt/packing/matrix-4x3-std430.comp @@ -0,0 +1,85 @@ +#version 450 +layout(local_size_x = 1) in; + +#define T mat4x3 +#define PACKING std430 + +layout(set = 0, binding = 0, PACKING) buffer SSBOCol +{ + layout(column_major) T col_major0; + layout(column_major) T col_major1; +}; + +layout(set = 0, binding = 1, PACKING) buffer SSBORow +{ + layout(row_major) T row_major0; + layout(row_major) T row_major1; +}; + +void load_store_to_variable_col_major() +{ + // Load to variable. + T loaded = col_major0; + + // Store from variable. + col_major1 = loaded; +} + +void load_store_to_variable_row_major() +{ + // Load to variable. + T loaded = row_major0; + + // Store to variable. + row_major0 = loaded; +} + +void copy_col_major_to_col_major() +{ + // Copy col -> col + col_major0 = col_major1; +} + +void copy_row_major_to_col_major() +{ + // Copy row -> col + col_major0 = row_major0; +} + +void copy_col_major_to_row_major() +{ + // Copy col -> row + row_major0 = col_major0; +} + +void copy_row_major_to_row_major() +{ + // Copy row -> row + row_major0 = row_major1; +} + +void copy_columns() +{ + // Copy columns/rows. + col_major0[1] = row_major0[1]; + row_major0[1] = col_major0[1]; +} + +void copy_elements() +{ + // Copy individual elements. + col_major0[0][1] = row_major0[0][1]; + row_major0[0][1] = col_major0[0][1]; +} + +void main() +{ + load_store_to_variable_col_major(); + load_store_to_variable_row_major(); + copy_col_major_to_col_major(); + copy_col_major_to_row_major(); + copy_row_major_to_col_major(); + copy_row_major_to_row_major(); + copy_columns(); + copy_elements(); +} diff --git a/shaders-msl-no-opt/packing/matrix-4x4-std430.comp b/shaders-msl-no-opt/packing/matrix-4x4-std430.comp new file mode 100644 index 00000000..3a1eb9f0 --- /dev/null +++ b/shaders-msl-no-opt/packing/matrix-4x4-std430.comp @@ -0,0 +1,85 @@ +#version 450 +layout(local_size_x = 1) in; + +#define T mat4 +#define PACKING std430 + +layout(set = 0, binding = 0, PACKING) buffer SSBOCol +{ + layout(column_major) T col_major0; + layout(column_major) T col_major1; +}; + +layout(set = 0, binding = 1, PACKING) buffer SSBORow +{ + layout(row_major) T row_major0; + layout(row_major) T row_major1; +}; + +void load_store_to_variable_col_major() +{ + // Load to variable. + T loaded = col_major0; + + // Store from variable. + col_major1 = loaded; +} + +void load_store_to_variable_row_major() +{ + // Load to variable. + T loaded = row_major0; + + // Store to variable. + row_major0 = loaded; +} + +void copy_col_major_to_col_major() +{ + // Copy col -> col + col_major0 = col_major1; +} + +void copy_row_major_to_col_major() +{ + // Copy row -> col + col_major0 = row_major0; +} + +void copy_col_major_to_row_major() +{ + // Copy col -> row + row_major0 = col_major0; +} + +void copy_row_major_to_row_major() +{ + // Copy row -> row + row_major0 = row_major1; +} + +void copy_columns() +{ + // Copy columns/rows. + col_major0[1] = row_major0[1]; + row_major0[1] = col_major0[1]; +} + +void copy_elements() +{ + // Copy individual elements. + col_major0[0][1] = row_major0[0][1]; + row_major0[0][1] = col_major0[0][1]; +} + +void main() +{ + load_store_to_variable_col_major(); + load_store_to_variable_row_major(); + copy_col_major_to_col_major(); + copy_col_major_to_row_major(); + copy_row_major_to_col_major(); + copy_row_major_to_row_major(); + copy_columns(); + copy_elements(); +}