* 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>
* Add SetFlushInterval op to the model tester
Changes the collection's flush_interval_sec mid-run through the same path
update_collection takes (persist the optimizer-config diff, then recreate
the optimizers in the background). The model is untouched: what it perturbs
is the flush cadence, so how much of the workload is still WAL-only when a
restart hits, plus the worker stop/start race in on_optimizer_config_update.
Kept in FORCE_OFF for now: with the optimizer on it makes stale point state
visible within a few ops of the config change. Narrowed to
recreate_optimizers_background, see the comment on Swarm::BASE.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_011AjmS5GFeGztfP3JqutnXj
* Keep SetFlushInterval enabled in the swarm
Drops it from FORCE_OFF so the divergence it surfaces is reachable without
--enable-force-off (which would also enable the broken vector-name ops).
The evidence moves from the FORCE_OFF comment onto the op's own doc.
The two optimizer-on harness gates now fail whenever the swarm draws the op.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_011AjmS5GFeGztfP3JqutnXj
* Propagate proxied changes when unwrapping proxies on optimization failure
unwrap_proxy puts the wrapped segments back into the segment holder, so the
changes recorded on the proxy while the optimization ran (deleted points,
index and vector-name changes) have to reach the wrapped segment first. They
did not, so every point deleted or overwritten during the optimization kept
its pre-optimization copy live next to the new copy in the write segment, and
reads saw both: counts too high, scroll and search returning the stale copy.
The snapshot unproxy path already does this; the optimizer failure path was
the only place putting a wrapped segment back without it. It is reachable
whenever the shard outlives the cancellation, in particular an update_collection
that recreates the optimizers while an optimization is in flight.
Lock order is holder-then-updates, matching try_unproxy_segment: updates-then-
holder-write deadlocks against the snapshot path.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_011AjmS5GFeGztfP3JqutnXj
* Drop the stale failure note from the SetFlushInterval doc
The divergence it described is fixed in this branch.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_011AjmS5GFeGztfP3JqutnXj
* test as well with 0s as flushing interval
* Fail optimization unwrapping when proxy propagation fails
Losing proxied deletes and index changes is data corruption, so return the
error instead of logging it: no proxy is unwrapped and the changes stay
served by the proxies. The cancelled-segment cleanup moves ahead of
unwrap_proxy so the orphan is still removed when that error fires.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01CpoaAtGbAQAEuxEi8ScHc5
* Drop the model tester --flush-interval-sec flag
SetFlushInterval covers the interval now, so the run starts at the shipped
5s default (fixture::INITIAL_FLUSH_INTERVAL_SEC, still traced in the header)
and the ops move it from there. Also documents what 0 does now that it is a
generated value.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01CpoaAtGbAQAEuxEi8ScHc5
* Fail snapshot unproxying when proxy propagation fails
Both paths logged the error and unwrapped anyway, dropping the deletes and
index changes that never reached the wrapped segment. Same reasoning as
unwrap_proxy in the optimizer.
try_unproxy_segment hands the lock back and leaves the proxy installed, the
failure mode its doc already describes: the caller keeps it in `proxies` and
unproxy_all_segments retries the propagation right after. unproxy_all_segments
returns before touching the holder, so the temp segment the surviving proxies
write into stays in place (remove_segment_if_not_needed only checks whether it
is empty and appendable, not whether a proxy still references it).
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Claude-Session: https://claude.ai/code/session_01CpoaAtGbAQAEuxEi8ScHc5
---------
Co-authored-by: Claude Opus 5 (1M context) <noreply@anthropic.com>
ast-grep 0.45 no longer parses a leading `::` fragment as a pattern, so
`pattern: ::$MOD` stopped matching and every `::common::` / `::wal::` path
survived into the generated qdrant-edge crate, failing `just rs-check`.
Matching the node text instead keeps the rule working on 0.44 and 0.45: the
amalgamation output is byte identical to what 0.44 produced before.
Claude-Session: https://claude.ai/code/session_01CpoaAtGbAQAEuxEi8ScHc5
Co-authored-by: Claude Opus 5 (1M context) <noreply@anthropic.com>
`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>
* fix(strict-mode): enforce max_query_limit on scroll requests when limit is omitted (fixes#10373)
* style: rustfmt scroll query_limit for CI lint
---------
Co-authored-by: qdrant-cloud-bot <111755117+qdrant-cloud-bot@users.noreply.github.com>
* Fix flaky shard snapshot API CI readiness race
Run the prebuilt binary and poll /readyz instead of cargo run + fixed sleep, which can miss startup when cargo recompiles.
* Move shard snapshot API CI runner into a dedicated script
Keep workflow YAML thin by starting Qdrant, waiting for /readyz, and invoking shard-snapshot-api.sh from tests/shard-snapshot-api-tests.sh.
* fixup! Implement `ChangeAliases` operation
* fixup! Add `ChangeAliases` to replay-safety tests
* fixup! Add `ChangeAliases` tests
* De-slop ⛷️
* De-slop 🏂
* Add `TestSlowDown` and `TestTransientError` actions
These would have to be implemented on `TableOfContent` when switching
to `ConsensusStateMachine` as main consensus impl
* Handle more stupid corner-cases for `ChangeAliases` prop tests
* Add `AliasMapping::remove` and `AliasMapping::rename` methods
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Implement `ChangeAliases` operation
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Add `ChangeAliases` to replay-safety tests
Multi-action operations that rename an alias are skipped in the convergence
property: the current implementation does not replay them convergently, and
the machine reproduces that.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Add `ChangeAliases` tests
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Add test-only `PeerMetadata::new` constructor
Test state needs peers at a version other than this build. The `version` field
is crate-private and `current()` is the only constructor, so gate the new one
on the `testing` feature and enable it for the `storage` test build.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Implement `UpdatePeerMetadata` operation
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Add `UpdatePeerMetadata` to replay-safety tests
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Add `UpdatePeerMetadata` tests
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Implement `UpdateClusterMetadata` operation
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Add `UpdateClusterMetadata` to replay-safety tests
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Add `UpdateClusterMetadata` tests
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Implement `SetQuotaConfig` operation
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Add `SetQuotaConfig` to replay-safety tests
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Add `SetQuotaConfig` tests
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Implement `TestSlowDown` and `TestTransientError` operations
Both are node-local: one sleeps, the other fails at random. They plan no
actions, like `Nop`, so the replay-safety properties have nothing to add.
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* Add `TestSlowDown` and `TestTransientError` tests
Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
* fixup! Add `UpdatePeerMetadata` to replay-safety tests
* fixup! Add `UpdateClusterMetadata` tests
---------
Co-authored-by: Claude Opus 5 (1M context) <noreply@anthropic.com>
Unindexed Match::Text and Match::Phrase previously shared a String::contains
arm, so phrase order was ignored and queries matched across token boundaries.
Use the default Word tokenizer for best-effort parity with indexed fields.
Fixes#10182
qdrant-cloud-bot