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>
The test asserts live WAL contents after applied updates, but the default
fixture uses flush_interval_sec = 0 so the flush worker can truncate
earlier rewritten records before the assertion runs.
Fixes#10423
* Add optional dial9 Tokio telemetry behind a `dial9` feature
Integrate dial9 so storage runtimes can emit production-friendly Tokio
traces. Recording is off unless the crate is built with `--features dial9`
and DIAL9_ENABLED=true is set at runtime; with the feature off, runtime
construction is byte-for-byte unchanged.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01Y7X6MkjY3P7wpP2MfHdTDY
* Enable dial9 CPU and schedule profiling
Turn on cpu-profiling and sched events behind the same `dial9` feature,
add the DIAL9_CPU_* / DIAL9_SCHEDULE_* env knobs, and document the frame
pointer rustflags the stack unwinder needs.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01Y7X6MkjY3P7wpP2MfHdTDY
* Harden dial9 env parsing and the writer-failure path
- Reset Cargo.lock to the branch point and re-resolve, so the diff is
additive instead of re-resolving unrelated packages. This drops the
heck 0.5.0 -> 0.4.1 downgrade, which sat in the default build graph and
would have changed proto codegen identifier casing. The remaining
non-additive entry, toml_parser 1.0.9 -> 1.1.3, is forced by
proc-macro-crate via dial9-trace-format-derive.
- Parse DIAL9_* booleans the way dial9 does, accepting 1/y/yes/on and
0/n/no/off and warning on anything else. `str::parse::<bool>` took only
exact lowercase true/false, so DIAL9_CPU_PROFILE_ENABLED=0 silently left
99 Hz sampling on and DIAL9_ENABLED=1 silently left recording off.
- Require the numeric knobs to be positive. A zero disk budget made dial9
evict everything and stop recording within seconds while the log still
reported telemetry enabled.
- Treat a set-but-empty DIAL9_TRACE_DIR as unset. It skipped the /tmp
fallback and wrote up to the full budget into the working directory,
which is /qdrant next to storage/ in the official image.
- Return a disabled guard as soon as the trace writer fails, before
with_cpu_profiling and with_sched_events run. Those start their profilers
eagerly, opening a perf event per thread and installing a process-global
signal handler that build() would then discard.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01Y7X6MkjY3P7wpP2MfHdTDY
* Correct the dial9 docs and give them their own section
- `--cfg tokio_unstable` is required for any task data at all, not merely
for fuller coverage: dial9's poll, spawn and terminate hooks are all
`#[cfg(tokio_unstable)]`, and nothing in the repo sets the flag. Without
it there is no task timeline and DIAL9_TASK_TRACKING_ENABLED does nothing.
- Document `-C debuginfo=2`. `[profile.perf]` inherits `release` and sets no
`debug` key, so the documented build symbolized off the ELF symtab with
inlined callees collapsed and no file or line, unlike `[profile.bench]`
which sets `debug = true` for this reason.
- Move the dial9 material out from between the feature list and the prose
that belongs to it. Those paragraphs describe `tracing` instrumentation
and read as dial9's when the example is wedged in front of them, which
points readers at `#[tracing::instrument]` for a tool that records Tokio
runtime events and no tracing spans.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01Y7X6MkjY3P7wpP2MfHdTDY
* Use cfg_select!
---------
Co-authored-by: Claude Opus 5 (1M context) <noreply@anthropic.com>
Co-authored-by: timvisee <tim@visee.me>
* 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>
* 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>
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.
Remove the root CONTRIBUTING.md, which only contained a pointer to the
full guide plus a dev-branch notice that is already covered (more
thoroughly) at the top of docs/CONTRIBUTING.md. This leaves a single
source of truth for contribution guidelines.
Co-authored-by: Cursor <cursoragent@cursor.com>
* fix: reject mismatched dense dims in recommend average
Stop silently truncating oversized negative examples during
average_vector merge. Validate dense dimensions within each example
group and between positive/negative averages before zip-merge.
Fixes#10369
* Simplify: keep only the merge-time dimension check
The zip truncation in merge_positive_and_negative_avg is the only place
an oversized negative can silently pass the downstream dimension check;
within-group mismatches already grow the average to the max length and
fail the segment-entry check.
Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>
* style(query): make recommendation conversion explicit
* test: assert recommendation dimension errors
Issue: #10369
Make the regression test verify the exact WrongVectorDimension payload for mismatched recommendation vectors.
---------
Co-authored-by: qdrant-cloud-bot <111755117+qdrant-cloud-bot@users.noreply.github.com>
Co-authored-by: generall <andrey@vasnetsov.com>
Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
* 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>
* feat: create appendable segments when the write target is optimizing or the shard is empty
* review: SegmentManifestState::is_writable, caller-supplied temp dir, uuid from token
- `SegmentManifestState::is_writable` with a full match replaces the ad-hoc
`matches!` in the manifest enumerator.
- `ListedSegment` is destructured in `open` so every field is accounted for.
- `create_appendable_from` is test-only; `create_appendable` is the API.
- `create_appendable` builds the scratch segment in a caller-supplied local
`temp_path` (conventionally `<shard>/temp_segments`) instead of the system
temp dir, and takes the uuid from the build token instead of parsing the path.
Upload speed of `copy_dir_via` is tracked in #10433.
Co-Authored-By: Claude Fable 5.1 <noreply@anthropic.com>
---------
Co-authored-by: generall <andrey@vasnetsov.com>
Co-authored-by: Claude Fable 5.1 <noreply@anthropic.com>
* perf: skip retrieval in scroll when no payload or vectors are requested
Every scroll variant went through SegmentsSearcher::retrieve to build its
records, even when neither payload nor vectors were asked for. That is a
has_point lookup per id per segment plus a version and id resolution per
hit, only to yield records holding nothing but the id. The universal query
API always scrolls this way and fetches payload separately afterwards.
Build the bare records from the ids directly in that case. The retrieve
could only have dropped ids deleted in between, which the update lock held
across the scroll rules out.
* perf: fetch payload and vectors in the leaf of plain query requests
A query without prefetches and without rescoring is served by a single
leaf search or scroll whose result is returned as is. The planner still
built that leaf without payload or vectors and filled them in afterwards
through SegmentsSearcher::retrieve, which resolves every result id in
every segment again: the same cost #10312 removed from the search API,
paid once more at the end of each query.
Let the leaf carry the requested payload and vectors instead, so the
segment attaches them to the results it already holds by offset, and
clear the root plan so the fill step is skipped. Prefetch leaves and
rescored roots (MMR) are unchanged. As with the search API, this fetches
payload for each segment's candidates rather than for the merged top
`limit` alone.
* Use new_empty function
* fix: fetch payload and vectors in scroll leaves only (#10384)
A search leaf hydrates every segment's local top-k before merging, so
`with_payload` there multiplies payload I/O by the segment count — the
regression #6279 fixed and `test_payload_io_read_is_within_limit[query]`
guards. Scroll leaves retrieve once for the merged page, so they keep
fetching directly; search leaves stay bare and the root plan retrieves
for the final result.
Claude-Session: https://claude.ai/code/session_01SUWh5PqUeSefrUqUwXxU3E
Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
---------
Co-authored-by: Andrey Vasnetsov <andrey@vasnetsov.com>
Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
* 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>
* Split async IO into extension traits; only async-capable backends implement them
Move `read_bytes_async` / `open_async` off the universal `UniversalRead` /
`UniversalReadFs` traits into dedicated extension traits, `UniversalReadAsync`
and `UniversalReadFsAsync` (traits/async_io.rs). Only backends with a genuine
async story implement them — the blob family, the disk caches layered over it,
and a trivial ready-impl for mmap (tests and the mmap lookup path) — each in a
dedicated async_io.rs next to its sync impl.
`CachedFs` now requires its inner filesystem to be `UniversalReadFsAsync`; the
requirement reaches segment code through one supertrait bound on
`UniversalReadExt`. io_uring implements no async surface anymore: the
tokio_uring bridge thread, its tests, the musl-gated tokio-uring dependency,
and the `IoUringFile` read-only-segment wiring (`UniversalReadExt` impl and
the *RoIoUring condition-checker variants) are deleted — io_uring is not a
read-only-segment backend.
The payoff for live reload: `CachedFs::resolve_prefetched` awaits every parked
prefetch, and the edge refresh flow now runs preload -> resolve -> reload, so
the per-segment write locks never wait on IO.
Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>
* Decouple UniversalReadExt from the async filesystem requirement
UniversalReadExt is condition-checker dispatch; it never consumed the async
surface itself. Drop its `Fs: UniversalReadFsAsync` supertrait bound and relax
CachedFs's struct-level bound back to `UniversalReadFs` — the async requirement
now lives on the one impl that consumes it, `CachedReadFs for CachedFs`
(schedule_open parks the inner filesystem's `open_async` futures).
The bound then surfaces only on the lifecycle/preload impl blocks that go
through CachedReadFs (segment open, live-preload/reload, config reload, edge
load/refresh); the search path carries no async bounds at all.
Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>
---------
Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
* existing segments: wait for IO outside of search pool
* new segments: wait for IO outside of search pool
* extract reload into separate function
* Update lib/edge/Cargo.toml
---------
Co-authored-by: Tim Visée <tim+github@visee.me>
* rename `reopen`->`live_reload` and `schedule_reopen`->`live_preload`
* `UniversalRead::live_preload` returns a shared future
* assert snapshot-miss eagerly on `live_preload`
`live_reload` cannot see the failed preload: its blocking fallback
re-resolves the length from the remote and succeeds. The error
surfaces at preload time, as callers (`ok_not_found`) expect.
Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>
---------
Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
* [CachedFs] new `schedule` and `wait_all` primitives
* [AppendableIdTracker] don't reopen if just opened
* eager NotFound in `schedule_open`
* add traces for async reads
* finish `preopen`/`preload` with `wait_all`
* lock all segments in parallel for `live_reload`
* LIST before everything
to do: we don't have whole-fetch in async mode. to prevent sequential
`len`, we won't overlap static files with LIST.
* `wait_all` returns nothing
* Add repro test for Gridstore stale-gaps allocation panic
The region gaps (gaps.dat) are an acceleration structure derived from
the bitmask (bitmask.dat), persisted to a separate file without
ordering guarantees. After an unclean shutdown (power loss, kernel
crash) the gaps can claim free space where the bitmask has the blocks
marked used. An allocation in that state panics with "New page has
just been created", seen in production during WAL replay on startup.
This test simulates the torn state and expects it to be recovered; it
fails with that panic until the next commit.
* Rebuild Gridstore region gaps once on detected inconsistency
Offsets returned by the block search always come from scanning the
bitmask itself; the region gaps only steer where to look. Stale gaps
can therefore only cause missed allocations, never a wrong allocation:
every torn state funnels into the allocation failure that used to
panic with "New page has just been created".
Instead of paying for gaps validation on every open, detect the
inconsistency at that failure point, log a warning, rebuild the gaps
from the bitmask (repairing content and length), and retry. This is
allowed at most once per instance: after a rebuild the gaps are kept
consistent in memory, so a second failure would be a logic bug and
still panics. Also clamp proposed search windows to the bitmask length
so a length-diverged gaps file reaches the recoverable path instead of
an out-of-bounds panic.
* Fix typo
* Add repro test for gaps length divergence breaking page creation
BitmaskGaps::extend grows the file with zeroes before writing the new
all-free entries through the mmap. After an unclean shutdown the growth
can be persisted while the entry contents are lost, leaving phantom
all-zero entries beyond the bitmask, each claiming a full region.
Phantom full entries are invisible to the gap search, but they force
trailing_free_blocks to report zero, so the next allocation always
tries to create a new page and cover_new_page panics on its "Bitmask
length mismatch" assertion — before the lazy gaps rebuild from the
previous commit can detect anything.
The test expects opening the storage to repair the divergence; it
fails with that panic until the next commit.
* Repair gaps-to-bitmask length divergence when opening Gridstore
The number of regions the gaps file covers must match the bitmask, but
an unclean shutdown can break that: a lost extend writeback leaves
phantom all-zero entries beyond the bitmask, and a lost file growth
leaves the gaps file short. Phantom full entries force page creation
(they zero out trailing_free_blocks) and cover_new_page then panics on
its length assertion — before the lazy content rebuild can detect
anything, so that path cannot recover from this state.
Comparing the lengths is cheap, so do it on every open: on divergence,
log a warning, rebuild the gaps from the bitmask right away, and
consume the once-per-instance rebuild allowance. Allocation behavior
is unchanged on consistent storages.
* Reference to pull request
* Make gaps rebuild safe on Windows
Windows refuses to resize a file with a live user mapping, so the gaps
reset that recreated the file under its own mapping failed there with
OS error 1224 (ERROR_USER_MAPPED_FILE).
Split the rebuild along that constraint. The lazy content rebuild
keeps the mapping and overwrites the entries in place: it never needs
to resize, because a length divergence is repaired when the storage is
opened, and refuses with an error if it encounters one anyway. The
open-time length repair consumes the Bitmask by value so it can drop
the gaps mapping, atomically replace the file with the rebuilt
entries, and map it again — no resize of a mapped file on any
platform.
* Simplify gaps rebuild code
Cleanups from a review pass, no behavior change:
- compute_gaps: one read_all pass over region chunks instead of a
read_bit_range call per region, which also removes the loop body
duplicated from update_region_gaps
- BitmaskGaps::overwrite: take a slice instead of collecting an
iterator the only caller already holds as a Vec
- find_available_blocks: gate the divergence clamp on the O(1)
bit_len instead of hoisting read_all above it
- Gridstore::open: flatten the match-to-tuple into an if let, and
shorten the rebuild warning to match the runtime one
- tests: shared bitmask setup and value read-back helpers; drop the
length-divergence scenario from test_rebuild_gaps that
test_gaps_length_mismatch already covers (its search assertion
moved there)
- fix garbled log and comment wording
* 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>
* Add tools/callgraph: interactive call-graph reports via rust-analyzer
Generates a self-contained HTML report for one function: pan/zoom graphviz
graph of callers and callees, per-node docs and source snippets, exact call
sites with context, GitHub/editor links.
- rust-analyzer call hierarchy over LSP gives resolved (not textual) edges;
trait declarations and impls are bridged via goto-declaration /
goto-implementation so dispatch through a trait doesn't dead-end the walk
- test code excluded by running rust-analyzer with cfg(test) disabled, plus
path filters for tests/, benches/, examples/ targets
- no dependencies beyond rust-analyzer and graphviz on PATH
Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01RgxVoa6hfvgB7r6FVbcGmg
* Add screenshot to tools/callgraph README
Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01RgxVoa6hfvgB7r6FVbcGmg
---------
Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
The debug-tools workflow cross-compiles for x86_64-unknown-linux-musl,
but tokio-uring 0.5.0 requires libc::statx which is unavailable on musl.
Skip the tokio-uring dependency on musl and fall back to sync io_uring reads.
* `UniversalReadFs::open_async`
* `schedule_open` polls once
Scheduled opens must start eagerly: sync backends complete their
`open_async` on the first poll, preserving the prefetch contract
(handles outlive later file deletions/replacements). Moved down from
the integration branch so this PR stays green.
Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>
---------
Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
* Read vector runs that straddle a chunk boundary
Resolve a run into per-chunk parts instead of a single range, borrowing
when it lands in one chunk and copying when it spans two. The read
pipeline schedules one range per read, so a straddling run is read
outside it.
No writer produces such a run yet, so this changes nothing on its own.
It is what a reader needs before one does — including edge and
live-reload readers, which read files a different version wrote.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Place multivector runs without regard to chunk boundaries
Writers appended a multivector's inner vectors at the end of the row
space unless the run would cross a chunk boundary, in which case they
skipped the chunk tail — the batch writers padding the skipped rows with
explicit zero rows. That made chunk geometry part of the interface every
multivector storage had to reuse.
Runs now go at the end unconditionally and the chunked storage splits
the write across chunks, as it already did for a batch of single
vectors.
What is left of the geometry is a size cap: a multivector may not exceed
one chunk. It is fill-independent, so it constrains nothing about
placement, and it is what the volatile storage needs anyway — that one
returns a plain slice and so cannot serve a straddling run.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Split a run at chunk boundaries in one place
Reading, writing in place and appending each derived the split from
`remaining_chunk_capacity`, so every one of them had to know that a run
does not necessarily fit where it starts.
`split_run` hands out the parts instead: one per chunk the run covers,
each carrying where it goes and how much of the run it takes. Nothing
asks how much room is left any more, and `get_chunk_offset` goes with it.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Keep straddling runs on the read pipeline
Reading a straddling run outside the pipeline blocked the scheduling
loop on one read, which costs a round trip on a backend that fetches
remotely and drops the batch back to sequential.
A run is now scheduled as one read per chunk it covers. Parts complete
in any order, so each run holds what has landed until the last part
does, then hands the callback the stitched vectors. Runs taking a single
read carry the caller's data in the tag and never touch that table.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Stop capping a multivector at one chunk
The cap outlived its reason on disk, but the volatile storage still
needed it: its `get_many` handed out a slice of one chunk, so a run that
crossed a boundary had nowhere to come from. And since a volatile
storage is a target of the batched copy that builds a segment, dropping
the cap only on disk would have turned a rejected write into a failed
merge.
So the volatile storage splits and stitches too. Both are a few lines
each, and placing a run no longer skips a chunk tail, so `extend` is now
`insert_many` at the end of the storage.
Nothing user-facing moves: `MAX_MULTIVECTOR_FLATTENED_LEN` caps a
multivector at 1M elements, far inside a 32 MiB chunk, so the storages
only ever rejected what reached them unvalidated.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Schedule a single-read run without the queue
The scheduling loop resolved every run into the queue and then took it
straight back out, so the overwhelmingly common run — one that fits a
chunk — paid a push and a pop for nothing. It now goes to the pipeline
directly, and the queue holds only what a straddling run leaves behind.
Worth ~10% on the multivector read benchmark, and it collapses the
"top up, then take" pair into one decision. Extracting that bookkeeping
into helpers instead was measured and is much worse: the mmap pipeline
alternates one schedule with one wait, so the loop body is a few dozen
nanoseconds, and a helper carrying the cold map and stitching paths is
too big for the compiler to inline back into it.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Repoint the multivector WAL-replay test at a live rejection
The test upserted a multivector too large for a storage chunk, which no
longer fails: the storages stopped capping one at a chunk. Nothing else
covered a multivector operation that only the apply path rejects.
A raw blob that is not a whole number of quantized records still does,
so the test now uses that, alongside its dense and sparse siblings.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Test reading multivectors with legacy chunk-tail padding
Locks the compatibility contract that pre-straddle files — runs that
skip a chunk's leftover slots — still reopen as single-chunk borrows.
* chore: retrigger CI after flaky test-consensus-compose
* Move ReadTag into for_each_vector, its only user
Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01AN4Hgbd65gDhesthJk5bUY
---------
Co-authored-by: Claude Opus 5 (1M context) <noreply@anthropic.com>
Co-authored-by: qdrant-cloud-bot <111755117+qdrant-cloud-bot@users.noreply.github.com>
Tim Visée