MSL: Add std430 matrix access test.
This commit is contained in:
Hans-Kristian Arntzen
Родитель
249f8e5180
Коммит
6471236652
|
|
@ -0,0 +1,76 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
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);
|
||||
}
|
||||
|
||||
|
|
@ -0,0 +1,77 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
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);
|
||||
}
|
||||
|
||||
|
|
@ -0,0 +1,78 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
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);
|
||||
}
|
||||
|
||||
|
|
@ -0,0 +1,76 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
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);
|
||||
}
|
||||
|
||||
|
|
@ -0,0 +1,77 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
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);
|
||||
}
|
||||
|
||||
|
|
@ -0,0 +1,78 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
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);
|
||||
}
|
||||
|
||||
|
|
@ -0,0 +1,76 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
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);
|
||||
}
|
||||
|
||||
|
|
@ -0,0 +1,77 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
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);
|
||||
}
|
||||
|
||||
|
|
@ -0,0 +1,78 @@
|
|||
#pragma clang diagnostic ignored "-Wmissing-prototypes"
|
||||
|
||||
#include <metal_stdlib>
|
||||
#include <simd/simd.h>
|
||||
|
||||
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);
|
||||
}
|
||||
|
||||
|
|
@ -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();
|
||||
}
|
||||
|
|
@ -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();
|
||||
}
|
||||
|
|
@ -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();
|
||||
}
|
||||
|
|
@ -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();
|
||||
}
|
||||
|
|
@ -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();
|
||||
}
|
||||
|
|
@ -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();
|
||||
}
|
||||
|
|
@ -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();
|
||||
}
|
||||
|
|
@ -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();
|
||||
}
|
||||
|
|
@ -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();
|
||||
}
|
||||
Загрузка…
Ссылка в новой задаче