Skip to content

64-bit hygiene: string length into the base module, wider memcpy/memcmp, and three rules that catch the truncation - #3576

Merged
borisbat merged 10 commits into
masterfrom
bbatkin/lints-and-64bit
Jul 27, 2026
Merged

64-bit hygiene: string length into the base module, wider memcpy/memcmp, and three rules that catch the truncation#3576
borisbat merged 10 commits into
masterfrom
bbatkin/lints-and-64bit

Conversation

@borisbat

Copy link
Copy Markdown
Collaborator

Five related pieces of 64-bit hygiene, plus the tooling fix that made the rest safe to work on.

1. length on a string moves to the base module

This did not compile:

options gen2
[export] def main() {
    let s = "hello"
    print("{length(s)}\n")   // error: no matching function
}

...while empty(s) right next to it did. empty was already a base-module builtin; length was in strings. That split is arbitrary — the two belong together. Both string length overloads now bind in the base module beside empty, and the snippet above compiles.

The declarations move from aot_builtin_string.h to aot_builtin.h. That part is forced: a script can now reach length with no require at all, so AOT output for it does not include the strings header, and the declaration has to live in one that is always pulled in via aot.h. Verified by generating AOT for a script with and without require strings and diffing the include list.

require strings keeps working unchanged — base-module symbols are always visible, so nothing needed a re-export, and adding one would have made every call ambiguous.

2. The 64-bit surface those expose

  • long_length(string) / long_length(das_string). long_length previously covered only array and table. Without this, moving length would have left int64(length(str)) with no correct replacement for the new lint to suggest. Note stringLength() is itself uint32_t(strlen(...)), so these use strlen directly rather than reusing it.
  • memcpy and memcmp over uint, int64, uint64, plus the const-source twins AOT needs. das has no implicit promotion, so without these a size that is already 64-bit could only reach memcpy through a narrowing cast.

The int-returning string lengths now carry the same always-on DAS_VERIFYF guard that array and table length already use, so a string past 2^31 stops instead of returning a wrapped negative. No Context threading was needed — the existing array/table guards showed the pattern.

Signatures of the two pre-existing C++ functions are unchanged, so embedders calling them directly keep compiling.

3. LINT017 — 64-bit cast of a call with a long_ counterpart

let n = int64(length(arr))   // wraps at 2^31 BEFORE the cast runs
let n = long_length(arr)     // 64-bit throughout

The inner call returns int, so the truncation has already happened by the time the widening cast executes. The cast looks like it buys range and buys nothing.

The pair table is hardcoded deliberately. An "exists" check is not meaningful for user functions — a macro can add or remove them mid-compile — so the only well-defined domain is the builtin/daslib set, which is enumerable. Each pair is additionally gated on the receiver type, which is what keeps a same-named user overload silent, along with the fixed-array length generic in daslib/builtin.das that genuinely has no long_ twin.

4. LINT018 — narrowed memcpy / memcmp size

Now that (2) added the wider overloads, an int(...) cast on the size argument is pure loss:

unsafe(memcpy(dst, src, int(nbytes)))   // nbytes : int64 — truncates past 2GB
unsafe(memcpy(dst, src, nbytes))

Both LINT017 and LINT018 are LINT rather than PERF: this is code not handling 64 bits correctly, not code that is merely slow.

5. STYLE036 — inert type contract on a cast target

-const, -&, -[], -#, ==const and ==& are substitution contracts. They do work only while a generic binds — void? is void? regardless of -const.

What makes this exact rather than heuristic: inference clears a contract when it consumes one (ast_typedecl.cpp:223,422). So a flag still set at lint time is itself proof the contract was never consumed. The generic case removes its own evidence.

The one exclusion is an auto or still-unresolved alias target, where substitution has not run yet — linq's reinterpret over ARGT -const really does strip const. Adding that exclusion took the daslib hit count from 179 to 5. A concrete typedef is not such a case: with typedef CI = int const, casting to CI? -const keeps the const, so the contract is inert there too and the rule correctly fires. That was established by probing describe() output post-inference, not assumed.

6. MCP grep_usage fix (why it is in this PR)

resolve_path(".") returned the repo root with a trailing /. appended, and ast-grep cannot relativize candidate paths against a root spelled that way. Every --globs pattern containing a / silently matched nothing and reported "Found 0 matches" — indistinguishable from a genuine no-usages answer. Basename and **/-prefixed globs were unaffected, which is why no existing test caught it.

Three of the rules above are search-heavy, so this was fixed first rather than working on top of a search that silently lies.

What the new rules found

Everything they flagged in daslib/ is fixed here: 5 STYLE036 (dead -const on concrete reinterpret targets), 2 LINT017 (genuine truncation in debug.das / debug_eval.das), and 2 STYLE030 — require strings entries that existed only for length(str) and became unused once it moved. daslib/ lints clean.

Repo-wide the rules find 172 LINT017, 79 LINT018, and only 3 STYLE036 outside daslib. Spot-checked and real, e.g. memcpy(..., int(dim)) with dim : int64 in dasLLAMA and int64(length(take)) in the dictate example. These are pre-existing and untouched here — CI lints changed files only, and for scale the same sweep reports 4908 pre-existing PERF020. They will be picked up as those files are next edited.

Testing

Fixtures under utils/lint/tests/ for each rule (auto-skipped by the lint runner, so intentional violations do not trip the gate), and a grep_usage regression test that fails on the exact silent-zero glob.

preflight --full came back 13 passed / 3 failed; all three were local-environment artifacts, fixed and re-validated with the targeted gates:

  • docs/sphinx-latex + docs/sphinx-htmlmyst_parser was missing from the local Sphinx env (and installed against the wrong interpreter on the first attempt). Both builders re-run clean with -W --keep-going.
  • tests-aoterror[50101], and the hash list in the diagnostic named the very methods this PR adds to LintVisitor. Editing daslib does not invalidate per-test _aot_generated stubs, so they still encoded the old class hash. Wiped all 91 stub dirs and rebuilt the full test_aot (1344 TUs, 0 errors) — tests-aot and tests-interp then both pass. CI builds from a fresh checkout and cannot hit this; skills/aot_hash_desync_debugging.md gains a note that preflight does not wipe for you, plus the tell for distinguishing staleness from a real desync.

Pushed with --no-verify, which is the documented path in skills/make_pr.md for a failed run (fix → targeted re-validation → announced no-verify) rather than burning a second full run.

Known gap

STYLE035 has no section in lint.rst — pre-existing drift, not introduced here, and left alone to keep this diff on topic.

🤖 Generated with Claude Code

borisbat and others added 8 commits July 26, 2026 15:55
resolve_path(".") returned "<root>/." (path_join does not normalize), and
ast-grep cannot relativize candidate paths against a root spelled that way.
Every --globs pattern containing a "/" therefore matched nothing and the tool
reported "Found 0 matches" -- indistinguishable from a genuine no-usages
answer, which is the dangerous part. Basename globs and "**/"-prefixed ones
were unaffected, so the breakage was invisible in the existing tests.

grep_usage now takes das_root directly when no directory is given, and
resolve_path normalizes its result so any caller passing "." (or a path with
doubled separators) is covered too. path_join and normalize both round-trip
through std::filesystem::path::string(), so Windows separator behavior is
unchanged.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
length(str) needed `require strings` while empty(str) did not, which is an
arbitrary split -- the two belong together. Both string length overloads now
bind in the base module next to empty(). The declarations move to
aot_builtin.h rather than staying in aot_builtin_string.h: a script can now
reach length() with no require at all, so AOT output for it does not include
the strings header, and the decl has to live in one that is always pulled in
via aot.h.

Adds the missing 64-bit surface these expose:
  - long_length(string) / long_length(das_string) -- long_length previously
    covered only array and table, so after the move int64(length(str)) would
    have had no correct replacement to suggest.
  - memcpy / memcmp over uint, int64 and uint64, plus the const-source twins
    AOT needs. das has no implicit promotion, so without these a size that is
    already 64-bit could only reach memcpy through an int(...) cast, which
    truncates above 2GB.

The int-returning string lengths now carry the same always-on DAS_VERIFYF
guard that array and table length already use, so a >2GB string stops instead
of silently returning a wrapped negative. Note stringLength() itself is
uint32_t(strlen(...)), so long_length uses strlen directly rather than
reusing it.

Signatures of the two pre-existing functions are unchanged (builtin_string_length
keeps its now-unused Context*) so embedders calling them directly keep compiling.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
LINT017 flags int64(length(x)) / uint64(capacity(x)) and friends: the inner
call returns int, so the wrap past 2^31 already happened before the widening
cast ran. The cast looks like it buys 64-bit range and buys nothing.

LINT018 flags int(...) on a memcpy/memcmp size argument, now that both carry
uint/int64/uint64 overloads and the narrowing is pure loss.

Both are LINT rather than PERF -- this is code not handling 64 bits
correctly, not code that is merely slow.

The LINT017 pair table is hardcoded on purpose. An exists-check is not
meaningful for user functions (a macro can add or remove them mid-compile),
so the only well-defined domain is the builtin/daslib set, which is
enumerable. Each pair is additionally gated on the receiver type, which is
what keeps a same-named user overload silent -- and the fixed-array length()
generic in daslib/builtin.das, which genuinely has no long_ twin.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
-const / -& / -[] / -# / ==const / ==& are substitution contracts. They do
work only while a generic binds, and infer clears them when it consumes them
(ast_typedecl.cpp:223,422). A cast target that is already concrete has nothing
to consume the contract, so a flag still set at lint time is proof it did
nothing -- void? is void? regardless of -const.

That "infer clears what it consumes" property is what makes the rule precise
rather than heuristic: the generic case removes its own evidence. The one
remaining exclusion is an auto or still-unresolved alias target, where
substitution has not happened yet -- linq's reinterpret<ARGT -const> really
does strip const from whatever ARGT binds, and skipping those took the daslib
hit count from 179 to 5.

Note a concrete typedef is NOT such a case: reinterpret<CI? -const> where
CI = int const keeps the const, so the contract is inert there too and the
rule correctly fires.

Also fixes everything the three new rules found in daslib:
  - 5 x STYLE036, all dead -const on concrete reinterpret targets
  - 2 x LINT017, genuine 32-bit truncation in debug.das / debug_eval.das
  - 2 x STYLE030, `require strings` that existed only for length(str) and
    became unused when length moved to the base module

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
lint.rst gains a section per rule. skills/strings.md said `strings` provides
`length` -- that is now false, so it is corrected in place along with a note
that the int-returning form panics rather than wrapping.

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

install/CLAUDE.md gets the three new rule rows, and both CLAUDE.md files
record that length/empty on a string need no require -- that is the fact the
old split made confusing.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
The other pairs are all receiver-type gated; fread/fwrite were matched on
arity alone, so a user-defined 2-arg fread would have been told to call a
long_ twin that does not exist. A repo-wide sweep shows only length (170) and
capacity (2) ever fire in practice, so this costs no real coverage.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
A daslib edit made after the last build gives preflight's tests-aot gate a
50101 that CI, building from a fresh checkout, will never reproduce. Adding a
method to a widely-required class is enough to trigger it -- the class's
semantic hash is an input to every stub requiring it. Records the tell (the
diagnostic's hash list names functions you just added) and the wipe.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Copilot AI review requested due to automatic review settings July 26, 2026 23:55

Copilot AI left a comment

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Pull request overview

This PR improves 64-bit correctness and ergonomics across daslang’s core surface area by moving string length into the base module, adding 64-bit memcpy/memcmp overloads, and introducing new lint/style rules (with docs) to detect common truncation and inert-type-contract patterns. It also fixes an MCP grep_usage path-normalization bug that could silently zero results when using slash-anchored glob filters.

Changes:

  • Move string length/long_length builtins into the base runtime module (next to empty) and adjust AOT headers accordingly.
  • Add wider memcpy/memcmp overloads (uint/int64/uint64) and new lint rules (LINT017/LINT018) + style rule (STYLE036) with documentation updates.
  • Fix MCP grep_usage root handling and path normalization; add a regression test.

Reviewed changes

Copilot reviewed 26 out of 26 changed files in this pull request and generated 1 comment.

Show a summary per file
File Description
utils/mcp/tools/grep_usage.das Stops using resolve_path(".") as default search root; uses get_das_root() to avoid /. root spelling.
utils/mcp/tools/common.das Normalizes resolved paths to avoid /. and duplicate separators impacting ast-grep glob matching.
utils/mcp/test_tools.das Adds regression test ensuring slash-anchored glob filters don’t silently zero results.
utils/lint/tests/style036_inert_cast_modifier.das Adds a fixture file demonstrating STYLE036 cases (lint fixture directory is skipped by the lint tool).
utils/lint/tests/lint018_mem_size_narrowing.das Adds a fixture file demonstrating LINT018 cases.
utils/lint/tests/lint017_long_counterpart_cast.das Adds a fixture file demonstrating LINT017 cases.
src/builtin/module_builtin_string.cpp Removes registration/defs for string length from the strings module.
src/builtin/module_builtin_runtime.cpp Registers string length/long_length in the base module; adds wider memcpy/memcmp extern overloads.
skills/style_lint.md Documents the new STYLE036 rule in the style lint skill page.
skills/strings.md Updates string-skill guidance to reflect length/empty living in the base module.
skills/aot_hash_desync_debugging.md Adds guidance about local _aot_generated staleness vs real AOT hash desync.
install/CLAUDE.md Updates shipped instructions to reflect base-module string length/empty and new lint/style rules.
include/daScript/simulate/aot.h Adds AOT helper overloads for das_memcpy/das_memcmp with uint/int64/uint64 sizes (+ const-source variants).
include/daScript/simulate/aot_builtin.h Declares base-module string length/long_length functions for AOT output.
include/daScript/simulate/aot_builtin_string.h Removes string length decls from strings-specific AOT header.
doc/source/reference/language/lint.rst Documents LINT017, LINT018, and STYLE036 in the reference docs.
daslib/style_lint.das Implements STYLE036 detection for inert cast-target contracts.
daslib/lint.das Implements LINT017 (prefer long_ twins over widening casts) and LINT018 (avoid narrowing memcpy/memcmp sizes).
daslib/flatten_opt_preshade.das Removes inert -const from reinterpret target per STYLE036 intent.
daslib/flatten_opt_common.das Removes inert -const from reinterpret targets per STYLE036 intent.
daslib/delegate.das Drops require strings now that length is base-module accessible.
daslib/debug.das Fixes an identified truncation pattern by using long_length instead of widening length.
daslib/debug_eval.das Fixes an identified truncation pattern by using long_length instead of widening length.
daslib/async_boost.das Drops require strings now that length is base-module accessible.
daslib/array_boost.das Removes inert -const from reinterpret target per STYLE036 intent.
CLAUDE.md Updates repo instructions to reflect base-module string length/empty and new lint/style rules.

💡 Add Copilot custom instructions for smarter, more guided reviews. Learn how to get started.

Comment thread utils/mcp/tools/common.das Outdated
…call site

The normalize() I added to resolve_path broke extended_checks (windows, 64):
lexically_normal rewrites every separator to the platform preferred one, so on
Windows it turned the forward slashes of get_das_root() into backslashes and
test_resolve_load_modules_absolutizes_relative stopped finding the root as a
substring of the result. path_join only ever *appends* a separator, which is
why the two are not interchangeable -- my original reasoning that both merely
round-trip through std::filesystem::path::string() was wrong.

resolve_path goes back to exactly what it was, with a comment recording why it
must not normalize. grep_usage instead maps both spellings of "the repo root"
("" and ".") to get_das_root() directly, which is all the original bug needed;
the shared helper's other 25 call sites are untouched. Adds a test for the "."
spelling alongside the existing empty-directory one.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Copilot AI review requested due to automatic review settings July 27, 2026 00:37
@borisbat

Copy link
Copy Markdown
Collaborator Author

CI round 1: 36/37 green, one real failure — extended_checks (windows, 64), and it was mine.

test_resolve_load_modules_absolutizes_relative broke because of the normalize() I had added to the shared resolve_path. lexically_normal() rewrites every separator to the platform-preferred one, so on Windows it turned the forward slashes of get_das_root() into backslashes and the test's "result contains the root as a substring" invariant stopped holding. path_join only ever appends a separator — my original reasoning that the two merely round-trip through std::filesystem::path::string() was wrong, and it happened to be invisible on macOS where both spellings agree.

Fixed in b57426e by shrinking the change rather than patching around it: resolve_path reverts to exactly its previous behaviour (with a comment recording why it must not normalize), and grep_usage maps both spellings of "the repo root" — "" and "." — straight to get_das_root(). That is all the original silent-zero bug ever needed; the helper's other 25 call sites are untouched now.

Added a test for the "." spelling next to the existing empty-directory one. Locally: test_tools 419/419, test_crosstree_guard 11/11, and the original failing glob still returns its 100 matches.

Copilot AI left a comment

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Pull request overview

Copilot reviewed 26 out of 26 changed files in this pull request and generated 1 comment.

Comments suppressed due to low confidence (1)

utils/mcp/tools/common.das:528

  • resolve_path()’s new comment says normalize() is avoided because it would rewrite separators on Windows and break comparisons against get_das_root(), but resolve_path already calls path_join(), whose C++ implementation uses std::filesystem::path(...).string() (native separators). If get_das_root() is in generic / form, resolve_path still risks returning \ paths on Windows. Consider converting the joined path back to generic form (and update the comment accordingly).
    // Deliberately NOT normalize()d: lexically_normal rewrites every separator to the
    // platform preferred one, so on Windows it would turn the forward slashes of
    // get_das_root() into backslashes and break callers that compare against it.
    return path_join(get_das_root(), path)

Comment thread daslib/lint.das
Copilot review catch. The diagnostic said the value "still wraps past 2^31",
but array/table/string length now stop on an always-on DAS_VERIFYF guard --
added in this same PR -- so for the most common pairs it panics rather than
wraps. The accurate statement is that the cast widens an already-32-bit
result, so the limit is reached inside the callee before the cast runs;
whether that shows up as a wrap or as a panic is secondary.

Reworded the diagnostic, the fixture header, lint.rst and both CLAUDE.md rows.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Copilot AI review requested due to automatic review settings July 27, 2026 00:45

Copilot AI left a comment

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Pull request overview

Copilot reviewed 26 out of 26 changed files in this pull request and generated no new comments.

@borisbat

Copy link
Copy Markdown
Collaborator Author

CI round 2 on 32b3efa2: 35 pass / 0 fail / 4 skip — green. The 4 skips are the nightly-only Windows toolchains (mingw, clang-cl, and the two *_nightly cells), which are gated off per-PR by design.

Two corrections to the description above, both from this round:

  • §3's code comment reads // wraps at 2^31 BEFORE the cast runs. More precisely: the cast widens an already-32-bit result, so the limit is reached inside the callee — as a wrap for count / find_index, or as the always-on panic guard for array/table/string length/capacity that §2 adds. Copilot caught the diagnostic saying "wraps" for cases this PR makes panic; fixed in 32b3efa2 across the message, fixture, lint.rst and both CLAUDE.md rows.
  • §6 describes the bug correctly, but the fix is narrower than the first push: resolve_path is unchanged, and only grep_usage maps "" / "." to get_das_root(). The intermediate version hardened resolve_path with normalize() and broke extended_checks (windows, 64) — see the comment above.

@borisbat
borisbat merged commit d41cf11 into master Jul 27, 2026
39 checks passed
borisbat added a commit that referenced this pull request Jul 27, 2026
…oE expert chain, fleet boards (#3579)

* skills/babysit: don't re-run full preflight per review round — CI is the gate

The babysit loop was re-running full preflight (~20 min) before every push,
even for a 6-line review-fix commit. CI runs the whole matrix on every push
for free, so per round only the fast local gates that CI is unforgiving about
(compile/format/lint on the changed files, + docs/JIT smoke only if that class
was touched) are worth running — everything else is left to CI. Full preflight
stays only for a genuinely large change. Documents `git push --no-verify` as
the sanctioned bypass of the per-round preflight token (deliberate + CI-covered,
not the silent skip the make_pr rule bans), and how to tell an infra-canceled
runner red from a real one.

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

* dasLLAMA metal prefill: gptoss fusion lane — bias folds into the moe mm seed, act ghost-tail cut, prefix folds into bucket

Three structural cuts on the gptoss pp512 critical path (Wave 1 fusion lane):

- MetalMoeMulMmMx4 takes wb/hasb and seeds its accumulator tiles with stride-0
  simdgroup_loads of the expert's bias row — the three per-layer MetalMoeBiasRows
  dispatches (the 196K-tg early-exit launch storm, 15.1ms of window) fold away;
  the moe_bias chase arm now knocks the fold out (hasb=0) for attribution.
  MetalMoeBiasRows deleted.
- swiglu_oai prefill twin (MetalSwigluOaiPf) exits ghost rows past the last
  bucket's padded end on a uniform-broadcast liveness bound (~24% of mpad at 512;
  -1.6ms). The 3-binding MetalSwigluOai stays as-is — the decode rail's enc_ew2
  dispatches it with slots 0-2 (the fam-gptoss forced-step cell caught the shared
  source the first time around).
- MetalMoePrefix (1-tg serial walk) folds into MetalMoeBucket: each bucket tg
  computes its expert's padded-prefix base in-tg and publishes basep[e] — one
  dispatch and one serial barrier level per layer gone, bit-identical CSR.

gptoss metal pp512: GPU window 512.2 -> 495.2ms, 998 -> 1030.9 t/s best-window;
adjacent lcpp_bench pairs read 1032.6 vs 1033.8 (0.999, statistical tie, was
0.966) with tg128 1.045-1.064 held. Prefill dispatches 650 -> 626. 30B near-tie
re-paired clean (pp 1.018, tg 1.138). Gates: kernels suite, prefill base/kq/qkv,
all five MoE fam parity arms (PARITY_FULL) green.

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

* dasLLAMA records: gptoss metal card re-paired on the fusion lane — pp512 tie

das metal 1031.5 +/- 2.1 vs lcpp stock 1033.8 +/- 2.4 => pp512 0.998 (board 1.00),
tg128 88.1 vs 84.4 => 1.044 held; official rail, adjacent warm pair, tripwire silent
(cv 0.2%). Also re-merges site/files/dasllama/bench_records.json, which was stale
since before #3564 (still carried the 945.4 / 0.92 card).

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

* dasLLAMA records: first M4 Pro board — 8 models x 3 configs, 45/48 green

Anton's M4 Pro mini (10P+4E, 64GB, macOS 15.2), full pub_catalog sweep with
adjacent llama-bench refs at the pinned sha (ebd048f), fresh full tune
(sdot kstep2 family holds; smmla measured live and declined — 147.5 vs
165.9 GMAC/s, still one MMA pipe on M4). Reds, all pp: gpt-oss metal 0.93,
E4B metal 0.98, E4B accel 0.99. Standouts: 35B metal tg 1.27, 30B accel
pp 1.78, gpt-oss cpu pp 1.76.

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

* dasLLAMA records: M3 Air subset board — 16GB-class sanity pass (12B/E4B/gpt-oss)

Irina's M3 Air (4P+4E, 16GB, macOS 26.4.1), native build + native pinned refs.
CPU legs all green (gpt-oss accel pp 1.71); metal slightly red on the small GPU
(12B 0.97/0.94 — M1-tuned tiles + passive thermal); gpt-oss metal legs honestly
absent (12.1GB exceeds the 16GB box's Metal working set).

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

* idot P2: native aarch64 sdot lowering + dot_q8q8 GEMV migrated to plain-das idot4 — hand-NEON sdot4x4 leg retired

The P1 assumption "backends pattern-match the generic idot4 lowering to native
dots" fails on AArch64: the stride-4-shuffle widen-multiply form compiles to a
zip/uzp/smull expansion, zero SDOTs (probe-disassembled), while the plain
autovectorized loop DOES get sdot from the vectorizer. So the ledgered per-ISA
strengthening lands now:

- llvm_jit_intrin: idot/idot4 signed x signed emit ONE @llvm.aarch64.neon.sdot
  (idot adds a v4i32 reduce -> addv) on the host-features rail —
  g_target_host_features is exactly when +dotprod is force-appended, so the
  generic/cross rail (kernel-free exes) keeps the portable IR and can never hit
  "Cannot select". unsigned x signed needs USDOT (i8mm, absent on M1) — rides
  the i8mm un-gate follow-up. LLVM_JIT_CODEGEN_VERSION 0x48 -> 0x49.
- dasllama_math_default: new dot_q8q8_idot4x4/_f16s — the 4-accumulator GEMV
  dot in plain das (byte16 loads + idot4), bit-identical fold to the hand-NEON
  sdot4x4 leg (exact int sums, same float order).
- dasllama_math_aarch64_neon: the hand-NEON dot_q8q8_sdot4x4/_f16s are DELETED;
  the arm64-sdot backend's q8q8 kernels dispatch the plain-das dots. NEON
  intrinsics remain only for the mx4 + laneq tiles. dasllama_math_gen row tails
  switch with it.

Interleaved A/B (M1 Max, 8 P-cores, DRAM-rotated GEMV, warm adjacent cells):
single-thread idot4x4 50.9-51.8 GB/s vs hand-NEON 50.2-50.9 vs tuned-autovec
vec16 35.9-38.5 (n=2048 d=8192 and n=2816 d=1408); threaded all cells tie at
the bandwidth ceiling (~102/112 GB/s) — board-neutral, kernel-parity proven.
1-acc idot rewrites of the template measured NO gain (per-block hsum drains
the dot chain) — the measured reason the autovec template stays for x64.

Gates: tests/type_lattice/test_lattice_idot interp+jit; dasLLAMA test_matmul,
test_matmul_batch, test_kernel_backend, test_groupn, test_kquant,
test_fused_decode, test_mxfp4, test_batch_decode, test_forward, test_prefill,
test_parity (45/45 with DASLLAMA_CPU_PREFILL=1; the 4 tripwire errors without
it reproduce identically on master) — all -jit. Riders: two pre-existing
LINT017s in llvm_jit_run.das (uint64(length) -> long_length).

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

* lattice per-lane >> + native tbl1 shuffle; dot_mx4q8 migrated to plain das (shuffle+idot4)

Audit: the 18 integer lattice vector types (byte2-16, ubyte2-16, short2-8,
ushort2-8) carried ctors/converts/equality/swizzles and NO shift or bitwise
ops (fp16 family has closed arithmetic; workhorse int/uint vectors have the
full bit surface). This adds exactly `>>` — shift by a scalar int, count
masked to the LANE width, signed lanes arithmetic / unsigned logical — the
nibble-unpack shape (`shuffle(lut, nib >> 4)`). No <<, no >>=, no & | ^
(deliberate minimal scope).

- vectypes.h das_sv_shr template + 18 addExterns (the addEquNeqVal rail:
  extern for interp/AOT, JIT overrides natively).
- JIT: is_lattice_op2_native routes int-lattice ">>" to visitExprOp2_NonMonade
  (else the builtin-address gate would drop the enclosing function to interp);
  the shift preamble masks/truncs/splats the count at lane width; AShr/LShr
  arms accept tInt8/tInt16/tUInt8/tUInt16 vector bases.
- shuffle: aarch64 now emits AND + one @llvm.aarch64.neon.tbl1 (baseline NEON,
  same gate as the tbl16 intrinsics) instead of 16 dynamic extract/inserts.
  LLVM_JIT_CODEGEN_VERSION 0x49 -> 0x4a.
- dot_mx4q8_template rewritten in plain das: ubyte16 nibble load, shuffle LUT
  lookups (low via the &15 mask, high via the new per-byte >> 4), idot4
  accumulate — replaces tbl16_lo/hi + sdot4_w. Bit-exact vs the scalar oracle;
  disassembly shows the IDENTICAL per-block instruction mix (2 tbl + 2 sdot +
  1 and + 1 ushr); isolated GEMV at gptoss shape (n=d=2880) 16.1 vs 16.3 GB/s
  (-1.5%, fallback-u2 hint tuned for the old body — per-box tune re-crowns).
  The [tuned] grid and sidecar keys are unchanged.
- conformance: gen_type_conformance gains has_shr + a "shift right" section
  (arithmetic vs logical, count masking, shift-by-zero) on the 18 lattice
  types + int2/uint4 controls; 20 files regenerated, 438/438 interp + jit.
- datatypes.rst: the lattice int families are now storage + converts + >>.

Measured reason NOT to migrate dot_q8q8kv: its vec16 template already
compiles to sdot via the LoopVectorizer and hits ~39-40 GB/s single-thread at
KV row lengths (n=64/128/256); the 4-acc idot4x4 shape loses 2x there
(per-call fixed cost at 3.5ns/row rows), and the 1-acc idot form measured
exactly at vec16 parity in the GEMV race - no headroom. KV dots stay on the
tuned templates.

Riders: pre-existing LINT017 sweep (uint64(length) -> long_length) in
llvm_jit.das + llvm_boost.das, and the llvm_boost describe() LINT016
(':=' -> '= clone_string').

Gates: tests/type_lattice 438/438 interp+jit; dasLLAMA test_mxfp4,
test_kernel_backend, test_groupn -jit green.

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

* lattice bit surface: parity with the 32-bit vector families (<< & | ^ + compound assigns)

Completes the >> commit to full addFunctionVecBit parity on the 18 integer
lattice vector types: per-lane << (lanes wrap like the C cast), bitwise
& | ^, and the five compound assigns (<<= >>= &= |= ^=). Same rail as >>:
das_sv_* templates in vectypes.h + addExterns (interp/AOT), native JIT
lowering (is_lattice_op2_native now routes the whole 10-op surface;
isBinary accepts int-lattice vectors — the ~ Op1 arm stays unreachable
for lattice, matching the workhorse families where ~ is also absent).
Shift counts mask to the lane width on every form.

Conformance: the shift-right section grows into shift-right/shift-left/
bitwise sections (arithmetic vs logical >>, << lane wrap pinned via
two's-complement expected values computed in the generator, count
masking, all five compound assigns) on the 18 lattice types + int2/uint4
controls — 478/478 interp + jit. datatypes.rst updated.

Real-.dlim A/B of the migration commits (branch vs master worktree, same
binary, lcpp_bench, tune/warm discard + timed adjacent pairs, M1 Max t8):
Llama-3.2-1B Q8 pp512 887/861 vs 874/844 tok/s (+1.4/+2.0% lean),
tg128 82.4/82.4 vs 82.4/82.3 (tie); gpt-oss-20b mx4 pp512 ~200 both
sides, tg128 within +-1.5% with the sign following RUN ORDER (a reversed
pair flipped it) — statistical parity, no regression on either model.

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

* reconcile the merged branches: LINT017/018 sweep in the fusion-touched metal files

Post-merge lint of the gptoss-fusion surface under the post-#3576 rules:
14 int(...) memcpy-size casts dropped (int64 overloads exist) and 3
int64(length(...)) -> long_length in dasllama_metal_prefill.das — all
pre-existing on master, swept because the fusion lane owns these files
in this PR. bench_metal_moe_lab was already clean. Lint 0/0, format
verified, test_kernel_backend -jit green in the integration worktree.

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

* decode_prof: permanent per-bucket decode profiler (the position-0 detector)

One measurement window of -n single-token forwards at --depth, engine prof_add
buckets reset after prefill+warmup, pace directly comparable to a lcpp_bench tg
cell. --accel mirrors lcpp_bench's flag for profiling the accel flavor's decode.
The previous incarnation was a session tool that got cleaned up — only its tune
sidecar survived in the main tree; this one is a tracked benchmark so the
detector stops evaporating between decode chases.

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

* fused MoE expert chain (mx4): one team rendezvous for gate+up+act+requant -> down+reduce

The routed-expert groupn sequence ran 3 fork/joins per layer with act/requant/reduce
single-lane between them; the chain runs both stages inside ONE team_parallel_stages
rendezvous — stage 0 = (expert, 32-row span) gate+up rows + bias + act + requant, stage 1 =
dim spans computing every expert's down rows + bias + the weighted sum in slot order.
Bit-identical to the groupn path (same rows cores, same dot->bias->act element order, same
32-block requant) — proven by the new test_fused_decode_moe_chain gpt-oss arm (ids + logits
bit-exact, chain vs per-op).

New backend slot: mx4_rows (MatmulMx4RowsFn) — arm64/x64-gen register mx4q8_gemv_gen
directly (the [tune] GEMV is the rows core, the mm_rows precedent), arm64-sdot gets a
row-major dot_mx4q8 walk. The chain gates on kernel_backend_has_mx4_rows() + team mode +
32-divisible spans, so every other backend keeps the per-op dispatch path.

decode_prof gains --no-fused (the chain-vs-per-op A/B lever). Rider: LINT017/018 sweep in
dasllama_common.das (43 long_length + 4 memcpy casts, the touched-file sweep discipline).

Measured M1 note: gpt-oss cpu decode wall is FLAT under the fusion (blk_ffn 12493 vs 12468
us/token, 3 joins/layer -> 1) — a real datum against the join-cost theory of the remaining
lcpp gap; the chain still wins the structure (fewer syncs, threaded act/requant) and is the
substrate for further expert-chain work.

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

* decode_prof: report the fused-decode state on the config line

The --no-fused A/B is only trustworthy when the log carries proof of which path ran (the
matrix that motivated this silently benched chain-vs-chain through a zsh quoting bug).

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

* decode_prof: --trace saves a jobque lane trace of the window (archived-tool parity)

Lifted from history/dasLLAMA/benchmarks/decode_prof.das (the archived predecessor — which
this file should have credited instead of assuming deletion). First use settled the lane
question: at -t N all N lanes (N-1 workers + the computing caller) run 95-97% busy in the
decode window; t7 and t8 do equal wall work, so the t7==t8 pace plateau is shared-resource
saturation, not a parked or half-productive lane.

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

---------

Co-authored-by: Claude Fable 5 <noreply@anthropic.com>
pull Bot pushed a commit to forksnd/daScript that referenced this pull request Jul 27, 2026
LINT017/LINT018 shipped in GaijinEntertainment#3576 but dasLLAMA was never swept, so the engine
carried the debt the rules exist to catch. CI lints changed files only, which
meant it surfaced one file at a time forever.

  int64(length(x))   -> long_length(x)          193 sites
  uint64(length(x))  -> uint64(long_length(x))  (keeps the result type)
  memcpy(d, s, int(n)) -> memcpy(d, s, n)       142 sites

Both are real 32-bit truncations, not style: `length` returns int, so the
2^31 limit is hit INSIDE it — as a wrap, or as the panic guard on array/table/
string length — before the widening cast ever runs. memcpy now carries uint /
int64 / uint64 size overloads, so the int() narrowing was pure loss.

Also drops five requires the env-registry migration orphaned (daslib/fio and
daslib/strings_convert, where the last consumer moved to dasllama_env).

modules/dasLLAMA now lints clean end to end: 185 files, 0 issues.

Two sites are deliberately untouched — `uint64(length(words) * 4)` in the
coopmat port and the MoE lab. LINT017 matches a bare wrap only, so the cast
around an EXPRESSION does not fire. The 32-bit multiply inside is the same
class of truncation and the rule does not see it; noted rather than silently
widened, since both arrays are small by construction.

Co-Authored-By: Claude Opus 5 (1M context) <noreply@anthropic.com>
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

None yet

Development

Successfully merging this pull request may close these issues.

2 participants