remove mat panel, mat_task goes on span like

StgSharp.Mathematics/Numeric/Runtime/MatrixParallelTask.cs

@@ -48,13 +48,16 @@ namespace StgSharp.Mathematics.Numeric.Runtime
     internal unsafe struct MatrixParallelTask
     {
 
+        [FieldOffset(104)] private ulong reversed;
+
         [FieldOffset(0)] public L4.Handle Buffer0;
         [FieldOffset(16)] public L4.Handle Buffer1;
         [FieldOffset(32)] public L4.Handle Buffer2;
         [FieldOffset(48)] public L4.Handle Buffer3;
+        [FieldOffset(112)] public M128 ScalarPack;
         [FieldOffset(64)] public M512 Profile;
-        [FieldOffset(120)] public MatrixElementType ElementType;
+        [FieldOffset(64)] public MatrixElementType ElementType;
-        [FieldOffset(124)] public MatrixIntrinsicHandle ComputeHandle;
+        [FieldOffset(68)] public MatrixIntrinsicHandle ComputeHandle;
 
         public MatrixParallelTask()
         {
@@ -71,5 +74,17 @@ namespace StgSharp.Mathematics.Numeric.Runtime
                               sizeof(MatrixParallelTask));
         }
 
+        #region offset and length settings
+
+        [FieldOffset(72)] public int Offset0;
+        [FieldOffset(76)] public int Count0;
+        [FieldOffset(80)] public int Offset1;
+        [FieldOffset(84)] public int Count1;
+        [FieldOffset(88)] public int Offset2;
+        [FieldOffset(92)] public int Count2;
+        [FieldOffset(96)] public int Offset3;
+        [FieldOffset(100)] public int Count3;
+
+    #endregion
     }
 }

StgSharp.Native/sn-head/sn_internal.h

@@ -40,4 +40,6 @@
 #define PRIVATE static
 #define INTERNAL extern
 
+#define u64 uint64_t
+
 #endif //

StgSharp.Native/sn-head/sn_intrinsic.context.x86.h

@@ -128,7 +128,7 @@
 
 #pragma region matix function
 
-INTERNAL void f32_normalize(sn_vec_f32 * source, sn_vec_f32 * target);
+INTERNAL void f32_normalize(VEC_SEGMENT(float) * source, VEC_SEGMENT(float) * target);
 
 SN_MK_PROC_DECL_STD(float, sse, , k_quality);
 

StgSharp.Native/sn-head/sn_intrinsic.h

@@ -65,11 +65,37 @@ typedef union uint64_u {
 
 typedef struct mat_kernel mat_kernel;
 
-typedef struct mat_panel mat_panel;
+typedef union M128{
+        #if SN_IS_ARCH(x86_64)
+        __m128 m128;
+        #endif
+        float f32[4];
+        double f64[2];
+        int32_t i32[4];
+        int64_t i64[2];
+}M128;
 
-typedef union scalar_pack {
+typedef union M256{
-        uint64_t data[8];
+        #if SN_IS_ARCH(x86_64)
-} scalar_pack;
+        __m256 m256;
+        #endif
+        float f32[8];
+        double f64[4];
+        int32_t i32[8];
+        int64_t i64[4];
+}M256;
+
+typedef union M512{
+        #if SN_IS_ARCH(x86_64)
+        __m512 m512;
+        #endif
+        uint64_t u64[8];
+        uint32_t u32[16];
+        float f32[16];
+        double f64[8];
+        int32_t i32[16];
+        int64_t i64[8];
+}M512;
 
 typedef struct sn_L4_cache_head {
         uint64_t reverse_mask_0;
@@ -81,42 +107,33 @@ typedef struct sn_L4_cache_head {
 
 typedef struct sn_L4_handle {
         void *ptr;
-        sn_L4_cache_head head;
+        sn_L4_cache_head* head;
 } sn_L4_handle;
 
-typedef struct matrix_task_pack {
+typedef struct matrix_task {
         sn_L4_handle buffer_0;
         sn_L4_handle buffer_1;
         sn_L4_handle buffer_2;
         sn_L4_handle buffer_3;
-        uint64_t reverse_0;
+        union {
-        uint64_t reverse_1;
+                struct {
-        uint64_t reverse_2;
+                        int offset_0;
-        uint64_t reverse_3;
+                        int count_0;
-        uint64_t reverse_4;
+                        int offset_1;
-        uint64_t reverse_5;
+                        int count_1;
-        uint64_t reverse_6;
+                        int offset_2;
-        uint64_t reverse_7;
+                        int count_2;
+                        int offset_3;
+                        int count_3;
+                        uint64_t reverse;
+                        M128 scalar_pack;
+                };
+                M512 profile;
+        };
 
-} matrix_task_pack;
-
-typedef struct matrix_task {
-        void *mat_0;
-        void *mat_1;
-        void *mat_2;
-        void *scalar;
-        int32_t x;
-        int32_t y;
-        int32_t z;
-        int32_t common_x;
-        int32_t common_y;
-        int32_t common_z;
-        uint32_t mask;
-        uint32_t reverse;
-        scalar_pack padding;
 } matrix_task;
 
-#define MAT_TASK(T) matrix_task_pack
+#define MAT_TASK(T) matrix_task
 
 #if defined(__cplusplus)
 static_assert(sizeof(matrix_task) == 128, "matrix_parallel_task_package size mismatch");
@@ -128,7 +145,7 @@ typedef void(SN_DECL *UNI_MK_PROC)(matrix_task *task);
 typedef void(SN_DECL *UNI_SK_PROC)(mat_kernel const *left, mat_kernel const *right,
                                    mat_kernel *restrict ans, uint64_t scalar);
 
-typedef void(SN_DECL *VECTORNORMALIZEPROC)(sn_vec_f32 *source, sn_vec_f32 *target);
+typedef void(SN_DECL *VECTORNORMALIZEPROC)(VEC_SEGMENT(float) *source, VEC_SEGMENT(float) *target);
 typedef void(SN_DECL *DOTPROC)(void *transpose, __m128 *vector, __m128 *ans);
 typedef uint64_t(SN_DECL *FACTORIALROC)(int n);
 

StgSharp.Native/sn-head/sn_intrinsic.matkernel.h

@@ -12,13 +12,14 @@
 #define YMMCOUNT(T) (sizeof(T) / 2)
 #define ZMMCOUNT(T) (sizeof(T) / 4)
 
-#define SN_MAT_KERNEL_MEMBER(T) {\
+#define SN_MAT_KERNEL_MEMBER(T)        \
-        __m128 x[XMMCOUNT(T)];  \
+        {                              \
-        __m256 y[YMMCOUNT(T)];  \
+                __m128 x[XMMCOUNT(T)]; \
-        __m512 z[ZMMCOUNT(T)];  \
+                __m256 y[YMMCOUNT(T)]; \
-        T m[4][4];\
+                __m512 z[ZMMCOUNT(T)]; \
-}
+                T m[4][4];             \
-
+        }
+/*
 #define SN_DEFINE_MAT_KERNEL_TYPE(name, T) \
         typedef union name SN_MAT_KERNEL_MEMBER(T) name
 
@@ -26,120 +27,36 @@
 SN_DEFINE_MAT_KERNEL_TYPE(sn_mat_kernel_float, float);
 SN_DEFINE_MAT_KERNEL_TYPE(sn_mat_kernel_double, double);
 SN_DEFINE_MAT_KERNEL_TYPE(sn_mat_kernel_uint64_t, uint64_t);
-
+/* */
-#define MAT_KERNEL_DEFINE(T) \
+#define SN_MAT_KERNEL_DEFINE(T) \
         typedef union SN_MAT_KERNEL_TYPE_##T SN_MAT_KERNEL_MEMBER(T) SN_MAT_KERNEL_TYPE_##T
 
-#define SN_MAT_KERNEL_TYPE(T) SN_MAT_KERNEL_TYPE_##T
-MAT_KERNEL_DEFINE(float);
-MAT_KERNEL_DEFINE(double);
-MAT_KERNEL_DEFINE(uint64_t);
-
 #define MAT_KERNEL(T) SN_MAT_KERNEL_TYPE(T)
+#define SN_MAT_KERNEL_TYPE(T) SN_MAT_KERNEL_TYPE_##T
 
-#define VEC(T)                                  \
+SN_MAT_KERNEL_DEFINE(float);
-        union {                                 \
+SN_MAT_KERNEL_DEFINE(double);
-                __m128 f32_x[XMMCOUNT(T) / 4];  \
+SN_MAT_KERNEL_DEFINE(uint64_t);
-                __m256 f32_y[YMMCOUNT(T) / 4];  \
+
-                __m512 f32_z[ZMMCOUNT(T) / 4];  \
+
-                                                \
+#define SN_VEC_SEGMENT_MEMBER(T)               \
-                __m128d f64_x[XMMCOUNT(T) / 4]; \
+        {                                \
-                __m256d f64_y[YMMCOUNT(T) / 4]; \
+                __m128 f32_x[XMMCOUNT(T) / 4]; \
-                __m512d f64_z[ZMMCOUNT(T) / 4]; \
+                __m256 f32_y[YMMCOUNT(T) / 4]; \
-                T v[4];                         \
+                __m512 f32_z[ZMMCOUNT(T) / 4]; \
+                T v[4];                        \
         }
 
-typedef VEC(float) sn_vec_f32;
+#define SN_VEC_SEGMENT_DEFINE(T) \
+        typedef union SN_VEC_SEGMENT_TYPE_##T SN_VEC_SEGMENT_MEMBER(T) SN_VEC_SEGMENT_TYPE_##T
 
-/* Provide a named alias for the zero kernel type to ensure consistent linkage in C++. */
+#define VEC_SEGMENT(T) SN_VEC_SEGMENT_TYPE(T)
-typedef sn_mat_kernel_uint64_t sn_zero_kernel_u64;
+#define SN_VEC_SEGMENT_TYPE(T) SN_VEC_SEGMENT_TYPE_##T
 
-/* Helpers (all compile-time): */
+SN_VEC_SEGMENT_DEFINE(float);
-#define SN_LANES(T_VEC, T) (sizeof(T_VEC) / sizeof(T))
+SN_VEC_SEGMENT_DEFINE(double);
-#define SN_MAX(a, b) (((a)) > ((b)) ? (a) : (b))
-#define SN_CEIL_DIV(a, b) (((a) + (b) - 1) / (b))
-#define SN_BITS(x) ((int)(sizeof(x) * 8))
-#define PANEL_MINQ(T_VEC, T) SN_MAX(4, SN_MAX(SN_BITS(T_VEC) / 64, SN_BITS(T_VEC) / SN_BITS(T)))
-
-/* Panel geometry helpers (all compile-time): */
-#define PANEL_Q(T, T_VEC, MINQ) SN_MAX(SN_LANES(T_VEC, T), (MINQ))
-#define PANEL_CHUNKS(T, T_VEC, MINQ) SN_CEIL_DIV(PANEL_Q(T, T_VEC, MINQ), SN_LANES(T_VEC, T))
-
-/*
-        Column-major square panel with minimum side Q:
-        - Let R = lanes per vector = sizeof(T_VEC)/sizeof(T).
-        - Side length Q = max(R, MINQ) to enforce the requested minimum extent.
-        - Each column is stored contiguously via CH = ceil(Q/R) vector chunks.
-        - Memory layout:
-                        vec[col][chunk] is a T_VEC holding up to R scalars of column 'col'
-                        spanning rows chunk*R .. chunk*R + (R - 1).
-                        flat[] mirrors the same column-major order (column stride = Q).
-        - SIMD view remains integer-typed; cast inside micro-kernels to the
-                appropriate floating-point/integer vector type on demand.
-*/
-#define MAT_PANEL_SQ_MIN_NOALIGN(T, T_VEC, MINQ)                                  \
-        union {                                                                   \
-                T_VEC vec[PANEL_Q(T, T_VEC, MINQ)][PANEL_CHUNKS(T, T_VEC, MINQ)]; \
-                T flat[PANEL_Q(T, T_VEC, MINQ) * PANEL_Q(T, T_VEC, MINQ)];        \
-        }
-
-/* ISA presets with minimum square side:
-        - SSE  : MINQ=4   (avoid 2x2 for double)
-        - AVX2 : MINQ computed as max(4, vector_bits/64, vector_bits/element_bits)
-        - AVX-512: same rule (float/i32 naturally expand to 16)
-*/
-
-/* SSE panel layout: column-major buffer holding 4 logical columns with MINQ=4.
-        Columns: each column packs up to four scalars so ALIGN equals sizeof(__m128).
-        Alignment: every column boundary lands on a 16-byte multiple for XMM loads. */
-#define MAT_PANEL_sse(T) SN_ALIGNAS(16) MAT_PANEL_SQ_MIN_NOALIGN(T, __m128, PANEL_MINQ(__m128, T))
-/* AVX2 panel layout: column-major buffer holding 8 logical columns with MINQ per PANEL_MINQ.
-        Columns: each column stores up to lanes elements via contiguous chunks.
-        Alignment: column starts are 32-byte aligned so AVX2 loads/stores remain contiguous. */
-#define MAT_PANEL_avx(T) SN_ALIGNAS(32) MAT_PANEL_SQ_MIN_NOALIGN(T, __m256, PANEL_MINQ(__m256, T))
-/* AVX2/FMA panel layout shares the same geometry as AVX2. */
-#define MAT_PANEL_avx_fma(T) \
-        SN_ALIGNAS(32) MAT_PANEL_SQ_MIN_NOALIGN(T, __m256, PANEL_MINQ(__m256, T))
-/* AVX-512 panel layout: column-major buffer holding 8 or more columns depending on T.
-        Columns: each column expands to ceil(Q/lanes) ZMM chunks.
-        Alignment: column starts sit on 64-byte boundaries to match AVX-512 streaming access. */
-#define MAT_PANEL_512(T) SN_ALIGNAS(64) MAT_PANEL_SQ_MIN_NOALIGN(T, __m512, PANEL_MINQ(__m512, T))
 
 /* Global zero kernel declaration (defined in dllmain.c), aligned to 64 bytes */
-#ifdef __cplusplus
+extern __declspec(align(64)) MAT_KERNEL(u64) SN_ZERO_KERNEL;
-extern "C" {
-#endif
-extern __declspec(align(64)) sn_zero_kernel_u64 SN_ZERO_KERNEL;
-#ifdef __cplusplus
-}
-#endif
-
-/* Access helpers (all compile-time friendly):
-   Given Q = PANEL_Q(...), CH = PANEL_CHUNKS(...), R = lanes per vector:
-   - Column c occupies vec[c][0..CH-1].
-   - Row r inside column c is located at vec[c][r / R] with lane (r % R).
-*/
-#define PANEL_COL_VEC(panel_obj, c, chunk) ((panel_obj).vec[(int)(c)][(int)(chunk)])
-/* Scalar flat index helper for column-major addressing */
-#define PANEL_ELEM(panel_obj, T, T_VEC, MINQ, r, c) \
-        ((panel_obj).flat[(int)(c) * PANEL_Q(T, T_VEC, MINQ) + (int)(r)])
-
-/* API macros */
-#define BUILD_PANEL(arch, element) build_panel_##arch##_##element
-#define STORE_PANEL(arch, element) store_panel_##arch##_##element
-
-#define DECLARE_BUILD_PANEL(arch, element)                                               \
-        INTERNAL void SN_DECL BUILD_PANEL(arch, element)(                                \
-                MAT_PANEL_##arch(element) *restrict panel,                               \
-                MAT_KERNEL(element) const *const matrix, int col_length, int row_length, \
-                int col_index, int row_index)
-#define DECLARE_STORE_PANEL(arch, element)                                                     \
-        INTERNAL void SN_DECL STORE_PANEL(arch, element)(MAT_PANEL_##arch(element)             \
-                                                                 const *const panel,           \
-                                                         MAT_KERNEL(element) *restrict matrix, \
-                                                         int col_length, int row_length,       \
-                                                         int col_index, int row_index)
-
-
 
 #endif /* SN_MATKERNEL */

StgSharp.Native/sn-head/sn_intrinsic.std.h

@@ -46,73 +46,54 @@
 #define _SN_MATPROC_RIGHT_ANS_SCALAR 1
 #define _SN_MATPROC_ANS_SCALAR 1
 
-#define _SN_MK_EXPAND_RIGHT_ANS                                                      \
+#define _SN_MK_EXPAND_BUFFER_SIZE(i)\
-        SN_MAT_KERNEL_LOCAL const *right = (SN_MAT_KERNEL_LOCAL const *)task->buffer_1.ptr; \
+                int offset_##i = task->offset_##i;\
-        SN_MAT_KERNEL_LOCAL *restrict ans = (SN_MAT_KERNEL_LOCAL *)task->buffer_2.ptr;
+                int count_##i = task->count_##i;\
 
-#define _SN_MK_EXPAND_LEFT_RIGHT_ANS                                                 \
+#define _SN_MK_EXPAND_RIGHT_ANS                                                             \
+        SN_MAT_KERNEL_LOCAL const *right = (SN_MAT_KERNEL_LOCAL const *)task->buffer_1.ptr; \
+        SN_MAT_KERNEL_LOCAL *restrict ans = (SN_MAT_KERNEL_LOCAL *)task->buffer_2.ptr;\
+        _SN_MK_EXPAND_BUFFER_SIZE(1); \
+        _SN_MK_EXPAND_BUFFER_SIZE(2);
+
+#define _SN_MK_EXPAND_LEFT_RIGHT_ANS                                                        \
         SN_MAT_KERNEL_LOCAL const *left = (SN_MAT_KERNEL_LOCAL const *)task->buffer_0.ptr;  \
         SN_MAT_KERNEL_LOCAL const *right = (SN_MAT_KERNEL_LOCAL const *)task->buffer_1.ptr; \
-        SN_MAT_KERNEL_LOCAL *restrict ans = (SN_MAT_KERNEL_LOCAL *)task->buffer_2.ptr;
+        SN_MAT_KERNEL_LOCAL *restrict ans = (SN_MAT_KERNEL_LOCAL *)task->buffer_2.ptr;\
+        _SN_MK_EXPAND_BUFFER_SIZE(0); \
+        _SN_MK_EXPAND_BUFFER_SIZE(1);\
+        _SN_MK_EXPAND_BUFFER_SIZE(2);
 
-#define _SN_MK_EXPAND_RIGHT_ANS_SCALAR                                               \
+#define _SN_MK_EXPAND_RIGHT_ANS_SCALAR                                                      \
         SN_MAT_KERNEL_LOCAL const *right = (SN_MAT_KERNEL_LOCAL const *)task->buffer_1.ptr; \
         SN_MAT_KERNEL_LOCAL *restrict ans = (SN_MAT_KERNEL_LOCAL *)task->buffer_2.ptr;      \
-        scalar_pack *restrict scalar = (scalar_pack *restrict)task->scalar;
+        scalar_pack *restrict scalar = (scalar_pack *restrict)task->scalar;\
+        _SN_MK_EXPAND_BUFFER_SIZE(1);\
+        _SN_MK_EXPAND_BUFFER_SIZE(2);
 
-#define _SN_MK_EXPAND_LEFT_RIGHT_ANS_SCALAR                                          \
+#define _SN_MK_EXPAND_LEFT_RIGHT_ANS_SCALAR                                                 \
         SN_MAT_KERNEL_LOCAL const *left = (SN_MAT_KERNEL_LOCAL const *)task->buffer_0.ptr;  \
         SN_MAT_KERNEL_LOCAL const *right = (SN_MAT_KERNEL_LOCAL const *)task->buffer_1.ptr; \
         SN_MAT_KERNEL_LOCAL *restrict ans = (SN_MAT_KERNEL_LOCAL *)task->buffer_2.ptr;      \
-        scalar_pack *restrict scalar = (scalar_pack *restrict)task->scalar;
+        scalar_pack *restrict scalar = (scalar_pack *restrict)task->scalar;\
+        _SN_MK_EXPAND_BUFFER_SIZE(0);\
+        _SN_MK_EXPAND_BUFFER_SIZE(1);\
+        _SN_MK_EXPAND_BUFFER_SIZE(2);
 
-#define _SN_MK_EXPAND_ANS_SCALAR                                                \
+#define _SN_MK_EXPAND_ANS_SCALAR                                                       \
         SN_MAT_KERNEL_LOCAL *restrict ans = (SN_MAT_KERNEL_LOCAL *)task->buffer_2.ptr; \
-        scalar_pack *restrict scalar = (scalar_pack *restrict)&(task->padding);
+        scalar_pack *restrict scalar = (scalar_pack *restrict)&(task->padding);\
+        _SN_MK_EXPAND_BUFFER_SIZE(2);
 
 #define _SN_PARAM_STYLE(T) _SN_MATPROC_##T
 #define _SN_MK_EXPAND_PARAM(style) _SN_MK_EXPAND_##style
 
 #pragma endregion
 
-#pragma region mem layout
+#define __mk_param_std(param, T)                               \
-
+        typedef MAT_KERNEL(T) SN_MAT_KERNEL_LOCAL;             \
-#define _SN_MATPROC_LAYOUT_BUFFER 1
+        matrix_task const *_ = (_SN_PARAM_STYLE(param), task); \
-#define _SN_MATPROC_LAYOUT_TILE 1
+        _SN_MK_EXPAND_PARAM(param)
-#define _SN_MATPROC_LAYOUT_KERNEL 1
-
-#define _MK_MKPROC_EXPAND_INDEX_BUFFER                            \
-        size_t offset = UNSAFE_AS(size_t, task->padding.data[0]); \
-        size_t count = UNSAFE_AS(size_t, task->padding.data[1]);
-
-#define _MK_MKPROC_EXPAND_INDEX_BUFFER                            \
-        size_t offset = UNSAFE_AS(size_t, task->padding.data[0]); \
-        size_t count = UNSAFE_AS(size_t, task->padding.data[1]);
-
-#define _MK_MKPROC_EXPAND_INDEX_KERNEL                       \
-        size_t x = UNSAFE_AS(size_t, task->x);               \
-        size_t y = UNSAFE_AS(size_t, task->y);               \
-        size_t common_x = UNSAFE_AS(size_t, task->common_x); \
-        size_t common_y = UNSAFE_AS(size_t, task->common_y); \
-        size_t common_z = UNSAFE_AS(size_t, task->common_z);
-
-#define _MK_MKPROC_EXPAND_INDEX_TILE                         \
-        size_t x = UNSAFE_AS(size_t, task->x);               \
-        size_t y = UNSAFE_AS(size_t, task->y);               \
-        size_t common_x = UNSAFE_AS(size_t, task->common_x); \
-        size_t common_y = UNSAFE_AS(size_t, task->common_y); \
-        size_t common_z = UNSAFE_AS(size_t, task->common_z);
-
-#define _SN_MK_EXPAND_MEM(mem) _MK_MKPROC_EXPAND_INDEX_##mem
-#define _SN_MEM_LAYOUT(T) _SN_MATPROC_LAYOUT_##T
-
-#pragma endregion
-
-#define __mk_param_std(mem, param, T)                                                 \
-        typedef MAT_KERNEL(T) SN_MAT_KERNEL_LOCAL;                                    \
-        matrix_task_pack const *_ = (_SN_PARAM_STYLE(param), (_SN_MEM_LAYOUT(mem)), task); \
-        _SN_MK_EXPAND_PARAM(param);                                                   \
-        _SN_MK_EXPAND_MEM(mem)
 
 #define __sk_param_std(T)                          \
         typedef MAT_KERNEL(T) SN_MAT_KERNEL_LOCAL; \

StgSharp.Native/sn-src/mk.fma.x86.c

@@ -17,28 +17,28 @@
 
 SN_MK_PROC_DECL_STD(float, sse, , fma_pack)
 {
-        __mk_param_std(TILE, RIGHT_ANS, float);
+        __mk_param_std(RIGHT_ANS, float);
 
         memcpy(ans, right, sizeof(MAT_KERNEL(float)) * common_y * common_z);
 }
 
 SN_MK_PROC_DECL_STD(float, avx, , fma_pack)
 {
-        __mk_param_std(TILE, RIGHT_ANS, float);
+        __mk_param_std(RIGHT_ANS, float);
 
         memcpy(ans, right, sizeof(MAT_KERNEL(float)) * common_y * common_z);
 }
 
 SN_MK_PROC_DECL_STD(float, 512, , fma_pack)
 {
-        __mk_param_std(TILE, RIGHT_ANS, float);
+        __mk_param_std(RIGHT_ANS, float);
 
         memcpy(ans, right, sizeof(MAT_KERNEL(float)) * common_y * common_z);
 }
 
 SN_MK_PROC_DECL_STD(float, sse, , fma)
 {
-        __mk_param_std(TILE, LEFT_RIGHT_ANS, float);
+        __mk_param_std(LEFT_RIGHT_ANS, float);
 
         SN_MAT_KERNEL_LOCAL *restrict ans_cur =
                 ans + GET_INDEX(common_y, x, y); // GetKernelAddressUnsafe
@@ -95,7 +95,7 @@ SN_MK_PROC_DECL_STD(float, sse, , fma)
 
 SN_MK_PROC_DECL_STD(float, avx, , fma)
 {
-        __mk_param_std(TILE, LEFT_RIGHT_ANS, float);
+        __mk_param_std(LEFT_RIGHT_ANS, float);
         SN_MAT_KERNEL_LOCAL *restrict ans_cur = ans + GET_INDEX(common_y, x, y);
         register __m256 c0 = _mm256_load_ps((float *)&ans_cur->y[0]);
         register __m256 c1 = _mm256_load_ps((float *)&ans_cur->y[1]);
@@ -130,7 +130,7 @@ SN_MK_PROC_DECL_STD(float, avx, , fma)
 
 SN_MK_PROC_DECL_STD(float, avx, _fma, fma)
 {
-        __mk_param_std(TILE, LEFT_RIGHT_ANS, float);
+        __mk_param_std(LEFT_RIGHT_ANS, float);
         SN_MAT_KERNEL_LOCAL *restrict ans_cur = ans + GET_INDEX(common_y, x, y);
         register __m256 c0 = _mm256_load_ps((float *)&ans_cur->y[0]);
         register __m256 c1 = _mm256_load_ps((float *)&ans_cur->y[1]);
@@ -166,7 +166,7 @@ SN_MK_PROC_DECL_STD(float, avx, _fma, fma)
 
 SN_MK_PROC_DECL_STD(float, 512, , fma)
 {
-        __mk_param_std(TILE, LEFT_RIGHT_ANS, float);
+        __mk_param_std(LEFT_RIGHT_ANS, float);
         SN_MAT_KERNEL_LOCAL *restrict ans_cur = ans + GET_INDEX(common_y, x, y);
         register __m512 c0 = _mm512_load_ps((float *)&ans_cur->z[0]);
         register __m512 c1 = _mm512_load_ps((float *)&ans_cur->z[0]);
@@ -202,7 +202,7 @@ SN_MK_PROC_DECL_STD(float, 512, , fma)
 
 SN_MK_PROC_DECL_STD(double, sse, , fma)
 {
-        __mk_param_std(TILE, LEFT_RIGHT_ANS, double);
+        __mk_param_std(LEFT_RIGHT_ANS, double);
 
         SN_MAT_KERNEL_LOCAL *restrict ans_cur =
                 ans + GET_INDEX(common_y, x, y); // GetKernelAddressUnsafe
@@ -243,7 +243,7 @@ SN_MK_PROC_DECL_STD(double, sse, , fma)
 
 SN_MK_PROC_DECL_STD(double, avx, , fma)
 {
-        __mk_param_std(TILE, LEFT_RIGHT_ANS, double);
+        __mk_param_std(LEFT_RIGHT_ANS, double);
 
         SN_MAT_KERNEL_LOCAL *restrict ans_cur = ans + GET_INDEX(common_y, x, y); // tKernelAddressUnsafe
 
@@ -315,7 +315,7 @@ SN_MK_PROC_DECL_STD(double, avx, , fma)
 
 SN_MK_PROC_DECL_STD(double, avx, _fma, fma)
 {
-        __mk_param_std(TILE, LEFT_RIGHT_ANS, double);
+        __mk_param_std(LEFT_RIGHT_ANS, double);
 
         SN_MAT_KERNEL_LOCAL *restrict ans_cur = ans + GET_INDEX(common_y, x, y); // GetKernelAddressUnsafe
 
@@ -371,7 +371,7 @@ SN_MK_PROC_DECL_STD(double, avx, _fma, fma)
 
 SN_MK_PROC_DECL_STD(double, 512, , fma)
 {
-        __mk_param_std(TILE, LEFT_RIGHT_ANS, double);
+        __mk_param_std(LEFT_RIGHT_ANS, double);
         SN_MAT_KERNEL_LOCAL *restrict ans_cur = ans + GET_INDEX(common_y, x, y);
 
         register __m512d c0 = _mm512_load_pd((double *)&ans_cur->z[0]);

future/ngra-parser-design.md

@@ -0,0 +1,578 @@
+# NGRA → HIR → SSA → LLVM 设计笔记
+
+## 当前结论
+
+NGRA 不应直接承担完整 SSA 构建与数据流分析职责。
+
+NGRA 的职责是：
+
+* Token 分析
+* 岛屿压缩（Island Compression）
+* 结构识别
+* 局部语义绑定
+* 元数据传播
+* AST 构建
+
+SSA 的职责是：
+
+* CFG 构建
+* Def-Use 分析
+* Phi 插入
+* Value Versioning
+* 全局优化
+
+---
+
+# 总体架构
+
+推荐架构：
+
+```text
+Source Code
+    ↓
+NGRA
+    ↓
+五角星AST
+    ↓
+HIR（寄存器式）
+    ↓
+CFG
+    ↓
+SSA
+    ↓
+LLVM IR
+    ↓
+LLVM Backend
+    ↓
+Native Code / SPIR-V / CUDA
+```
+
+而不是：
+
+```text
+AST
+ ↓
+直接LLVM IR
+```
+
+原因：
+
+LLVM IR 过于底层。
+
+NGRA 与语言语义高度相关。
+
+中间需要一个：
+
+```text
+Language HIR
+```
+
+作为桥梁。
+
+---
+
+# 五角星AST职责
+
+五角星AST承担：
+
+```text
+Syntax
++
+Metadata
++
+Partial Semantic
+```
+
+信息。
+
+节点包含：
+
+```csharp
+Parent
+Prev
+Next
+Left
+Right
+```
+
+同时可附带：
+
+```csharp
+Type
+Symbol
+Attribute
+SourceSpan
+Flags
+```
+
+因此无需再构造独立 Bound AST。
+
+---
+
+# SCG（Semi Context Grammar）
+
+SCG 不等于 SSA。
+
+SCG 只负责：
+
+## 1. 声明语义
+
+例如：
+
+```csharp
+class Foo : Bar
+```
+
+直接得到：
+
+```text
+Foo
+Bar
+Inheritance
+```
+
+---
+
+## 2. Token级语义
+
+例如：
+
+```csharp
+123
+```
+
+立即得到：
+
+```text
+Int32
+```
+
+例如：
+
+```csharp
+int a = "abc";
+```
+
+立即报错。
+
+---
+
+## 3. 局部表达式语义
+
+例如：
+
+```csharp
+int + float
+```
+
+归约时：
+
+```text
+Implicit Conversion
+Operator Resolution
+```
+
+直接完成。
+
+---
+
+## 4. 上下文元数据
+
+允许读取：
+
+```text
+当前岛屿
+父岛屿
+全局声明表
+```
+
+中的信息。
+
+例如：
+
+```csharp
+namespace A
+{
+    class B
+}
+```
+
+Class Island 自动继承：
+
+```text
+CurrentNamespace=A
+```
+
+---
+
+# SCG边界
+
+SCG允许：
+
+```text
+当前节点
+当前归约栈
+当前岛屿元数据
+父岛屿元数据
+全局声明表
+```
+
+SCG禁止：
+
+```text
+CFG
+Phi
+Def-Use
+Alias Analysis
+Call Graph
+Loop Analysis
+```
+
+这些属于 SSA 阶段。
+
+---
+
+# HIR设计
+
+最终决定：
+
+不采用栈机IR。
+
+原因：
+
+AST 已经表达了操作数关系。
+
+如果先降成：
+
+```text
+load
+load
+add
+store
+```
+
+属于：
+
+```text
+丢失结构
+再恢复结构
+```
+
+没有意义。
+
+---
+
+# HIR采用寄存器式设计
+
+推荐形式：
+
+```text
+%0 = const 2
+%1 = mul c, %0
+%2 = add b, %1
+
+a := %2
+```
+
+而不是：
+
+```text
+load b
+load c
+push 2
+mul
+add
+store a
+```
+
+---
+
+# Value模型
+
+HIR中的表达式全部返回 Value。
+
+例如：
+
+```csharp
+HirValue EmitExpr(AstNode node)
+```
+
+返回：
+
+```csharp
+ValueId
+```
+
+---
+
+# Symbol与Value分离
+
+保留：
+
+```text
+SymbolId
+ValueId
+```
+
+两套体系。
+
+例如：
+
+```text
+a
+b
+c
+```
+
+属于：
+
+```text
+SymbolId
+```
+
+表达式结果：
+
+```text
+%0
+%1
+%2
+```
+
+属于：
+
+```text
+ValueId
+```
+
+---
+
+# Assignment处理
+
+例如：
+
+```csharp
+a = b + c;
+```
+
+生成：
+
+```text
+%0 = add b,c
+
+a := %0
+```
+
+此时：
+
+```text
+a
+```
+
+仍不是 SSA。
+
+只是：
+
+```text
+Symbol → Current Value
+```
+
+映射。
+
+---
+
+# SSA阶段
+
+SSA阶段负责：
+
+```text
+a
+↓
+a_0
+a_1
+a_2
+...
+```
+
+版本化。
+
+例如：
+
+```csharp
+a = 1;
+a = 2;
+```
+
+变：
+
+```text
+a_0 = 1
+a_1 = 2
+```
+
+---
+
+# CFG构建
+
+真正的分界线
+
+不是：
+
+```text
+AST
+↓
+HIR
+```
+
+而是：
+
+```text
+HIR
+↓
+CFG
+```
+
+因为：
+
+```text
+树
+↓
+图
+```
+
+在这里发生。
+
+例如：
+
+```csharp
+if(cond)
+{
+}
+else
+{
+}
+```
+
+生成：
+
+```text
+BB0
+  br cond
+
+BB1
+  ...
+
+BB2
+  ...
+
+BB3
+```
+
+---
+
+# LLVM IR定位
+
+LLVM IR不是语言IR。
+
+LLVM IR是：
+
+```text
+SSA Optimization IR
+```
+
+职责：
+
+```text
+跨平台
+优化
+后端接口
+```
+
+不负责：
+
+```text
+泛型
+属性
+高级语义
+```
+
+---
+
+# 最终路线
+
+```text
+NGRA
+ ↓
+SCG
+ ↓
+五角星AST
+ ↓
+寄存器式HIR
+ ↓
+CFG
+ ↓
+SSA
+ ↓
+LLVM IR
+ ↓
+x86/ARM
+
+或
+
+SPIR-V
+
+或
+
+CUDA/NVVM
+```
+
+---
+
+# 对蓝图系统的启发
+
+SSA最重要的启发：
+
+程序可以拆成：
+
+```text
+Control Flow Graph
++
+Data Flow Graph
+```
+
+两张图。
+
+因此未来节点管线系统可考虑：
+
+```text
+State Graph
++
+Value Graph
+```
+
+双图结构。
+
+从：
+
+```text
+DAG
+```
+
+演化为：
+
+```text
+Parallel State Machine
++
+Data Dependency Graph
+```
+
+这样能够自然支持：
+
+```text
+Branch
+Loop
+Coroutine
+Async
+Event
+```
+
+等复杂控制流。