v5: More UTF-8! More Bio! More Hardware!#318
Open
ashvardanian wants to merge 352 commits into
Open
Conversation
…ll ISAs
Replace the single-match `sz_utf8_find_newline`/`_whitespace` API with a unified
multistep boundary family that emits `(offset, length)` batches per call and
resumes via `bytes_consumed`, and rebuild every backend's peel on a single-pass,
native left-pack design with no per-match `ctz` and no nested drain loop.
Core API
- One canonical `sz_utf8_find_boundaries_t` typedef replaces the four single-match
function-pointer types; the single-match kernels (including serial) are deleted.
- `sz_iterators_default_steps_k = 16` is the shared default batch size.
Newline / whitespace peels (classify -> compact -> emit, per window)
- Ice Lake: `vpcompressb` packs the matching lane indices and per-lane lengths;
one masked `vpmovzxbq` widen-store emits up to 8 `(offset, length)` pairs in a
single straight-line wave. A window with more than 8 matches resumes past the
8th and re-classifies, so there is no register-shift loop.
- Haswell / NEON / LASX / PowerVSX / WASM: the native left-pack
(`vpermd` / `vqtbl2q` / `xvperm` / `vec_perm` / `i8x16_swizzle`) plus per-file
compact LUTs replace the scalar `ctz` index-find and the `group==4`/scalar-tail
store branch; bounded by a masked store where available, else a small fixed
stack scratch.
- SVE2 / RVV keep their native `svcompact` / `vcompress`. Tails are masked /
predicated in-SIMD on Ice Lake / SVE2 / RVV, serial-scalar on the fixed-width ISAs.
UAX-29 word-boundary kernels (forward + reverse, every ISA)
- Replace the nested `while (boundary) { ctz/clz }` drain - and Ice Lake's
`boundary_lanes[64]` stack round-trip - with one flat window loop: an
early-`continue` scalar fallback, then a single fixed sub-block loop that folds
compaction and the shifted-difference (`lengths = boundary - shift1(boundary,
carry)`, carrying the open word's start/end). Reverse walks sub-blocks
high-to-low through a descending compaction LUT. Ice Lake emits up to 8 and
re-classifies; the wider-mask ISAs emit-all per window. A re-classified window
clears the open word's own boundary (lane 2 forward, lane 62 reverse) so it is
not re-emitted as a zero-length word.
- `sz_u64_bits_reverse` (explicit SWAR over `sz_u64_bytes_reverse`) drives the
descending `vpcompressb`; x86 has no scalar bit-reverse instruction.
Bindings (C++ / Rust / Python / Swift)
- Split iterators carry no `done`/`exhausted` state bool (count==0 + sentinel),
gain compile-time `steps`/`STEPS` and `skip_empty` knobs, and default to KEEPING
empty segments (N+1) with opt-in skipping across every split range. Removes the
deprecated Rust split aliases and the duplicate `WORD_BOUNDARIES_BATCH` constant;
word splitters are const-generic over `STEPS`.
Validated byte-exact against the serial reference on every backend (native,
QEMU for sve2/rvv/lasx/ppc, wasmtime for wasm) over a full capacity tiling with
dense-boundary and EOF-truncation fixtures, plus the C++20 suite and `cargo test`.
Land the case-insensitive find correctness fix and its test coverage atop a project-wide vocabulary cleanup. Semantic changes: - Fix a real offset bug in the serial 2/3-folded uncased find: a match whose folded runes cross an expansion boundary (needle ſß folding to "sss" inside haystack ßß folding to "ssss") reconstructed its start from per-codepoint byte lengths and returned a negative/underflowed offset. It now tracks the source- codepoint begin pointer per history slot, so it reports offset 0, length 4. - Redesign `uncased_violation` (was `case_invariant`) to return a pointer to the first cased codepoint, or NULL when fully caseless, mirroring `norm_violation`. The invariant-needle find shortcut becomes a `== SZ_NULL_CHAR` check. - Add an independent fold-subset reference oracle and a three-way checker, a generative cross-expansion enumerator, and a uniform per-backend battery, so the serial-vs-SIMD differential is also validated against ground truth (it was blind to bugs the SIMD kernels share through their serial fallback). Rust gains the same coverage (exercising the v128 kernel under Wasmtime) and Python a crossing test. Mechanical: rename `case_insensitive`/`case_fold`/`case_invariant` to `uncased`/`uncased_fold`/`uncased_violation` and purge the `ci` abbreviation across the C core, all SIMD backends, the Rust/Python/Swift/JS bindings, build files, and tests; move the `utf8_case_insensitive/` and `utf8_case_fold/` trees to `utf8_uncased/` and `utf8_uncased_fold/`. External Unicode references (`CaseFolding.txt`, "Unicode Case Folding", UAX #21) are preserved.
Split the monolithic entry file into one translation unit per domain (hash, utf8, uncased, string, find, sort) that compile into a single binary, so the standards matrix rebuilds incrementally instead of recompiling one 5K-line file. Unify each domain on a fixed aspect taxonomy: one known-answer `test_<domain>_unit` (the former golden + API battery, merged), a `test_<domain>_all` per-ISA driver that replaces the monolithic `test_equivalence`, and differential/safety helpers reshaped to take templated backend wrappers (`<op>_from_sz_<func_>`, mirroring the bench files) instead of loose kernel-pointer lists. main() registers every `_unit` and `_all` uniformly through `run_test`. Native focus (serial/haswell/icelake); every `#if SZ_USE_*` branch for other ISAs is preserved. Restore two Kelvin/Angstrom test literals that an editing pass had NFKC-normalized (U+212A->K, U+212B->Å), now pinned as byte-exact \x escapes.
…odec Decoding now goes through a single validating primitive (sz_rune_parse, ex-_strict) that returns the codepoint length or sz_utf8_invalid_k: well-formed runes fold/search/sort/normalize exactly as before, while any byte that does not begin a well-formed codepoint is passed through unchanged as a one-byte maximal subpart (resync at the next byte). This kills the data-mutating behaviour on malformed input - serial silently dropped surrogates and collapsed overlongs - and makes serial byte-identical to every SIMD backend. Also fixes a folded-iterator out-of-bounds read and a reverse-iterator byte-loss on truncated input. The rune codec (parse/parse_unchecked/export/valid) moves out of the foundational types.h into utf8_runes.h, the bare name sz_rune_parse goes to the validating variant (unsafe one keeps _unchecked), the dead sz_rune_rparse/sz_runes_parse are removed, and every decoder now returns the length with a single rune out-param. Carries the supporting types.h changes: SZ_MAY_ALIAS on the SIMD vector unions (fixes strict-aliasing UB), the sz_u*_nth_set_bit / bit-reverse helpers, and the SVE2-AES feature-macro gate for Clang. Valid-UTF-8 output is byte-identical across all paths.
sz_hash_state_update only absorbed a 64-byte block once it filled, and sz_hash_state_digest finalized without consuming the buffered remainder - so two inputs differing only past the last block boundary collided, and streaming disagreed with the one-shot sz_hash. Defer the filling block and fold it in finalize, so one-shot and streaming share one finalization and pick the minimal/full construction by total length identically. The per-ISA buffering is rewritten with masked/predicated loads (AVX-512 mask, RVV vl, contiguous copy elsewhere) instead of byte-cursor loops and head/body/tail bookkeeping. Canonical sz_hash values are unchanged across all backends.
find_nth becomes a flat early-continue using the shared sz_u*_nth_set_bit helper (no nested branch or bit-clear loop); the multistep newline/whitespace peels and word-boundary kernels drop the per-ISA head/body/tail bookkeeping; the RVV newline/whitespace twins and word-boundary fwd/rev paths are folded into shared drivers; unpack_chunk is un-nested; popcount/ctz use native intrinsics where the mask is not already materialized; and the essay-length docstrings are trimmed. Also flattens the AVX2 sort 3-way partition. Output stays byte-exact vs serial across all ISAs.
Restructures the per-domain test binaries and harness, and migrates the rune-codec call sites (sz_rune_parse signature, malformed-input guards) to match the consolidated codec.
`sz_copy`/`sz_move` operate on arbitrary byte buffers, so their head/tail scalar copies and small-length SIMD stores routinely run on misaligned addresses. The Haswell kernel reached them through `*(sz_u64_t *)ptr` casts and `&ptr->ymm` union-member access, and the LASX kernel through the same `*(sz_u64_t *)ptr` casts - all undefined behavior under `-fsanitize=alignment`, even though they emit correct unaligned moves on x86/LoongArch. Add symmetric `sz_u16/32/64_store` helpers mirroring `sz_u16/32/64_load` (both honor `SZ_USE_MISALIGNED_LOADS`, so production builds still emit a single unaligned access), route the scalar head/tail through the load/store helpers, and store the SIMD lanes straight to the `__m128i*`/`__m256i*` byte pointer instead of a vec-union member. A full ASan+UBSan sweep of the suite is now clean.
Read and write the packed public `sz_hash_state_t` through an aligned, vector-typed twin loaded and stored with each backend's widest native unaligned op, instead of casting the packed bytes to an aligned vector type (which `vmovdqa`-faulted on an unaligned streaming state). One-shot hashing builds the twin directly; streaming loads it at entry and stores it back at exit. The AES round math is untouched, so the hash output is byte-identical on every backend.
`split`/iterate/slice churn mints a fixed-size `Str` (or `Strs`) header per element and frees it moments later. Cache dead headers on an intrusive singly-linked free-list - the `next` link is threaded through the dead object's own storage (`Str::parent`, the `Strs::data` union), so the state is just a head pointer and a counter per type, no array. The allocation helpers pop a cached header back, collapsing the per-element `PyObject_Malloc`/`Free` round-trip (~7x faster header churn in isolation). The lists live in per-interpreter module state reached via `PyState_FindModule`, so the types stay static and it is sub-interpreter / free-threading clean. Capacity is bounded at 64 headers per type; the existing module `m_free` hook drains both lists at teardown.
Pivot the four alignment engines (Levenshtein, Levenshtein-UTF8, Needleman-Wunsch, Smith-Waterman) from a pairwise batch (results[i] = dist(first[i], second[i])) to a cross-product: results[query][candidate] = dist(query, candidate) over all pairs, row-major and query-major into a strided matrix. Passing no candidates requests symmetric self-similarity (lower triangle + diagonal, mirrored on write). This is the shape clustering, dedup, retrieval reranking, and all-vs-all bioinformatics need, and it is what unlocks the per-lane work below. Add an inter-sequence kernel family, candidate_lane_walker, that packs one shared query against many candidates one-per-SIMD-lane and walks the DP row-by-row, so every lane stays busy independent of string length (the intra-sequence anti-diagonal wastes lanes at short lengths). Specialized for Ice Lake (AVX-512, 64xu8 / 32xu16 / 32xi16), Haswell (AVX2, 16xu16 / 16xi16), and NEON (16xu8 / 8xu16 / 8xi16), with a scalar serial reference oracle. Levenshtein reuses the query's Myers match-bitmasks across its candidate run; weighted NW/SW reuse the query's 32-class substitution profile. Each engine owns grow-only scratch sized once per call and threaded down as a span view, so scoring leaves never allocate. On the GPU, replace the per-call host grouping sort with cub::DeviceMergeSort over a hoisted grow-only scratch member, and fix a latent out-of-bounds read in the warp tile-scorers (the fixed-width SIMD window over-read one char past the string ends on the final odd tail) with predicated masked tail loads rather than padding the inputs. Bindings, C ABI, and tests move to the matrix shape: C signals symmetric with a NULL candidates pointer; Rust returns UnifiedMat<T> with row()/dimensions(); Python returns a 2-D Q-by-C ndarray with candidates=None for symmetric.
Add an optional `out=` writable buffer-protocol argument to `Strs.argsort`,
mirroring the `hash_multiseed` precedent. When given a contiguous `uint64`
buffer (e.g. `numpy.uintp` or `array('Q')`), the permutation is written into
it and the buffer is returned, avoiding the per-element `PyLong` tuple. A full
sort writes straight into `out`; top-k sorts into scratch and copies only the
leading `top` indices, never clobbering past `result_count`.
Inline the previously shared sort-option parser into `sorted` and `argsort`
since their argument sets now differ, and convert both docstrings to the
structured `Args:`/`Returns:` form used elsewhere.
Consolidate the inter-sequence cross-product onto one shared host-side candidate-lane driver with dyadic length-bucket tiling, reused by every byte NW/SW (linear+affine) and the new non-unit/affine byte Levenshtein SIMD batch engines across haswell/icelake/neon; the GPU weighted engines share the Levenshtein dyadic tier router. Replace the parallel pointer-to-spans arguments of the inter-sequence Myers batch kernels with a single lane_pairs_view batch type and give the multi-word kernels snake_case names. Redesign the similarity benchmark to feed independent, randomly-sampled query and candidate batches in all-pairs and pairwise modes over the exact aggregate cell-update denominator. Strip duplicated overload comments and prose across the similarity engines.
Add 32-bit-cell inter-sequence candidate-lane kernels per CPU backend (haswell 8-lane, icelake 16-lane, neon 4-lane): i32 weighted Needleman-Wunsch/Smith-Waterman (linear+affine, global+local) and u32 uniform Levenshtein (linear+affine). These are strict width-mirrors of the existing 16-bit kernels and will back the width-tiered driver so weighted pairs whose accumulated score exceeds the i16 range stay vectorized instead of dropping to the scalar diagonal walker. Kernels are added but not yet wired into the engines. Also convert the Ice Lake AVX-512 score load/stores to the width-typed _mm*_loadu/storeu_epi* forms (no reinterpret_cast).
Restructure the shared host driver to dispatch each length bucket to a narrow (i16/u16) or wide (i32/u32) candidate-lane kernel by its score reach, via a reusable per-width tier loop, so weighted Needleman-Wunsch / Smith-Waterman and non-unit/affine Levenshtein cells whose accumulated score exceeds the 16-bit range stay vectorized instead of dropping to the scalar diagonal walker. Wire every CPU candidate-lane engine to the two-kernel driver with narrow+wide fits predicates. Add cross-product tests at lengths and costs that overflow the narrow tier, pinning the new kernels cell-by-cell against the serial oracle on haswell, icelake, and neon.
Extend the per-thread weighted Needleman-Wunsch / Smith-Waterman register and batch kernels to compute their recurrence with Hopper's three-input DPX add-max intrinsics under a compile-time capability guard, matching the device-tier scorer. Results stay bit-identical to the scalar path (the Smith-Waterman zero-clamp remains a post-fold max to preserve exactness); the non-Hopper path is unchanged. u32/u64 Levenshtein and i64 NW/SW keep inheriting the scalar base, as Hopper has no DPX for those widths.
Give the Haswell (4-lane) and NEON (2-lane) UTF-8 Levenshtein engines a bit-parallel rune Myers fast path, mirroring the Ice Lake rune Myers and its rune-keyed open-addressing Peq hash, so unit-cost UTF-8 edit distance batches in SIMD on every CPU backend instead of falling back to the per-pair rune diagonal walker. The per-pair rune fallback for non-unit and over-long cells now binds the scalar serial scorer (no per-backend rune tile_scorer exists). Also add the delegating single-pair layout()/operator() shim to the Haswell byte Myers for the per-pair engine path.
Pack the dyadic bucket bit_width(shorter-1) into the GPU tier-sort key (Levenshtein and weighted alike) instead of the raw saturated length, so the radix sort bands tasks into power-of-two length groups within each priority tier and launches run at more uniform depth - matching the CPU candidate-lane dyadic tiling. Pure re-ordering; results unchanged.
Template the shared host driver, scratch sizing, and transpose buffer over the kernel's element type (char for byte engines, rune_t for UTF-8) instead of hardcoding char, so the UTF-8 engines can route rune candidates through the same width-tiered candidate-lane path. Backward-compatible: byte engines deduce char and are unchanged.
Remove the unreachable score_range_lane_walker_ method (and its orphaned fits_lane_range_ / transcode_runes_ / reach-limit helpers) from the haswell, icelake, and neon UTF-8 engines. Superseded by the rune Myers fast path and the shared candidate-lane driver; grep-confirmed zero callers.
At exactly 16 bytes `cuda_status_t` is returned in the `RAX:RDX` register pair; NVCC 12.x and host g++ miscompile the leading `status`/`cuda_error` eightbyte inside large translation units (the elapsed-time eightbyte survives), so a GPU engine that genuinely succeeds reports a garbage status to its caller. This surfaced as the `stringzillas_bench_similarities_cu20` benchmark aborting with an "unrecognized" status even though the kernels ran and produced correct results. Padding the struct past 16 bytes moves it into the SysV memory-return class (sret), which both toolchains compile correctly, fixing every GPU engine's status read without touching any call site.
…shtein Non-unit and affine UTF-8 Levenshtein previously fell back to a per-pair serial rune DP on every CPU backend - the cross-product never packed independent rune pairs across SIMD lanes. Add explicit `rune_t` candidate-lane walkers (u16 narrow + u32 wide, linear and affine) on haswell, icelake and neon, and route the linear-non-unit and affine UTF-8 engines through the width-tiered cross-product driver, leaving the serial rune DP only as the can't-fill-lanes per-pair fallback. Each kernel is bit-exact to the serial UTF-8 oracle; tests fill the inter-lane group with uniform >512-rune rows so the batched path is actually exercised rather than collapsing to 1x1.
The warp-tier GPU aligners pack the DP diagonals back-to-back at `k * bytes_per_diagonal`, each band holding `max_diagonal_length` cells. The DP only indexes `[0, max_diagonal_length)`, so the bands themselves fit exactly - but GPU shared memory is serviced in `register_width`-byte transaction words, and a cell narrower than that word (`u8`/`u16`) is read by fetching the enclosing word. Reading the last cell of the final band therefore touches up to `register_width - bytes_per_cell` bytes past it. Between bands that overhang lands in the next band; past the final band it runs off the end of the tightly-packed per-warp shared allocation and faults - an `Invalid __shared__` access (cudaErrorLaunchFailure) that hit the affine warp kernel on short, length-diverse inputs (tokenized words). Account for it in `diagonal_memory_requirements` by reserving exactly one widened-read overhang (`register_width - bytes_per_cell`, zero once cells reach the word width). The CPU walkers were already immune - they cache- line-pad between bands. Verified compute-sanitizer-clean and bit-exact on cpp20 + cu20.
…eries The per-query Myers reuse kernel mapped one warp per query, so a single-query / many-candidate cross-product (the retrieval regime) lit up exactly one warp: ncu showed 1.6% occupancy, 0.1% SM throughput, a 113 ms launch where the full-device case finishes in 145 us. Decompose the work as 2-D `(query, candidate-segment)` units instead: each query's candidate row is split into `ceil(warps_per_device / queries)` contiguous segments (clamped to the candidate count), and warps grid-stride over `queries * segments` units. When queries already fill the device this reduces to one segment per query (unchanged); when they don't, the spare warps take candidate slices and the device fills regardless of query count. The Peq build (rebuilt once per query-segment - cheap for single-word queries) and the Myers recurrence are byte-identical, so results are unchanged. q=1, len=64: 37 -> 29,286 GCUPS (~790x), now matching the q=4096 rate and compute-bound at 81% SM. Bit-exact on cu20.
Match the StringWars knob: score `STRINGWARS_BATCH_PER_CORE` pairs per core (default 256), where a CPU logical core and a GPU streaming multiprocessor each count as one core, so the per-device pair budget auto-scales as `STRINGWARS_BATCH_PER_CORE * device_parallelism` with no fixed CPU/GPU multiplier. Because the workload is a 2-D cross-product, that budget is split as a square - each axis (queries, candidates) grows with the square root of the per-device budget - so the balanced shape is `side x side` and the retrieval corner (one query) can grow its candidate axis with the knob instead of the old hard 256K ceiling. The query sweep is now anchored on that square side (retrieval, candidate- skew, balanced, query-skew); `STRINGWARS_BATCH` still overrides with explicit per-query counts. On an H100 the default tile is `256 * 132 SMs`, and e.g. `STRINGWARS_BATCH_PER_CORE=16384` grows the q=1 shape from q1xc33792 to q1xc1048576.
…queries Unit-cost Levenshtein on queries longer than 64 bytes fell off a cliff: the size-generic thread-per-pair kernel keeps its multi-word Peq in GLOBAL scratch, so the inner loop's per-(word x text-char) match-table load is L2-bound (profiled lts 82%, sm 5%, 0.04 inst/cycle) - ~1 TCUPS vs the single-word warp-reuse kernel's ~28. Generalize the warp-per-query reuse design to a compile-time `words_count_` (1/2/4 -> shorter <= 64/128/256): one warp owns a query, builds its multi-word Peq once into SHARED, reuses it across the query's candidates (2-D query/candidate-segment spread), with per-lane VP/VN in registers. The recurrence is copied verbatim from the proven generic kernel, so it stays bit-exact. Single-word (<= 64) still routes to the original hand-tuned kernel - its scalar recurrence beats the one-word-specialized generic body (~2.3x), so warp1 keeps it and only warp2/warp4 use the new kernel. len128: 998 -> ~11,800 GCUPS (12x); len256: 853 -> ~12,100 (14x); len64 unchanged (~28,800). ncu len128: sm 5% -> 48%, Peq L2 traffic eliminated (residual L2 is candidate-text re-reads). Bit-exact on cu20.
The cu20 suite drove the device only through 1x1 make_pairwise tiles and a single-word (<= 24) reuse case, so the multi-word Peq-in-shared reuse path (65-256 byte queries) and the >256 generic fallback had no cell-by-cell coverage. Add W2/W4/mixed-W and generic GPU cases to `test_similarities_cross_product`, each filling a candidate row wider than a warp and pinned against the serial oracle.
The thread-per-pair "batch tier" held the per-thread DP row in local memory once a pair exceeded the register tier, so device-timed blosum62 scoring collapsed to ~60-190 GCUPS over 160-512 chars while the lane-split warp anti-diagonal kernel (already used above 512) sustains ~420-660 there. Narrow the batch tier to the register width (batch_text_limit_k 512 -> 128) so its DP row can never spill, and send longer pairs straight to the warp kernel. Measured on H100 (NW): len160 60 -> 424 (7x), len256 96 -> 527, len512 193 -> 657; no change <=128 or >512. Added weighted NW/SW GPU cross-product cases over the rerouted 129-512 band, bit-exact vs the serial oracle (cu20).
…UTF-8 OOM The GPU similarity engines gain a codepoint-level UTF-8 Levenshtein spanning all three tiers (register thread-per-pair, warp anti-diagonal, device-tiled wavefront). Data stays UTF-8-encoded on the device: a rune-offset index prefix-scans each string once, and the scorers decode codepoints on the fly. The branchless decode clamps each continuation read to the rune's own length, matching sz_rune_parse_unchecked while never reading past a byte tape. Levenshtein no longer dispatches slower than Needleman-Wunsch: unit-cost edit distance is gated onto bit-parallel Myers only within levenshtein_myers_max_shorter_k, and the device-tiled tier scores uniform substitution cost branchlessly instead of through the matrix substituter. The affine register kernels (unit_gotoh / weighted_gotoh) relax to __launch_bounds__(256, 1) so ptxas stops spilling (weighted_gotoh 136 B -> 0, unit_gotoh 8 B -> 0 of local spill traffic on the hot recurrence). The Ice Lake UTF-8 cross-product no longer aborts at 100k-byte inputs: the worst_cell_scratch_ estimate caps the rune-Myers Peq reservation by the runtime fits_myers gate (26 GB -> 34 MB) for a kernel that never runs at that length. Also lands the GPU kernel reorganization (Hopper specializations out of cuda.cuh via device functors; per-tier dispatch tables; author-named recurrences) and the Haswell/Ice Lake weighted-walker class-hoist plus the affine E-track rolling register.
The uncased family had no scalable Arm kernels. The fold kernel walks 64-byte superchunks as vector-sized sub-chunks with carried after-lead views, classifies leads through the shared 64-entry family LUT, and routes each superchunk to a per-script handler - caseless, Latin, Cyrillic, Greek, Armenian, Georgian, or the guarded serial delegate - with predicated loads and stores replacing NEON's zero-padded stack staging. Stop lanes lower to a window mask and a shared serial helper finds the boundary. The lead-family and delta tables move to the serial header under ISA-neutral names; NEON keeps thin aliases. The search kernel mirrors NEON's probe-filtered design: a three-probe ASCII driver and a scripted driver whose fold and alarm callbacks receive driver-built neighbour views. Chunks span TWO sub-vectors - single-vector windows degenerated to byte-at-a-time stepping on long Cyrillic folded runs - and every lead rewrite (eszett, Greek promotes, micro, Armenian ligature leads) anchors on the carried next view so rewrites survive sub-vector edges; the shared central-Europe and Greek delta tables also migrate to the serial header. The case-invariant probe, the cased-rune finder, and the order delegate complete the API. At the minimal 128-bit vector length the predicate-setting compares serialize on one pipe and both kernels trail NEON - fold reaches ~70% and search 40-60% of its throughput while beating serial - so dispatch stays behind the wider-than-NEON gate where growing registers shift the balance. Validated bit-exact against serial over hundred-thousand mixed-script differentials and fuzz batteries at three vector lengths.
StringZilla now exports its header-only target through a `stringzillaTargets` export set with generated version & config files, so `find_package(stringzilla)` works after installation. The vendored ForkUnion becomes a fallback rather than a mandate: when a parent project already defines the `forkunion_header` target, the nested copy is skipped instead of colliding, and an unpopulated submodule fails with an actionable hint about `git submodule update`. Subproject consumption stops mutating global state: the build-type default is no longer forced into the shared cache, CTest machinery arms only for top-level builds, CUDA targets no longer auto-enable just because `nvcc` is on the PATH, the shared-library include paths wear `BUILD_INTERFACE` genexes, and the main-project check rides the builtin `PROJECT_IS_TOP_LEVEL` on a CMake 3.21 floor.
The first big-endian execution of the hash family - an s390x wheel built and tested locally under QEMU - caught `sz_hash_multiseed` diverging from `sz_hash` at every 17..63-byte length that is not a multiple of 16. The ladder de-overlaps its tail block with `sz_hash_shift_in_register_serial_`, whose 64-bit lane shifts realize the intended byte-array shift only in little-endian lane order; on big-endian the bytes moved the wrong way, so the ladder disagreed with the byte-wise lanes the multi-seed prepare and the streaming path build. The shift now mirrors under `SZ_IS_BIG_ENDIAN_` - a left shift of the big-endian-composed 128-bit value - and states the byte-array contract in a comment. Little-endian code is unchanged; the full s390x wheel suite now passes 6951 tests with zero failures, and POWER inherits the fix wherever big-endian delegates hashing to serial.
The SVE2 delimiter scan fed BMP and astral membership through chains of two- and three-level byte gathers - nine-cycle serialized loads that lost to NEON everywhere but Chinese at the minimal vector length. The rewrite keeps every verdict in registers and re-organizes the walk around 64-byte windows streamed as register chunks, each peeking one vector ahead: the shifted views cover every lane's continuations, so the trusted-margin re-scan disappears and the per-chunk resolution chains overlap in the out-of-order core instead of serializing per 16-byte tile. Membership needs no gathers. Lanes below U+0100 read bitmap row 0 through a pair of zero-padded table registers - the byteset kernel's wrapped two-TBL pattern. Other leads survive a 256-bit pre-filter over the high byte only if their 256-codepoint block holds any delimiter at all - a new 32-byte table; row 1 of the block table is the unique empty row and covers 198 of 256 highs, including the whole CJK range. Survivors resolve one DISTINCT high byte per round: running text shares one to three highs per chunk, and each round loads that block's 32-byte row once and settles every lane carrying it. Astral lanes ride the same distinct-value loop keyed on the (super, sub) pair. The 3-byte and 4-byte validity refinements run only when such leads are present, so 2-byte scripts skip them. Each window lowers its verdicts to one 64-bit lane mask and emits once through a shared ctz drain that rereads only the lead nibble for the length - no re-decode, no per-tile compaction fixed costs. The scan now beats NEON on all eleven corpora - English 1.30 GiB/s vs 692 MiB/s, German 1.07 GiB/s vs 435, mixed multilingual 800 vs 475, Vietnamese 725 vs 339 - and beats the scalar reference on six, so dispatch takes it unconditionally at every vector length.
The NEON delimiter scan read its bitmap bytes through a sixteen-lane scalar walk - indices stored to the stack, gathered byte by byte, reloaded into a register - once per quarter for BMP membership and again for astral lanes. The walk is gone. Lanes below U+0100 read bitmap row 0 directly through a vqtbl2q pair; other leads survive a 256-bit pre-filter over the high byte only if their block holds any delimiter (most of the plane, including all CJK letters, maps to the unique empty row); survivors resolve one DISTINCT high byte per round, loading that block's 32-byte row into a table pair and settling every lane carrying it in-register. vmaxvq extracts each round's high, which is never zero because sub-U+0100 lanes were split off. Astral lanes ride the same loop keyed on the (plane, sub) pair, with the plane kept one-biased so an empty set reads as zero and clamped to the sixteen addressable planes so a malformed lead can never index past the L1 table. Every corpus gains 5-60%: Russian 368 to 494 MiB/s, Chinese 345 to 462, Korean 349 to 557, Thai 407 to 564, German 434 to 572, mixed multilingual 475 to 619, English 692 to 726.
The SVE2 decoder's two clean-chunk lanes each prove one homogeneous shape - ASCII with 2-byte pairs, or ASCII with 3-byte triples - so scripts that interleave all three widths in a single chunk, the steady state of Vietnamese prose, fell through to the general gather-fed engine and trailed the scalar reference by a fifth. A third lane now catches the mix once both specialists decline: the same coverage algebra with the owed-lane smear built from both lead kinds at once, both-width completeness and the E0/ED bounds in one pass, codepoints assembled through a two-stage width select, and the trailing overhang combining a last-lane 2-byte lead (owes one), a last-lane 3-byte lead (owes two), and a second-to-last 3-byte lead (owes one) - mutually exclusive in a structure-clean chunk. Carries export the union: any multi-byte lead owes lane one, 3-byte leads owe lane two. Replacing the specialists with one generalized lane was measured first and rejected: Vietnamese gained but every pure script paid 5-12% for the width selects it never uses. As a third tier the mix decodes at 614 MiB/s - up 49%, ahead of the 520 MiB/s scalar reference - mixed multilingual gains 4%, and every other corpus stays within measurement noise.
…spaces The NEON whitespace scan built its shifted views by rotating the window onto itself, which poisons the last lanes and forced a two-lane trusted margin - 12.5% of every window re-scanned. The views now come from overlapping unaligned loads riding the free load pipes, so every lane is trusted and the step is the full register. A window is also probed for the four leads that can actually start a multi-byte whitespace - C2, E1, E2, E3 - and skips the entire fourteen-compare class block when none is present, which covers both ASCII prose and every non-Latin script whose letters use other leads: Cyrillic, Arabic, CJK, and Thai text never pays for spaces it cannot contain. The scan now beats the scalar reference on the mixed multilingual corpus for the first time - 737 vs 693 MiB/s (was 550) - and on Russian 770 vs 760 (was 580), English 626 vs 586 (was 483), and Chinese 1.41 vs 0.82 GiB/s (was 1.16).
A batched SVE2+AES intersection kernel - the 4-lane analog of sz_hash_sve2_upto16_, mirroring the Ice Lake donor with predicated loads and interleaved AES streams - was implemented, validated, and benchmarked on Graviton 5, then deleted for failing the 1.5x gate over serial on short-token workloads (~0.9-1.16x, tied). The intersection is probe-bound, not hash-bound: Ice Lake wins on x86 only because it also vectorizes the table probe with 64-bit gathers and scatters, and at the 128-bit vector length SVE gathers are a documented trap on this core, so batching the hash alone cannot clear the bar. The serial passthrough stays, with the design, the measured numbers, and the re-entry condition recorded in the stub.
Four dispatch decisions were accidents of code order or stale gates; all are now explicit and backed by Graviton 5 numbers, and the tie-break policy is documented beside sz_sve_wider_than_neon_: a scalable kernel dispatches unconditionally only when it wins at the minimal 128-bit length on the mixed corpus; width-dependent winners wait behind the gate; and families whose scalar walk beats every 128-bit front install no Arm SIMD at all. The SVE argsort block preceded NEON's unconditionally, so NEON silently won by ordering - now it wins by measurement (53 vs 43 MiB/s at 128 bits) and the scalable kernel waits for wider registers. Byte search keeps its unconditional SVE dispatch deliberately: predicated heads make short inputs nearly twice as fast as NEON (1.55 vs 0.82 GiB/s) while long scans tie. The whitespace scan returns to NEON at 128 bits - the load-view rework overtook the scalable front on the mixed corpus (737 vs 663 MiB/s). Grapheme clustering installs no Arm SIMD front: clusters are dense enough that the scalar walk leads every corpus at every capacity (mixed 185 vs 64 NEON / 101 SVE2 MiB/s), so the vector kernels stay compiled, tested reserves until wider registers flip the economics. The multiseed hash keeps NEONAES with the reason recorded inline.
The NEON decoder classified every window through the general engine regardless of shape, trailing the scalar reference on all multibyte scripts. Two fast lanes now sit in front of it, ported from the SVE2 decoder's design. On an E0..EF lead one vld3q_u8 structure load deinterleaves sixteen (lead, continuation, continuation) triples - the steady state of CJK, Hangul, and Thai prose - and a handful of range compares validates all of them at once, including the E0-overlong and ED-surrogate first-continuation bounds; the clean prefix assembles with zips and widening stores, no compaction. The tile accepts a prefix covering the whole attempt or at least four runes, so separator-dense scripts do not pay for failed attempts, and it declines entirely within 48 bytes of the buffer end so the unpredicated load never over-reads. Windows that mix ASCII with well-formed 2- and 3-byte sequences - every alphabetic script and their mixes - are proven clean by a few shifted-mask tests (the leads' owed-lane smear must equal the continuation pattern, every lead must see its continuations, and the E0/ED bounds must hold) and skip the general classification, with the 3-byte machinery built only when such leads are present. Both lanes leave the maximal-subpart U+FFFD handling to the unchanged general engine. Every multibyte corpus gains: Chinese 335 to 643 MiB/s, Japanese 336 to 658, Thai 378 to 650, mixed multilingual 528 to 653, Russian 394 to 481, Korean 346 to 435; English and pure ASCII hold at 1.44 and 2.8 GiB/s. The scalar reference keeps a lead on some dense scripts - the remaining gap is windows with astral content that decline both lanes - which only matters on pre-SVE2 cores where this path dispatches.
Re-measured every Arm row on an AWS Graviton5 c9g instance (Neoverse-V3, 128-bit SVE2), replacing the Graviton3 figures. Substring search reaches 10.9 GB/s (3.6x LibC strstr), byte-set line splitting 4.3 GB/s, argsort 0.92 s over eight million words, random generation 678 MB/s, and the new Unicode rows gain their Arm numbers - case folding 0.8 GB/s and uncased search 2.1 GB/s (30x ICU). The provenance note now records the exact instance and core. The parallel-layer StringZillas edit-distance and alignment rows are left at their prior figures: the similarities matrix benchmark returns an engine-status error on this box that needs a separate look, and that layer is out of the current release's scope.
Merges origin/main-v5 into the local main-v5, joining two independent
lines of work with no overlapping files.
From this branch - the Graviton 5 (Neoverse-V3, 128-bit SVE2) campaign,
eighteen commits:
- New SVE2 kernels completing the UTF-8 family: grapheme clusters,
sentence and line boundaries, case folding, and case-insensitive
search now have scalable Arm siblings alongside the existing runes,
tokens, norm, and word-break kernels.
- Re-vectorized the delimiter scan on both SVE2 (64-byte windowed
streaming, distinct-high in-register membership) and NEON (the same
distinct-high scheme replacing a per-lane stack walk), and closed
the decode residuals with a third clean-window lane and a NEON vld3
tile - purging the vectorized-detector/serial-handler shortcuts.
- A correctness fix for the cased-rune probe across five backends,
dispatch tie-breaks aligned with measured 128-bit throughput, and a
refreshed Arm benchmark column in the README.
From upstream - six commits: a byte-order fix in the serial short-hash
tail ladder, 32-bit-safe argsort indices in the wheel tests, POWER VSX
lead-classify and Clang 20 cast fixes, Debian arch recognition, and
package-export hygiene.
Coverage after the merge: every UTF-8 family carries both a NEON and an
SVE2 kernel. On this 128-bit core the SVE2 sibling is dispatched where
it wins - decode, count, seek, delimiters, sentences - while NEON is
kept for the compare-bound families (newlines, whitespaces, line
breaks, case folding and search) and the scalar walk is kept for dense
grapheme clusters; the scalable kernels stay gated for the wider
registers that flip the balance. All backends build clean and the
filtered suites pass.
…line-break chunks The sentence-break and line-break classifiers ran the eight-gather BMP class cascade on every chunk, even when the whole 16-byte window is ASCII - the common shape in markup, code, and English prose. A shared helper now reads the class bytes from page 0 of the flat table with a cheap svtbl LUT when the entire chunk is below U+0100 (bit- identical to the gather at high=0, low=byte) and takes the full gather otherwise, so a pure-ASCII chunk skips the cascade while a mixed or multibyte chunk pays only one compare and a predicate test. The win is modest at the 128-bit vector length - the shared rule engine, not the classify gather, dominates these kernels there (English +2-5%, multilingual flat, bit-exact everywhere) - but it is free, and the skip grows more valuable as the register widens and the gather's fixed latency rises relative to the LUT. A first attempt that always computed the LUT and blended per lane regressed multilingual text and was replaced by the whole-chunk branch that never adds cost to a non-ASCII chunk.
The RISC-V decoder had an ASCII run lane and a vlseg2e8 2-byte lane but sent every 3-byte sequence - the steady state of CJK, Hangul, Devanagari, and Thai prose - through the general classify/gather window path, even though vlseg3e8 was available and unused. A 3-byte tile now mirrors the 2-byte sibling and the Arm svld3/vld3 decode tiles: one segmented load deinterleaves the lead and two continuation streams, the triples validate entirely in-register (lead in E0..EF, both continuations well-formed, and the first continuation tightened for the E0 overlong and ED surrogate ranges), a vfirst bounds the maximal good prefix, and the kept run assembles and stores with no gather. Whatever ends the run defers to the general window path, which keeps the maximal-subpart U+FFFD contract. Validated bit-exact against the serial reference under qemu-riscv64 over 200k mixed-script fuzz iterations at full and capacity-limited output, and confirmed the tile decodes a pure-CJK run directly.
…assify The AVX2 decoder classified every lane of every window through the length LUT, overrun scan, range checks, and orphan promotion. A clean fast lane now sits in front of it, the AVX2 twin of the NEON and SVE2 clean lanes: when the whole decodable window is ASCII plus well-formed C2..DF 2-byte pairs and E0..EF 3-byte triples - the shape of Latin, Cyrillic, CJK, Hangul, Thai, and Vietnamese prose with ASCII separators - a handful of shifted-mask tests prove the structure (every start is ASCII or a 2-/3-byte lead, every lead owns its continuations, no continuation is orphaned, the E0-overlong and ED-surrogate first-continuation bounds hold) and the codepoints drain through the same path the general engine uses, skipping the per-lane classification that dominates its cost. Windows with any 4-byte or invalid lead, or any irregularity, decline to the unchanged general path at the same cursor. Validated bit-exact against the serial reference under qemu-x86_64 over 60k mixed-script fuzz iterations at full and capacity-limited output; the lane was confirmed to fire on Cyrillic, CJK, and mixed 2/3-byte inputs.
The shared byte-lane bridges the NEON and SVE/SVE2 kernels rely on - the predicate-to-u64 movemask multiplies and the vreinterpret-based lane extracts - assume little-endian lane-to-byte order and would mis-execute on a big-endian aarch64_be build, which the endianness probe already recognizes. The three dispatch macros now require !SZ_IS_BIG_ENDIAN_, so such a target falls back to the portable serial path instead of compiling in miscomputing kernels. Little-endian Arm (every shipping part, and MSVC AArch64) is unaffected; the merged serial short-hash byte-order fix is the precedent that this lane-order class of bug is real.
The AVX2 `sz_copy_haswell` used a bare `size_t` for its tail-skip counter, which is undeclared under `-ffreestanding -DSZ_AVOID_LIBC=1` (no `<stddef.h>`). It compiled only because the normal build pulls in libc; the no-libc configuration failed to build. Uses `sz_size_t`, the project's own width type, like every other counter in the file.
…lassify The AVX-512 twin of the NEON, Haswell, and SVE2 clean lanes: an ASCII + well-formed C2..DF 2-byte + E0..EF 3-byte window is proven with a handful of shifted 64-bit-mask tests and drained through the same path the general engine uses, skipping the per-lane validity, orphan, and maximal-subpart machinery. A window with any 4-byte or invalid lead, or any irregularity, declines to the unchanged general path at the same cursor. Bit-exact by construction: a genuinely clean window makes the general path build the identical emit_starts, zero ill_formed, and declared consumed_length this feeds to sz_utf8_rune_drain_icelake_. The instruction-identical Haswell twin is validated bit-exact under qemu-x86_64 over 60k fuzz iterations; Ice Lake's AVX-512-VBMI cannot be executed under emulation, so this is compile-verified under the full VBMI/VBMI2 feature set and covered at runtime by real-x86 CI.
… classify The LASX twin of the NEON, Haswell, Ice Lake, and SVE2 clean lanes: an ASCII + well-formed C2..DF 2-byte + E0..EF 3-byte window is proven with shifted 32-bit-mask tests and drained through the existing path, skipping the per-lane length and range classification. Any 4-byte or irregular window declines to the unchanged general path at the same cursor. Validated bit-exact against the serial reference under qemu-loongarch64 over 60k mixed-script fuzz iterations; the lane was confirmed to fire on Cyrillic, CJK, Hangul, and Greek prose.
…sify The VSX twin of the NEON, Haswell, Ice Lake, and SVE2 clean lanes, between the existing ASCII lane and the general classify: an ASCII + well-formed 2-byte + 3-byte window is proven with shifted lane-mask tests and drained through the existing path, skipping the length and range classification. Irregular windows decline unchanged; the big-endian serial path is untouched. Validated bit-exact against the serial reference under qemu-ppc64le over 60k mixed-script fuzz iterations, firing on essentially every window of clean prose.
…serial stride The v128 decoder vectorized only the ASCII prefix and handed the entire multibyte, malformed, and truncation remainder to sz_utf8_decode_serial over a span - the last live instance of the vectorized-front / serial- handler pattern the other backends had already shed. It now carries a real vector multibyte decoder built on the existing 64-byte window substrate, mirroring the POWER VSX and LASX general decoders: a single- source high-nibble length classification, branchless overrun defer, well-formed plus orphan-promotion mask algebra shared verbatim with the other backends, a maximal-subpart length table, and a gather-free drain that left-packs the emitted starts with swizzles and width-blends the 1/2/3/4-byte codepoints. The public entry keeps the ASCII prefix peel, loops the window decoder, and finalizes a declined window-edge truncation through the shared bounded maximal-subpart helper - no serial span call remains. Validated bit-exact against the serial reference under wasmtime: the 60k mixed-script fuzz differential (full and capacity-limited) plus an exhaustive sweep of all one- and two-byte strings, all ~16.8M three-byte strings, and 400k window-straddling inputs at every output capacity all matched.
`sz_utf8_fold_lead_family_t_` was defined three times with identical values: in `serial.h`, in `icelake.h`, and under a back-end prefix in `haswell.h`. `utf8_uncased_fold.h` includes `serial.h` and `icelake.h` unconditionally and back to back, so every translation unit reaching `stringzilla.h` failed to compile on x86-64, breaking nine CI jobs. Keep the definition in `serial.h`, the arch-neutral header every back-end already includes, and let Ice Lake and Haswell consume it alongside NEON and SVE2. The per-back-end rationale moves onto the code it explains: the `VPERMB` indexing note onto the Ice Lake table, the compare-tree and `VPMOVMSKB` note onto the Haswell flag construction. Rename the flags to the `_k` suffix every other enum constant in the tree uses and drop the private `_t_` marker from the type.
Twenty-one private types carried a trailing underscore after `_t` to mark them internal - four in `hash/`, sixteen in `utf8_uncased/`, one in `utf8_uncased_fold/`. The marker reads poorly and buys nothing: these types are already unexported implementation details, and the public enums have always used a plain `_t`. Rename them all to `_t`. No collisions with existing names, and the change is a pure identifier substitution.
Thirty `SZ_UTF8_NORM_*` object-like macros become grouped enums with the `_k` suffix. Six of them were hand-computed from others - the three-value `bits`/`size`/`mask` triples for both tries, and the two Hangul counts - and are now derived in the enum body, so retuning a stage means editing one number instead of keeping three in sync. Enum constants are also visible to a debugger. Twenty-one of the macros lived in the generated `tables.h`. They were never emitted by the generator recipe, which derives array contents only, so they move to `serial.h` beside the nine Hangul constants already there. That leaves `tables.h` holding nothing but generated data and lets the full Hangul family share one enum, where `S_COUNT = L_COUNT * N_COUNT` is expressible. Rescope the `do NOT edit by hand` banner accordingly: it now marks the array initializers as generated and states plainly that the surrounding code and comments are hand-written and editable, and that the recipe emits table data rather than the file. Add a Hangul sweep to the normalization oracles. Hangul is algorithmic and absent from the tables, so neither `NormalizationTest.txt` nor the ICU sweep pinned the syllable-block constants; the new sweep walks the block precomposed and decomposed against `unicodedata`, overrunning it on both sides so an off-by-one syllable count surfaces as an unassigned codepoint decomposing.
`push` and `pull_request` both filtered on `main*`, but the two filters match different things: `push` matches the branch being pushed, `pull_request` matches the PR's base. A branch like `main-v5` with a PR into `main` satisfied both, so every push built the same commit twice - around 45 redundant runner VMs each time, enough to leave the queue backed up for ten minutes. Restrict `push` to `main` alone. Branches with an open PR are still covered through `pull_request`, and fork PRs keep their coverage, which they would lose if `push` were kept as the primary trigger instead. Add a concurrency group keyed on the source branch. `head_ref` on a PR event and `ref_name` on a push both name that branch, so any number of PRs from one branch share a group. Cancellation is limited to PR events, leaving a `main` build that feeds the release pipeline to finish.
A single Linux wheel job built every architecture through cibuildwheel's `archs = ["all"]`, pulling around eleven manylinux and musllinux images onto one runner. That exhausts the ~14 GB of disk and dies with `no space left on device` after roughly two hours of QEMU emulation - the failure that surfaced once the compile was unblocked and the wheel jobs finally ran. Split the Linux wheels by architecture class in both workflows. Native x86_64 and native aarch64 each build on their own runner - aarch64 on `ubuntu-24.04-arm` with no emulation, matching the CUDA job that already does this - and the emulation-only arches (ppc64le, s390x, i686, armv7l, riscv64) get one runner each, so no runner pulls more than a single image. Every architecture still builds all five CPython versions, so PR coverage is unchanged. Under emulation the fuzz baselines run at `SZ_TESTS_MULTIPLIER=0.05`, the depth the C++ `test_cross_qemu` job already uses; native jobs keep full depth. The release publish and status jobs are rewired to the new split job keys, and the per-job artifact names stay within the `cibw-wheels-<pkg>-*` globs the publish steps download by. Refresh the actions while here: `setup-python` to v7.0.0 (drops only the unused `pip-install` input), `checkout` to v7, and the prerelease artifact actions up to the upload v7 / download v8 pairing release already uses. Also fix a `setup-python` reference that had lost its `actions/` owner prefix.
The Windows wheel jobs relied on cibuildwheel's implicit per-runner defaults, which left two gaps. `stringzilla` set no `CIBW_ARCHS_WINDOWS`, so on the Intel runner it built AMD64 and 32-bit x86 and shipped no ARM64 wheel at all. `stringzillas-cpus` requested `AMD64 ARM64` on the same Intel runner, so its ARM64 wheel was cross-compiled and never tested - the Windows-on-Arm platform cannot run ARM64 binaries on an Intel host. In `release.yml`, split both into an explicit x86 job on `windows-2022` and an ARM job on the native `windows-11-arm` runner, mirroring the Linux `_x86` / `_arm` layout and setting `CIBW_ARCHS_WINDOWS` explicitly rather than leaning on defaults. On the native ARM runner the ARM64 wheels are built and tested natively. The ARM matrix starts at 3.11, the first CPython with an official Windows ARM64 build. In `prerelease.yml`, add `windows-11-arm` to the `build_wheels_native` matrix - excluding 3.10 for the same reason - so every PR validates the Windows ARM64 wheel path release now ships, instead of it first running live at a release. `stringzilla` now ships a tested Windows ARM64 wheel it never produced before, and the `stringzillas-cpus` ARM64 wheel is tested instead of built blind. The publish and status jobs are rewired to the four new keys, and the artifact names stay within the `cibw-wheels-stringzilla-*` and `cibw-wheels-stringzillas-cpus-*` globs the publish steps download by.
ashvardanian
changed the title
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
1 participant
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
v4 vectorized sorting, hashing, and edit distances.
v5 goes after the part everyone agreed was too irregular for SIMD — Unicode — plus the hardware and packaging that make it portable.