Commit Graph
130 Commits
Author SHA1 Message Date
qdrant-cloud-bot dec9cd9310 Rename MmapFlusher to Flusher (#10496)
The type is used by RAM and other non-mmap storages (often as a no-op),
so the Mmap-prefixed name was misleading.
2026-09-07 14:56:46 +00:00
EduardoandClaude Opus 5 c2137f95c0 test: cover Bits1_5 in TurboQuant test matrices (#10448)
Nine bit-width matrices in `turboquant::quantization` omitted
`TQBits::Bits1_5`, so the variant went unexercised there. This is the gap
#10390 closed for `quantize_output_byte_length`, where a `quantized_size`
double-padding bug had slipped through a hardcoded [Bits1, Bits2, Bits4]
list.

Add `Bits1_5` to the seven matrices that accept it, including the rstest
cases of `score_precomputed_dispatches_all_bit_widths`, whose name already
claims to cover every bit width.

The two `Unpadded` rotation tests keep their lists: `TurboQuantizer::new`
asserts "Bits1_5 requires TQRotation::Padded", since `Bits1_5` rotates
into its x1.5 padding and an unpadded rotation would leave half the codes
carrying nothing. Each list now names that invariant, so the omission
reads as deliberate rather than as the same oversight.
`unpadded_rotation_matches_padded_for_padding_free_dims` also claimed
multiples of 8 are padding-free "for every supported bit width", which
`Bits1_5` falsified: `padded_dim(8)` is 16.

Co-authored-by: Claude Opus 5 <noreply@anthropic.com>
2026-09-03 11:33:15 +02:00
61ffec42ab Fall back to per-vector TQ scoring for scattered ids (#10381)
* Fall back to per-vector TQ scoring when runs are short

Run-batched scoring pays off on plain and dense filtered scans but
regresses HNSW, where neighbor ids rarely form consecutive runs and batch
setup dominates. Gate both EncodedVectorsTQ::score_points and Turbo
score_query_batch on offsets_worth_batch_scoring, which takes the run path
only when the ids split into runs averaging BATCH_SCORE_MIN_MEAN_RUN
vectors or more.

The average decides, not the longest run: a sorted id list -- what a
filtered scan hands the scorer -- already contains adjacent pairs at ~1%
density while its runs still average one vector, so a "contains a run of
>= 2" test sends those down the run path to pay setup per vector, measured
at up to +75% against the better path. An average also stays independent
of the batch size the driver slices ids into, which a longest-run test does
not. The threshold comes from the measured crossover -- mean run 2.0-3.6,
stable across dims 128/512/1024, both RAM storages and the 1/2/4-bit
widths -- and a fully contiguous block is recognized in O(1), so a plain
scan pays nothing for the gate.

* io_uring: never gate run-batched scoring

The gate exists because run batching costs setup that short runs do not
repay on RAM and mmap storages. io_uring is the opposite: one run-granular
read beats the batched per-vector path at every density measured -- 27% on
HNSW-shaped id lists, 43% at 25% filter density, 93% on a full scan --
because per-request submission and completion bookkeeping dominates once
the data sits in the page cache. Gating it costs 36% on HNSW-shaped lists.

Add EncodedStorage::prefers_run_reads, defaulting to false so every storage
keeps its current routing, and override it for the single-file quantized
storage when its backend is io_uring. Remote backends (object stores, a
gRPC peer) deliberately keep the per-vector path: their reads pipeline
across a batch, while run-granular reads would serialize the round trips.

QuantizedStorage::is_in_ram_or_mmap() still reports true for every backend,
which is what routes io_uring into the gate in the first place. Correcting
that would also change how the multivector storage picks between its
in-memory and uring scoring paths, so it is left to a separate change.

* Rename prefers_run_reads to prefers_contiguous_reads

"Run reads" is easy to misread as "execute reads"; contiguous makes the
storage I/O preference explicit.

* QuantizedStorage::for_each_run: pipeline run reads on async backends

With `prefers_contiguous_reads()` true for io_uring, every batch goes
through `for_each_run`, which read each run synchronously: a scattered
id list (HNSW neighbours) waited on one disk read per vector, where
`for_each_in_batch` kept the whole batch in flight through `read_batch`.
Submit all runs of a batch together, still one read per run, so
scattered reads stay pipelined while a scan still reads each run in one
request.  Backends without async reads keep the sequential loop.

`turbo_vector_search` (dim 1024, 200k vectors, 4096 shuffled ids per
iteration) against dev: cold scattered io_uring 187 ms -> 26.5 ms
(dev 29.5 ms); the warm scan keeps 198 ms -> 21 ms.  Warm scattered
lands at 5.06 ms (dev 4.67 ms), giving up the 3.68 ms of synchronous
reads, which only holds with the data already in the page cache.

Co-Authored-By: Claude Fable 5.1 <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01VRiRaiPsm5CEBgpQ7VAQab

* Iterate consecutive runs and share the run-scoring gate

`for_each_consecutive_run` becomes `consecutive_runs()`, a lazy iterator
over `Run { first, start, len }`, so a storage can feed runs straight into
a read pipeline. `QuantizedStorage::for_each_run` loses the intermediate
`Vec` and the duplicated `ReadRange` construction: the pipelined branch
maps the iterator into `read_batch`, the synchronous branch keeps the
per-run `Sequential`/`Random` hint that picks between mmap's two mappings.

The routing condition duplicated at both scoring call sites moves into
`EncodedStorage::prefers_run_scoring`: same expression, one place.

`for_each_run`'s contract no longer promises run order: pipelined
backends report reads as they complete, so callers address results by
`first`. Add a contract test over the mmap and disk-cache backends; the
latter is the async-capable backend that runs on every platform and
covers the `read_batch` branch io_uring takes on Linux, which no test
exercised before.

Measured on Apple M3 against 1f2d1264e, interleaved A/B/B/A: the run
path is unchanged on all four storages (-0.3%, +0.0%, +0.9%, -1.1%,
within replicate noise).

Co-Authored-By: Claude Fable 5.1 <noreply@anthropic.com>

* Rename Run to ConsecutiveRun

`Run` on its own reads as "execute", the same ambiguity that got
`prefers_run_reads` renamed earlier in this branch. `ConsecutiveRun`
names what the value is and pairs with `consecutive_runs()`, the iterator
that yields it.

Co-Authored-By: Claude Fable 5.1 <noreply@anthropic.com>

* io_uring pipeline: submit eagerly only while reads are outstanding

`IoUringPipeline::wait()` called `submit_and_wait(0)` whenever a completion
was ready and anything was enqueued. `read_batch` enqueues one entry per
consumed completion, so on a warm page cache — where reads complete inline
at submission — that was one `io_uring_enter` per read. Scattered quantized
scoring over io_uring ran at ~630 ns/point warm against ~460 for the same
reads issued synchronously (which, without direct_io, are plain `pread`s).

Now the eager submission happens only while the kernel still has reads
outstanding (`in_progress` minus the completions already waiting in the
queue). When everything submitted so far has completed — the warm case —
the enqueued entries wait and go down together once the ready completions
run out. On a cold device nothing changes: a completion is answered with a
submission as before, so the in-flight depth never sags. Two fixed rules
tried first (submit only when nothing is ready; submit once half the queue
piled up) both cost the cold path, +6 % and +3 %, in proportion to how long
the device sat idle while ready completions were drained.

turbo_vector_search / turbo_uring_ab, Zen 4, `taskset -c 7`, prebuilt
binaries run alternately, cold rows with the page cache dropped:

  warm scattered, uring hnsw:   634 -> 472 ns/point  (-25 %)
  warm scattered, uring p0.25:  477 -> 369 ns/point  (-23 %)
  warm sequential, uring p1.00:  45 ->  45           (flat)
  cold scattered, uring:        29.9 -> 29.9 ms/iter (flat, 4 reps each within 0.7 %)
  mmap rows (control):          within ±2 %

Co-Authored-By: Claude Fable 5.1 <noreply@anthropic.com>

---------

Co-authored-by: Ivan Pleshkov <pleshkov.ivan@gmail.com>
Co-authored-by: timvisee <tim@visee.me>
Co-authored-by: Claude Fable 5.1 <noreply@anthropic.com>
2026-09-02 17:57:27 +02:00
4d5fb7e6c7 TurboQuant AVX512 + Neon reduce four vectors at once (#10439)
* 4 way hsum for avx512 and neon

* Fix AVX512 regression

* Revert "Fix AVX512 regression"

This reverts commit 7da4676bdc.

* avx512: keep the per-vector reduction for two-byte queries

The shared group reduction is a clear win at QUERY_BYTES == 1 and a large
loss at 2. Measured on Zen 4 against this commit's parent: 4-bit batch
scoring +42% at dim 64 and +45% at 128, 2-bit +23%, while 1-bit gains -6%.

The cause is register pressure. Acc512<QUERY_BYTES> is
[[__m512i; 2]; QUERY_BYTES], so a group of four holds 8 ZMM at one query
byte but 16 at two, and the transpose needs every vector's lanes at once,
so nothing retires early. objdump on the batch kernel counts 0 ZMM spills
to the stack before this PR and 64 after -- and the two instantiations
that spill are exactly the two widths that regress.

Gate the shared reduction on QUERY_BYTES == 1 and let two-byte queries
reduce and release one accumulator at a time, as they did before. The
group width stays at four: shrinking it to two for those widths measured
much worse (+39...+73%), since four vectors x two chains is what covers
the VPDPBUSD latency.

ns per 512-vector run, medians of 4 interleaved reps, L2-resident pool,
against this commit's parent:

  width  dim     parent   this PR   with the gate
  1-bit    64      2680    -6.3 %          -6.3 %
  1-bit   128      2682    -6.3 %          -6.2 %
  1-bit   512      2525    -6.0 %          -7.2 %
  2-bit    64      2161   +23.5 %          +1.2 %
  2-bit   128      2162   +23.6 %          +1.2 %
  2-bit   512      3058   +24.7 %          +0.5 %
  4-bit    64      1486   +42.2 %          +3.4 %
  4-bit   128      1412   +45.0 %          +2.8 %
  4-bit   512      2971   +28.8 %          +1.2 %
  4-bit  1536      8681   +16.0 %          -1.0 %

Spills drop from 64 to 2. The few percent left at the smallest 4-bit dims
come from the `interleave` test now sitting inside the group loop instead
of outside it, where the compiler must keep the untaken branch live;
hoisting it back out measured +0.1...+0.2%, at the cost of duplicating
the loop.

NEON is untouched here and not measured on this machine.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>

* avx512: resolve the interleave choice outside the group loop

The gate on the previous commit left 4-bit batch scoring ~3% above the
parent at the smallest dims. The cause is the `interleave` test sitting
inside the group loop: the untaken arm stays live for the register
allocator, so the widths that never take it still pay for it.

Pass the choice as a const parameter instead, so the loop is compiled once
per shape and both arms fold away, and build the non-interleaved group with
array::from_fn rather than zero-initialising it first.

ns per 512-vector run, medians of 3 interleaved reps, L2-resident pool,
against the PR's parent:

  width  dim     parent   this PR   gate only   with this commit
  4-bit   64       1487   +42.6 %      +3.4 %             +0.0 %
  4-bit  128       1414   +44.9 %      +2.6 %             +0.0 %
  4-bit  512       2978   +28.5 %      +1.0 %             -0.2 %
  4-bit 1536       8814    +9.4 %      -3.6 %             -0.5 %
  2-bit   64       2160   +23.6 %      +1.2 %             +0.2 %
  2-bit  128       2156   +24.4 %      +0.9 %             +0.5 %
  2-bit  512       3059   +23.5 %      +0.8 %             +0.0 %
  1-bit   64       2679    -6.4 %      -6.5 %             -7.2 %
  1-bit  128       2682    -6.3 %      -6.5 %             -7.1 %
  1-bit  512       2518    -6.1 %      -6.2 %             -6.4 %
  1-bit 1536       5851    -2.9 %      -2.9 %             -2.7 %

The two-byte widths are back on the parent and the one-byte win grows a
little, since it was paying for the same branch.

One cell moves the other way: 2-bit at 1536 reads +2.5 % where the gate
alone read -3.0 %. That is the non-interleaved path this commit also
restructures, at the dim where this machine is bimodal; three reps do not
separate it from noise.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>

---------

Co-authored-by: Ivan Pleshkov <pleshkov.ivan@gmail.com>
Co-authored-by: Claude Opus 5 (1M context) <noreply@anthropic.com>
2026-09-02 16:40:27 +02:00
Ivan Pleshkov 6d973050c3 QuerySimd: hold the query planes in one allocation (#10441)
The planes were PLANES * QUERY_BYTES separate Vecs -- four for the 4-bit
width, sixteen for a 16-bit query at the 1-bit width. Scoring re-reads all
of them for every vector, so keeping them in that many allocations spreads
one small, permanently hot structure over as many pages, TLB entries and
prefetch streams as there are planes.

Put them in a single buffer instead, plane (b, k) at
(b * PLANES + k) * plane_len, and hand the kernels their slice through an
unchecked accessor: it is re-taken for every block of every vector, and a
bounds-checked one measured 28-36% slower than the previous layout.
2026-09-02 16:10:37 +02:00
Jojii bfc5428a8d 4-way reduction (#10437) 2026-09-02 12:13:21 +02:00
Ivan PleshkovandClaude Fable 5 4aa8066b75 Turbo4 batched scan (#10362)
* TurboQuantizer::score_precomputed_batch: score a contiguous run of vectors

Batch counterpart of `score_precomputed` for vectors stored back to
back at `quantized_size()`: the width's kernel scores the whole run of
codes in one `dotprod_batch` call, then a second pass applies each
vector's extras.  L1 dequantizes per vector and stays a plain loop.

Tested against per-vector `score_precomputed` for every width,
distance, and mode over run lengths that leave every group remainder.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* EncodedStorage::for_each_run: serve consecutive offsets as contiguous runs

`for_each_run(offsets, callback(first, count, bytes))` splits the
offsets into maximal runs of consecutive ids the storage can serve
from one contiguous slice, so a sequential scan resolves chunk lookups
and reads once per run instead of once per vector.  The default serves
every vector as its own run; `for_each_consecutive_run` is the shared
run detection for storages that override it, with a per-run cap for
chunk boundaries.  The test storage overrides it (its data is one flat
buffer).

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* EncodedVectors::score_points: batched scoring entry point, run-batched for TQ

`score_points(query, offsets, scores)` scores a batch of points.  The
default keeps the per-vector loop the scorers run today, so SQ/PQ/BQ
are unchanged.  TurboQuant overrides it: on RAM/mmap storages it walks
`for_each_run` and scores each contiguous run with one
`score_precomputed_batch` call, hoisting the score inversion out of
the loop; backends with async reads keep the pipelined per-vector
path.  Non-consecutive offsets degrade to single-vector runs, so
scattered access keeps its previous cost.

Integration test: `score_points` vs `score_point` for every bit width
and mode, Dot and inverted L2, over sequential, scattered and
descending id orders.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Quantized storages: for_each_run over their contiguous regions

The RAM storage and both chunked mmap storages cap runs at their chunk
boundary and serve each run with one `get_many`; the single-file mmap
storage serves any run as one sequential read.  Unit test on the RAM
storage: runs cover every offset once, in order, with bytes identical
to per-point reads, across the internal chunk boundary.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuantizedQueryScorer: score batches through EncodedVectors::score_points

Routes `score_stored_batch` through the batched entry point, so
TurboQuant-as-quantization scans score contiguous runs with one kernel
call per run; SQ/PQ/BQ keep the per-vector loop via the default.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* TurboScoring::score_query_batch: run-batched scoring for Turbo4 storages

Adds the batch counterpart of `score_query_bytes` to the trait, with
one shared implementation over the storage's `EncodedStorage`:
consecutive ids are coalesced into contiguous runs, each run scored by
a single `score_precomputed_batch` call, and the metric sign applied
once over the batch.  Backends with async reads keep the pipelined
per-vector path.  `TurboQueryScorer::score_stored_batch` now calls it.

The batch-vs-single storage test grows to 8192 vectors so a full
ascending scan crosses a chunk boundary of the chunked backend, and
runs that scan on the chunked, mmap and io_uring backends.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* score_precomputed_batch: keep the extras pass in L1

The kernel pass and the extras pass now alternate over sub-runs of 64
vectors instead of each covering the whole run: for a run of several
hundred vectors the second pass otherwise refetched every vector's
extras from L2.  Measured with 512-vector runs from the full-scan
driver at dim 512: the regression against 64-vector runs went from
+11 % to +2 %.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Bench: exhaustive search over Turbo4 storages through the plain-index driver

`turbo4_full_scan` runs `BatchFilteredSearcher::peek_top_visible` —
the exact path of a non-indexed search — over 200k normalized random
vectors for Turbo4 as datatype (appendable chunked, in RAM) and Turbo4
as quantization (over a RAM dense storage), at dims 64 to 1024, so the
fixed per-point cost of the scan driver is measured next to the kernel.
`TURBO_SCAN_DIMS=64,128` narrows the dims while iterating.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

---------

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
2026-09-02 11:41:07 +02:00
Ivan PleshkovandClaude Fable 5 94f3abeb22 Turbo4 dotprod batch (#10392)
* QuerySimd::dotprod_batch: score a contiguous run of vectors in one call

The entry point for scanning a contiguous run of encoded vectors at a
stride: `out[v]` ← score of the vector at `data[v * stride..]`.  It
scores vector by vector for now; the SIMD batch kernels that share the
query loads across vectors follow.

Bench: `query{4,2,1}bit_dotprod_scan` — a hot query against runs of 512
consecutive vectors streaming from DRAM at the TurboQuant stride, per
vector and through `dotprod_batch`.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuerySimd: interleaved AVX-512 batch kernel with a fused reduction

Vectors up to four cache lines are scored in groups of four that share
every query block load and tail mask; the group's independent
accumulators keep `VPDPBUSD` saturated while one vector's reduction
overlaps with the next group's loads.  Longer vectors keep the
per-vector walk, since the hardware prefetcher streams four interleaved
byte streams far worse than one (measured at the 4-bit width: +10 % at
dim 512, 2× slower at dim 1024).

The per-vector reduction fuses the query bytes before the horizontal
sum — `low + K · high` in i32 lanes, then one tree that widens to i64
at the end — for vectors within a per-width lane bound derived from the
encoding (2040 bytes at 4 bits, 1020 at 2, 255 for the wide 1-bit
query; unbounded for a one-byte query).  A test pins the derivation to
the hand-computed 4-bit value and drives every width to its bound with
the heaviest possible inputs.

Bench: `batch_avx512_vnni` rows in the scan groups.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuerySimd: AVX2 batch kernel; one fused reduction for AVX2 and AVX-512

The AVX2 batch kernel scores vectors one at a time: its loop-carried
chain is a single `vpaddd` per accumulator (the `maddubs → madd`
products hang off the loads), so interleaving vectors only adds
register pressure on the 16 YMM registers — measured 10–15 % slower
with groups of two or four at the 4-bit width.

The AVX2 per-vector reduction now uses the same fused tree as the
AVX-512 one, within the same per-width lane bound; the bound test
drives both kernels.

Bench: `batch_avx2` rows in the scan groups.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuerySimd: copy the tail block in constant-size pieces

The SSE, AVX2 and NEON kernels run their last partial block on a
zero-padded copy of the remaining bytes.  A `len`-byte copy compiles to
a `memcpy` call plus a `memset` for the padding — and the call forces
the accumulators out of their registers around it.  Copy in power-of-
two pieces of constant size instead: `len` is the same for every vector
of a query, so the piece branches predict perfectly.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuerySimd: interleaved NEON batch kernels

The SDOT and plain NEON block loops take `N` vectors at a stride, and
the batch entry points score vectors up to four cache lines in groups
of four — the same policy as the AVX-512 kernel, with the group
threshold carried over from the AVX-512 measurement rather than tuned
on ARM hardware.

Bench: `batch_neon` and `batch_neon_sdot` rows in the scan groups.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

---------

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
2026-09-02 09:39:32 +02:00
Luis Cossío af76cdfa8c [updater] Batch upsert quantized vectors (#10417)
* upsert quantized vectors in batch

* fix bounds
2026-09-01 10:30:06 -04:00
Ivan PleshkovandClaude Fable 5 0f2b02ee11 Turbo4 query simd (#10391)
* TQ SIMD: one backend ladder, resolved once per query

The 2- and 4-bit kernels share the same preference order (AVX-512 VNNI
→ AVX2 → SSE → NEON + SDOT → NEON → scalar), spelled out six times as
chains of `is_x86_feature_detected!` — and `Query{2,4}bitSimd::dotprod`
re-ran its chain for every vector scored.

Move the ladder into one `simd::SimdBackend` enum with a single `detect()`.
The query types resolve it in `new()` and dispatch on the stored value;
the symmetric `score_{2,4}bit_internal*` entry points dispatch on
`SimdBackend::detect()`.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuerySimd: one query layout and scalar reference for every packing width

`Query{1,2,4}bitSimd` are three copies of the same idea — quantize the
query into i8 halves, multiply them against an integer codebook — each
with its own query layout and its own set of SIMD kernels.  Introduce
`simd::query::QuerySimd<PLANES>`, generic over the number of codes per
packed byte (2, 4 or 8), which the three widths will share.

The query halves are stored as planes, one per code position within a
byte: plane `k` entry `j` is the half of query dim `PLANES · j + k`.
That is the order the codes come out of raw data bytes with a shift and
a mask, so a kernel never has to unpack them into dim order.  Planes are
zero-padded to the widest SIMD block, so a partial last block on the
data side multiplies against zeros.

The widths contribute only their integer encoding (`Encoding`: codebook
table, offset, scale and query range); the 1-bit width gets one here —
`{0, 128}` with offset 64 on x86_64, `∓127` on aarch64 — chosen so the
query keeps full i8 halves.  Only the scalar reference exists yet; the
SIMD kernels follow, and the width types switch over once they're in.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuerySimd: AVX-512 VNNI kernel on the query planes

One ZMM of packed codes per block; for each plane the codes are shifted
down by the code width, masked and looked up in the codebook with one
`vpshufb`, then `VPDPBUSD` folds them into the plane's low and high
accumulators.  Two accumulator pairs per vector keep the VNNI latency
off the critical path at every width.  The last partial block is a
masked load whose dead lanes multiply against the planes' zero padding.

The shift count is an immediate, so the shift-by-width helper spells
out the three widths in a `match` — the only place the kernel is not
literally generic over `PLANES`.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuerySimd: NEON SDOT kernel on the query planes

The AVX-512 kernel's shape on 128-bit registers: one `TBL` codebook
lookup per plane, `SDOT` (inline asm — `vdotq_s32` is still unstable)
into two accumulator pairs.  The last partial block runs on a
zero-padded copy of the remaining bytes.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuerySimd: AVX2 kernel on the query planes

One YMM of packed codes per block, one `vpshufb` lookup per plane and
`maddubs → madd` against ones into the same two accumulator pairs as
the VNNI kernel.  The `maddubs` pair sums stay inside i16 by the
per-width query bounds.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuerySimd: SSE and plain NEON kernels on the query planes

The 128-bit forms of the AVX2 and SDOT kernels: `maddubs → madd` on
XMM, `vmull_s8 → vpadalq_s16` on NEON without `dotprod`.  Every backend
of the shared query type now has its kernel.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Query4bitSimd: score through QuerySimd<2>

`Query4bitSimd` becomes an alias of the shared query type; its own
chunk-and-tail query layout and the per-backend kernels built on it go
away, along with the accuracy tests the shared module now runs for
every width.  What stays in `query4bit` is the 4-bit encoding and the
symmetric `score_4bit_internal*` paths.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Query2bitSimd: score through QuerySimd<4>

The 2-bit asymmetric kernels unpacked every 4 packed bytes into 16
centroid bytes through two `pshufb` / `TBL` pair-table lookups and a
zip before a single multiply-accumulate step — on AVX-512 that was four
128-bit unpacks and six lane inserts per pair of `VPDPBUSD`.  On the
query planes the same 16 codes cost one shift, one mask and one lookup
per plane, straight from a full-width load.

`Query2bitSimd` becomes an alias of the shared query type; its chunk
layout and per-backend kernels go away.  The pair-table unpack stays
for the symmetric `score_2bit_internal*` paths.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Query1bitSimd: score through QuerySimd<8>

The 1-bit asymmetric kernels bit-plane-transposed the query and scored
`Σ_b 2^b · popcount(data AND plane_b)` per 16-byte block — eight
AND + popcount + add steps on XMM (even with AVX-512, through the VL
forms) and a `BITS`-deep accumulator array.  On the query planes a
sign bit is just a one-bit code: shift, mask, a two-entry codebook
lookup and the same multiply-accumulate as the wider widths, on full
256-/512-bit registers.

`Query1bitSimd` becomes an alias of the shared query type.  Its query
width was a const parameter (8 bits by default, 16 for TQ+ through the
`Bits1Wide` variant); the shared encoding always carries 16-bit halves,
so the variant and the TQ+ special case go away.  The popcount kernels
stay for the symmetric `score_1bit_internal`, where XOR + popcount is
the right tool.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Bench: cold per-vector rows for every width

`query{4,2,1}bit_dotprod_cold` from one generic body — scalar reference,
public `dotprod` and each backend — so the widths can be compared on one
host.  `TURBO_SIMD_DIMS` narrows or widens the dims of a run and
`TURBO_SIMD_POOL_KB` shrinks the pool to L1 for hot-kernel numbers.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* QuerySimd: query bytes as a parameter; 8-bit queries for the 1-bit width

The shared kernels always carried two query bytes (a ~16-bit query),
which cost the 1-bit width its 8-bit-query speed: the old bit-plane
kernel scored a 16-byte vector in 4.9 ns hot (34.7 ns cold) against
9.1 ns (54.7 ns) through the planes, the difference being the second
byte's multiply-accumulates on a vector that fills a quarter of one
block.  Above 512 dims the planes win either way.

Make the number of query bytes a parameter: `QuerySimd<PLANES,
QUERY_BYTES>` with one plane per query byte and code position, and one
accumulator pair per query byte.  A one-byte query is scaled to the
range of a single byte, `RADIX / 2 − 1`.  `Query1bitSimd` is the
one-byte instance — at parity with the old kernel at small dims (cold
36.9 / 38.3 / 38.4 ns at d = 128 / 256 / 512) and 1.8× faster at 1536
(68 vs 121 ns) — and `Query1bitWideSimd` the two-byte one, which TQ+
selects through the `Bits1Wide` variant as before.  The 2- and 4-bit
widths keep two bytes.

Bench: `query1bit_wide_dotprod_cold` and a `query1bit_wide` row next
to BQ.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

---------

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
2026-09-01 14:15:17 +02:00
Ivan PleshkovandClaude Fable 5 24aa7faaeb Fix TurboQuantizer::quantized_size double-padding for Bits1_5 (#10390)
`quantized_size_for` pads its `dim` argument, so passing the already
padded `self.padded_dim` applied the x1.5 expansion of `Bits1_5` a
second time.  Nothing on disk depends on the value for that width: the
quantization path sizes its records with `quantized_size_for` from the
raw dim, and the Turbo datatype storages, which do use
`quantized_size()` as their record size, are fixed at Bits4, where the
padding is idempotent.  The wrong value only over-reserved the
`quantize` output buffer.

Compute the packed size from `padded_dim` directly and cover `Bits1_5`
in the byte-length test.

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
2026-08-31 09:17:11 +02:00
Anton Karpov b88becd3b5 docs: parameter names in doc comments that the signatures do not have (#10290)
11 names across 7 files. Renames that did not reach the comment above them
(further_searches for further_results, query_context for segment_query_context,
block_ranges for local_block_ranges, op for operation twice, request for
requests, max_threads for max_kmeans_threads), and 3 arguments that were
removed from a signature and left documented (is_on_disk, collection_params,
search_runtime_handle with timeout).

Documentation only, no behaviour change.
2026-08-21 15:36:27 +02:00
Arnaud GourlayandClaude Opus 5 0e33974697 Replace permutation_iterator with rand's index sampler (#10217)
`cargo audit` flags rand 0.7.3 as unsound (RUSTSEC-2026-0097), and
permutation_iterator 0.1.2 is its sole importer. The crate is
unmaintained, so the finding is permanent for as long as we depend on
it.

Everything we used it for is "pick k distinct random indices out of n",
which is exactly `rand::seq::index::sample` from the workspace rand.
Switch the three src call sites and the two benches over, and drop the
dependency. 11 crates leave Cargo.lock.

Also fix a comment in quantile.rs claiming the permutation was
deterministic per count: the old crate keyed itself from thread_rng on
every call, so it never was.

Co-authored-by: Claude Opus 5 (1M context) <noreply@anthropic.com>
2026-08-14 13:51:08 +02:00
liam 362131f926 fix(quantization): skip cc compile when no SIMD sources were added (#10209) 2026-08-13 01:30:56 +00:00
1a8c9a6397 [UpdateOnly] implement the appendable quantized-vector overlay (dense, Binary/Turbo) (#10161)
* [UpdateOnly] implement the appendable quantized-vector overlay (dense, Binary/Turbo)

Appendable/plain segments can carry live quantized vectors today: PlainVectorIndex::
update_vector calls quantized_vectors.upsert_vector alongside the raw vector on every
insert (lib/segment/src/index/plain_vector_index/lifecycle.rs), auto-created for a
fresh segment when appendable_quantization is on and the method supports it
(QuantizationConfig::supports_appendable — Binary and Turbo only; Scalar/Product are
policy-gated off regardless of storage backend). The update-only vector-storage stack
(this PR's base) had no equivalent: UpdateOnlyVectorStorage::open never read
quantization_config, and nothing under vector_storage/*/update_only/ mentioned
quantization at all — a segment configured with quantization would silently lose it
end-to-end once written through this path.

This adds UpdateOnlyQuantizedVectors, mirroring QuantizedVectors' auto-create/reopen
behavior but scoped to dense (single-vector) Binary/Turbo — the two methods that
support incremental appends, matching current capability exactly (multivector support
is a follow-up: it needs its own append-only offsets storage, mirroring
MultivectorOffsetsStorageChunked the same way this mirrors QuantizedChunkedStorage).

The only new machinery is UpdateOnlyQuantizedChunkedStorage, an EncodedStorage backed
by UpdateOnlyChunkedVectors (append-only, S: UniversalAppend) instead of
ChunkedVectors' positional writes (S: UniversalWrite) — everything else reuses the
quantization crate's EncodedVectorsBin::encode/load and EncodedVectorsTQ::encode/load
completely unchanged, since both are already generic over the storage backend. It
writes files in the exact layout QuantizedChunkedStorage reads, so a promoted segment's
quantized data reads through the existing, unmodified reader with no new reading code.
UpdateOnlyChunkedVectors gains one addition: a `get` method to read back a single
vector, needed because EncodedVectors::load validates the storage's vector size by
reading vector 0 (skipped when the store is still empty).

Verified: the update-only writer's persisted bytes, read back through the standard
(non-update-only) QuantizedChunkedStorage + EncodedVectorsBin/TQ::load, match a
RAM-backed reference fed the same vectors one at a time through upsert_vector,
byte-for-byte, for both Binary and Turbo.

Co-Authored-By: Claude Sonnet 5 <noreply@anthropic.com>

* [UpdateOnly] fix quantized reopen: resume writing shouldn't validate stored reads

The previous commit made reopening a non-empty quantized overlay panic
(EncodedVectorsBin/TQ::load validates a non-empty store by reading its
vector 0, which UpdateOnlyQuantizedChunkedStorage's write-only design
cannot serve) and worked around it with a redundant pre-check plus a
todo!(), narrowing the tests to single-session-only writes.

Both of those were the wrong fix. A writer resuming appends doesn't need
`load`'s read-and-validate — it only needs the fitted metadata (encoding,
stats) to keep encoding consistently, and that invariant already holds by
construction: every vector this writer ever encodes is sized from the same
`quantized_vector_size` `load` and the new path both read. Added
`EncodedVectorsBin`/`EncodedVectorsTQ::reopen_for_write` to the
quantization crate — identical to `load` minus the validating read — and
switched `open_existing` to it. `UpdateOnlyQuantizedChunkedStorage` stays
write-only as originally designed; no new read capability, no pre-check,
no todo.

Tests restored to the original two-writer split (write half, drop, reopen,
write the rest), now genuinely exercising resume-with-data instead of
avoiding it, and still passing byte-for-byte against the reference.

Co-Authored-By: Claude Sonnet 5 <noreply@anthropic.com>

* [UpdateOnly] split EncodedStorage into EncodedStorageWrite + EncodedStorage

A write-only storage (the update-only quantized overlay) had to fake a
full EncodedStorage impl with unreachable!() read stubs just to satisfy
EncodedVectorsBin/TQ's generic bound. Split the trait so a write-only
backend only needs to implement EncodedStorageWrite; EncodedStorage adds
the read methods on top. The overlay now implements EncodedStorageWrite
alone — no panicking stand-ins for methods that don't exist.

* [UpdateOnly] remove UpdateOnlyQuantizedVectors::create

Nothing in this stack builds the first appendable segment of a
collection yet (that's still a todo!() in edge/src/update_only), so
create() had no real caller and open() had to guess from file absence
whether to invoke it. open() now only reopens an overlay create()
already persisted; the bootstrap logic moved into tests.rs as a
private fixture helper, since tests still need it to build fixtures.

* [UpdateOnly] fix CI: codespell typo and lint dead-code on unwired write path

codespell flagged "implementors" (wants "implementers") in two doc
comments. Separately, CI's lint job runs clippy without --all-targets,
so the update-only quantized write path — genuinely unreachable from
any non-test code until #10152 wires it into a segment — trips
-D warnings dead-code. Scope #![allow(dead_code)] to the two files
that are only exercised by their own tests today, and allow the now
test-only UpdateOnlyQuantizedChunkedStorageBuilder re-export.

* [UpdateOnly] fix ast-grep: use expect(dead_code) instead of allow

* fix CI: remove unused EncodedStorageWrite import in gpu vector storage

Left over from splitting EncodedStorage into EncodedStorageWrite +
EncodedStorage; only caught under --all-features since gpu is gated
behind a feature flag.

---------

Co-authored-by: Claude Sonnet 5 <noreply@anthropic.com>
Co-authored-by: qdrant-cloud-bot <111755117+qdrant-cloud-bot@users.noreply.github.com>
2026-08-11 10:57:47 +02:00
xzfc 75385df69f Remove dead code (#10030)
* Remove dead code

* Remove unused dependencies

* `allow(dead_code)` -> `expect(dead_code)`

* ast-grep: rule-tests/*-test.yml => tests/*-test.yml

For brevity.

* ast-grep: forbid allow(dead_code)
2026-07-30 20:59:31 +00:00
Ivan PleshkovandClaude Opus 5 e06bee8d46 Fix TQ quantized vector layout alignment (#10005)
`EncodedVectorsTQ::layout()` declared `align_of::<f32>()`, but the size it
reports is the packed dimensions plus a 4-byte-multiple extras trailer, which
is not a multiple of 4 for three quarters of all dimensions (e.g. dim=756,
Bits4: 378 + 4 = 382 bytes).

The claim was never true — the encoded storage packs vectors at
`id * quantized_vector_size` with no per-vector padding — and nothing relies on
it: packed dimensions are read through unaligned SIMD loads (`loadu` / `vld1`)
and the extras trailer through `f32::from_le_bytes` on a byte slice.

It is also actively harmful. Inline HNSW storage packs link vectors
back-to-back using this layout and rejects one whose size is not a multiple of
its alignment, so building an index with `inline_storage` enabled fails for
those dimensions — and retries forever as an optimization crashloop.

Use `align_of::<u8>()`, matching scalar and product quantization. Old links
files stay readable: both layouts are persisted in the file header and the
reader takes size and alignment from there, never from the live quantizer.

Add a test covering the `size % align == 0` invariant across awkward
dimensions, bit widths, distances and modes — `layout()` had no coverage, which
is why the mismatch went unnoticed on the multiple-of-32 dimensions everyone
uses in practice.

Co-authored-by: Claude Opus 5 (1M context) <noreply@anthropic.com>
2026-07-28 11:34:11 +02:00
xzfc 7e99cdd86a UioResult (#9933) 2026-07-22 14:34:33 +00:00
Luis CossíoandClaude Fable 5 fca67539f4 Mandatory batch impls (#9935)
* make `EncodedStorage::for_each_batch` mandatory

* make `DenseVectorStorageRead::for_each_in_dense_batch` mandatory

* make `DenseTQVectorStorage::for_each_in_dense_batch` mandatory

* make `DenseTQVectorStorage::read_dense_tq_bytes` mandatory

* make `QueryScorer::score_stored_batch` mandatory

...and implement for tq multivectors

* [AI] make `IdTrackerRead::internal_versions_batch` mandatory

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* [AI] make `IdTrackerRead::external_ids_batch` mandatory

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* [AI] make `DiskMappingsSource::resolve_internal_batch` mandatory

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

---------

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
2026-07-22 09:21:48 -04:00
Arnaud GourlayandClaude Fable 5 c196d2eb1a Benches: use SmallRng instead of ChaCha12-based generators (#9887)
* Benches: use SmallRng instead of ChaCha12-based generators

All benchmarks used StdRng or rand::rng() (ThreadRng), both backed by the
ChaCha12 block cipher in rand 0.10. Benchmarks do not need crypto-strength
randomness, and several draw random values inside the timed closure, so
cipher work was included in the measurement itself.

Switch every bench target to SmallRng (Xoshiro256++), and key the HNSW
graph cache and sparse index cache by RNG algorithm so stale caches built
from the old generator are not reused against newly generated vectors.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Benches: replace free-function rand::random with local SmallRng

Addresses review: rand::random draws from the thread RNG (ChaCha12),
including inside the timed loop of the pq score benchmark.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

---------

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
2026-07-17 17:22:27 +02:00
Arnaud Gourlay 0d2f5c7e85 Miscellaneous cleanups (#9849) 2026-07-15 13:44:51 +02:00
Arnaud GourlayandClaude Fable 5 d4e0f354da Single inverse rotation in TurboQuant symmetric L1 scoring (#9835)
The inverse Hadamard rotation is linear, so it distributes over
subtraction: |R'd1 - R'd2| = |R'(d1 - d2)|. Subtract the dequantized
vectors in rotated space and inverse-rotate the difference once,
instead of inverse-rotating both vectors per score call.

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
2026-07-14 15:42:10 +02:00
Arnaud GourlayandClaude Fable 5 27672b8011 [AI] Speed up TurboQuant Hadamard rotation by ~3x (#9660)
* Speed up TurboQuant Hadamard rotation by ~3x

Profiling apply_inverse showed 44% of cycles in the Fisher-Yates swap
replay (serial LCG plus a 64-bit modulo per element), not in the WHT.

- Materialize the shared permutations as index maps once in
  HadamardRotation::new; each apply-time permutation becomes a flat
  gather pass ping-ponging between the vector and a thread-local
  scratch buffer. The replay path stays as the cfg(test) parity oracle.
- Fuse consecutive WHT outer butterfly stage pairs (h, 2h) into one
  pass over the array (AVX2), halving memory traffic for h >= 16.
- Fuse the normalization multiply into the transform's final-stage
  stores (wht_dispatch_scaled), removing the separate normalize pass.

Output is bit-identical on all paths: pinned by the existing
struct-vs-replay and SIMD-vs-scalar bit-equal tests plus a new
wht_dispatch_scaled parity test. NEON is unchanged.

Criterion hadamard bench (Zen 5): apply 2.8-3.7x faster across dims
128-4096 (1024: 6.98us -> 2.03us), apply_inverse 2.8-3.5x
(1024: 6.10us -> 2.05us).

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Address review: harden gather contract, small-dim parity, grow-only scratch

- Mark gather_permuted as unsafe fn: the get_unchecked justification relied
  on a caller invariant (map indices in range) that the signature did not
  surface; document it as a # Safety contract instead.
- Add dims 5 and 50 to static_rotation_matches_struct_and_roundtrips to pin
  the map path against the replay oracle on degenerate chunk splits.
- Make the thread-local gather scratch grow-only, so threads alternating
  between dims no longer shrink and re-zero the buffer on every call.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Hard-assert input length in apply/apply_inverse

Fail fast at the public boundary: with debug_assert only, a release
build would WHT-normalize a wrong-length slice before the gather's
hard length asserts panic. Flagged by CodeRabbit.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

---------

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
2026-07-14 11:39:12 +02:00
273bbe1729 Fix use-after-free in u8-quantized vector reads on owning storages (#9801)
* Fix use-after-free in u8-quantized vector reads on owning storages

EncodedVectorsU8::get_vec_ptr extracted a raw pointer from the buffer
returned by EncodedStorage::get_vector_data and dropped the buffer
before the pointer was dereferenced. For storages returning
Cow::Borrowed (mmap) this happened to be sound, but for storages that
return Cow::Owned (disk-cache misses, uring backends) every user of
get_vec_ptr read freed memory: the internal SIMD scorers,
encode_internal_vector, and get_quantized_vector_offset_and_code, which
exported the dangling buffer through a safe &[u8].

Make parse_vec_data return the code as a slice borrowing the input, so
the borrow checker forces callers to keep the storage buffer alive
while reading it, and drop get_vec_ptr entirely. This matches how
encoded_vectors_binary already binds the buffers at its scoring sites.
Change get_quantized_vector_offset_and_code to return the code as a
sub-view of the buffer Cow itself, and adapt its GPU caller.

The new regression test drives these paths through a storage that
always returns owned buffers; under Miri it reproduces the
use-after-free on the previous code and passes with this fix.

Fixes #9799

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* additional assertion

---------

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
Co-authored-by: Ivan Pleshkov <pleshkov.ivan@gmail.com>
2026-07-13 16:40:25 +02:00
Arnaud GourlayandClaude Fable 5 cad112bb1c Fix Clippy 1.97 (#9716)
* Remove from_iter_instead_of_collect from workspace lints

The lint was removed from clippy (beta) and now triggers
renamed_and_removed_lints warnings in every crate.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Fix clippy::chunks_exact_to_as_chunks

Replace chunks_exact with a constant chunk size by as_chunks.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Fix clippy::needless_late_init

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Fix clippy::useless_borrows_in_formatting

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Fix clippy::uninlined_format_args

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Fix clippy::for_kv_map

Iterate map values directly instead of discarding keys.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Allow clippy::result_large_err on QueueProxyShard::new_from_version

The Err variant intentionally hands the LocalShard back to the caller.
Same pattern as the existing allow on ForwardProxyShard::new.

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

* Allow clippy::result_unit_err on wait_for_consensus_commit

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

---------

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
2026-07-07 14:53:15 +02:00
Ivan Pleshkov 9c1b0ed9ee [TQDT] Quantization over tq strorage (#9507)
* quantization over tq strorage

are you happy fmt

are you happy clippy

rotation refactor

rotation refactor

fix tests

better test name

review remarks

dont rotate query in raw scorer

tq sources revert

quantized_scoring_datatype revert

simplify

are you happy fmt

* remove old comments

* review remarks
2026-07-01 15:03:37 +02:00
8c8a72d120 Remove read_multi_iter to fix macOS linker symbol overflow (#9643)
* remove unused iter_offsets

* Replace MultivectorOffsetsStorage::iter_offsets with callback-based for_each_offset

First step of removing the iterator-returning read API (whose deep,
composable generic types blow up mangled symbol size). Convert the
offsets read from an iterator to a callback the caller pushes into:

- trait method iter_offsets -> for_each_offset(ids, FnMut(usize, MultivectorOffset))
  returning common::universal_io::Result<()>
- Mmap impl now uses the callback read_batch (drops one read_iter use)
- Ram / Chunked impls push into the callback; Chunked still goes through
  iter_vectors for now (converted in a later step)
- the single caller (for_each_in_multi_batch) passes a closure

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Implement read_batch directly on ReadPipeline, not via read_iter

read_batch now drives the pipeline itself (refill-then-wait loop, like
read_multi_iter) and invokes the callback per result, instead of
consuming the iterator returned by read_iter. A step toward removing the
iterator-returning read API.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Remove the ReadMulti RPC from the StorageRead gRPC service

ReadMulti was the only real consumer of UniversalRead::read_multi (which
itself relies on read_multi_iter). Removing the RPC end-to-end clears the
path to dropping that read API. StorageReadService keeps all its other
RPCs (ListFiles, FileExists, FileLength, ReadBytes, ReadBytesStream,
ReadWhole, ReadBatch).

- proto: drop `rpc ReadMulti` + ReadMulti{Entry,Request,Response}
- regenerated lib/api + uio-client generated code; drop ReadMulti
  validation rules in lib/api/build.rs
- tonic: delete the read_multi handler + its 2 tests
- uio-client: delete Client::read_multi, the mock-server impl, and 2 tests

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Remove UniversalRead::read_multi

Its only real consumer was the StorageRead ReadMulti gRPC handler (removed
in the previous commit); the two wrapper forwarders had no callers. Drop
the trait method and both forwarders (typed/read_only), and remove the
io_uring test that only existed to compare read_multi vs read_multi_iter
(read_multi_iter stays covered by the other tests). Another step toward
removing the iterator-returning read API.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Drive ReadPipeline directly in gridstore read_from_pages

Replace the read_multi_iter call in Pages::read_from_pages with a direct
pipeline loop (refill-then-drain), scheduling each multi-page read on its
own page file. Behavior unchanged; another step toward removing the
iterator-returning read_multi_iter.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Drive ReadPipeline directly in gridstore read_batch_from_pages

Replace the second read_multi_iter call (in Pages::read_batch_from_pages)
with a direct pipeline loop, scheduling each (ReadMeta, page, range) on its
own page file and propagating errors via GridstoreError. Single/multi-page
buffering and out-of-order reassembly are unchanged. No more read_multi_iter
in gridstore.

Measured overhead on warm mmap (both paths are zero-copy borrows): ~0.3 ns
per read of fixed control cost, flat across read sizes — well under 0.1% of
a real payload read.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Drive ReadPipeline directly in on-disk postings with_posting_views

Replace read_iter in OnDiskPostings::with_posting_views with a direct
pipeline loop. wait_bytemuck yields a file-borrowed Cow, so postings are
still stored zero-copy in raw_postings (a read_batch swap would have forced
an owned copy of every posting list per query on the mmap backend).

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Read on-disk posting headers via read_batch, drop the HeadersBatch iterator

headers_iter now reads headers with the callback read_batch API: each header
is parsed (copied) out of the read bytes, so nothing borrows the file past the
read — no pipeline needed. Since the read is now eager, HeadersBatch holds the
collected Vec<HeaderResult> directly instead of a Box<dyn Iterator>, dropping
the boxing, the dynamic dispatch, and the struct's lifetime parameter.
with_posting_views takes the Vec and still pipelines the posting reads.

Removes the last read_iter use in this file.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Use read_batch in simple_disk_cache populate_from

populate_from reads one byte per block only to fault blocks into the local
cache, discarding the bytes — a no-op-callback read_batch fits exactly. Drives
the same DiskCachePipeline as before; one less read_iter caller.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Implement read_iter directly on ReadPipeline, not via read_multi_iter

read_iter now drives the pipeline itself (refill-then-wait loop, mirroring
read_bytes_iter) instead of mapping its ranges onto self and calling
read_multi_iter. Same signature and iterator contract, so all callers are
unchanged. Leaves iter_vectors as read_multi_iter's only remaining caller.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Revert "Drive ReadPipeline directly in on-disk postings with_posting_views"

This reverts commit c02c41f17b.

* Add callback-based for_each_vector next to iter_vectors

for_each_vector drives the ReadPipeline directly across chunk files and
invokes a fallible callback per flattened multi-vector, returning
OperationResult, instead of returning an iterator built on read_multi_iter.
Callers will migrate onto it so iter_vectors (read_multi_iter's last caller)
can be removed.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Make dense for_each_in_batch / for_each_in_dense_batch fallible

Thread OperationResult up the dense batch-read path so io_uring read
errors propagate instead of being .expect()ed deep inside the storage.
The for_each_in_dense_batch scorer path (custom/metric query scorers)
now carries the Result to the infallible score() boundary where it is
.expect()ed; read_vectors keeps its () signature and .expect()s locally.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Convert dense read_vectors to for_each_vector

The read-only and appendable dense storage read_vectors impls drove
ChunkedVectors::iter_vectors directly; switch them to the callback-based
for_each_vector and .expect() the result at the (infallible) read_vectors
boundary. Removes the last dense-path iter_vectors callers.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Convert quantized for_each_offset to for_each_vector + OperationResult

The two chunked MultivectorOffsetsStorage impls drove iter_vectors to read
the offset table; switch them to the callback-based for_each_vector.
for_each_vector returns OperationResult, so upgrade the for_each_offset
trait (and all four impls) from universal_io::Result to OperationResult
(the universal_io -> Operation direction, via ?). No error downgrade.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Convert EncodedStorage/EncodedVectors iter_batch to callback for_each_batch

The two chunked-mmap EncodedStorage impls drove ChunkedVectors::iter_vectors
to back iter_batch. Replace the iterator-returning iter_batch on both the
EncodedStorage and EncodedVectors traits (quantization crate) with a callback
for_each_batch(FnMut(usize, &[u8])), and switch the chunked impls to
for_each_vector. The callback is infallible: the chunked impls .expect() the
read internally, matching iter_vectors' prior panic-on-read-error behavior,
so no OperationError is downgraded. Scorers and the multivector readers adopt
the callback; the accumulating multivector path owns (to_vec) only when it
must buffer across components.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Convert multivector read paths to for_each_vector

The read_only multivector free fn chained two iter_vectors (offsets feeding
vectors) - the recursive iterator nesting behind the worst symbol bloat.
Replace it with a callback for_each_vector that resolves the per-point
offsets into a Vec first, then drives ChunkedVectors::for_each_vector over
the flattened vectors. Migrate both multivector storages' read_vectors and
for_each_in_batch_multi accordingly, .expect()ing at their infallible
boundaries.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Remove read_multi_iter and ChunkedVectors::iter_vectors

With every caller migrated to callback-based for_each_vector/for_each_batch,
delete the last iterator-returning multi-read APIs: ChunkedVectors::iter_vectors
(segment) and the read_multi_iter trait method plus its mmap override, the
TypedStorage/ReadOnly wrapper forwarders, and the two io_uring unit tests.

These deeply-nested monomorphized iterator types (read_multi_iter feeding
read_multi_iter) produced >1 MiB mangled drop_in_place symbols that overflowed
the macOS ld symbol-name limit; the callback rewrite eliminates them.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Pass owned Cow through for_each_vector/for_each_batch to avoid a copy

The callback-based readers handed the callback a borrowed `&[u8]`/`&[T]`,
forcing the quantized multivector batch read to `to_vec()` each sub-vector
into its per-point buffer. But io_uring-like backends already return a freshly
owned buffer per read (`ACow::Owned`), so that was a redundant second copy.

Change `ChunkedVectorsRead::for_each_vector` and the `EncodedStorage` /
`EncodedVectors` `for_each_batch` callbacks to receive `Cow<[..]>` by value.
The buffering path now `into_owned()`s it — a move when the backend returned
owned (the case this path targets), a copy only for a borrowed Cow (mmap),
which never reaches this path. Immediate-use callers (scorers,
score_point_max_similarity, the dense/multivector readers) just deref the Cow;
the dense readers drop their now-redundant `Cow::Borrowed` wraps.

Also clarifies the multivector reader: `SubVectorOwner`/`owners`/
`sub_vector_offsets` naming, docs, and a corrected comment noting the per-point
buffer is what makes regrouping order-independent under out-of-order completion.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Drop the removed ReadMulti JWT access test; fix clippy unwrap_or_default

The ReadMulti StorageRead RPC was removed earlier in this branch, so the
consensus JWT-access test (and its registry entry) for it must go too. Also
switch the multivector buffer's `or_insert_with(SmallVec::new)` to
`or_default()` per clippy::unwrap_or_default.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* Add MmapFile::read_batch

---------

Co-authored-by: Claude Opus 4.8 (1M context) <noreply@anthropic.com>
Co-authored-by: xzfc <xzfcpw@gmail.com>
2026-07-01 14:03:32 +02:00
Ritwij Aryan Parmar 014c158960 Fix scalar L2 quantized score shift (#9518)
* fix scalar L2 quantized score shift

* Format scalar quantization imports

* Fix quantization test clippy
2026-06-23 23:13:18 +02:00
Ivan Pleshkov 47767d19fd unpadded rotations (#9450) 2026-06-12 22:24:23 +02:00
c79ca1bdd1 feat: live_reload for read-only quantized vectors (#9375)
* feat: live_reload for read-only quantized vectors

* Propagate live_reload params through quantized chunked storage

Thread fs, deleted_points, new_points and hw_counter through the full
quantized live_reload chain instead of synthesizing empty deltas and a
disposable hardware counter at the leaf storages.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* fix: use LiveReload trait in chunked mmap

* fix: use LiveReload trait

* fix: compiler errors

---------

Co-authored-by: generall <andrey@vasnetsov.com>
Co-authored-by: Claude Opus 4.8 (1M context) <noreply@anthropic.com>
2026-06-09 16:58:02 +02:00
Jojii 617d44bde8 [TQDT] TurboVectorStorage Implementation (#9331)
* Partial function impls for TurboVectorStorage

* create/load functionality + tests

# Conflicts:
#	lib/segment/src/vector_storage/turbo/mod.rs
#	lib/segment/src/vector_storage/turbo/turbo_encoded_vectors.rs

* Fix vectors are not rotated back

* update_from impl

* Rebase fixes

* Proper update_from for ChunkedMmap + better tests

* Improve test coverage

* Move update_from to new DenseTQVectorStorage trait

* Review remark quantization::DistanceType

* Add populate and clear_cache to TurboVectorStorage and TurboEncodedVectorStorage
2026-06-09 13:53:32 +02:00
Andrey VasnetsovandClaude Opus 4.8 80acb6c55d feat: read-only QuantizedVectors generic over UniversalRead (#9346)
* feat: add read-only QuantizedVectors generic over UniversalRead

Introduce `QuantizedVectorsRead<S>` / `QuantizedVectorStorageRead<S>`, a
read-only counterpart of `QuantizedVectors` organized like
`VectorStorageReadEnum`: generic over the `UniversalRead` backend `S`,
opened from existing on-disk data, with no create/upsert/builder path and
no disk writes.

Highlights:
- Keep both in-RAM (`*Ram`) and read-only mmap (`*Mmap`) variants; drop the
  appendable `*ChunkedMmap` variants (the only mutable ones).
- All bulk reads go through `S`: add `QuantizedRamStorage::from_universal_read`
  and `MultivectorOffsetsStorageRam::open`, and make
  `MultivectorOffsetsStorageMmap` generic over `S` (default `MmapFile`).
- Share scorer construction between the read-write and read-only enums via a
  `QuantizedScorerDispatch` trait, so only the per-variant match is duplicated
  while the datatype/distance and per-query dispatch live once in the builder.

Tested: read-only vs read-write scorer parity (scalar/binary/product, single
and multivector, RAM and mmap), covering both `raw_scorer` and
`raw_internal_scorer`.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* refactor: route quantized RAM loading through UniversalRead

Follow-up to the read-only quantized work, removing direct-filesystem reads
and load-path duplication:

- `EncodedVectors{U8,PQ,Bin,TQ}::load` now take a `UniversalRead` filesystem
  and read metadata via `read_json_via` instead of `fs::read_to_string` —
  every read flows through universal IO. Single-vector `load` returns
  `common::universal_io::Result`; `validate_storage_vector_size` stays
  `std::io::Result`.
- Add `common::universal_io::OneshotFile<S>`: a thin RAII wrapper over any
  `UniversalRead` handle that evicts the data from cache via `clear_ram_cache`
  on drop (the universal-IO counterpart of `fs::OneshotFile`).
- Collapse `QuantizedRamStorage::{from_file, from_universal_read}` into one
  `from_file<S: UniversalRead>`. It reads the whole file in a single access
  (no separate `len()` round-trip — cheaper on S3-like backends) and loads via
  the new `VolatileChunkedVectors::extend`, which inserts one chunk per
  `copy_from_slice` instead of one vector at a time.
- RW callers pass the local `READ_FS` (mmap) backend; the read-only loader
  passes its `S`.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* feat: load appendable (chunked) quantization format read-only

`storage_type` only selects the on-disk layout, so the read-only quantized
storage must be able to load both — the immutable flat format and the
appendable chunked format (produced only by Binary/TurboQuant). Previously the
read-only loader rejected the Mutable layout outright.

- Add `QuantizedChunkedStorageRead<S>` and `MultivectorOffsetsStorageChunkedRead<S>`,
  read-only wrappers over the existing `ChunkedVectorsRead<_, S>` primitive
  (mirrors the dense read-view's chunked read storage). Generic over the
  `UniversalRead` backend, on-disk, no write path.
- Add `BinaryChunked`/`TQChunked` (+ multi) variants to `QuantizedVectorStorageRead`
  and wire them through every accessor and the scorer dispatch.
- Route `storage_type == Mutable` to the chunked read variants in the loader and
  drop the blanket rejection.
- Tests: parametrize the read-only/read-write parity tests over `storage_type`
  and add chunked (Mutable) cases for binary/turbo, single and multivector.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* refactor: split quantized chunked & multivector storage into modules

`quantized_chunked_mmap_storage.rs` and `quantized_multivector_storage.rs` had
grown to hold several unrelated structures, making them hard to navigate.
Convert each into a module (pure code movement, no behavior change):

- quantized_chunked_mmap_storage/{read_write,read_only}.rs — the appendable
  mmap storage + builder vs. the read-only chunked storage.
- quantized_multivector_storage/{mod,offsets}.rs — the core
  `QuantizedMultivectorStorage` + offset traits stay in mod.rs; the four
  `MultivectorOffsetsStorage*` backends move to offsets.rs.

Public paths are unchanged via re-exports.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* refactor: single source of truth for quantized vector size

The on-disk stride (bytes per quantized vector), including the binary
`u8`(multi)/`u128`(single) word-type choice, was independently re-derived in
every load/open site with no compile-time link — the kind of value-level
duplication that silently diverges (it already caused the binary multi
`u8`/`u128` bug).

Extract it into one place:
- `QuantizedVectors::quantized_vector_size(quantization_config, vector_parameters, is_multi)`
  and the `QuantizedVectorsConfig::quantized_vector_size(is_multi)` convenience.

Route every reader through it:
- read-only `open_single`/`open_multi` and the read-write `{scalar,pq,binary,turbo}`
  loaders now hoist `config.quantized_vector_size(is_multi)` once instead of
  recomputing the per-method formula per branch.

The create (write) path keeps its own computation for now; it stays guarded by
`validate_storage_vector_size` and the read-only/read-write parity tests.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* refactor: flat-match quantized loaders on a shared QuantizedStorageKind

The loaders chose the concrete storage variant via a nested
`method × (storage_type / is_ram)` match, re-derived independently in the
read-only and read-write paths — hard to follow and easy to drift.

- Add `QuantizedStorageKind` + `QuantizedVectorsConfig::storage_kind(on_disk)`
  (and `is_ram(on_disk)`): the single place the method × backend decision lives.
- Both loaders now compute the kind once and use a flat 10-arm match:
  - read-only `open_single`/`open_multi`;
  - read-write `load_single`/`load_multi` (consolidating the four per-method
    `{scalar,pq,binary,turbo}/load.rs` files, which are removed).

Both matches are exhaustive over the same enum, so the read and read-write
variants can no longer fall out of sync without a compile error.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* refactor: genericize write-capable chunked quantized storages over Fs

Drop the hardcoded `MmapFile` specialization from `QuantizedChunkedStorage`,
`QuantizedChunkedStorageBuilder`, and `MultivectorOffsetsStorageChunked`. They
now expose the `Fs` backend (defaulting to `MmapFile`) and accept the fs handle
as a parameter, matching the read-only variants.

Introduce a single shared `ReadFile` type alias and `READ_FS` value handle in
the `quantized_vectors` module root, used by the create, load, and storage enum
paths so the local-file backend is named in one place.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* refactor: move is_in_ram_or_mmap match into UniversalKind

Deduplicate the identical kind-to-residency match in the read-only and
write-capable chunked quantized storages by adding
UniversalKind::is_in_ram_or_mmap.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

* fix: adapt TurboVectorStorage to quantized rename + update_from move

Integration fix after rebasing onto dev's TurboQuant work:
- QuantizedChunkedMmapStorage -> QuantizedChunkedStorage<MmapFile>
- update_from moved off the VectorStorage trait onto an inherent method,
  matching the per-kind sub-trait refactor; TurboVectorStorage implements
  neither DenseVectorStorage<T> nor the other kind traits.

Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>

---------

Co-authored-by: Claude Opus 4.8 (1M context) <noreply@anthropic.com>
2026-06-08 12:56:57 +02:00
Tim Visée 4e54bbc94a Fix OOB heap read crash with malicious snapshot (#9268)
* Validate quantized u8 data size on load

* Validate other quantization types

* Add test

* Use fs_err
2026-06-03 15:17:04 +02:00
Arnaud Gourlay abc53d717f Remove unecesssary Clippy allows (#9267) 2026-06-02 16:54:27 +02:00
Jojii ac6b04a69e [TQDT] Static Rotations (#9106)
* static rotations

* Review remarks
2026-05-29 03:19:44 +02:00
Jojii e7d63ed8a4 Refactor TurboQuantizer (#9095)
* Refactor TurboQuant into common/

# Conflicts:
#	lib/common/turboquant/src/math.rs
#	lib/quantization/src/encoded_vectors_tq.rs
#	lib/quantization/tests/integration/test_tq.rs

* [ai] fix docs

* [ai] tighten function visibilities

* Review remark

* include turboquant in amalgamate

* Rebase

* Move common/turboqunat back (but keep the refactor)

* Move vector_stats.rs back
2026-05-28 00:32:08 +02:00
Roman Titov 65dda5ff45 Merge pull request #9034
* Add `iter_batch` method to `EncodedStorage` and `EncodedVectors`

* Remove `EncodedVectors::for_each_in_batch` method

* Add `for_each_in_multi_batch` and `score_vector_max_similarity` metho…

* Refactor `score_stored_batch` for quantized multi-vector scorers...

* Remove `QuantizedMultivectorStorage::score_multi`

* Implement `score_points_batch_mmap` and `score_points_batch_uring`...

* Implement runtime routing between mmap and io_uring batch scoring met…

* review: rename + comments
2026-05-25 17:00:47 +02:00
Andrey VasnetsovandClaude Opus 4.7 82176fd0f8 Fix TurboQuant heap memory under-reporting (#9099)
EncodedVectorsTQ was the only quantizer that did not override the
EncodedVectors::heap_size_bytes() trait method, so it fell back to the
default of 0. For the RAM-backed variants (TQRam/TQRamMulti) this meant
the entire resident quantized dataset was reported as 0 bytes and
misclassified as fully on-disk by the MemoryReporter; the always-resident
quantizer tables (rotation + TQ+ error-correction vectors) and encoding
buffer were also uncounted for every variant.

Make heap_size_bytes() a required trait method (remove the default impl)
so every quantizer must account for its own heap explicitly, then add the
missing TurboQuant implementation: storage backend + quantizer tables +
encoding buffer.

Co-authored-by: Claude Opus 4.7 (1M context) <noreply@anthropic.com>
2026-05-20 09:48:51 +02:00
d0b84ef907 Fix clippy wildcard_enum_match_arm lint in test_tq.rs (#9101)
Replace wildcard match `_` with explicit enum variants
`DistanceType::Dot | DistanceType::L1 | DistanceType::L2` to satisfy
clippy's wildcard_enum_match_arm lint.

Co-authored-by: Cursor Agent <agent@cursor.com>
Co-authored-by: Cursor <cursoragent@cursor.com>
2026-05-20 09:32:21 +02:00
Srimon Danguria 327551dec3 Validate TurboQuant internal scoring (#9080)
* Validate TurboQuant internal scoring

* feat: cargo +nightly fmt --all

* feat: enhance TurboQuant scoring with error correction and mode handling
2026-05-20 01:48:55 +02:00
xzfc faf2714f4b Warn on clippy::wildcard_enum_match_arm (#9096) 2026-05-19 18:14:15 +00:00
Luis Cossío b6744f6353 [UIO] Generic quantized storage (#9059)
* use generic in QuantizedMmapStorage

* rename to QuantizedStorage

* be explicit about S

* rename builder to `QuantizedStorageBuilder`

* rename file to `quantized_storage.rs`
2026-05-15 21:13:19 -04:00
Luis Cossío 757a168167 make quantized_vector_size method static (#9060) 2026-05-15 18:31:23 -04:00
Ivan Pleshkov 46f9fe5721 tq renormalization for l2 (#8989) 2026-05-11 15:02:01 +02:00
Roman Titov 2a47997654 Merge pull request #8791
* Add `EncodedStorage::for_each_in_batch` method

* Implement `for_each_in_batch` method for `QuantizedChunkedMmapStorage`

* Add `EncodedVectors::for_each_in_batch` method

* Add `EncodedVectors::score` method

* Implement `score_stored_batch` for `QuantizedQueryScorer` and `Quanti…

* Remove `TElement` and `TMetric` type parameters from `QuantizedMultiQ…

* Use `QuantizedMultiQueryScorer` when building `raw_internal_scorer`...
2026-05-07 20:20:54 +02:00
Ivan Pleshkov 665b91d587 Apply p square for TQ+ (dont use std+mean) (#8877)
* apply p square

* review remarks and reduce ram

* review remarks

* clean tmp logs

* review remarks

* review remarks

* trigger ci
2026-05-06 09:36:00 +02:00
Jojii 8765e33515 TQ Hadamard SIMD (#8883)
* [ai+manual] hadamard SIMD

* [ai] Neon

* Minor refactor

* fix import
2026-05-04 21:31:27 +02:00
Ivan Pleshkov 452943bb4b TQ+ SIMD (#8851)
* tq plus simd

* review remarks
2026-05-03 14:57:22 +02:00
Ivan PleshkovandClaude Opus 4.7 eb2e3b8416 Tq plus (#8836)
* Add TQ+ ErrorCorrection on top of renorm

Per-coordinate shift+scale fits each rotated, length-rescaled coord onto
the codebook's N(0, 1) grid before quantization. EncodedVectorsTQ::encode
runs a first pass to fit the stats when TQMode::Plus.

Scoring stays correct under renorm's `scaling_factor` framework:
- Asymmetric: precompute_query scales `Q .* D'` and stashes `qm = ⟨Q, M⟩`
  on EncodedQueryTQ; score_precomputed adds qm to raw_dot before applying
  scaling_factor.
- Symmetric: scalar slow path computes `Σ X+_a X+_b D'_i² + xm_a + xm_b
  − ⟨M, M⟩` (xm stored per vector in extras, mm_const cached on
  ErrorCorrection). Result feeds the existing `* v1_scale * v2_scale` arms.
  SIMD reuse for this path is a follow-up.

Storage layout: TQMode::Plus extras are 4 bytes longer (xm appended after
scaling_factor). Zero-vector inputs skip EC application so renorm's
existing zero-norm guard keeps producing score ≈ 0 within tolerance.

VectorStats refactor: streaming `VectorStatsBuilder` so the Plus first
pass can feed Welford with a reused buffer; `build` now takes `dim`
directly and is generic over `T: Into<f64>`.

Integration tests run on both Normal and Plus via rstest cases.

Co-Authored-By: Claude Opus 4.7 (1M context) <noreply@anthropic.com>

* Fix TQ+ recall regression on non-uniform per-coord variance data

Two compounding bugs in the renorm + TQ+ composition:

1. centroid_norm was measured on `X+` centroids, which have chi-squared
   norm distribution across vectors (~10% spread for d=256). renorm
   assumed `cn` should be deterministic `sqrt(d)` (just quantization
   drift), so the per-vector correction ended up amplifying intrinsic
   chi-squared noise into ranking error. Fix: revert EC per coord before
   measuring (`c · D' + M`), matching llama-turbo-quant. The reverted
   centroids approximate `rescaled` which has length `sqrt(d)` exactly
   by construction. dequantize follows the same convention so the
   stored `scaling_factor = l2/cn` round-trips back to the original l2.

2. Asymmetric query path pre-scaled `Q* = R_q · D'` before SIMD encoding.
   The SIMD encoder normalizes by `max(|input|)`, so a query whose coords
   span 5× magnitude (which `R_q · D'` does on real data) loses precision
   on the small-D' coords. Fix: keep `rotated` unscaled, store it as a
   side field on `EncodedQueryTQ`, and use a scalar decode-and-dot path
   for TQ+ (`Σ R_q_i · c_i · D'_i + qm`). SIMD support for this is a
   follow-up.

Catches both via `recall_skewed_data` test on data with 8 spike-variance
input coords. Without the fixes, Bits4 Plus dropped to 0.93 vs Normal
0.98; after the fixes, Plus tracks Normal within 2%.

Co-Authored-By: Claude Opus 4.7 (1M context) <noreply@anthropic.com>

* TQ+ asymmetric: skip the SIMD encoding entirely

The TQ+ asymmetric path doesn't use the SIMD-encoded query — it goes
through `score_precomputed_ec` with `rotated_query`. So building the
SIMD form was wasted work + memory. Make `data` an `Option` and only
populate it for the cases that actually use it (Normal mode any distance,
TQ+ L1 via dequantize fallback).

Co-Authored-By: Claude Opus 4.7 (1M context) <noreply@anthropic.com>

* Revert TQ+ to SIMD scoring path

The whole point of TQ+ is that scoring code-paths stay identical between
Normal and Plus modes — only the query precomputation changes. We
pre-scale `Q* = R_q · D'` so the existing SIMD raw_dot computes
`⟨Q · D', X+⟩` directly, then add `qm` and apply renorm's scaling_factor.

Drops `score_precomputed_ec` and `EncodedQueryTQ::rotated_query`. The
recall regression that motivated the scalar fallback was entirely from
the `compute_centroid_norm` bug (measuring `‖X+‖` instead of `‖rescaled‖`)
fixed in the prior commit; SIMD precision was a red herring.

`recall_skewed_data` confirms: Bits4 Normal=0.984 / Plus=0.978, Bits2
Normal=0.902 / Plus=0.908.

Co-Authored-By: Claude Opus 4.7 (1M context) <noreply@anthropic.com>

* TQ+ Bits1: widen query encoding to 12 bits

For 1-bit storage with TQ+, the per-coord `D' = 1/scale` pre-scaling on
the query can push some coords toward the small end of the SIMD encoder's
integer range (which normalizes by `max(|input|)`). At 8 bits those small
coords lose precision; at 12 bits the rounding error drops ~10× per the
existing `test_query_dotprod_matches_reference` parity test.

`Query1bitSimd` is already generic over BITS so this is just a new
`EncodedQueryTQData::Bits1Wide(Query1bitSimd<12>)` variant + a TQ+/Bits1
dispatch in `precompute_query`. Bits2/Bits4 don't need this — their
storage is fine-grained enough that query precision isn't the bottleneck,
and their SIMD encoders aren't generic over BITS today.

Co-Authored-By: Claude Opus 4.7 (1M context) <noreply@anthropic.com>

* are you happy fmt

* TQ+ Bits1Wide: bump to 16-bit query quantization (kernel max)

12 bits helped on real datasets but not enough — push to the kernel's
ceiling of 16. `Query1bitSimd<BITS>` asserts `BITS ∈ [2, 16]`, so this
is the most precision the existing SIMD path can give us before needing
a wider integer kernel.

Co-Authored-By: Claude Opus 4.7 (1M context) <noreply@anthropic.com>

* TQ+: shift by per-coord median, not mean

For 1-bit storage the codebook boundary sits at 0, so post-shift values
are quantized purely by sign. Median is the sign-balance point of the
distribution; mean isn't (skewed coords pull mean off the median).

On anisotropic embeddings — dbpedia-openai being the reference case —
mean-based shift produced a ~60/40 biased sign distribution per coord,
losing 1-bit's representational capacity. Median-based shift restores
50/50 and matches llama-turbo-quant's behavior. Higher bit-widths are
less sensitive but still benefit; the codebook boundaries still lie at
distribution-percentile-aware positions when the data is centered on
the median.

Median requires per-coord samples in memory, so cap the stats pass at
10K vectors. Estimates converge fast (~√N) — 10K is plenty even for
million-vector indexes. The encoding pass still processes every vector.

Co-Authored-By: Claude Opus 4.7 (1M context) <noreply@anthropic.com>

* TQ+ Bits1Wide: revert to 12-bit query quantization

The recall regression on anisotropic data was the mean-vs-median shift,
not query precision. 12 bits is enough headroom for the per-coord D'
pre-scaling and avoids the extra storage of the 16-bit form.

Co-Authored-By: Claude Opus 4.7 (1M context) <noreply@anthropic.com>

* TQ+ Bits1Wide: bump back to 16-bit query quantization

12 bits helped a bit but not enough on the real dataset. Bump to the
kernel's ceiling. If 16 still isn't enough, the next step is checking
whether the gap is real (re-measure llama branch) before widening the
SIMD integer kernel itself.

Co-Authored-By: Claude Opus 4.7 (1M context) <noreply@anthropic.com>

* are you happy clippy

* use mean

* 1bit error correction

* review remarks

* are you happy clippy

* review remarks

---------

Co-authored-by: Claude Opus 4.7 (1M context) <noreply@anthropic.com>
2026-04-30 13:39:05 +02:00
Jojii 6c7442d210 [ai + manual] Add renorm (#8837) 2026-04-29 16:16:23 +02:00