TurboQuant: FastScan layout would make it faster than INT8, but changes the on-disk format #57

Summary

TurboQuant's lookup-table scan is at the limit of what its current storage layout allows.
Measured on an Apple M5 Pro with make benchmark (1M vectors, dim 768, cosine, k=20,
file-backed database), all indexes preloaded:

TurboQuant beats brute force comfortably. What it does not beat is INT8, which is 4x
faster than TURBO4 at twice the index size — so TurboQuant's argument today is memory,
not throughput. Making it competitive on both is possible, but it is a storage-format
change rather than a kernel optimisation, which is what this issue is for deciding.

Note. An earlier version of this issue said TURBO4 was slower than the exact
scan. That was an artifact of benchmarking against an in-memory database, where the
exact scan does not pay to read its 3 GB. The benchmark is now file-backed and the
numbers above are the corrected ones. The case for FastScan is unchanged, but the gap it
has to close is 4x against INT8, not the 20x+ the in-memory figures implied.

Why the current scan cannot go faster

Each quantized row is stored contiguously:

[ rowid 8B ][ scale 4B ][ packed codes ceil(dim*bits/8) B ]

At dim 768 / 4 bits that is 396 bytes per row. The scan builds a query-dependent LUT
(turbo_build_query_lut(): 384 rows x 256 entries x 4 B = 384 KB for bits=4) and
then, per stored vector, performs 384 dependent gathers — query_lut[r*256 + packed[r]].

Those gathers already run at roughly one lookup per cycle, which is the load unit's
limit. There is no headroom:

• NEON has no gather instruction at all.
• AVX2's vgatherdps is not faster than scalar loads for this access pattern.

This was measured while investigating #55: plain C with four independent double
accumulators matched every hand-written SIMD version to within 2%, which is why that PR
collapsed five near-identical implementations into one. The parallelism was never the
bottleneck; the memory gathers are.

What FastScan changes

The FastScan family (Quicker ADC, Bolt, FAISS IVFPQFastScan, SCANN) moves the lookup
table from memory into registers, using the byte-shuffle instruction as a 16-entry
LUT: pshufb / _mm256_shuffle_epi8 on x86, vqtbl1q_u8 on NEON. One instruction
performs 16, 32 or 64 lookups in parallel with no memory access.

Three constraints follow:

1. 4-bit codes only. 16 possible values is exactly a pshufb table. TURBO2 and
   TURBO3 cannot use this path.
2. The LUT must be quantized to uint8. The shuffle looks up bytes, not floats, so
   partial products are scaled to 8 bits and accumulated in 16 bits with periodic
   widening, plus a per-block scale. This costs some recall — bounding that loss is
   well covered in the literature but needs validating on real embeddings.
3. The codes must be interleaved, which is the format change.

The format change

today     [v0: codes 0..383][v1: codes 0..383][v2: ...]
FastScan  [32 rowids][32 scales][code position 0 for all 32 vectors][position 1 for all 32]...

Blocks of 32 vectors, transposed. One 32-byte register then holds the same code position
for 32 different vectors, and a single shuffle computes 32 partial products.

The instruction count changes shape entirely: from ~384 memory gathers per vector to
roughly 5 instructions (1 load + 2 shuffles + 2 adds) per 32 vectors per byte position —
about 60 register operations per vector instead of 384 memory accesses.

What it costs

• vector0_<table>_<column> becomes unreadable across versions, in both directions.
  It needs a format marker so an old index is detected and rejected with a clear message
  rather than decoded as garbage.
• Upgrading requires re-running vector_quantize(). No data is lost — the raw vectors
  are still in the user's table — but it is a full pass over the data: minutes for
  millions of rows.
• 19 sites in sqlite-vector.c assume a fixed per-row stride (total_stride,
  INT64_FROM_INT8PTR(current)) and would move to block iteration. That includes the
  bounds checks added in Source audit: 13 defects fixed, SIMD kernels actually built and tested #53, which are expressed as counter * total_stride and would
  need reworking for the block layout.
• A partial final block (rows not a multiple of 32) needs padding or a scalar tail.
• Per-lane details: _mm256_shuffle_epi8 and _mm512_shuffle_epi8 operate within
  128-bit lanes, so the 16-entry table has to be replicated per lane. vqtbl1q_u8 is a
  clean 16-byte table lookup.

What it buys

The literature reports 5–10x over LUT-in-memory scanning. TURBO4 is at 151.5 ms/query,
and INT8 at 37.6 ms; closing a 4x gap is comfortably inside that range, so FastScan
would plausibly make TURBO4 as fast as or faster than INT8 at half the index size
(378 MB against 740 MB for 1M x 768). TurboQuant would then have both arguments instead of
one, rather than being the mode you pick when memory is what binds.

Worth noting what the 4x gap is not: it is not a gap against brute force, which
TurboQuant already wins by 3.2x. This work is about displacing INT8 as the default
recommendation, not about making quantization worthwhile.

The decision

This is worth doing if TurboQuant is meant to become the recommended mode. It is not worth
doing if TurboQuant stays "the option for when memory is the binding constraint", which is
what it honestly is today and what the README says.

Reproduce any of the numbers above with make benchmark HARDWARE="<your CPU>".

Open questions if it goes ahead:

• Keep TURBO2/TURBO3 on the current layout and give only TURBO4 the FastScan path,
  or drop the other widths?
• Reject old indexes and require re-quantization, or read both layouts for one release?
• How much recall does the 8-bit LUT quantization actually cost on real embeddings — worth
  measuring with test/recall_turboquant_real.py before committing to the work.

Measurements above are reproducible with make benchmark.