Q8_0 tied embeddings

hexstate_quantize.c

@@ -633,6 +633,7 @@ static int is_attention_tensor(const char *gguf_name)
  * conservative too" — creating coherent precision allocation.
  * ═══════════════════════════════════════════════════════════════════════════ */
 
+
 /* ── Multi-quhit expanded scale table ──
  * Search grid: 24×24 = 576 (d, dmin) candidates
  * Quhit encoding: bin 24 → 6 for D=6 quhits (BP operates on 6-state marginals)
@@ -2246,6 +2247,345 @@ static void quantize_tensor_q4_0_hpc(const float *weights, int64_t n_elements,
     free(best_candidate);
 }
 
+/* ════════════════════════════════════════════════════════════════════════
+ * Q8_0 HPC QUANTIZER — Shor pipeline at 8 bits
+ *
+ * Same pipeline as Q4_0: WLS scale + tight candidate grid scored on the
+ * extended objective (weighted SSE + DC + vesica/wave), triality-quhit
+ * graph with Boltzmann-encoded candidate errors, CZ chain entanglement,
+ * Shor Griffiths-Niu sequential measurement for bin consensus, greedy
+ * override (HEX_GREEDY_OVERRIDE_RATIO), then per-block ULP polish, the
+ * vesica/DC error-shaping descent with an extended-objective guard, and
+ * the candidate floor. Intended for embedding / LM-head tensors (tied
+ * embeddings especially), where 2-4 bit codes destroy logit precision.
+ * At 8 bits the candidate grid is tight (±1.5%) — the win over naive
+ * amax/127 rounding comes from WLS + ULP + spectral selection, not from
+ * coarse scale exploration.
+ * ════════════════════════════════════════════════════════════════════════ */
+
+#ifndef QK8_0
+#define QK8_0 32
+#endif
+typedef struct { uint16_t d; int8_t qs[QK8_0]; } hex_block_q8_0;
+
+#define Q8_N_CAND 24
+static const float Q8_NEIGHBOR_MULTS[Q8_N_CAND] = {
+    0.9850f, 0.9865f, 0.9880f, 0.9895f, 0.9910f, 0.9925f,
+    0.9940f, 0.9952f, 0.9964f, 0.9976f, 0.9988f, 1.0000f,
+    1.0010f, 1.0020f, 1.0030f, 1.0040f, 1.0052f, 1.0064f,
+    1.0076f, 1.0088f, 1.0100f, 1.0115f, 1.0130f, 1.0150f,
+};
+/* 24 candidates → 6 quhit states (4 per bin), same folding as Q4_0 */
+static const int Q8_CAND_TO_QUHIT[Q8_N_CAND] = {
+    0,0,0,0, 1,1,1,1, 2,2,2,2, 3,3,3,3, 4,4,4,4, 5,5,5,5
+};
+
+static inline float q8_block_ext_err(const float *bw, const float *iw,
+                                     float d, int8_t *qs_out)
+{
+    float e_arr[QK8_0];
+    float id = (fabsf(d) > 1e-20f) ? 1.0f / d : 0.0f;
+    float err = 0.0f;
+    for (int j = 0; j < QK8_0; j++) {
+        int q = gguf_nearest_int(bw[j] * id);
+        if (q < -127) q = -127; if (q > 127) q = 127;
+        if (qs_out) qs_out[j] = (int8_t)q;
+        float e = bw[j] - (float)q * d;
+        e_arr[j] = e;
+        float w = iw ? iw[j] : 1.0f;
+        err += e * e * w;
+    }
+    return err + hex_spectral_penalty(e_arr, QK8_0);
+}
+
+static void quantize_tensor_q8_0_hpc(const float *weights, int64_t n_elements,
+                                     hex_block_q8_0 *output,
+                                     float *out_total_error,
+                                     const float *imat_importance, int verbose)
+{
+    int64_t n_blocks = n_elements / QK8_0;
+    float total_err = 0.0f;
+    (void)verbose;
+
+    float (*cand_errors)[Q8_N_CAND] = (float (*)[Q8_N_CAND])
+        calloc(n_blocks, sizeof(float[Q8_N_CAND]));
+    uint16_t (*cand_d16)[Q8_N_CAND] = (uint16_t (*)[Q8_N_CAND])
+        calloc(n_blocks, sizeof(uint16_t[Q8_N_CAND]));
+    int *best_candidate = (int *)malloc(n_blocks * sizeof(int));
+    if (!cand_errors || !cand_d16 || !best_candidate) {
+        free(cand_errors); free(cand_d16); free(best_candidate);
+        if (out_total_error) *out_total_error = -1.0f;
+        return;
+    }
+
+    /* ── Phase 1+2: WLS-refined scale + tight candidate grid ── */
+    #pragma omp parallel for schedule(dynamic, 256)
+    for (int64_t blk = 0; blk < n_blocks; blk++) {
+        const float *bw = weights + blk * QK8_0;
+        const float *iw = imat_importance ? imat_importance + blk * QK8_0 : NULL;
+
+        float amax = 0.0f;
+        for (int j = 0; j < QK8_0; j++) {
+            float av = fabsf(bw[j]);
+            if (av > amax) amax = av;
+        }
+        float wls_d = amax / 127.0f;
+
+        /* ggml-style fixed-point WLS with DC rank-1 augmentation */
+        for (int it = 0; it < 3 && wls_d > 1e-20f; it++) {
+            float inv_d = 1.0f / wls_d;
+            float num = 0.0f, den = 0.0f, dcS = 0.0f, dcQ = 0.0f;
+            for (int j = 0; j < QK8_0; j++) {
+                int q = gguf_nearest_int(bw[j] * inv_d);
+                if (q < -127) q = -127; if (q > 127) q = 127;
+                float qf = (float)q;
+                float w  = iw ? iw[j] : 1.0f;
+                num += w * bw[j] * qf;
+                den += w * qf * qf;
+                dcS += bw[j];
+                dcQ += qf;
+            }
+            num += (HEX_DC_LAMBDA / (float)QK8_0) * dcS * dcQ;
+            den += (HEX_DC_LAMBDA / (float)QK8_0) * dcQ * dcQ;
+            if (den > 1e-15f) {
+                float d_new = num / den;
+                if (d_new > 1e-20f) wls_d = d_new;
+            }
+        }
+
+        for (int ci = 0; ci < Q8_N_CAND; ci++) {
+            float    trial_d  = wls_d * Q8_NEIGHBOR_MULTS[ci];
+            uint16_t d16      = gguf_fp32_to_fp16(trial_d);
+            float    actual_d = gguf_fp16_to_fp32(d16);
+            cand_d16  [blk][ci] = d16;
+            cand_errors[blk][ci] = q8_block_ext_err(bw, iw, actual_d, NULL);
+        }
+        best_candidate[blk] = 11;   /* ×1.0000 neutral seed */
+    }
+
+    /* ── Phase 3: Shor graph — triality quhits, CZ chain, GN measurement ── */
+    int shor_ran = 0;
+    if (n_blocks >= 2) {
+        int64_t graph_blocks = (n_blocks > 200) ? 200 : n_blocks;
+        int64_t stride = n_blocks / graph_blocks;
+
+        HPCGraph *graph = hpc_create(graph_blocks);
+        if (graph) {
+            shor_ran = 1;
+
+            /* Adaptive temperature from the candidate-error landscape */
+            float temperature = 1e-10f;
+            {
+                double err_accum = 0.0;
+                int err_count = 0;
+                for (int64_t gi = 0; gi < graph_blocks && gi < 100; gi++) {
+                    int64_t blk = gi * stride;
+                    float max_e = 0.0f;
+                    for (int c = 0; c < Q8_N_CAND; c++)
+                        if (cand_errors[blk][c] > max_e)
+                            max_e = cand_errors[blk][c];
+                    err_accum += (double)max_e;
+                    err_count++;
+                }
+                if (err_count > 0) {
+                    temperature = (float)(err_accum / err_count) * 0.1f;
+                    if (temperature < 1e-10f) temperature = 1e-10f;
+                }
+            }
+
+            /* Boltzmann-encode stride-aggregated candidate errors as
+             * quhit amplitudes (24 candidates folded into 6 states) */
+            for (int64_t i = 0; i < graph_blocks; i++) {
+                float agg_errors[Q8_N_CAND];
+                for (int c = 0; c < Q8_N_CAND; c++) agg_errors[c] = 0.0f;
+                int64_t blk_start = i * stride;
+                int64_t blk_end   = blk_start + stride;
+                if (blk_end > n_blocks) blk_end = n_blocks;
+                for (int64_t b = blk_start; b < blk_end; b++)
+                    for (int c = 0; c < Q8_N_CAND; c++)
+                        agg_errors[c] += cand_errors[b][c];
+                float min_err = 1e30f;
+                for (int c = 0; c < Q8_N_CAND; c++)
+                    if (agg_errors[c] < min_err) min_err = agg_errors[c];
+
+                double amp_re[6] = {0,0,0,0,0,0};
+                double amp_norm = 0.0;
+                for (int ci = 0; ci < Q8_N_CAND; ci++)
+                    amp_re[Q8_CAND_TO_QUHIT[ci]] +=
+                        exp(-(double)(agg_errors[ci] - min_err) /
+                             (2.0 * (double)temperature));
+                for (int v = 0; v < 6; v++) amp_norm += amp_re[v] * amp_re[v];
+                if (amp_norm > 1e-30) {
+                    double inv = 1.0 / sqrt(amp_norm);
+                    for (int v = 0; v < 6; v++) amp_re[v] *= inv;
+                }
+                for (int v = 0; v < 6; v++) {
+                    graph->locals[i].edge_re[v] = amp_re[v];
+                    graph->locals[i].edge_im[v] = 0.0;
+                }
+                graph->locals[i].primary = VIEW_EDGE;
+                graph->locals[i].dirty = DIRTY_VERTEX | DIRTY_DIAGONAL | DIRTY_FOLDED;
+                graph->locals[i].delta_valid = 0;
+                triality_update_mask(&graph->locals[i]);
+            }
+
+            for (int64_t i = 0; i < graph_blocks - 1; i++)
+                hpc_cz(graph, i, i + 1);
+
+            double (*marg)[6] = (double (*)[6])calloc(graph_blocks, sizeof(double[6]));
+            int *measured = (int *)calloc(graph_blocks, sizeof(int));
+            if (marg && measured) {
+                shor_measure_graph(graph, graph_blocks, marg, measured, 1);
+
+                /* Per-block selection: best candidate inside the Shor-
+                 * measured bin, then greedy override against the global
+                 * argmin — identical Step-F semantics to Q2_K/Q4_0. */
+                for (int64_t i = 0; i < graph_blocks; i++) {
+                    int bin = measured[i];
+                    if (bin < 0 || bin > 5) {
+                        double bm = -1.0; bin = 0;
+                        for (int v = 0; v < 6; v++)
+                            if (marg[i][v] > bm) { bm = marg[i][v]; bin = v; }
+                    }
+                    int64_t blk_start = i * stride;
+                    int64_t blk_end   = blk_start + stride;
+                    if (blk_end > n_blocks) blk_end = n_blocks;
+                    for (int64_t b = blk_start; b < blk_end; b++) {
+                        float bin_best = 1e30f; int bin_cand = -1;
+                        float g_best   = 1e30f; int g_cand   = 0;
+                        for (int c = 0; c < Q8_N_CAND; c++) {
+                            float e = cand_errors[b][c];
+                            if (e < g_best) { g_best = e; g_cand = c; }
+                            if (Q8_CAND_TO_QUHIT[c] == bin && e < bin_best) {
+                                bin_best = e; bin_cand = c;
+                            }
+                        }
+                        int sel = (bin_cand >= 0) ? bin_cand : g_cand;
+                        if (g_best < cand_errors[b][sel] * HEX_GREEDY_OVERRIDE_RATIO)
+                            sel = g_cand;
+                        best_candidate[b] = sel;
+                    }
+                }
+            }
+            free(marg); free(measured);
+            hpc_destroy(graph);
+        }
+    }
+    if (!shor_ran) {
+        for (int64_t blk = 0; blk < n_blocks; blk++) {
+            float g_best = cand_errors[blk][0]; int g_cand = 0;
+            for (int c = 1; c < Q8_N_CAND; c++)
+                if (cand_errors[blk][c] < g_best) {
+                    g_best = cand_errors[blk][c]; g_cand = c;
+                }
+            best_candidate[blk] = g_cand;
+        }
+    }
+
+    /* ── Phase 4: ULP polish + vesica/DC shaping guard + floor ── */
+    #pragma omp parallel for schedule(dynamic, 256) reduction(+:total_err)
+    for (int64_t blk = 0; blk < n_blocks; blk++) {
+        const float *bw = weights + blk * QK8_0;
+        const float *iw = imat_importance ? imat_importance + blk * QK8_0 : NULL;
+        int cidx = best_candidate[blk];
+
+        uint16_t best_d16 = cand_d16[blk][cidx];
+        float    best_err = cand_errors[blk][cidx];
+
+        /* ±8 fp16 ULP joint search on the extended objective */
+        for (int du = -8; du <= 8; du++) {
+            if (du == 0) continue;
+            int c16 = (int)cand_d16[blk][cidx] + du;
+            if (c16 <= 0 || c16 > 0x7BFF) continue;
+            float td  = gguf_fp16_to_fp32((uint16_t)c16);
+            float err = q8_block_ext_err(bw, iw, td, NULL);
+            if (err < best_err) { best_err = err; best_d16 = (uint16_t)c16; }
+        }
+
+        /* Candidate floor: final ≤ best raw grid candidate (by construction
+         * the ULP search already starts from it, so this is implicit). */
+        float  d = gguf_fp16_to_fp32(best_d16);
+        int8_t qs[QK8_0];
+        (void)q8_block_ext_err(bw, iw, d, qs);
+
+        /* Vesica/DC greedy shaping with extended-objective guard */
+        {
+            int8_t qs_shaped[QK8_0];
+            memcpy(qs_shaped, qs, QK8_0);
+            float e_live[QK8_0], v_live[QK8_0 / 2];
+            float vesica_cur = 0.0f, dc_cur = 0.0f;
+            for (int k = 0; k < QK8_0; k++)
+                e_live[k] = bw[k] - (float)qs_shaped[k] * d;
+            for (int p = 0; p < QK8_0 / 2; p++) {
+                v_live[p] = e_live[p] + e_live[p + QK8_0 / 2];
+                vesica_cur += v_live[p] * v_live[p];
+                dc_cur     += v_live[p];
+            }
+            float metric_cur = 4.0f * vesica_cur + dc_cur * dc_cur;
+            for (int pass = 0; pass < QK8_0; pass++) {
+                int best_k = -1, best_q_alt = 0;
+                float best_delta = 0.0f;
+                for (int k = 0; k < QK8_0; k++) {
+                    int q_try = (e_live[k] >= 0.0f) ? qs_shaped[k] + 1
+                                                    : qs_shaped[k] - 1;
+                    if (q_try < -127 || q_try > 127) continue;
+                    float e_new = bw[k] - (float)q_try * d;
+                    float de    = e_new - e_live[k];
+                    int   pi    = (k < QK8_0 / 2) ? k : k - QK8_0 / 2;
+                    float v_new = v_live[pi] + de;
+                    float ves_a = vesica_cur - v_live[pi] * v_live[pi]
+                                             + v_new * v_new;
+                    float dc_a  = dc_cur + de;
+                    float delta = metric_cur - (4.0f * ves_a + dc_a * dc_a);
+                    if (delta > best_delta) {
+                        best_delta = delta; best_k = k; best_q_alt = q_try;
+                    }
+                }
+                if (best_k < 0) break;
+                {
+                    float e_new = bw[best_k] - (float)best_q_alt * d;
+                    float de    = e_new - e_live[best_k];
+                    int   pi    = (best_k < QK8_0 / 2) ? best_k
+                                                       : best_k - QK8_0 / 2;
+                    float v_new = v_live[pi] + de;
+                    vesica_cur += v_new * v_new - v_live[pi] * v_live[pi];
+                    dc_cur     += de;
+                    metric_cur  = 4.0f * vesica_cur + dc_cur * dc_cur;
+                    v_live[pi]      = v_new;
+                    e_live[best_k]  = e_new;
+                    qs_shaped[best_k] = (int8_t)best_q_alt;
+                }
+            }
+            /* Guard on the extended objective vs originals */
+            float e_b[QK8_0], e_s[QK8_0];
+            float err_b = 0.0f, err_s = 0.0f;
+            for (int k = 0; k < QK8_0; k++) {
+                float w = iw ? iw[k] : 1.0f;
+                e_b[k] = bw[k] - (float)qs[k]        * d;
+                e_s[k] = bw[k] - (float)qs_shaped[k] * d;
+                err_b += e_b[k] * e_b[k] * w;
+                err_s += e_s[k] * e_s[k] * w;
+            }
+            err_b += hex_spectral_penalty(e_b, QK8_0);
+            err_s += hex_spectral_penalty(e_s, QK8_0);
+            if (err_s < err_b) memcpy(qs, qs_shaped, QK8_0);
+        }
+
+        output[blk].d = best_d16;
+        for (int k = 0; k < QK8_0; k++) {
+            output[blk].qs[k] = qs[k];
+            float e = bw[k] - (float)qs[k] * d;
+            total_err += e * e;          /* pure reconstruction SSE report */
+        }
+    }
+
+    free(cand_errors);
+    free(cand_d16);
+    free(best_candidate);
+    if (out_total_error) *out_total_error = total_err;
+}
+
+
 /* Re-derive the 4-bit sub-scale codes (Ls, Lm) for a candidate (d, dmin)
  * pair from the Phase-1 float scales/mins. Bit-identical to the Phase-2b
  * candidate generation, so stored codes are unnecessary. */
@@ -4619,6 +4959,18 @@ void hexstate_quantize_tensor_q4_0_hpc(const float *weights, int64_t n_elements,
     if (out_error) *out_error = err;
 }
 
+int hexstate_q8_0_block_bytes(void)    { return (int)sizeof(hex_block_q8_0); }
+int hexstate_q8_0_block_elements(void) { return QK8_0; }
+
+void hexstate_quantize_tensor_q8_0_hpc(const float *weights, int64_t n_elements,
+                                       void *output, float *out_error,
+                                       const float *imat_importance, int verbose)
+{
+    quantize_tensor_q8_0_hpc(weights, n_elements,
+                             (hex_block_q8_0 *)output, out_error,
+                             imat_importance, verbose);
+}
+
 #ifndef HEXSTATE_LIBRARY
 /* ═══════════════════════════════════════════════════════════════════════════
  * MAIN