Hugging Face's logo

HelloSun
/
flint_cpu_opt

Text Generation
GGUF
llama.cpp
cpu
avx512-vnni
amx
Mixture of Experts
flint
nvfp4
conversational
Model card Files
xet
 Community
flint_cpu_opt / patches /flint-cpu-opt.patch
HelloSun's picture
HelloSun
Add NVFP4 CPU kernels results, integrate model card, add handoff
45f1d0b verified
Raw History Blame Contribute Delete
96.2 kB
diff --git a/ggml/include/ggml.h b/ggml/include/ggml.h
index d62a63104..0cc5ed9cb 100644
--- a/ggml/include/ggml.h
+++ b/ggml/include/ggml.h
@@ -2678,6 +2678,25 @@ extern "C" {
struct ggml_tensor * state,
int64_t K);
+ // Cache variant of ggml_gated_delta_net (single-snapshot, K == 1): instead of a gathered copy of the
+ // initial state, the state of sequence i is read straight from row ids[i] of the state cache, and the
+ // final state is written in place to row i of state_dst (a view of consecutive cache rows), so the caller
+ // needs no get_rows before and no cpy after the op.
+ // state_cache : [S_v*S_v*H_v, n_cells] f32, contiguous rows
+ // ids : [n_seqs] i32, source cell of every sequence
+ // state_dst : [S_v*S_v*H_v, n_seqs] f32 view into the cache (destination cells)
+ // The packed output keeps the layout of ggml_gated_delta_net; its snapshot part is left unwritten.
+ GGML_API struct ggml_tensor * ggml_gated_delta_net_cache(
+ struct ggml_context * ctx,
+ struct ggml_tensor * q,
+ struct ggml_tensor * k,
+ struct ggml_tensor * v,
+ struct ggml_tensor * g,
+ struct ggml_tensor * beta,
+ struct ggml_tensor * state_cache,
+ struct ggml_tensor * ids,
+ struct ggml_tensor * state_dst);
+
// DSA lightning indexer
//
// q: [n_embd_idx, n_head_idx, n_batch, ne3 ]
diff --git a/ggml/src/ggml-cpu/amx/amx.cpp b/ggml/src/ggml-cpu/amx/amx.cpp
index e90c2797f..e8b61a2eb 100644
--- a/ggml/src/ggml-cpu/amx/amx.cpp
+++ b/ggml/src/ggml-cpu/amx/amx.cpp
@@ -161,6 +161,17 @@ class extra_buffer_type : ggml::cpu::extra_buffer_type {
if (op->ne[0] % (TILE_N * 2)) {
return false;
}
+ // machine-specific knobs: keep selected types out of the AMX buffer (decode then uses the VNNI vec_dot kernels)
+ {
+ // Q6_K (lm_head): the AVX512-VNNI vec_dot is ~9% faster for decode than the AMX M=1 kernel; opt back in with GGML_AMX_Q6K=1
+ static const char * keep_q6 = getenv("GGML_AMX_Q6K");
+ if (src0->type == GGML_TYPE_Q6_K && !(keep_q6 && keep_q6[0] == '1')) { return false; }
+ static const char * skip = getenv("GGML_AMX_SKIP");
+ if (skip) {
+ if (src0->type == GGML_TYPE_Q6_K && strstr(skip, "q6_K")) { return false; }
+ if (src0->type == GGML_TYPE_Q8_0 && strstr(skip, "q8_0")) { return false; }
+ }
+ }
int alignment;
switch (src0->type) {
case GGML_TYPE_Q4_0:
diff --git a/ggml/src/ggml-cpu/amx/mmq.cpp b/ggml/src/ggml-cpu/amx/mmq.cpp
index 9f3a744b5..d23e4d971 100644
--- a/ggml/src/ggml-cpu/amx/mmq.cpp
+++ b/ggml/src/ggml-cpu/amx/mmq.cpp
@@ -1989,6 +1989,91 @@ struct tinygemm_kernel_vnni<block_q8_K, block_iq4_xs, float, BLOCK_M, BLOCK_N, B
(const char *)src0->data + src0_offset + PACKED_INDEX(nb * kTilesN, 0, KB, TILE_SIZE), \
(float *) dst->data + dst_offset + nb_start, ldc)
+// Q8_0 x Q8_0 AMX gemm with register accumulation: one 16x16 output tile at a time, its 16 float row
+// accumulators live in zmm registers across all K blocks (the generic kernel read-modify-writes C through memory
+// every 32-wide K block and converts the A scale per row). The AMX product of block i+1 is issued before the AVX
+// epilogue of block i so the two overlap. Per-block math (cvt(int32) * (dW*dA) accumulated with an fma, blocks in
+// order) is identical to acc_C, so results are bit-identical. Returns false when KB is too large for the scale cache.
+// tile registers are pasted into the asm string by the intrinsics, so they must be literals: one function per slot
+#define Q8_REGACC_ISSUE_FN(NAME, TB_, TA_, TC_, SLOT) \
+static ALWAYS_INLINE void NAME(const block_q8_0 * RESTRICT A, const char * RESTRICT B, int KB, int u, int t, \
+ int mrows, int i, int8_t (*TileA)[TILE_M * TILE_K], int32_t (*TileC)[TILE_M * TILE_N]) { \
+ const int TILE_SIZE = get_tile_size<block_q8_0>(); /* 16 x 34 B + 16 x int32 compensation */ \
+ const char * Bb = B + PACKED_INDEX(t, i, KB, TILE_SIZE); \
+ _tile_loadd(TB_, Bb, TILE_N * VNNI_BLK); \
+ const block_q8_0 * Ab = A + (size_t) (u * TILE_M) * KB + i; \
+ if (mrows == TILE_M) { \
+ _tile_loadd(TA_, Ab->qs, KB * sizeof(block_q8_0)); \
+ } else { \
+ unpack_A(TileA[SLOT], Ab, KB, mrows); \
+ _tile_loadd(TA_, TileA[SLOT], TILE_K); \
+ } \
+ _tile_zero(TC_); \
+ _tile_dpbssd(TC_, TA_, TB_); \
+ _tile_stored(TC_, TileC[SLOT], TILE_N * sizeof(int32_t)); \
+}
+Q8_REGACC_ISSUE_FN(q8_regacc_issue0, 0, 2, 4, 0)
+Q8_REGACC_ISSUE_FN(q8_regacc_issue1, 1, 3, 5, 1)
+#undef Q8_REGACC_ISSUE_FN
+
+static bool tinygemm_kernel_amx_q8_0_regacc(int M, int KB, const block_q8_0 * RESTRICT A, const char * RESTRICT B,
+ float * RESTRICT C, int ldc) {
+ constexpr int MAX_KB = 256;
+ if (KB > MAX_KB || KB < 1) {
+ return false;
+ }
+ const int TILE_SIZE = get_tile_size<block_q8_0>(); /* 16 x 34 B + 16 x int32 compensation */
+ constexpr int SCALE_OFF = TILE_N * TILE_K;
+
+ alignas(64) static thread_local int8_t TileA[2][TILE_M * TILE_K];
+ alignas(64) static thread_local int32_t TileC[2][TILE_M * TILE_N];
+ alignas(64) static thread_local float sA[2 * TILE_M][MAX_KB];
+
+ const int rows = std::min(M, 2 * TILE_M);
+ for (int r = 0; r < 2 * TILE_M; ++r) {
+ for (int i = 0; i < KB; ++i) {
+ sA[r][i] = r < rows ? GGML_CPU_FP16_TO_FP32(A[(size_t) r * KB + i].d) : 0.0f;
+ }
+ }
+
+ const int nu = rows > TILE_M ? 2 : 1;
+ for (int u = 0; u < nu; ++u) {
+ const int mrows = std::min(TILE_M, rows - u * TILE_M);
+ for (int t = 0; t < 2; ++t) {
+ __m512 acc[TILE_M];
+ #pragma GCC unroll 16
+ for (int m = 0; m < TILE_M; ++m) {
+ acc[m] = _mm512_setzero_ps();
+ }
+
+ q8_regacc_issue0(A, B, KB, u, t, mrows, 0, TileA, TileC);
+ for (int i = 0; i < KB; ++i) {
+ const int slot = i & 1;
+ if (i + 1 < KB) {
+ if (slot == 0) {
+ q8_regacc_issue1(A, B, KB, u, t, mrows, i + 1, TileA, TileC);
+ } else {
+ q8_regacc_issue0(A, B, KB, u, t, mrows, i + 1, TileA, TileC);
+ }
+ }
+ const __m512 vd0 = _mm512_cvtph_ps(_mm256_loadu_si256((const __m256i *) (B + PACKED_INDEX(t, i, KB, TILE_SIZE) + SCALE_OFF)));
+ const int32_t * tile = TileC[slot];
+ const float * sa = &sA[u * TILE_M][0];
+ #pragma GCC unroll 16
+ for (int m = 0; m < TILE_M; ++m) {
+ const __m512 vtile = _mm512_cvtepi32_ps(_mm512_loadu_si512(tile + m * TILE_N));
+ const __m512 vd1 = _mm512_set1_ps(sa[(size_t) m * MAX_KB + i]);
+ acc[m] = _mm512_fmadd_ps(vtile, _mm512_mul_ps(vd0, vd1), acc[m]);
+ }
+ }
+ for (int m = 0; m < mrows; ++m) {
+ _mm512_storeu_ps(C + (size_t) (u * TILE_M + m) * ldc + t * TILE_N, acc[m]);
+ }
+ }
+ }
+ return true;
+}
+
template <typename TA, typename TB, typename TC, int BLOCK_K,
typename std::enable_if<!is_type_qkk<TB>::value, int>::type = 0>
void tinygemm_kernel_amx(int M, int N, int KB, const void * RESTRICT _A, const void * RESTRICT _B, TC * RESTRICT C, int ldc) {
@@ -1997,6 +2082,12 @@ void tinygemm_kernel_amx(int M, int N, int KB, const void * RESTRICT _A, const v
const bool need_unpack = do_unpack<TB>::value;
GGML_ASSERT(M <= 2 * TILE_M && N == 2 * TILE_N);
+ if constexpr (std::is_same<TB, block_q8_0>::value && std::is_same<TA, block_q8_0>::value) {
+ static const bool regacc = [] { const char * e = getenv("GGML_AMX_Q8_REGACC"); return !(e && e[0] == '0'); }();
+ if (regacc && tinygemm_kernel_amx_q8_0_regacc(M, KB, static_cast<const block_q8_0 *>(_A), static_cast<const char *>(_B), C, ldc)) {
+ return;
+ }
+ }
const TA * RESTRICT A = static_cast<const TA *>(_A);
const char * RESTRICT B = static_cast<const char *>(_B);
@@ -2434,8 +2525,9 @@ void ggml_backend_amx_mul_mat(const ggml_compute_params * params, struct ggml_te
if (M == 1) {
// MB = 1 and handle 8 tiles in each block
- constexpr int kTilesN = 4;
- constexpr int BLOCK_N = TILE_N * kTilesN;
+ // small N (e.g. 512) would give fewer tasks than threads with 64-wide blocks; use 32-wide blocks
+ const int BLOCK_N = (N <= 1024) ? 32 : 64;
+ const int kTilesN = BLOCK_N / TILE_N;
const int NB = div_up(N, BLOCK_N);
parallel_for_ggml(params, n_batch * NB, [&](int begin, int end) {
diff --git a/ggml/src/ggml-cpu/arch/x86/quants.c b/ggml/src/ggml-cpu/arch/x86/quants.c
index 21cd6312a..37efb7fae 100644
--- a/ggml/src/ggml-cpu/arch/x86/quants.c
+++ b/ggml/src/ggml-cpu/arch/x86/quants.c
@@ -941,6 +941,49 @@ void ggml_vec_dot_q4_1_q8_1(int n, float * GGML_RESTRICT s, size_t bs, const voi
__m256 acc = _mm256_setzero_ps();
float summs = 0;
+ float vtot = 0.0f;
+
+#if defined(__AVX512F__) && defined(__AVX512BW__) && defined(__AVX512VL__) && defined(__AVX512DQ__) && defined(__AVX512VNNI__) && defined(__F16C__)
+ // AVX512-VNNI path: two 32-wide blocks per iteration. x codes are 0..15 (unsigned), y codes are int8, so
+ // vpdpbusd gives the exact int32 sums; per-block d*dy is applied per 8-lane half, m*sy is accumulated separately.
+ {
+ static int q41pf = -1;
+ if (q41pf < 0) { const char * e = getenv("GGML_Q41_PF"); q41pf = e ? atoi(e) : 192; }
+ const __m256i m4 = _mm256_set1_epi8(0x0F);
+ __m256 acc0 = _mm256_setzero_ps();
+ __m256 acc1 = _mm256_setzero_ps();
+ __m128 sm = _mm_setzero_ps();
+ const __m256i sidx = _mm256_setr_epi32(0,0,0,0, 2,2,2,2);
+ for (; ib + 3 < nb; ib += 4) {
+ _mm_prefetch((const char *)(x + ib + q41pf), _MM_HINT_T1);
+ _mm_prefetch((const char *)(x + ib + q41pf) + 64, _MM_HINT_T1);
+#define Q41_PAIR(ACC, I) { \
+ const block_q4_1 * xa = x + (I); const block_q4_1 * xb = x + (I) + 1; \
+ const block_q8_1 * ya = y + (I); const block_q8_1 * yb = y + (I) + 1; \
+ /* [block a | block b] in one ymm, built with memory-source vinserti128 (no shuffle-port work) */ \
+ const __m256i v = _mm256_inserti128_si256(_mm256_castsi128_si256(_mm_loadu_si128((const __m128i *) xa->qs)), \
+ _mm_loadu_si128((const __m128i *) xb->qs), 1); \
+ const __m256i lo = _mm256_and_si256(v, m4); \
+ const __m256i hi = _mm256_and_si256(_mm256_srli_epi16(v, 4), m4); \
+ const __m256i ylo = _mm256_inserti128_si256(_mm256_castsi128_si256(_mm_loadu_si128((const __m128i *) ya->qs)), \
+ _mm_loadu_si128((const __m128i *) yb->qs), 1); \
+ const __m256i yhi = _mm256_inserti128_si256(_mm256_castsi128_si256(_mm_loadu_si128((const __m128i *) (ya->qs + 16))), \
+ _mm_loadu_si128((const __m128i *) (yb->qs + 16)), 1); \
+ const __m256i dot = _mm256_dpbusd_epi32(_mm256_dpbusd_epi32(_mm256_setzero_si256(), lo, ylo), hi, yhi); \
+ /* headers: x = (d, m), y = (d, s) -> pp = (d0*dy0, m0*s0, d1*dy1, m1*s1) */ \
+ const __m128 hx = _mm_cvtph_ps(_mm_unpacklo_epi32(_mm_cvtsi32_si128(*(const int *) xa), _mm_cvtsi32_si128(*(const int *) xb))); \
+ const __m128 hy = _mm_cvtph_ps(_mm_unpacklo_epi32(_mm_cvtsi32_si128(*(const int *) ya), _mm_cvtsi32_si128(*(const int *) yb))); \
+ const __m128 pp = _mm_mul_ps(hx, hy); \
+ sm = _mm_add_ps(sm, pp); /* lanes 1 and 3 carry m*s */ \
+ const __m256 sc = _mm256_permutevar8x32_ps(_mm256_castps128_ps256(pp), sidx); \
+ ACC = _mm256_fmadd_ps(sc, _mm256_cvtepi32_ps(dot), ACC); }
+ Q41_PAIR(acc0, ib)
+ Q41_PAIR(acc1, ib + 2)
+#undef Q41_PAIR
+ }
+ vtot = hsum_float_8(_mm256_add_ps(acc0, acc1)) + _mm_cvtss_f32(_mm_shuffle_ps(sm, sm, 0x01)) + _mm_cvtss_f32(_mm_shuffle_ps(sm, sm, 0x03));
+ }
+#endif
// Main loop
for (; ib < nb; ++ib) {
@@ -969,7 +1012,7 @@ void ggml_vec_dot_q4_1_q8_1(int n, float * GGML_RESTRICT s, size_t bs, const voi
#endif
}
- *s = hsum_float_8(acc) + summs;
+ *s = hsum_float_8(acc) + summs + vtot;
#else
UNUSED(nb);
UNUSED(x);
@@ -1080,7 +1123,43 @@ void ggml_vec_dot_nvfp4_q8_0(int n, float * GGML_RESTRICT s, size_t bs, const vo
int ib = 0;
float sumf = 0;
-#if defined(__AVX2__)
+#if defined(__AVX512F__) && defined(__AVX512VL__) && defined(__AVX512VNNI__) && defined(__AVX2__)
+ // AVX512-VNNI path: one 64-element block per iteration. |w| (0..12) is the unsigned vpdpbusd operand and the
+ // sign of w moves onto the int8 activations. Int32 lane k sums sub-block k/2 (lo + hi nibbles of 4 bytes).
+ {
+ const __m256i lut = _mm256_broadcastsi128_si256(_mm_loadu_si128((const __m128i *) kvalues_fp4));
+ const __m256i m4b = _mm256_set1_epi8(0x0f);
+ const __m512i perm_lo = _mm512_setr_epi64(0, 2, 4, 6, 0, 0, 0, 0);
+ const __m512i perm_hi = _mm512_setr_epi64(1, 3, 5, 7, 0, 0, 0, 0);
+ __m256 acc = _mm256_setzero_ps();
+ for (; ib < nb; ++ib) {
+ const __m256i qbits = _mm256_loadu_si256((const __m256i *) x[ib].qs);
+ const __m256i wlo = _mm256_shuffle_epi8(lut, _mm256_and_si256(qbits, m4b));
+ const __m256i whi = _mm256_shuffle_epi8(lut, _mm256_and_si256(_mm256_srli_epi16(qbits, 4), m4b));
+
+ // activation bytes in weight byte order: lo -> elems 16s+j, hi -> elems 16s+8+j
+ const __m512i z = _mm512_inserti64x4(_mm512_castsi256_si512(_mm256_loadu_si256((const __m256i *) y[2*ib + 0].qs)),
+ _mm256_loadu_si256((const __m256i *) y[2*ib + 1].qs), 1);
+ const __m256i alo = _mm512_castsi512_si256(_mm512_permutexvar_epi64(perm_lo, z));
+ const __m256i ahi = _mm512_castsi512_si256(_mm512_permutexvar_epi64(perm_hi, z));
+
+ const __m256i dot = _mm256_dpbusd_epi32(_mm256_dpbusd_epi32(_mm256_setzero_si256(),
+ _mm256_abs_epi8(wlo), _mm256_sign_epi8(alo, wlo)),
+ _mm256_abs_epi8(whi), _mm256_sign_epi8(ahi, whi));
+
+ const float dy0 = GGML_CPU_FP16_TO_FP32(y[2*ib + 0].d);
+ const float dy1 = GGML_CPU_FP16_TO_FP32(y[2*ib + 1].d);
+ const float s0 = GGML_CPU_UE4M3_TO_FP32(x[ib].d[0]) * dy0;
+ const float s1 = GGML_CPU_UE4M3_TO_FP32(x[ib].d[1]) * dy0;
+ const float s2 = GGML_CPU_UE4M3_TO_FP32(x[ib].d[2]) * dy1;
+ const float s3 = GGML_CPU_UE4M3_TO_FP32(x[ib].d[3]) * dy1;
+ const __m256 scales = _mm256_set_ps(s3, s3, s2, s2, s1, s1, s0, s0);
+
+ acc = _mm256_fmadd_ps(_mm256_cvtepi32_ps(dot), scales, acc);
+ }
+ sumf = hsum_float_8(acc);
+ }
+#elif defined(__AVX2__)
const __m128i values128 = _mm_loadu_si128((const __m128i*)kvalues_fp4);
const __m128i m4b = _mm_set1_epi8(0x0f);
@@ -2118,7 +2197,89 @@ void ggml_vec_dot_q4_K_q8_K(int n, float * GGML_RESTRICT s, size_t bs, const voi
uint32_t utmp[4];
-#if defined __AVX2__
+#if defined(__AVX512F__) && defined(__AVX512BW__) && defined(__AVX512VL__) && defined(__AVX512VNNI__)
+ // AVX512-VNNI path (Q4_K x Q8_K): 64 weights per vector, two 32-wide sub-blocks per vector.
+ const __m512i m4 = _mm512_set1_epi8(0xF);
+ #define K4_SCALE_IDX(a, b) _mm512_inserti64x4( \
+ _mm512_castsi256_si512(_mm256_set_m128i(_mm_set1_epi16((short)((2*(a)) | ((2*(a)+1) << 8))), _mm_set1_epi16((short)((2*(a)) | ((2*(a)+1) << 8))))), \
+ _mm256_set_m128i(_mm_set1_epi16((short)((2*(b)) | ((2*(b)+1) << 8))), _mm_set1_epi16((short)((2*(b)) | ((2*(b)+1) << 8)))), 1)
+ const __m512i sidx0 = K4_SCALE_IDX(0, 2);
+ const __m512i sidx1 = K4_SCALE_IDX(1, 3);
+ const __m512i sidx2 = K4_SCALE_IDX(4, 6);
+ const __m512i sidx3 = K4_SCALE_IDX(5, 7);
+ #undef K4_SCALE_IDX
+
+ __m512 acc = _mm512_setzero_ps();
+ __m128 acc_m = _mm_setzero_ps();
+
+ for (int i = 0; i < nb; ++i) {
+#ifndef K4_PF_DIST
+#define K4_PF_DIST 40 // ~9 KB ahead, tuned for Xeon 8559C (prefetch into L2)
+#endif
+ static int k4pf = -1;
+ if (k4pf < 0) { const char * e = getenv("GGML_K4_PF"); k4pf = e ? atoi(e) : K4_PF_DIST; }
+ {
+ const char * pf = (const char *)(x + i + k4pf);
+ _mm_prefetch(pf, _MM_HINT_T1);
+ _mm_prefetch(pf + 64, _MM_HINT_T1);
+ _mm_prefetch(pf + 128, _MM_HINT_T1);
+ }
+
+ const float d = y[i].d * GGML_CPU_FP16_TO_FP32(x[i].d);
+ const float dmin = -y[i].d * GGML_CPU_FP16_TO_FP32(x[i].dmin);
+
+ memcpy(utmp, x[i].scales, 12);
+ utmp[3] = ((utmp[2] >> 4) & kmask2) | (((utmp[1] >> 6) & kmask3) << 4);
+ const uint32_t uaux = utmp[1] & kmask1;
+ utmp[1] = (utmp[2] & kmask2) | (((utmp[0] >> 6) & kmask3) << 4);
+ utmp[2] = uaux;
+ utmp[0] &= kmask1;
+
+ const uint8_t * GGML_RESTRICT q4 = x[i].qs;
+ const int8_t * GGML_RESTRICT q8 = y[i].qs;
+
+ const __m256i mins_and_scales = _mm256_cvtepu8_epi16(_mm_set_epi32(utmp[3], utmp[2], utmp[1], utmp[0]));
+
+ const __m256i q8sums = _mm256_loadu_si256((const __m256i*)y[i].bsums);
+ const __m128i q8s = _mm_hadd_epi16(_mm256_extracti128_si256(q8sums, 0), _mm256_extracti128_si256(q8sums, 1));
+ const __m128i prod = _mm_madd_epi16(_mm256_extracti128_si256(mins_and_scales, 1), q8s);
+ acc_m = _mm_fmadd_ps(_mm_set1_ps(dmin), _mm_cvtepi32_ps(prod), acc_m);
+
+ const __m512i scales = _mm512_broadcast_i32x4(_mm256_extracti128_si256(mins_and_scales, 0));
+
+ const __m512i v0 = _mm512_loadu_si512((const void *)(q4));
+ const __m512i v1 = _mm512_loadu_si512((const void *)(q4 + 64));
+
+ const __m512i Q0 = _mm512_loadu_si512((const void *)(q8));
+ const __m512i Q1 = _mm512_loadu_si512((const void *)(q8 + 64));
+ const __m512i Q2 = _mm512_loadu_si512((const void *)(q8 + 128));
+ const __m512i Q3 = _mm512_loadu_si512((const void *)(q8 + 192));
+
+ const __m512i a0 = _mm512_and_si512(v0, m4);
+ const __m512i a1 = _mm512_and_si512(_mm512_srli_epi16(v0, 4), m4);
+ const __m512i a2 = _mm512_and_si512(v1, m4);
+ const __m512i a3 = _mm512_and_si512(_mm512_srli_epi16(v1, 4), m4);
+
+ const __m512i b0 = _mm512_shuffle_i64x2(Q0, Q1, 0x44);
+ const __m512i b1 = _mm512_shuffle_i64x2(Q0, Q1, 0xEE);
+ const __m512i b2 = _mm512_shuffle_i64x2(Q2, Q3, 0x44);
+ const __m512i b3 = _mm512_shuffle_i64x2(Q2, Q3, 0xEE);
+
+ __m512i sumi = _mm512_setzero_si512();
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(a0, b0), _mm512_shuffle_epi8(scales, sidx0));
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(a1, b1), _mm512_shuffle_epi8(scales, sidx1));
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(a2, b2), _mm512_shuffle_epi8(scales, sidx2));
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(a3, b3), _mm512_shuffle_epi8(scales, sidx3));
+
+ acc = _mm512_fmadd_ps(_mm512_set1_ps(d), _mm512_cvtepi32_ps(sumi), acc);
+ }
+
+ acc_m = _mm_add_ps(acc_m, _mm_movehl_ps(acc_m, acc_m));
+ acc_m = _mm_add_ss(acc_m, _mm_movehdup_ps(acc_m));
+
+ *s = _mm512_reduce_add_ps(acc) + _mm_cvtss_f32(acc_m);
+
+#elif defined __AVX2__
const __m256i m4 = _mm256_set1_epi8(0xF);
@@ -2296,7 +2457,104 @@ void ggml_vec_dot_q5_K_q8_K(int n, float * GGML_RESTRICT s, size_t bs, const voi
uint32_t utmp[4];
-#if defined __AVX2__
+#if defined(__AVX512F__) && defined(__AVX512BW__) && defined(__AVX512VL__) && defined(__AVX512VNNI__)
+ // AVX512-VNNI path (Q5_K x Q8_K), same structure as the Q4_K kernel: 64 weights per vector, two 32-wide
+ // sub-blocks per vector. The 5th bit of sub-block s is bit s of qh[l]; it is merged in as +16 before maddubs.
+ const __m512i m4 = _mm512_set1_epi8(0xF);
+ const __m512i m10 = _mm512_set1_epi8(0x10);
+ #define K5_SCALE_IDX(a, b) _mm512_inserti64x4( \
+ _mm512_castsi256_si512(_mm256_set_m128i(_mm_set1_epi16((short)((2*(a)) | ((2*(a)+1) << 8))), _mm_set1_epi16((short)((2*(a)) | ((2*(a)+1) << 8))))), \
+ _mm256_set_m128i(_mm_set1_epi16((short)((2*(b)) | ((2*(b)+1) << 8))), _mm_set1_epi16((short)((2*(b)) | ((2*(b)+1) << 8)))), 1)
+ const __m512i sidx0 = K5_SCALE_IDX(0, 2);
+ const __m512i sidx1 = K5_SCALE_IDX(1, 3);
+ const __m512i sidx2 = K5_SCALE_IDX(4, 6);
+ const __m512i sidx3 = K5_SCALE_IDX(5, 7);
+ #undef K5_SCALE_IDX
+ // per-vector bit selectors: low 32 bytes test bit s, high 32 bytes test bit s+2 (sub-block pairs (0,2),(1,3),(4,6),(5,7))
+ #define K5_BM(a, b) _mm512_inserti64x4(_mm512_castsi256_si512(_mm256_set1_epi8((char)(1 << (a)))), _mm256_set1_epi8((char)(1 << (b))), 1)
+ const __m512i bm0 = K5_BM(0, 2);
+ const __m512i bm1 = K5_BM(1, 3);
+ const __m512i bm2 = K5_BM(4, 6);
+ const __m512i bm3 = K5_BM(5, 7);
+ #undef K5_BM
+
+ static int k5pf = -1;
+ if (k5pf < 0) { const char * e = getenv("GGML_K5_PF"); k5pf = e ? atoi(e) : 48; }
+
+ __m512 acc = _mm512_setzero_ps();
+ __m128 acc_m = _mm_setzero_ps();
+
+ for (int i = 0; i < nb; ++i) {
+ {
+ const char * pf = (const char *)(x + i + k5pf);
+ _mm_prefetch(pf, _MM_HINT_T1);
+ _mm_prefetch(pf + 64, _MM_HINT_T1);
+ _mm_prefetch(pf + 128, _MM_HINT_T1);
+ }
+
+ const float d = y[i].d * GGML_CPU_FP16_TO_FP32(x[i].d);
+ const float dmin = -y[i].d * GGML_CPU_FP16_TO_FP32(x[i].dmin);
+
+ memcpy(utmp, x[i].scales, 12);
+ utmp[3] = ((utmp[2] >> 4) & kmask2) | (((utmp[1] >> 6) & kmask3) << 4);
+ const uint32_t uaux = utmp[1] & kmask1;
+ utmp[1] = (utmp[2] & kmask2) | (((utmp[0] >> 6) & kmask3) << 4);
+ utmp[2] = uaux;
+ utmp[0] &= kmask1;
+
+ const uint8_t * GGML_RESTRICT q5 = x[i].qs;
+ const int8_t * GGML_RESTRICT q8 = y[i].qs;
+
+ const __m256i mins_and_scales = _mm256_cvtepu8_epi16(_mm_set_epi32(utmp[3], utmp[2], utmp[1], utmp[0]));
+
+ const __m256i q8sums = _mm256_loadu_si256((const __m256i*)y[i].bsums);
+ const __m128i q8s = _mm_hadd_epi16(_mm256_extracti128_si256(q8sums, 0), _mm256_extracti128_si256(q8sums, 1));
+ const __m128i prod = _mm_madd_epi16(_mm256_extracti128_si256(mins_and_scales, 1), q8s);
+ acc_m = _mm_fmadd_ps(_mm_set1_ps(dmin), _mm_cvtepi32_ps(prod), acc_m);
+
+ const __m512i scales = _mm512_broadcast_i32x4(_mm256_extracti128_si256(mins_and_scales, 0));
+
+ const __m512i qhv = _mm512_broadcast_i64x4(_mm256_loadu_si256((const __m256i *)x[i].qh));
+
+ const __m512i v0 = _mm512_loadu_si512((const void *)(q5));
+ const __m512i v1 = _mm512_loadu_si512((const void *)(q5 + 64));
+
+ const __m512i Q0 = _mm512_loadu_si512((const void *)(q8));
+ const __m512i Q1 = _mm512_loadu_si512((const void *)(q8 + 64));
+ const __m512i Q2 = _mm512_loadu_si512((const void *)(q8 + 128));
+ const __m512i Q3 = _mm512_loadu_si512((const void *)(q8 + 192));
+
+ // a_k = nibble + 16 * bit(qh[l], sub-block). qhv holds qh (32 B) duplicated in both 256-bit halves;
+ // lanes 0-31 of a vector are sub-block s (low half) and 32-63 are sub-block s+2 (high half) -> per-half bit masks.
+ const __m512i n0 = _mm512_and_si512(v0, m4);
+ const __m512i a0 = _mm512_mask_add_epi8(n0, _mm512_test_epi8_mask(qhv, bm0), n0, m10);
+ const __m512i n1 = _mm512_and_si512(_mm512_srli_epi16(v0, 4), m4);
+ const __m512i a1 = _mm512_mask_add_epi8(n1, _mm512_test_epi8_mask(qhv, bm1), n1, m10);
+ const __m512i n2 = _mm512_and_si512(v1, m4);
+ const __m512i a2 = _mm512_mask_add_epi8(n2, _mm512_test_epi8_mask(qhv, bm2), n2, m10);
+ const __m512i n3 = _mm512_and_si512(_mm512_srli_epi16(v1, 4), m4);
+ const __m512i a3 = _mm512_mask_add_epi8(n3, _mm512_test_epi8_mask(qhv, bm3), n3, m10);
+
+ const __m512i b0 = _mm512_shuffle_i64x2(Q0, Q1, 0x44);
+ const __m512i b1 = _mm512_shuffle_i64x2(Q0, Q1, 0xEE);
+ const __m512i b2 = _mm512_shuffle_i64x2(Q2, Q3, 0x44);
+ const __m512i b3 = _mm512_shuffle_i64x2(Q2, Q3, 0xEE);
+
+ __m512i sumi = _mm512_setzero_si512();
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(a0, b0), _mm512_shuffle_epi8(scales, sidx0));
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(a1, b1), _mm512_shuffle_epi8(scales, sidx1));
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(a2, b2), _mm512_shuffle_epi8(scales, sidx2));
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(a3, b3), _mm512_shuffle_epi8(scales, sidx3));
+
+ acc = _mm512_fmadd_ps(_mm512_set1_ps(d), _mm512_cvtepi32_ps(sumi), acc);
+ }
+
+ acc_m = _mm_add_ps(acc_m, _mm_movehl_ps(acc_m, acc_m));
+ acc_m = _mm_add_ss(acc_m, _mm_movehdup_ps(acc_m));
+
+ *s = _mm512_reduce_add_ps(acc) + _mm_cvtss_f32(acc_m);
+
+#elif defined __AVX2__
const __m256i m4 = _mm256_set1_epi8(0xF);
const __m128i mzero = _mm_setzero_si128();
@@ -2500,7 +2758,81 @@ void ggml_vec_dot_q6_K_q8_K(int n, float * GGML_RESTRICT s, size_t bs, const voi
const int nb = n / QK_K;
-#if defined __AVX2__
+#if defined(__AVX512F__) && defined(__AVX512BW__) && defined(__AVX512VL__) && defined(__AVX512VNNI__)
+ // AVX512-VNNI path (Q6_K x Q8_K): each 64-byte vector holds two 32-wide chunks.
+ const __m512i m15 = _mm512_set1_epi8(0xF);
+ const __m512i m30 = _mm512_set1_epi8(0x30);
+ const __m512i shl_cnt = _mm512_inserti64x4(_mm512_castsi256_si512(_mm256_set1_epi16(4)), _mm256_set1_epi16(2), 1);
+ const __m512i shr_cnt = _mm512_inserti64x4(_mm512_castsi256_si512(_mm256_set1_epi16(0)), _mm256_set1_epi16(2), 1);
+ #define Q6_SIDX(c1, c2) _mm512_inserti64x4( \
+ _mm512_castsi256_si512(_mm256_set_m128i(_mm_set1_epi16((short)(2*(c1)+1)), _mm_set1_epi16((short)(2*(c1))))), \
+ _mm256_set_m128i(_mm_set1_epi16((short)(2*(c2)+1)), _mm_set1_epi16((short)(2*(c2)))), 1)
+ const __m512i sidx0 = Q6_SIDX(0, 1);
+ const __m512i sidx1 = Q6_SIDX(2, 3);
+ const __m512i sidx2 = Q6_SIDX(4, 5);
+ const __m512i sidx3 = Q6_SIDX(6, 7);
+ #undef Q6_SIDX
+
+ __m512 acc = _mm512_setzero_ps();
+
+ for (int i = 0; i < nb; ++i) {
+#ifndef Q6_PF_DIST
+#define Q6_PF_DIST 44 // ~9 KB ahead, tuned for Xeon 8559C (prefetch into L2)
+#endif
+ static int q6pf = -1;
+ if (q6pf < 0) { const char * e = getenv("GGML_Q6_PF"); q6pf = e ? atoi(e) : Q6_PF_DIST; }
+ {
+ const char * pf = (const char *)(x + i + q6pf);
+ _mm_prefetch(pf, _MM_HINT_T1);
+ _mm_prefetch(pf + 64, _MM_HINT_T1);
+ _mm_prefetch(pf + 128, _MM_HINT_T1);
+ _mm_prefetch(pf + 192, _MM_HINT_T1);
+ }
+
+ const float d = y[i].d * GGML_CPU_FP16_TO_FP32(x[i].d);
+
+ const uint8_t * GGML_RESTRICT ql = x[i].ql;
+ const uint8_t * GGML_RESTRICT qh = x[i].qh;
+ const int8_t * GGML_RESTRICT q8 = y[i].qs;
+
+ // -32 offset handled with bsums
+ const __m256i q8sums = _mm256_loadu_si256((const __m256i*)y[i].bsums);
+ const __m256i scales_16 = _mm256_cvtepi8_epi16(_mm_loadu_si128((const __m128i*)x[i].scales));
+ const __m256i q8sclsub = _mm256_slli_epi32(_mm256_madd_epi16(q8sums, scales_16), 5);
+ const __m512i scales512 = _mm512_zextsi256_si512(scales_16);
+
+ const __m512i QL0 = _mm512_loadu_si512((const void *)(ql));
+ const __m512i QL1 = _mm512_loadu_si512((const void *)(ql + 64));
+ const __m512i H0 = _mm512_broadcast_i64x4(_mm256_loadu_si256((const __m256i*)(qh)));
+ const __m512i H1 = _mm512_broadcast_i64x4(_mm256_loadu_si256((const __m256i*)(qh + 32)));
+
+ const __m512i X0 = _mm512_loadu_si512((const void *)(q8));
+ const __m512i X1 = _mm512_loadu_si512((const void *)(q8 + 64));
+ const __m512i X2 = _mm512_loadu_si512((const void *)(q8 + 128));
+ const __m512i X3 = _mm512_loadu_si512((const void *)(q8 + 192));
+
+ // 6-bit values (unsigned 0..63): low nibble / high nibble of ql combined with 2 bits of qh
+ const __m512i q6_0 = _mm512_or_si512(_mm512_and_si512(QL0, m15),
+ _mm512_and_si512(_mm512_sllv_epi16(H0, shl_cnt), m30));
+ const __m512i q6_1 = _mm512_or_si512(_mm512_and_si512(_mm512_srli_epi16(QL0, 4), m15),
+ _mm512_and_si512(_mm512_srlv_epi16(H0, shr_cnt), m30));
+ const __m512i q6_2 = _mm512_or_si512(_mm512_and_si512(QL1, m15),
+ _mm512_and_si512(_mm512_sllv_epi16(H1, shl_cnt), m30));
+ const __m512i q6_3 = _mm512_or_si512(_mm512_and_si512(_mm512_srli_epi16(QL1, 4), m15),
+ _mm512_and_si512(_mm512_srlv_epi16(H1, shr_cnt), m30));
+
+ __m512i sumi = _mm512_zextsi256_si512(_mm256_sub_epi32(_mm256_setzero_si256(), q8sclsub));
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(q6_0, X0), _mm512_permutexvar_epi16(sidx0, scales512));
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(q6_1, X1), _mm512_permutexvar_epi16(sidx1, scales512));
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(q6_2, X2), _mm512_permutexvar_epi16(sidx2, scales512));
+ sumi = _mm512_dpwssd_epi32(sumi, _mm512_maddubs_epi16(q6_3, X3), _mm512_permutexvar_epi16(sidx3, scales512));
+
+ acc = _mm512_fmadd_ps(_mm512_set1_ps(d), _mm512_cvtepi32_ps(sumi), acc);
+ }
+
+ *s = _mm512_reduce_add_ps(acc);
+
+#elif defined __AVX2__
const __m256i m3 = _mm256_set1_epi8(3);
const __m256i m15 = _mm256_set1_epi8(15);
@@ -4078,7 +4410,75 @@ void ggml_vec_dot_iq4_xs_q8_K(int n, float * GGML_RESTRICT s, size_t bs, const v
const int nb = n / QK_K;
-#if defined __AVX2__
+#if defined(__AVX512F__) && defined(__AVX512BW__) && defined(__AVX512VL__) && defined(__AVX512VNNI__)
+ // AVX512-VNNI path (IQ4_XS x Q8_K). 64 qs bytes = 4 sub-blocks (16 B each, low nibble -> values 0..15, high
+ // nibble -> values 16..31 of the same sub-block). The kvalues LUT is stored as unsigned (v + 128) so that
+ // vpdpbusd can be used with 32-bit accumulation; the +128 bias is removed once per super-block through bsums.
+ const __m512i lut = _mm512_broadcast_i32x4(_mm_xor_si128(_mm_loadu_si128((const __m128i *)kvalues_iq4nl), _mm_set1_epi8((char)0x80)));
+ const __m512i m4 = _mm512_set1_epi8(0xF);
+ const __m512i idxA = _mm512_setr_epi32(0,0,0,0, 1,1,1,1, 2,2,2,2, 3,3,3,3);
+ const __m512i idxB = _mm512_setr_epi32(4,4,4,4, 5,5,5,5, 6,6,6,6, 7,7,7,7);
+ const __m128i hsh = _mm_setr_epi16(0, 2, 4, 6, 8, 10, 12, 14);
+
+ static int ipf = -1;
+ if (ipf < 0) { const char * e = getenv("GGML_IQ4XS_PF"); ipf = e ? atoi(e) : 40; }
+
+ __m512 acc = _mm512_setzero_ps();
+
+ for (int ibl = 0; ibl < nb; ++ibl) {
+ {
+ const char * pf = (const char *)(x + ibl + ipf);
+ _mm_prefetch(pf, _MM_HINT_T1);
+ _mm_prefetch(pf + 64, _MM_HINT_T1);
+ _mm_prefetch(pf + 128, _MM_HINT_T1);
+ }
+
+ // 6-bit sub-block scales (-32), 8 per super-block
+ uint32_t sl; memcpy(&sl, x[ibl].scales_l, 4);
+ const __m128i xl = _mm_cvtsi32_si128((int)sl);
+ const __m128i l8 = _mm_unpacklo_epi8(_mm_and_si128(xl, _mm_set1_epi8(0x0f)),
+ _mm_and_si128(_mm_srli_epi16(xl, 4), _mm_set1_epi8(0x0f)));
+ const __m128i h16 = _mm_slli_epi16(_mm_and_si128(_mm_srlv_epi16(_mm_set1_epi16((short)x[ibl].scales_h), hsh), _mm_set1_epi16(3)), 4);
+ const __m128i ls16 = _mm_sub_epi16(_mm_or_si128(_mm_cvtepu8_epi16(l8), h16), _mm_set1_epi16(32));
+ const __m512i ls512 = _mm512_castsi256_si512(_mm256_cvtepi16_epi32(ls16));
+ const __m512i sc0 = _mm512_permutexvar_epi32(idxA, ls512);
+ const __m512i sc1 = _mm512_permutexvar_epi32(idxB, ls512);
+
+ const uint8_t * qs = x[ibl].qs;
+ const int8_t * q8 = y[ibl].qs;
+
+ const __m512i q0 = _mm512_loadu_si512((const void *)(qs));
+ const __m512i q1 = _mm512_loadu_si512((const void *)(qs + 64));
+ const __m512i Y0 = _mm512_loadu_si512((const void *)(q8));
+ const __m512i Y1 = _mm512_loadu_si512((const void *)(q8 + 64));
+ const __m512i Y2 = _mm512_loadu_si512((const void *)(q8 + 128));
+ const __m512i Y3 = _mm512_loadu_si512((const void *)(q8 + 192));
+
+ const __m512i xl0 = _mm512_shuffle_epi8(lut, _mm512_and_si512(q0, m4));
+ const __m512i xh0 = _mm512_shuffle_epi8(lut, _mm512_and_si512(_mm512_srli_epi16(q0, 4), m4));
+ const __m512i xl1 = _mm512_shuffle_epi8(lut, _mm512_and_si512(q1, m4));
+ const __m512i xh1 = _mm512_shuffle_epi8(lut, _mm512_and_si512(_mm512_srli_epi16(q1, 4), m4));
+
+ __m512i p0 = _mm512_dpbusd_epi32(_mm512_setzero_si512(), xl0, _mm512_shuffle_i64x2(Y0, Y1, 0x88));
+ p0 = _mm512_dpbusd_epi32(p0, xh0, _mm512_shuffle_i64x2(Y0, Y1, 0xDD));
+ __m512i p1 = _mm512_dpbusd_epi32(_mm512_setzero_si512(), xl1, _mm512_shuffle_i64x2(Y2, Y3, 0x88));
+ p1 = _mm512_dpbusd_epi32(p1, xh1, _mm512_shuffle_i64x2(Y2, Y3, 0xDD));
+
+ __m512i sumi = _mm512_add_epi32(_mm512_mullo_epi32(p0, sc0), _mm512_mullo_epi32(p1, sc1));
+
+ // remove the +128 bias: 128 * sum_k ls_k * (bsums[2k] + bsums[2k+1])
+ const __m256i bs = _mm256_loadu_si256((const __m256i *)y[ibl].bsums);
+ const __m256i lsd = _mm256_set_m128i(_mm_unpackhi_epi16(ls16, ls16), _mm_unpacklo_epi16(ls16, ls16));
+ const __m512i corr = _mm512_slli_epi32(_mm512_zextsi256_si512(_mm256_madd_epi16(bs, lsd)), 7);
+ sumi = _mm512_sub_epi32(sumi, corr);
+
+ const float d = GGML_CPU_FP16_TO_FP32(x[ibl].d) * y[ibl].d;
+ acc = _mm512_fmadd_ps(_mm512_set1_ps(d), _mm512_cvtepi32_ps(sumi), acc);
+ }
+
+ *s = _mm512_reduce_add_ps(acc);
+
+#elif defined __AVX2__
const __m128i values128 = _mm_loadu_si128((const __m128i*)kvalues_iq4nl);
const __m128i m4b = _mm_set1_epi8(0x0f);
diff --git a/ggml/src/ggml-cpu/ops.cpp b/ggml/src/ggml-cpu/ops.cpp
index 85e17018c..119d3c997 100644
--- a/ggml/src/ggml-cpu/ops.cpp
+++ b/ggml/src/ggml-cpu/ops.cpp
@@ -5316,6 +5316,29 @@ static void ggml_compute_forward_get_rows_f32(
const int ith = params->ith;
const int nth = params->nth;
+ // few rows: parallelize across the row instead of across rows so that a single
+ // large gather (e.g. the recurrent state of a single sequence) uses all threads
+ if (nr < nth && nc >= 4096) {
+ const int64_t per = (nc + nth - 1)/nth;
+ const int64_t c0 = per*ith;
+ const int64_t c1 = MIN(c0 + per, nc);
+ if (c0 < c1) {
+ for (int64_t i = 0; i < nr; ++i) {
+ const int64_t i12 = i/(ne11*ne10);
+ const int64_t i11 = (i - i12*ne11*ne10)/ne10;
+ const int64_t i10 = (i - i12*ne11*ne10 - i11*ne10);
+ const int64_t i01 = *(int32_t *) ((char *) src1->data + i10*nb10 + i11*nb11 + i12*nb12);
+
+ GGML_ASSERT(i01 >= 0 && i01 < ne01);
+
+ memcpy((char *) dst->data + i10*nb1 + i11*nb2 + i12*nb3 + c0*sizeof(float),
+ (char *) src0->data + i01*nb01 + i11*nb02 + i12*nb03 + c0*sizeof(float),
+ (c1 - c0)*sizeof(float));
+ }
+ }
+ return;
+ }
+
// rows per thread
const int dr = (nr + nth - 1)/nth;
@@ -11108,6 +11131,9 @@ static void ggml_compute_forward_gated_delta_net_one_chunk(
ggml_tensor * src_g = dst->src[3];
ggml_tensor * src_beta = dst->src[4];
ggml_tensor * src_state = dst->src[5];
+ // cache mode (see ggml_gated_delta_net_cache): src[6] = source cell ids, src[7] = in-place destination view
+ const ggml_tensor * src_ids = dst->src[6];
+ const ggml_tensor * src_sdst = dst->src[7];
const int64_t S_v = src_v->ne[0];
const int64_t H = src_v->ne[1];
@@ -11119,7 +11145,7 @@ static void ggml_compute_forward_gated_delta_net_one_chunk(
GGML_ASSERT(ggml_is_contiguous_rows(src_v));
GGML_ASSERT(ggml_is_contiguous(src_g));
GGML_ASSERT(ggml_is_contiguous(src_beta));
- GGML_ASSERT(ggml_is_contiguous(src_state));
+ GGML_ASSERT(src_ids ? src_state->nb[1] == (size_t) (S_v * S_v * H) * sizeof(float) : ggml_is_contiguous(src_state));
GGML_ASSERT(src_g->ne[0] == 1 || src_g->ne[0] == S_v);
GGML_ASSERT(src_beta->ne[0] == 1);
@@ -11139,6 +11165,7 @@ static void ggml_compute_forward_gated_delta_net_one_chunk(
// K (snapshot slot count) is an op param; state holds s0 only [S_v, S_v, H, n_seqs].
const int64_t K = ggml_get_op_params_i32(dst, 0);
GGML_ASSERT(K >= 1);
+ GGML_ASSERT(!src_ids || K == 1);
// per-seq stride in floats (seq s starts at state + s * seq_stride)
const int64_t state_seq_stride = src_state->nb[3] / sizeof(float);
@@ -11182,13 +11209,18 @@ static void ggml_compute_forward_gated_delta_net_one_chunk(
// For K>1, work in scratch and copy out per-token when the slot is in range.
float * s_out = (K > 1)
? state_work
- : state_out_base + (iv3 * H + iv1) * S_v * S_v;
+ : (src_sdst
+ ? (float *) src_sdst->data + iv3 * (src_sdst->nb[1] / sizeof(float)) + iv1 * S_v * S_v
+ : state_out_base + (iv3 * H + iv1) * S_v * S_v);
// copy input state into the working buffer and operate in-place
// state layout [S_v, S_v, H, n_seqs]: seq iv3 starts at iv3 * state_seq_stride.
- const float * s_in = state_in_base + iv3 * state_seq_stride + iv1 * S_v * S_v;
- memcpy(s_out, s_in, S_v * S_v * sizeof(float));
-
+ const float * s_in = src_ids
+ ? state_in_base + (int64_t) ((const int32_t *) src_ids->data)[iv3] * (src_state->nb[1] / sizeof(float)) + iv1 * S_v * S_v
+ : state_in_base + iv3 * state_seq_stride + iv1 * S_v * S_v;
+ if (s_out != s_in) {
+ memcpy(s_out, s_in, S_v * S_v * sizeof(float));
+ }
// attn output pointer for first token of this (head, seq)
float * attn_data = attn_out_base + (iv3 * n_tokens * H + iv1) * S_v;
@@ -11203,6 +11235,56 @@ static void ggml_compute_forward_gated_delta_net_one_chunk(
// state is stored transposed: s_out[j*S_v + i] = S[i][j]
// so row j of s_out = column j of S (contiguous access)
+ bool gdn_done = false;
+#if defined(__AVX512F__)
+ if (!kda && S_v == 128) {
+ // Fused 2-pass update for the scalar-gate case (non-KDA), S_v == 128:
+ // delta[j] = (v[j] - g * dot(S_j, k)) * beta
+ // S_j <- g * S_j + delta[j] * k ; out[j] = dot(S_j, q) * scale
+ // Same element-wise fma as the reference path; only the dot rounding order differs.
+ const float g = expf(g_d[0]);
+ __m512 kv[8], qv[8];
+ for (int i = 0; i < 8; ++i) {
+ kv[i] = _mm512_loadu_ps(k_d + 16 * i);
+ qv[i] = _mm512_loadu_ps(q_d + 16 * i);
+ }
+ for (int64_t j = 0; j < 128; j += 4) {
+ float sums[4];
+ for (int r = 0; r < 4; ++r) {
+ const float * row = s_out + (j + r) * 128;
+ __m512 a0 = _mm512_mul_ps(_mm512_loadu_ps(row), kv[0]);
+ __m512 a1 = _mm512_mul_ps(_mm512_loadu_ps(row + 16), kv[1]);
+ a0 = _mm512_fmadd_ps(_mm512_loadu_ps(row + 32), kv[2], a0);
+ a1 = _mm512_fmadd_ps(_mm512_loadu_ps(row + 48), kv[3], a1);
+ a0 = _mm512_fmadd_ps(_mm512_loadu_ps(row + 64), kv[4], a0);
+ a1 = _mm512_fmadd_ps(_mm512_loadu_ps(row + 80), kv[5], a1);
+ a0 = _mm512_fmadd_ps(_mm512_loadu_ps(row + 96), kv[6], a0);
+ a1 = _mm512_fmadd_ps(_mm512_loadu_ps(row + 112), kv[7], a1);
+ sums[r] = _mm512_reduce_add_ps(_mm512_add_ps(a0, a1));
+ }
+ for (int r = 0; r < 4; ++r) {
+ delta[j + r] = (v_d[j + r] - g * sums[r]) * beta_val;
+ }
+ }
+ const __m512 gv = _mm512_set1_ps(g);
+ for (int64_t j = 0; j < 128; ++j) {
+ float * row = s_out + j * 128;
+ const __m512 dv = _mm512_set1_ps(delta[j]);
+ __m512 aq0 = _mm512_setzero_ps(), aq1 = _mm512_setzero_ps();
+ for (int i = 0; i < 8; i += 2) {
+ const __m512 s0 = _mm512_fmadd_ps(dv, kv[i], _mm512_mul_ps(gv, _mm512_loadu_ps(row + 16 * i)));
+ const __m512 s1 = _mm512_fmadd_ps(dv, kv[i + 1], _mm512_mul_ps(gv, _mm512_loadu_ps(row + 16 * (i + 1))));
+ _mm512_storeu_ps(row + 16 * i, s0);
+ _mm512_storeu_ps(row + 16 * (i + 1), s1);
+ aq0 = _mm512_fmadd_ps(s0, qv[i], aq0);
+ aq1 = _mm512_fmadd_ps(s1, qv[i + 1], aq1);
+ }
+ attn_data[j] = _mm512_reduce_add_ps(_mm512_add_ps(aq0, aq1)) * scale;
+ }
+ gdn_done = true;
+ }
+#endif
+ if (!gdn_done) {
if (kda) {
// precompute exp(g) into delta scratch (reused below)
for (int64_t i = 0; i < S_v; ++i) {
@@ -11235,6 +11317,8 @@ static void ggml_compute_forward_gated_delta_net_one_chunk(
attn_data[j] = sum * scale;
}
+ }
+
attn_data += S_v * H; // advance to next token
if (K > 1) {
diff --git a/ggml/src/ggml-cpu/repack.cpp b/ggml/src/ggml-cpu/repack.cpp
index 9282ca299..0431b7ada 100644
--- a/ggml/src/ggml-cpu/repack.cpp
+++ b/ggml/src/ggml-cpu/repack.cpp
@@ -5009,7 +5009,10 @@ static const ggml::cpu::tensor_traits * ggml_repack_get_optimal_repack_type(cons
#endif
}
} else if (cur->type == GGML_TYPE_Q4_K) {
- if (ggml_cpu_has_avx2()) {
+ const char * rp_q4k = getenv("GGML_REPACK_Q4K");
+ // Xeon 8559C: the AVX512-VNNI Q4_K vec_dot beats the AVX2 8x8 repack on decode+prefill; opt in with GGML_REPACK_Q4K=1
+ const bool q4k_repack = (rp_q4k && rp_q4k[0] == '1');
+ if (q4k_repack && ggml_cpu_has_avx2()) {
if (cur->ne[1] % 8 == 0) {
return &q4_K_8x8_q8_K;
}
diff --git a/ggml/src/ggml-cpu/tiled/tiled-kernel.cpp b/ggml/src/ggml-cpu/tiled/tiled-kernel.cpp
index 5819ac07c..b28255b79 100644
--- a/ggml/src/ggml-cpu/tiled/tiled-kernel.cpp
+++ b/ggml/src/ggml-cpu/tiled/tiled-kernel.cpp
@@ -204,6 +204,147 @@ static void tiled_run_microtile_vnni(const tiled_tile_src0 & src0, const tiled_t
tiled_run_micro_vnni_8x16<SUBBLK, HAS_MIN, BIAS, NK>(src0, src1, i0 + 8, j0, num_k, slab, buf, buf_stride);
}
+// q4_1: every 32-wide block has its own f32 scale and min on both sides (weights: d,m; activations: d).
+// weight = d_w*q + m_w (q in 0..15), act = d_a*code (code int8), so one 32-block contributes
+// d_a * (d_w * dot(q, code) + m_w * sum(code)).
+// The integer dot and the integer sum are exact; the four per-block f32 scale factors are applied in f32,
+// which only changes the accumulation order (same as the reference vec_dot, which also accumulates in f32).
+template <int NK>
+static void tiled_run_micro_q41_vnni_8x16(const tiled_tile_src0 & src0, const tiled_tile_src1 & src1,
+ int i0, int j0, int num_k, int slab, float * buf, int buf_stride) {
+ constexpr int SUBBLK = 32;
+ constexpr int NB = TILED_TILE_K / SUBBLK; // 8
+ constexpr int NS = SUBBLK / 16; // 2
+ constexpr int NG = SUBBLK / 4; // 8
+ constexpr int NUM_ROWS = 8;
+
+ const int nkr = (NK > 0) ? NK : num_k;
+ const int seff = (NK > 0) ? 0 : slab;
+ const int qk_stride = nkr * TILED_TILE_K;
+ const int qk_off = seff * TILED_TILE_K;
+ const int nb_stride = NB * nkr;
+ const int nb_off = seff * NB;
+ const int bs_stride = (nkr == 1) ? TILED_TILE_ROWS : TILED_MICRO;
+
+ __m512 facc[NUM_ROWS];
+ for (int t = 0; t < NUM_ROWS; t++) { facc[t] = _mm512_setzero_ps(); }
+
+ const int32_t * bsums = src1.bsums;
+ for (int s = 0; s < NB; s++) {
+ __m512i bsums32 = _mm512_load_si512((const __m512i *) &bsums[(seff * TILED_MICRO + s * NS) * bs_stride + j0]);
+ for (int u = 1; u < NS; u++) {
+ bsums32 = _mm512_add_epi32(bsums32, _mm512_load_si512((const __m512i *) &bsums[(seff * TILED_MICRO + s * NS + u) * bs_stride + j0]));
+ }
+ const __m512 bterm = _mm512_cvtepi32_ps(bsums32);
+
+ __m512i acc16[NUM_ROWS];
+ for (int t = 0; t < NUM_ROWS; t++) { acc16[t] = _mm512_setzero_si512(); }
+
+#ifdef __GNUC__
+ #pragma GCC unroll 8
+#endif
+ for (int g = 0; g < NG; g++) {
+ const int kg = s * NG + g;
+ const __m512i codes = _mm512_load_si512((const __m512i *) &src1.q[(j0 / TILED_MICRO) * (TILED_MICRO * qk_stride)
+ + qk_off + (kg % TILED_MICRO) * qk_stride + (kg / TILED_MICRO) * (TILED_MICRO * 4)]);
+ for (int t = 0; t < NUM_ROWS; t++) {
+ const uint32_t u4 = *(const uint32_t *) &src0.q[(i0 + t) * qk_stride + qk_off + kg * 4];
+ const __m512i u4b = _mm512_set1_epi32((int) u4);
+ acc16[t] = _mm512_dpbusd_epi32(acc16[t], u4b, codes);
+ }
+ }
+
+ const __m512 dav = _mm512_loadu_ps(&src1.sd[(seff * NB + s) * bs_stride + j0]);
+ for (int t = 0; t < NUM_ROWS; t++) {
+ const int ar = i0 + t;
+ const float dw = src0.sd[ar * nb_stride + nb_off + s];
+ const float mw = src0.sm[ar * nb_stride + nb_off + s];
+ const __m512 raw = _mm512_cvtepi32_ps(acc16[t]);
+ const __m512 tv = _mm512_fmadd_ps(_mm512_set1_ps(dw), raw, _mm512_mul_ps(_mm512_set1_ps(mw), bterm));
+ facc[t] = _mm512_fmadd_ps(tv, dav, facc[t]);
+ }
+ }
+
+ for (int t = 0; t < NUM_ROWS; t++) {
+ float * p = &buf[(i0 + t) * buf_stride + j0];
+ _mm512_store_ps(p, _mm512_add_ps(_mm512_load_ps(p), facc[t]));
+ }
+}
+
+template <int NK>
+static void tiled_run_microtile_q41_vnni(const tiled_tile_src0 & src0, const tiled_tile_src1 & src1,
+ int i0, int j0, int num_k, int slab, float * buf, int buf_stride) {
+ tiled_run_micro_q41_vnni_8x16<NK>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
+ tiled_run_micro_q41_vnni_8x16<NK>(src0, src1, i0 + 8, j0, num_k, slab, buf, buf_stride);
+}
+
+// nvfp4: 16 subblocks of 16 per 256-slab, each with its own f32 weight scale (sd16); q8_0 activations carry one
+// f32 scale per 32-block (sd). Codes are LUT + 12 (unsigned), so the bias is removed per subblock with the
+// int32 code sums (bsums): sum(w*a) = sum((w+12)*a) - 12*sum(a). Integer sums are exact.
+template <int NK>
+static void tiled_run_micro_nv4_vnni_8x16(const tiled_tile_src0 & src0, const tiled_tile_src1 & src1,
+ int i0, int j0, int num_k, int slab, float * buf, int buf_stride) {
+ constexpr int NB32 = TILED_TILE_K / 32; // 8 activation blocks
+ constexpr int NB16 = TILED_TILE_K / 16; // 16 weight subblocks
+ constexpr int NG = 4; // 4-element groups per 16-subblock
+ constexpr int NUM_ROWS = 8;
+
+ const int nkr = (NK > 0) ? NK : num_k;
+ const int seff = (NK > 0) ? 0 : slab;
+ const int qk_stride = nkr * TILED_TILE_K;
+ const int qk_off = seff * TILED_TILE_K;
+ const int nb_stride = NB16 * nkr;
+ const int nb_off = seff * NB16;
+ const int bs_stride = (nkr == 1) ? TILED_TILE_ROWS : TILED_MICRO;
+
+ __m512 facc[NUM_ROWS];
+ for (int t = 0; t < NUM_ROWS; t++) { facc[t] = _mm512_setzero_ps(); }
+
+ const int32_t * bsums = src1.bsums;
+ const __m512i twelve = _mm512_set1_epi32(12);
+ for (int sb = 0; sb < NB16; sb++) {
+ const __m512i bsum12 = _mm512_mullo_epi32(twelve,
+ _mm512_load_si512((const __m512i *) &bsums[(seff * TILED_MICRO + sb) * bs_stride + j0]));
+
+ __m512i acc16[NUM_ROWS];
+ for (int t = 0; t < NUM_ROWS; t++) { acc16[t] = _mm512_setzero_si512(); }
+
+#ifdef __GNUC__
+ #pragma GCC unroll 4
+#endif
+ for (int g = 0; g < NG; g++) {
+ const int kg = sb * NG + g;
+ const __m512i codes = _mm512_load_si512((const __m512i *) &src1.q[(j0 / TILED_MICRO) * (TILED_MICRO * qk_stride)
+ + qk_off + (kg % TILED_MICRO) * qk_stride + (kg / TILED_MICRO) * (TILED_MICRO * 4)]);
+ for (int t = 0; t < NUM_ROWS; t++) {
+ const uint32_t u4 = *(const uint32_t *) &src0.q[(i0 + t) * qk_stride + qk_off + kg * 4];
+ const __m512i u4b = _mm512_set1_epi32((int) u4);
+ acc16[t] = _mm512_dpbusd_epi32(acc16[t], u4b, codes);
+ }
+ }
+
+ const __m512 dav = _mm512_loadu_ps(&src1.sd[(seff * NB32 + sb / 2) * bs_stride + j0]);
+ for (int t = 0; t < NUM_ROWS; t++) {
+ const int ar = i0 + t;
+ const float dw = src0.sd16[ar * nb_stride + nb_off + sb];
+ const __m512 raw = _mm512_cvtepi32_ps(_mm512_sub_epi32(acc16[t], bsum12));
+ facc[t] = _mm512_fmadd_ps(raw, _mm512_mul_ps(_mm512_set1_ps(dw), dav), facc[t]);
+ }
+ }
+
+ for (int t = 0; t < NUM_ROWS; t++) {
+ float * p = &buf[(i0 + t) * buf_stride + j0];
+ _mm512_store_ps(p, _mm512_add_ps(_mm512_load_ps(p), facc[t]));
+ }
+}
+
+template <int NK>
+static void tiled_run_microtile_nv4_vnni(const tiled_tile_src0 & src0, const tiled_tile_src1 & src1,
+ int i0, int j0, int num_k, int slab, float * buf, int buf_stride) {
+ tiled_run_micro_nv4_vnni_8x16<NK>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
+ tiled_run_micro_nv4_vnni_8x16<NK>(src0, src1, i0 + 8, j0, num_k, slab, buf, buf_stride);
+}
+
#endif // __AVX512VNNI__ && __AVX512VL__
#if defined(__AVX2__)
@@ -496,23 +637,88 @@ static void tiled_run_microtile_avx(const tiled_tile_src0 & src0, const tiled_ti
#endif // __AVX__ && !__AVX2__
+// q4_1 reference/scalar path (non-VNNI builds): same math as the VNNI kernel, per 32-block f32 scales.
+template <int NK>
+static void tiled_run_microtile_q41_scalar(const tiled_tile_src0 & src0, const tiled_tile_src1 & src1,
+ int i0, int j0, int num_k, int slab, float * buf, int buf_stride) {
+ constexpr int SUBBLK = 32;
+ constexpr int NB = TILED_TILE_K / SUBBLK;
+ constexpr int NS = SUBBLK / 16;
+
+ const int nkr = (NK > 0) ? NK : num_k;
+ const int seff = (NK > 0) ? 0 : slab;
+ const int qk_stride = nkr * TILED_TILE_K;
+ const int qk_off = seff * TILED_TILE_K;
+ const int nb_stride = NB * nkr;
+ const int nb_off = seff * NB;
+ const int bs_stride = (nkr == 1) ? TILED_TILE_ROWS : TILED_MICRO;
+
+ for (int t = 0; t < TILED_MICRO; t++) {
+ const int ar = i0 + t;
+ const uint8_t * q0 = &src0.q[ar * qk_stride + qk_off];
+ for (int j = 0; j < TILED_MICRO; j++) {
+ const int br = j0 + j;
+ const int8_t * q1 = &src1.q[br * qk_stride + qk_off];
+ float acc = 0.0f;
+ for (int s = 0; s < NB; s++) {
+ int32_t bsum = 0;
+ for (int u = 0; u < NS; u++) { bsum += src1.bsums[(seff * TILED_MICRO + s * NS + u) * bs_stride + br]; }
+ int32_t raw = 0;
+ for (int e = 0; e < SUBBLK; e++) { raw += (int32_t) q0[s * SUBBLK + e] * (int32_t) q1[s * SUBBLK + e]; }
+ const float da = src1.sd[(seff * NB + s) * bs_stride + br];
+ const float dw = src0.sd[ar * nb_stride + nb_off + s];
+ const float mw = src0.sm[ar * nb_stride + nb_off + s];
+ acc += da * (dw * (float) raw + mw * (float) bsum);
+ }
+ buf[ar * buf_stride + j0 + j] += acc;
+ }
+ }
+}
+
// main microtile entry point. num_k = K-blocks per row (the narrow path reads num_k slabs at row
// stride num_k*256 so each weight row is one long stream); slab = this call's slab. num_k=1,
// slab=0 is the standard single-slab path.
// num_k==1 dispatches to the NK=1 instantiation, whose strides/offsets are compile-time
// constants (shifts/scaled-leas). num_k>1 (the narrow path) uses the NK=0 runtime-strided
// version; it is memory-bound, so one version serves all of it.
-template <int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS>
+// Q41: q4_1 weights + q8_1 activations (per-32-block f32 scales) use the dedicated kernels above.
+template <int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS, bool Q41>
void tiled_run_microtile(const tiled_tile_src0 & src0, const tiled_tile_src1 & src1,
int i0, int j0, int num_k, int slab, float * buf, int buf_stride) {
const bool standard = (num_k == 1);
#if defined(__AVX512VNNI__) && defined(__AVX512VL__) && defined(__AVX512DQ__)
+ if constexpr (Q41) {
+ if constexpr (SUBBLK == 16) {
+ // nvfp4 (per-16 weight scales), see tiled_run_micro_nv4_vnni_8x16
+ if (standard) {
+ tiled_run_microtile_nv4_vnni<1>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
+ } else {
+ tiled_run_microtile_nv4_vnni<0>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
+ }
+ return;
+ }
+ if (standard) {
+ tiled_run_microtile_q41_vnni<1>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
+ } else {
+ tiled_run_microtile_q41_vnni<0>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
+ }
+ return;
+ }
if (standard) {
tiled_run_microtile_vnni<SUBBLK, HAS_MIN, BIAS, 1>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
} else {
tiled_run_microtile_vnni<SUBBLK, HAS_MIN, BIAS, 0>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
}
-#elif defined(__AVX2__)
+#else
+ if constexpr (Q41) {
+ if (standard) {
+ tiled_run_microtile_q41_scalar<1>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
+ } else {
+ tiled_run_microtile_q41_scalar<0>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
+ }
+ return;
+ }
+#if defined(__AVX2__)
if (standard) {
tiled_run_microtile_avx2<SUBBLK, HAS_MIN, BIAS, 1, ACTBIAS>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
} else {
@@ -531,6 +737,7 @@ void tiled_run_microtile(const tiled_tile_src0 & src0, const tiled_tile_src1 & s
tiled_run_microtile_scalar<SUBBLK, HAS_MIN, BIAS, 0>(src0, src1, i0, j0, num_k, slab, buf, buf_stride);
}
#endif
+#endif
}
// explicit instantiations for the in-use formats (q4_K and q5_K share the constants)
@@ -556,6 +763,12 @@ template void tiled_run_microtile<16, false, 4, true>(const tiled_tile_src0 & sr
// q2_K: BIAS = 0
template void tiled_run_microtile<16, true, 0, false>(const tiled_tile_src0 & src0, const tiled_tile_src1 & src1,
int i0, int j0, int num_k, int slab, float * buf, int buf_stride);
+// q4_1: per-32-block f32 scales, q8_1 activations
+template void tiled_run_microtile<32, false, 0, false, true>(const tiled_tile_src0 & src0, const tiled_tile_src1 & src1,
+ int i0, int j0, int num_k, int slab, float * buf, int buf_stride);
+// nvfp4: per-16-subblock f32 scales, q8_0 activations
+template void tiled_run_microtile<16, false, 0, false, true>(const tiled_tile_src0 & src0, const tiled_tile_src1 & src1,
+ int i0, int j0, int num_k, int slab, float * buf, int buf_stride);
#define MIN(a, b) ((a) < (b) ? (a) : (b))
diff --git a/ggml/src/ggml-cpu/tiled/tiled-kernel.h b/ggml/src/ggml-cpu/tiled/tiled-kernel.h
index 1a97f406a..c5d081420 100644
--- a/ggml/src/ggml-cpu/tiled/tiled-kernel.h
+++ b/ggml/src/ggml-cpu/tiled/tiled-kernel.h
@@ -29,12 +29,18 @@
// SUBBLK=32 formats index at stride 8 and leave the slack unused.
struct tiled_tile_src0 {
static constexpr int NB_MAX = TILED_TILE_K / 16; // max subblocks per 256-elem block
+ static constexpr int NB32 = TILED_TILE_K / 32; // 32-wide subblocks (q4_1)
alignas(64) uint8_t q[TILED_TILE_ROWS * TILED_TILE_K]; // unsigned quants, widened to uint8
float d[TILED_TILE_ROWS]; // One d from each input block, widened to f32
float dmin[TILED_TILE_ROWS]; // dmin from each input block (if applicable), widened to F32
int32_t scales[TILED_TILE_ROWS * NB_MAX]; // per-subblock scale, stored as int32_t
int32_t mins[TILED_TILE_ROWS * NB_MAX]; // per-subblock min, used when HAS_MIN
+ // q4_1: no super-block scale, each 32-wide block carries its own f32 scale/min
+ float sd[TILED_TILE_ROWS * NB32]; // per-32-block scale
+ float sm[TILED_TILE_ROWS * NB32]; // per-32-block min
+ // nvfp4: one f32 scale per 16-wide subblock (UE4M3 decoded), codes are LUT + 12 (0..24)
+ float sd16[TILED_TILE_ROWS * NB_MAX];
};
// src1 tile: built from q8_K (wdata)
@@ -45,6 +51,8 @@ struct tiled_tile_src1 {
alignas(64) int32_t bsums[(TILED_TILE_K / 16) * TILED_TILE_ROWS];
// f32 (not f16): q8_k stores fp16, the unpack converts once
float d[TILED_TILE_ROWS];
+ // q4_1: q8_1 activations carry a scale per 32-wide block (no super-block d)
+ float sd[(TILED_TILE_K / 32) * TILED_TILE_ROWS];
};
// per-thread workspace: all tiled state lives here, allocated in wdata (one slot per thread)
@@ -71,6 +79,26 @@ inline void tiled_unpk_nib4(const uint8_t * src, uint8_t * lo, uint8_t * hi) {
_mm256_storeu_si256((__m256i *) lo, _mm256_and_si256(v, _mm256_set1_epi8(0x0F)));
_mm256_storeu_si256((__m256i *) hi, _mm256_srli_epi32(_mm256_and_si256(v, _mm256_set1_epi8((int8_t) 0xF0)), 4));
}
+// nvfp4: 32 packed bytes (one 64-element block) -> 64 codes, LUT(kvalues_fp4) + 12 in 0..24.
+// byte B = 8s+j: lo -> elem 16s+j, hi -> elem 16s+8+j
+inline void tiled_unpk_fp4_64(const uint8_t * src, uint8_t * dst) {
+ const __m256i lut = _mm256_setr_epi8(12, 13, 14, 15, 16, 18, 20, 24, 12, 11, 10, 9, 8, 6, 4, 0,
+ 12, 13, 14, 15, 16, 18, 20, 24, 12, 11, 10, 9, 8, 6, 4, 0);
+ const __m256i v = _mm256_loadu_si256((const __m256i *) src);
+ const __m256i lo = _mm256_shuffle_epi8(lut, _mm256_and_si256(v, _mm256_set1_epi8(0x0F)));
+ const __m256i hi = _mm256_shuffle_epi8(lut, _mm256_and_si256(_mm256_srli_epi16(v, 4), _mm256_set1_epi8(0x0F)));
+ // x = [s0 | s2], y = [s1 | s3] (128-bit halves)
+ const __m256i x = _mm256_unpacklo_epi64(lo, hi);
+ const __m256i y = _mm256_unpackhi_epi64(lo, hi);
+ _mm256_storeu_si256((__m256i *) dst, _mm256_permute2x128_si256(x, y, 0x20));
+ _mm256_storeu_si256((__m256i *) (dst + 32), _mm256_permute2x128_si256(x, y, 0x31));
+}
+// 16-byte source (one q4_1 block: lo -> elems 0..15, hi -> elems 16..31)
+inline void tiled_unpk_nib4_16(const uint8_t * src, uint8_t * lo, uint8_t * hi) {
+ const __m128i v = _mm_loadu_si128((const __m128i *) src);
+ _mm_storeu_si128((__m128i *) lo, _mm_and_si128(v, _mm_set1_epi8(0x0F)));
+ _mm_storeu_si128((__m128i *) hi, _mm_srli_epi32(_mm_and_si128(v, _mm_set1_epi8((int8_t) 0xF0)), 4));
+}
// 2-bit values at bit offset S
template <int S> inline void tiled_unpk_2bit(const uint8_t * src, uint8_t * dst) {
_mm256_storeu_si256((__m256i *) dst, _mm256_and_si256(
@@ -129,6 +157,17 @@ inline void tiled_unpk_tern8(const uint8_t * src, int8_t delta, uint8_t * dst) {
inline void tiled_unpk_nib4(const uint8_t * src, uint8_t * lo, uint8_t * hi) {
for (int l = 0; l < 32; l++) { lo[l] = (uint8_t) (src[l] & 0xF); hi[l] = (uint8_t) (src[l] >> 4); }
}
+inline void tiled_unpk_fp4_64(const uint8_t * src, uint8_t * dst) {
+ static const uint8_t lut[16] = { 12, 13, 14, 15, 16, 18, 20, 24, 12, 11, 10, 9, 8, 6, 4, 0 };
+ for (int B = 0; B < 32; B++) {
+ const int s = B / 8, j = B % 8;
+ dst[16 * s + j] = lut[src[B] & 0xF];
+ dst[16 * s + 8 + j] = lut[src[B] >> 4];
+ }
+}
+inline void tiled_unpk_nib4_16(const uint8_t * src, uint8_t * lo, uint8_t * hi) {
+ for (int l = 0; l < 16; l++) { lo[l] = (uint8_t) (src[l] & 0xF); hi[l] = (uint8_t) (src[l] >> 4); }
+}
template <int S>
inline void tiled_unpk_2bit(const uint8_t * src, uint8_t * dst) {
for (int l = 0; l < 32; l++) { dst[l] = (uint8_t) ((src[l] >> S) & 3); }
@@ -164,7 +203,9 @@ inline void tiled_unpk_tern8(const uint8_t * src, int8_t delta, uint8_t * dst) {
// ACTBIAS (AVX2 only): the activation is pre-biased +128 by tiled_repack_src1,
// num_k = K-blocks per row: the tile holds num_k slabs at row stride num_k*256
// (Default case is num_k=1, 256x256 tiles, we go to longer num_k to improve memory bandwidth when num_rows is small)
-template <int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS>
+// Q41: src0 is q4_1 (per-32-block f32 scale/min in sd/sm) and src1 is q8_1 (per-32-block f32 scale in sd);
+// the other format constants (SUBBLK=32, HAS_MIN=false, BIAS=0, ACTBIAS=false) are unused in that mode.
+template <int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS, bool Q41 = false>
void tiled_run_microtile(const tiled_tile_src0 & src0, const tiled_tile_src1 & src1,
int i0, int j0, int num_k, int slab, float * buf, int buf_stride);
diff --git a/ggml/src/ggml-cpu/tiled/tiled.cpp b/ggml/src/ggml-cpu/tiled/tiled.cpp
index 77e1a9945..96640b50d 100644
--- a/ggml/src/ggml-cpu/tiled/tiled.cpp
+++ b/ggml/src/ggml-cpu/tiled/tiled.cpp
@@ -534,11 +534,57 @@ static void tiled_unpack_src0(const block_iq1_m * rows, int64_t row_stride, int
}
}
+// q4_1: 8 x 32-wide blocks per 256-slab, each with its own f32 d (scale) and m (min).
+static void tiled_unpack_src0(const block_q4_1 * rows, int64_t row_stride, int n_rows, tiled_tile_src0 * tile, int num_k) {
+ GGML_ASSERT(n_rows <= TILED_TILE_ROWS);
+ constexpr int NB = TILED_TILE_K / 32; // 8 x 32-wide blocks
+
+ const int qk_stride = num_k * TILED_TILE_K;
+ const int nb_stride = NB * num_k;
+ for (int slab = 0; slab < num_k; slab++) {
+ for (int r = 0; r < n_rows; r++) {
+ const int q_off = r * qk_stride + slab * TILED_TILE_K;
+ const int s_off = r * nb_stride + slab * NB;
+ const block_q4_1 * x = rows + r * row_stride + slab * NB;
+ for (int s = 0; s < NB; s++) {
+ tile->sd[s_off + s] = ggml_fp16_to_fp32(x[s].data.data.d);
+ tile->sm[s_off + s] = ggml_fp16_to_fp32(x[s].data.data.m);
+ // 16 bytes -> elems 0..15 (low nibble) and 16..31 (high nibble)
+ tiled_unpk_nib4_16(x[s].qs, &tile->q[q_off + s * 32], &tile->q[q_off + s * 32 + 16]);
+ }
+ }
+ }
+}
+
+// nvfp4: 4 x 64-element blocks per 256-slab, each block 4 UE4M3 subblock scales (16 elems each) + 32 nibble bytes.
+// Scales go to sd16 (one per 16-wide subblock); codes are LUT + 12 (see tiled_unpk_fp4_64).
+static void tiled_unpack_src0(const block_nvfp4 * rows, int64_t row_stride, int n_rows, tiled_tile_src0 * tile, int num_k) {
+ GGML_ASSERT(n_rows <= TILED_TILE_ROWS);
+ constexpr int NBLK = TILED_TILE_K / 64; // 4 nvfp4 blocks per 256-slab
+ constexpr int NB16 = TILED_TILE_K / 16; // 16 subblocks per 256-slab
+
+ const int qk_stride = num_k * TILED_TILE_K;
+ const int nb_stride = NB16 * num_k;
+ for (int slab = 0; slab < num_k; slab++) {
+ for (int r = 0; r < n_rows; r++) {
+ const int q_off = r * qk_stride + slab * TILED_TILE_K;
+ const int s_off = r * nb_stride + slab * NB16;
+ const block_nvfp4 * x = rows + r * row_stride + slab * NBLK;
+ for (int b = 0; b < NBLK; b++) {
+ for (int s = 0; s < 4; s++) {
+ tile->sd16[s_off + 4 * b + s] = ggml_ue4m3_to_fp32(x[b].d[s]);
+ }
+ tiled_unpk_fp4_64(x[b].qs, &tile->q[q_off + b * 64]);
+ }
+ }
+ }
+}
+
// unpack src1 tile from q8_K rows. num_k = K-blocks per row: each row decodes num_k consecutive
// source blocks (rows[r][kblk + slab]) into the tile at row stride num_k*256, so each activation
// row is one long stream. num_k=1 is the standard single-slab path (byte-identical to before).
-static void tiled_unpack_src1_q8_K(const block_q8_K * const * rows, int n_rows, tiled_tile_src1 * tile,
- int kblk, int num_k) {
+static void tiled_unpack_src1(const block_q8_K * const * rows, int n_rows, tiled_tile_src1 * tile,
+ int kblk, int num_k) {
GGML_ASSERT(n_rows <= TILED_TILE_ROWS);
const int n_padded = (n_rows + TILED_MICRO - 1) & ~(TILED_MICRO - 1);
const int bs_stride = (num_k == 1) ? TILED_TILE_ROWS : TILED_MICRO;
@@ -569,6 +615,71 @@ static void tiled_unpack_src1_q8_K(const block_q8_K * const * rows, int n_rows,
}
}
+// q4_1 src1: q8_1 activation rows. 8 blocks of 32 per 256-slab, each block has its own f32 scale;
+// the integer per-16 sums are recomputed from the codes (q8_1 only stores d and d*sum).
+static void tiled_unpack_src1(const block_q8_1 * const * rows, int n_rows, tiled_tile_src1 * tile,
+ int kblk, int num_k) {
+ GGML_ASSERT(n_rows <= TILED_TILE_ROWS);
+ const int n_padded = (n_rows + TILED_MICRO - 1) & ~(TILED_MICRO - 1);
+ const int bs_stride = (num_k == 1) ? TILED_TILE_ROWS : TILED_MICRO;
+ constexpr int NBLK = TILED_TILE_K / 32; // 8 q8_1 blocks per 256-slab
+ constexpr int n16 = TILED_TILE_K / TILED_MICRO;
+ for (int slab = 0; slab < num_k; slab++) {
+ const int q_off = slab * TILED_TILE_K;
+ for (int r = 0; r < n_padded; r++) {
+ if (r < n_rows) {
+ const block_q8_1 * xb = rows[r] + (kblk + slab) * NBLK;
+ for (int b = 0; b < NBLK; b++) {
+ memcpy(&tile->q[r * (num_k * TILED_TILE_K) + q_off + b * 32], xb[b].qs, 32);
+ int32_t s0 = 0, s1 = 0;
+ for (int e = 0; e < 16; e++) { s0 += xb[b].qs[e]; s1 += xb[b].qs[16 + e]; }
+ tile->bsums[(slab * n16 + 2 * b + 0) * bs_stride + r] = s0;
+ tile->bsums[(slab * n16 + 2 * b + 1) * bs_stride + r] = s1;
+ tile->sd[(slab * NBLK + b) * bs_stride + r] = ggml_fp16_to_fp32(xb[b].data.data.d);
+ }
+ tile->d[slab * TILED_MICRO + r] = 1.0f; // per-32 scales are applied in the kernel
+ } else {
+ memset(&tile->q[r * (num_k * TILED_TILE_K) + q_off], 0, TILED_TILE_K);
+ for (int i = 0; i < n16; i++) { tile->bsums[(slab * n16 + i) * bs_stride + r] = 0; }
+ for (int b = 0; b < NBLK; b++) { tile->sd[(slab * NBLK + b) * bs_stride + r] = 0.0f; }
+ tile->d[slab * TILED_MICRO + r] = 1.0f;
+ }
+ }
+ }
+}
+
+// nvfp4 src1: q8_0 activation rows (one f32 scale per 32-wide block, per-16 sums recomputed from the codes).
+static void tiled_unpack_src1(const block_q8_0 * const * rows, int n_rows, tiled_tile_src1 * tile,
+ int kblk, int num_k) {
+ GGML_ASSERT(n_rows <= TILED_TILE_ROWS);
+ const int n_padded = (n_rows + TILED_MICRO - 1) & ~(TILED_MICRO - 1);
+ const int bs_stride = (num_k == 1) ? TILED_TILE_ROWS : TILED_MICRO;
+ constexpr int NBLK = TILED_TILE_K / 32; // 8 q8_0 blocks per 256-slab
+ constexpr int n16 = TILED_TILE_K / TILED_MICRO;
+ for (int slab = 0; slab < num_k; slab++) {
+ const int q_off = slab * TILED_TILE_K;
+ for (int r = 0; r < n_padded; r++) {
+ if (r < n_rows) {
+ const block_q8_0 * xb = rows[r] + (kblk + slab) * NBLK;
+ for (int b = 0; b < NBLK; b++) {
+ memcpy(&tile->q[r * (num_k * TILED_TILE_K) + q_off + b * 32], xb[b].qs, 32);
+ int32_t s0 = 0, s1 = 0;
+ for (int e = 0; e < 16; e++) { s0 += xb[b].qs[e]; s1 += xb[b].qs[16 + e]; }
+ tile->bsums[(slab * n16 + 2 * b + 0) * bs_stride + r] = s0;
+ tile->bsums[(slab * n16 + 2 * b + 1) * bs_stride + r] = s1;
+ tile->sd[(slab * NBLK + b) * bs_stride + r] = ggml_fp16_to_fp32(xb[b].d);
+ }
+ tile->d[slab * TILED_MICRO + r] = 1.0f; // per-32 scales are applied in the kernel
+ } else {
+ memset(&tile->q[r * (num_k * TILED_TILE_K) + q_off], 0, TILED_TILE_K);
+ for (int i = 0; i < n16; i++) { tile->bsums[(slab * n16 + i) * bs_stride + r] = 0; }
+ for (int b = 0; b < NBLK; b++) { tile->sd[(slab * NBLK + b) * bs_stride + r] = 0.0f; }
+ tile->d[slab * TILED_MICRO + r] = 1.0f;
+ }
+ }
+ }
+}
+
// GGML_CPU_TILED_MM: master switch, on by default. If off, we fast return false and normal vec_dot mul_mat resumes
static bool ggml_tiled_matmul_enabled(void) {
static bool enabled = true;
@@ -617,6 +728,8 @@ static bool ggml_tiled_supported(const struct ggml_tensor * src0,
case GGML_TYPE_Q6_K:
case GGML_TYPE_Q5_K:
case GGML_TYPE_Q4_K:
+ case GGML_TYPE_Q4_1:
+ case GGML_TYPE_NVFP4:
case GGML_TYPE_Q3_K:
case GGML_TYPE_Q2_K:
case GGML_TYPE_IQ4_XS:
@@ -716,7 +829,9 @@ static void tiled_store_window(const float * buf, int n_src0, int n_src1, int bu
// unpack and microtile are reused unchanged from mul_mat.
// src0 rows per g group; finer than the TILED_TILE_K kernel tile so all threads stay busy at small ne01
-#define TILED_MMID_GROUP 64
+#ifndef TILED_MMID_GROUP
+#define TILED_MMID_GROUP 32
+#endif
// like tiled_store_window, but the dst columns are not contiguous: col_ptrs[j] points at the
// start of dst column j, whose rows are contiguous (dim 0, nb0 == 4 bytes).
@@ -756,11 +871,11 @@ static void tiled_store_window_scatter(const float * buf, int n_src0, int n_src1
// one (g, k) macrotile: zero the acc window, sweep K in 256 element slabs, scatter the result rows
// rows points at this k window's routed src1 rows, row r at its base block
-template <typename B, int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS>
+template <typename B, int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS, typename BA = block_q8_K, bool Q41 = false>
static void tiled_mmid_gemm_window(struct ggml_tensor * dst, const struct ggml_tensor * src0,
const char * src0_cur, int64_t r, int64_t k, int64_t nrows,
const int32_t * expert_rows,
- const block_q8_K * const * rows,
+ const BA * const * rows,
tiled_ws * ws) {
const int64_t ne00 = src0->ne[0];
const int64_t ne01 = src0->ne[1];
@@ -770,6 +885,8 @@ static void tiled_mmid_gemm_window(struct ggml_tensor * dst, const struct ggml_t
const size_t src0_bs = ggml_type_size(src0->type);
const int64_t src0_stride = src0->nb[1] / src0_bs;
+ // source blocks per 256-element slab (1 for K-quants, 8 for q4_1's 32-wide blocks)
+ const int kblk_scale = (int) (TILED_TILE_K / ggml_blck_size(src0->type));
// the window is at most TILED_MMID_GROUP x TILED_TILE_K, so zero only that region of acc
for (int64_t i = 0; i < n_src0; i++) {
@@ -792,19 +909,19 @@ static void tiled_mmid_gemm_window(struct ggml_tensor * dst, const struct ggml_t
const int kstart = (int) (k0 / TILED_TILE_K);
// activation: one 16-row band (nrows <= 16). Unpack the whole k_extent chunk (all
// slabs) so each activation row is a long stream, then repack each slab in place
- tiled_unpack_src1_q8_K(rows, nrows, &ws->src1, kstart, num_k);
+ tiled_unpack_src1(rows, nrows, &ws->src1, kstart, num_k);
// repack the whole k_extent chunk in place: num_k slabs x 4 tiles, 16 rows, row stride num_k*256
tiled_repack_src1(&ws->src1, 0, num_k, ACTBIAS);
// weight groups (16 at a time): unpack the k_extent chunk of the group (the long per-row
// read), then one standard MAC per slab accumulating into the same acc rows
for (int64_t ir0 = r; ir0 < r_end; ir0 += TILED_MICRO) {
const int n0 = (int) MIN(TILED_MICRO, r_end - ir0);
- const B * wbase = (const B *) (src0_cur + ir0 * src0->nb[1] + kstart * src0_bs);
+ const B * wbase = (const B *) (src0_cur + ir0 * src0->nb[1] + kstart * kblk_scale * src0_bs);
tiled_unpack_src0(wbase, src0_stride, n0, &ws->src0, num_k);
tiled_repack_src0<SUBBLK>(&ws->src0, n0, num_k, BIAS, ACTBIAS);
float * buf = ws->acc + (ir0 - r) * TILED_TILE_ROWS;
for (int slab = 0; slab < num_k; slab++) {
- tiled_run_microtile<SUBBLK, HAS_MIN, BIAS, ACTBIAS>(ws->src0, ws->src1, 0, 0, num_k, slab, buf, TILED_TILE_ROWS);
+ tiled_run_microtile<SUBBLK, HAS_MIN, BIAS, ACTBIAS, Q41>(ws->src0, ws->src1, 0, 0, num_k, slab, buf, TILED_TILE_ROWS);
}
}
}
@@ -812,15 +929,15 @@ static void tiled_mmid_gemm_window(struct ggml_tensor * dst, const struct ggml_t
// standard per-slab path: K is stepped in 256-element slabs
for (int64_t ib = 0; ib < ne00; ib += TILED_TILE_K) {
const int kblk = (int) (ib / TILED_TILE_K);
- tiled_unpack_src0((const B *) (src0_cur + r * src0->nb[1] + kblk * src0_bs), src0_stride, n_src0, &ws->src0, 1);
+ tiled_unpack_src0((const B *) (src0_cur + r * src0->nb[1] + kblk * kblk_scale * src0_bs), src0_stride, n_src0, &ws->src0, 1);
tiled_repack_src0<SUBBLK>(&ws->src0, n_src0, 1, BIAS, ACTBIAS);
- tiled_unpack_src1_q8_K(rows, nrows, &ws->src1, kblk, 1);
+ tiled_unpack_src1(rows, nrows, &ws->src1, kblk, 1);
// 16x16 microtiles sweeping the window; repack each src1 band just-in-time
// (j0-outer) so only the bands actually used are repacked and each is L1-hot
for (int64_t ir1 = 0; ir1 < nrows; ir1 += TILED_MICRO) {
tiled_repack_src1(&ws->src1, (int) ir1, 1, ACTBIAS);
for (int64_t ir0 = r; ir0 < r_end; ir0 += TILED_MICRO) {
- tiled_run_microtile<SUBBLK, HAS_MIN, BIAS, ACTBIAS>(ws->src0, ws->src1,
+ tiled_run_microtile<SUBBLK, HAS_MIN, BIAS, ACTBIAS, Q41>(ws->src0, ws->src1,
(int) (ir0 - r), (int) ir1, 1, 0,
ws->acc, TILED_TILE_ROWS);
}
@@ -834,7 +951,7 @@ static void tiled_mmid_gemm_window(struct ggml_tensor * dst, const struct ggml_t
// one expert of MUL_MAT_ID: each k window's dispatched rows are pointed at by a per-window
// row pointer list and swept over the thread's row windows, the dst rows are scattered back
// per window
-template <typename B, int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS>
+template <typename B, int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS, typename BA = block_q8_K, bool Q41 = false>
static void ggml_compute_forward_mul_mat_id_tiled_one_expert(
const struct ggml_compute_params * params,
struct ggml_tensor * dst,
@@ -877,22 +994,22 @@ static void ggml_compute_forward_mul_mat_id_tiled_one_expert(
for (int64_t k = 0; k < cne1; k += TILED_TILE_K) {
const int64_t nrows = MIN(TILED_TILE_K, cne1 - k);
- const block_q8_K * rows[TILED_TILE_ROWS];
+ const BA * rows[TILED_TILE_ROWS];
for (int64_t i = 0; i < nrows; i++) {
const int64_t i11 = expert_rows[2 * (k + i) + 0] % ne11;
const int64_t i12 = expert_rows[2 * (k + i) + 1];
- rows[i] = (const block_q8_K *) ((const char *) wdata + (i11 + i12 * ne11) * nbw1);
+ rows[i] = (const BA *) ((const char *) wdata + (i11 + i12 * ne11) * nbw1);
}
for (int64_t g = g0; g < g1; g++) {
const int64_t r = g * TILED_MMID_GROUP;
- tiled_mmid_gemm_window<B, SUBBLK, HAS_MIN, BIAS, ACTBIAS>(dst, src0, src0_cur, r, k, nrows, expert_rows, rows, ws);
+ tiled_mmid_gemm_window<B, SUBBLK, HAS_MIN, BIAS, ACTBIAS, BA, Q41>(dst, src0, src0_cur, r, k, nrows, expert_rows, rows, ws);
}
}
}
-template <typename B, int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS>
+template <typename B, int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS, typename BA = block_q8_K, bool Q41 = false>
static void ggml_compute_forward_mul_mat_tiled_one_chunk(
const struct ggml_compute_params * params,
struct ggml_tensor * dst,
@@ -928,6 +1045,8 @@ static void ggml_compute_forward_mul_mat_tiled_one_chunk(
const int64_t src0_stride = nb01 / src0_bs; // blocks between src0 rows
const int64_t src1_stride = (src1->type == vec_dot_type ? src1->nb[1] : row_size) / src1_bs;
+ // source blocks per 256-element slab (1 for K-quants, 8 for q4_1's 32-wide blocks)
+ const int kblk_scale = (int) (TILED_TILE_K / ggml_blck_size(src0->type));
const int64_t TILE = 256;
const int64_t MICRO = 16;
@@ -960,9 +1079,9 @@ static void ggml_compute_forward_mul_mat_tiled_one_chunk(
char * dst_col = (char *) dst->data + i12 * nb2 + i13 * nb3;
// rows[r] is the window's row r base block; the k-slab offset is applied in the unpack
- const block_q8_K * rows[TILED_TILE_ROWS];
+ const BA * rows[TILED_TILE_ROWS];
for (int r = 0; r < n_src1; r++) {
- rows[r] = (const block_q8_K *) ((const char *) wdata + (iir1 + r) * src1_stride * src1_bs);
+ rows[r] = (const BA *) ((const char *) wdata + (iir1 + r) * src1_stride * src1_bs);
}
for (int64_t iir0 = ir0_start; iir0 < ir0_end; iir0 += TILE) {
@@ -981,19 +1100,19 @@ static void ggml_compute_forward_mul_mat_tiled_one_chunk(
const int kstart = (int) (k0 / TILED_TILE_K);
// activation: one 16-row band (n_src1 <= 16). Unpack the whole k_extent chunk
// (all slabs) so each activation row is a long stream, then repack each slab in place
- tiled_unpack_src1_q8_K(rows, n_src1, &ws->src1, kstart, num_k);
+ tiled_unpack_src1(rows, n_src1, &ws->src1, kstart, num_k);
// repack the whole k_extent chunk in place, if kernel wants to: num_k slabs x 4 tiles, 16 rows, row stride num_k*256
tiled_repack_src1(&ws->src1, 0, num_k, ACTBIAS);
// weight groups: unpack the k_extent chunk of each 16-row group (the long per-row
// read), then one standard MAC per slab accumulating into the same acc rows
for (int64_t ir0 = iir0; ir0 < iir0_end; ir0 += MICRO) {
const int n0 = (int) MIN(MICRO, iir0_end - ir0);
- const B * wbase = (const B *) (src0_row + ir0 * nb01 + kstart * src0_bs);
+ const B * wbase = (const B *) (src0_row + ir0 * nb01 + kstart * kblk_scale * src0_bs);
tiled_unpack_src0(wbase, src0_stride, n0, &ws->src0, num_k);
tiled_repack_src0<SUBBLK>(&ws->src0, n0, num_k, BIAS, ACTBIAS);
float * buf = ws->acc + (ir0 - iir0) * TILED_TILE_ROWS;
for (int slab = 0; slab < num_k; slab++) {
- tiled_run_microtile<SUBBLK, HAS_MIN, BIAS, ACTBIAS>(ws->src0, ws->src1, 0, 0, num_k, slab, buf, TILED_TILE_ROWS);
+ tiled_run_microtile<SUBBLK, HAS_MIN, BIAS, ACTBIAS, Q41>(ws->src0, ws->src1, 0, 0, num_k, slab, buf, TILED_TILE_ROWS);
}
}
}
@@ -1001,15 +1120,15 @@ static void ggml_compute_forward_mul_mat_tiled_one_chunk(
// standard per-slab path: K is stepped in 256-element chunks
for (int64_t ib = 0; ib < ne00; ib += TILE) {
const int kblk = (int) (ib / TILE);
- tiled_unpack_src0((const B *) (src0_row + iir0 * nb01 + kblk * src0_bs), src0_stride, n_src0, &ws->src0, 1);
+ tiled_unpack_src0((const B *) (src0_row + iir0 * nb01 + kblk * kblk_scale * src0_bs), src0_stride, n_src0, &ws->src0, 1);
tiled_repack_src0<SUBBLK>(&ws->src0, n_src0, 1, BIAS, ACTBIAS);
- tiled_unpack_src1_q8_K(rows, n_src1, &ws->src1, kblk, 1);
+ tiled_unpack_src1(rows, n_src1, &ws->src1, kblk, 1);
// 16x16 microtiles sweeping the window; repack each src1 band before first use
for (int64_t ir1 = iir1; ir1 < iir1_end; ir1 += MICRO) {
tiled_repack_src1(&ws->src1, (int) (ir1 - iir1), 1, ACTBIAS);
for (int64_t ir0 = iir0; ir0 < iir0_end; ir0 += MICRO) {
- tiled_run_microtile<SUBBLK, HAS_MIN, BIAS, ACTBIAS>(ws->src0, ws->src1,
+ tiled_run_microtile<SUBBLK, HAS_MIN, BIAS, ACTBIAS, Q41>(ws->src0, ws->src1,
(int) (ir0 - iir0), (int) (ir1 - iir1), 1, 0,
ws->acc, TILED_TILE_ROWS);
}
@@ -1024,7 +1143,7 @@ static void ggml_compute_forward_mul_mat_tiled_one_chunk(
}
}
-template <typename B, int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS>
+template <typename B, int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS, typename BA = block_q8_K, bool Q41 = false>
static void ggml_compute_forward_mul_mat_tiled_driver(
const struct ggml_compute_params * params,
struct ggml_tensor * dst) {
@@ -1133,7 +1252,7 @@ static void ggml_compute_forward_mul_mat_tiled_driver(
const int64_t ir1_start = dr1 * ith1;
const int64_t ir1_end = MIN(ir1_start + dr1, nr1);
- ggml_compute_forward_mul_mat_tiled_one_chunk<B, SUBBLK, HAS_MIN, BIAS, ACTBIAS>(params, dst, ir0_start, ir0_end, ir1_start, ir1_end, ws);
+ ggml_compute_forward_mul_mat_tiled_one_chunk<B, SUBBLK, HAS_MIN, BIAS, ACTBIAS, BA, Q41>(params, dst, ir0_start, ir0_end, ir1_start, ir1_end, ws);
if (nth >= nchunk0 * nchunk1) {
break;
@@ -1145,7 +1264,7 @@ static void ggml_compute_forward_mul_mat_tiled_driver(
// src0 type dispatch, shared by the MUL_MAT and MUL_MAT_ID entries: one expert for
// MUL_MAT_ID (expert_rows != NULL), the full op for MUL_MAT
-template <typename B, int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS>
+template <typename B, int SUBBLK, bool HAS_MIN, int BIAS, bool ACTBIAS, typename BA = block_q8_K, bool Q41 = false>
static bool tiled_matmul_dispatch(const struct ggml_compute_params * params,
struct ggml_tensor * dst,
const int32_t * expert_rows,
@@ -1153,9 +1272,9 @@ static bool tiled_matmul_dispatch(const struct ggml_compute_params * params,
int64_t cne1,
char * scratch) {
if (expert_rows == NULL) {
- ggml_compute_forward_mul_mat_tiled_driver<B, SUBBLK, HAS_MIN, BIAS, ACTBIAS>(params, dst);
+ ggml_compute_forward_mul_mat_tiled_driver<B, SUBBLK, HAS_MIN, BIAS, ACTBIAS, BA, Q41>(params, dst);
} else {
- ggml_compute_forward_mul_mat_id_tiled_one_expert<B, SUBBLK, HAS_MIN, BIAS, ACTBIAS>(params, dst, cur_a, cne1, expert_rows, scratch);
+ ggml_compute_forward_mul_mat_id_tiled_one_expert<B, SUBBLK, HAS_MIN, BIAS, ACTBIAS, BA, Q41>(params, dst, cur_a, cne1, expert_rows, scratch);
}
return true;
}
@@ -1174,6 +1293,12 @@ static bool ggml_tiled_matmul_type_dispatch(const struct ggml_compute_params * p
return tiled_matmul_dispatch<block_q5_K, 32, true, 0, false>(params, dst, expert_rows, cur_a, cne1, scratch);
case GGML_TYPE_Q4_K:
return tiled_matmul_dispatch<block_q4_K, 32, true, 0, false>(params, dst, expert_rows, cur_a, cne1, scratch);
+ case GGML_TYPE_Q4_1:
+ return tiled_matmul_dispatch<block_q4_1, 32, false, 0, false, block_q8_1, true>(params, dst, expert_rows, cur_a, cne1, scratch);
+#if defined(__AVX512VNNI__) && defined(__AVX512VL__) && defined(__AVX512DQ__)
+ case GGML_TYPE_NVFP4:
+ return tiled_matmul_dispatch<block_nvfp4, 16, false, 0, false, block_q8_0, true>(params, dst, expert_rows, cur_a, cne1, scratch);
+#endif
case GGML_TYPE_Q3_K:
return tiled_matmul_dispatch<block_q3_K, 16, false, 4, true>(params, dst, expert_rows, cur_a, cne1, scratch);
case GGML_TYPE_Q2_K:
diff --git a/ggml/src/ggml.c b/ggml/src/ggml.c
index 5d8449a88..203f15e43 100644
--- a/ggml/src/ggml.c
+++ b/ggml/src/ggml.c
@@ -6416,6 +6416,59 @@ struct ggml_tensor * ggml_solve_tri(
// ggml_gated_delta_net
+struct ggml_tensor * ggml_gated_delta_net_cache(
+ struct ggml_context * ctx,
+ struct ggml_tensor * q,
+ struct ggml_tensor * k,
+ struct ggml_tensor * v,
+ struct ggml_tensor * g,
+ struct ggml_tensor * beta,
+ struct ggml_tensor * state_cache,
+ struct ggml_tensor * ids,
+ struct ggml_tensor * state_dst) {
+ GGML_ASSERT(ggml_is_contiguous_rows(q));
+ GGML_ASSERT(ggml_is_contiguous_rows(k));
+ GGML_ASSERT(ggml_is_contiguous_rows(v));
+ GGML_ASSERT(ggml_is_contiguous(g));
+ GGML_ASSERT(ggml_is_contiguous(beta));
+ GGML_ASSERT(q->type == GGML_TYPE_F32 && k->type == GGML_TYPE_F32 && v->type == GGML_TYPE_F32);
+ GGML_ASSERT(g->type == GGML_TYPE_F32 && beta->type == GGML_TYPE_F32);
+
+ const int64_t S_v = v->ne[0];
+ const int64_t H = v->ne[1];
+ const int64_t n_tokens = v->ne[2];
+ const int64_t n_seqs = v->ne[3];
+
+ GGML_ASSERT(g->ne[0] == 1 || g->ne[0] == S_v);
+ GGML_ASSERT(beta->ne[0] == 1);
+
+ GGML_ASSERT(state_cache->type == GGML_TYPE_F32);
+ GGML_ASSERT(state_cache->ne[0] == S_v * S_v * H);
+ GGML_ASSERT(state_cache->nb[1] == (size_t) (S_v * S_v * H) * sizeof(float));
+ GGML_ASSERT(ids->type == GGML_TYPE_I32);
+ GGML_ASSERT(ids->ne[0] == n_seqs);
+ GGML_ASSERT(state_dst->type == GGML_TYPE_F32);
+ GGML_ASSERT(state_dst->ne[0] == S_v * S_v * H && state_dst->ne[1] == n_seqs);
+ GGML_ASSERT(state_dst->nb[1] == state_cache->nb[1]);
+
+ const int64_t ne[4] = { S_v * H, n_tokens * n_seqs + S_v * n_seqs, 1, 1 };
+ struct ggml_tensor * result = ggml_new_tensor(ctx, GGML_TYPE_F32, 4, ne);
+
+ ggml_set_op_params_i32(result, 0, 1); // K == 1
+
+ result->op = GGML_OP_GATED_DELTA_NET;
+ result->src[0] = q;
+ result->src[1] = k;
+ result->src[2] = v;
+ result->src[3] = g;
+ result->src[4] = beta;
+ result->src[5] = state_cache;
+ result->src[6] = ids;
+ result->src[7] = state_dst;
+
+ return result;
+}
+
struct ggml_tensor * ggml_gated_delta_net(
struct ggml_context * ctx,
struct ggml_tensor * q,
diff --git a/src/llama-graph.cpp b/src/llama-graph.cpp
index 7309705dc..de5403dbe 100644
--- a/src/llama-graph.cpp
+++ b/src/llama-graph.cpp
@@ -3915,6 +3915,13 @@ ggml_tensor * llm_graph_context::build_rs(
// {state_size, rs_size} -> {state_size, n_rs - i0}
const int64_t i0 = n_rs - state_copy->ne[0];
+ // custom getter + no extra states: nothing to gather or copy
+ if (get_state_rows && n_rs == (uint32_t) n_seqs) {
+ ggml_tensor * output_states = get_state_rows(ctx0, states, state_copy_main);
+ ggml_build_forward_expand(gf, output_states);
+ return output_states;
+ }
+
ggml_tensor * states_all = ggml_get_rows(ctx0, states, state_copy);
ggml_tensor * output_states = get_state_rows ?
diff --git a/src/models/delta-net-base.cpp b/src/models/delta-net-base.cpp
index 2e15777ee..efc8c0e8e 100644
--- a/src/models/delta-net-base.cpp
+++ b/src/models/delta-net-base.cpp
@@ -521,6 +521,46 @@ ggml_tensor * llm_build_delta_net_base::build_conv_state(
return conv_input;
}
+ggml_tensor * llm_build_delta_net_base::build_recurrent_attn_cache(
+ llm_graph_input_rs * inp,
+ ggml_tensor * ssm_states_all,
+ ggml_tensor * state_cache,
+ ggml_tensor * q,
+ ggml_tensor * k,
+ ggml_tensor * v,
+ ggml_tensor * g,
+ ggml_tensor * b,
+ int il) {
+ const auto * mctx_cur = inp->mctx;
+ const auto kv_head = mctx_cur->get_head();
+
+ const int64_t S_v = v->ne[0];
+ const int64_t H_v = v->ne[1];
+ const int64_t n_seq_tokens = q->ne[2];
+ const int64_t n_seqs = q->ne[3];
+
+ GGML_ASSERT(n_seqs == 1 && cparams.n_rs_seq == 0);
+
+ ggml_tensor * state_dst = ggml_view_2d(ctx0, ssm_states_all, hparams.n_embd_s(), n_seqs, ssm_states_all->nb[1],
+ kv_head * hparams.n_embd_s() * ggml_element_size(ssm_states_all));
+
+ ggml_tensor * result = ggml_gated_delta_net_cache(ctx0, q, k, v, g, b, state_cache, inp->s_copy_main, state_dst);
+ if (n_seq_tokens == 1) {
+ res->add_fused_node({LLM_FUSED_OP_GDN_AR, result, il});
+ } else {
+ res->add_fused_node({LLM_FUSED_OP_GDN_CH, result, il});
+ }
+
+ ggml_tensor * output = ggml_view_4d(ctx0, result,
+ S_v, H_v, n_seq_tokens, n_seqs,
+ ggml_row_size(result->type, S_v),
+ ggml_row_size(result->type, S_v * H_v),
+ ggml_row_size(result->type, S_v * H_v * n_seq_tokens), 0);
+ cb(output, "attn_output", il);
+
+ return output;
+}
+
ggml_tensor * llm_build_delta_net_base::build_recurrent_attn(
llm_graph_input_rs * inp,
ggml_tensor * ssm_states_all,
diff --git a/src/models/models.h b/src/models/models.h
index 5d9b6077a..d0ba683fc 100644
--- a/src/models/models.h
+++ b/src/models/models.h
@@ -88,6 +88,19 @@ struct llm_build_delta_net_base : public llm_graph_context {
// run delta-net attention and write the new recurrent state(s) back to ssm_states_all
// s: (head_v_dim, head_v_dim, num_v_heads, n_seqs); returns output: (head_v_dim, num_v_heads, n_seq_tokens, n_seqs)
+ // single-sequence fast path: reads the state straight from the cache and updates it in place
+ // (no get_rows before, no cpy after); only valid for n_rs_seq == 0, n_seqs == 1 and the fused GDN ops
+ ggml_tensor * build_recurrent_attn_cache(
+ llm_graph_input_rs * inp,
+ ggml_tensor * ssm_states_all,
+ ggml_tensor * state_cache,
+ ggml_tensor * q,
+ ggml_tensor * k,
+ ggml_tensor * v,
+ ggml_tensor * g,
+ ggml_tensor * b,
+ int il);
+
ggml_tensor * build_recurrent_attn(
llm_graph_input_rs * inp,
ggml_tensor * ssm_states_all,
diff --git a/src/models/qwen35moe.cpp b/src/models/qwen35moe.cpp
index 939f10f7f..ffb396c1c 100644
--- a/src/models/qwen35moe.cpp
+++ b/src/models/qwen35moe.cpp
@@ -1,3 +1,4 @@
+#include <cstdlib>
#include "models.h"
#include "llama-memory-recurrent.h"
@@ -406,9 +407,24 @@ ggml_tensor * llama_model_qwen35moe::graph::build_layer_attn_linear(
ggml_tensor * conv_input = build_conv_state(inp, conv_states_all, qkv_mixed, conv_kernel_size, conv_channels, il);
- ggml_tensor * state = build_rs(inp, ssm_states_all, hparams.n_embd_s(), n_seqs);
- state = ggml_reshape_4d(ctx0, state, head_v_dim, head_v_dim, num_v_heads, n_seqs);
- cb(state, "state_predelta", il);
+ // Single-sequence decode/prefill fast path (CPU GDN): read the recurrent state straight from the cache and
+ // update it in place, instead of get_rows + gdn + cpy. Disable with LLAMA_GDN_NO_INPLACE=1.
+ static const bool gdn_inplace_allowed = getenv("LLAMA_GDN_NO_INPLACE") == nullptr;
+ const bool gdn_inplace = gdn_inplace_allowed && n_seqs == 1 && cparams.n_rs_seq == 0 &&
+ (n_seq_tokens == 1 ? cparams.fused_gdn_ar : cparams.fused_gdn_ch) &&
+ mctx_cur->get_n_rs() == 1;
+
+ ggml_tensor * state = nullptr;
+ ggml_tensor * state_cache = nullptr;
+ if (gdn_inplace) {
+ // the custom getter hands back the cache itself; build_rs still orders the optional state clearing
+ state_cache = build_rs(inp, ssm_states_all, hparams.n_embd_s(), n_seqs,
+ [](ggml_context *, ggml_tensor * states, ggml_tensor *) { return states; });
+ } else {
+ state = build_rs(inp, ssm_states_all, hparams.n_embd_s(), n_seqs);
+ state = ggml_reshape_4d(ctx0, state, head_v_dim, head_v_dim, num_v_heads, n_seqs);
+ cb(state, "state_predelta", il);
+ }
ggml_tensor * conv_output_proper = ggml_ssm_conv(ctx0, conv_input, conv_kernel);
cb(conv_output_proper, "conv_output_raw", il);
@@ -467,7 +483,9 @@ ggml_tensor * llama_model_qwen35moe::graph::build_layer_attn_linear(
cb(k_conv, "k_conv_predelta", il);
cb(v_conv, "v_conv_predelta", il);
- ggml_tensor * output = build_recurrent_attn(inp, ssm_states_all, q_conv, k_conv, v_conv, gate, beta, state, il);
+ ggml_tensor * output = gdn_inplace
+ ? build_recurrent_attn_cache(inp, ssm_states_all, state_cache, q_conv, k_conv, v_conv, gate, beta, il)
+ : build_recurrent_attn(inp, ssm_states_all, q_conv, k_conv, v_conv, gate, beta, state, il);
// z: [head_dim, n_heads, n_tokens, n_seqs] -> [n_heads * n_tokens * n_seqs, head_dim]
ggml_tensor * z_2d = ggml_reshape_4d(ctx0, z, head_v_dim, num_v_heads, n_seq_tokens, n_seqs);