Skip to content
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension


Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
76 changes: 76 additions & 0 deletions .github/workflows/forkdiff-check.yml
Original file line number Diff line number Diff line change
@@ -0,0 +1,76 @@
name: Fork Diff Check

# Fails a PR whose `fork.yaml` no longer describes this fork.
#
# The fork-diff page (https://layr-labs.github.io/mlx-c/, rendered by
# protolambda/forkdiff from `fork.yaml`, deployed by forkdiff-pages.yml) is
# only useful while it is true, and two things make it go stale silently:
#
# 1. A rebase onto newer upstream. `base.hash` keeps pointing at the old
# upstream commit, so the page shows upstream's own changes as the fork's.
# `scripts/check_forkdiff.py` requires base.hash == merge-base(HEAD,
# upstream/main); a rebase moves the merge-base and the gate stays red
# until the hash — and the sections — are brought up to date.
# 2. A fork change nobody described. Every file in the base..HEAD diff must
# match a section glob (or a global ignore), and every glob must still
# match something.
#
# The page is also rendered here so a fork.yaml that forkdiff itself rejects
# cannot merge.

on:
pull_request:
push:
branches: [main]
workflow_dispatch:

permissions:
contents: read

jobs:
analysis-up-to-date:
name: fork.yaml describes the fork
runs-on: ubuntu-latest
timeout-minutes: 20
steps:
- uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2
with:
fetch-depth: 0 # full history: merge-base with upstream, and forkdiff diffs against base.hash

- name: Fetch upstream main
run: |
git fetch --no-tags --quiet https://github.com/ml-explore/mlx-c.git \
main:refs/remotes/upstream/main
echo "upstream/main = $(git rev-parse --short refs/remotes/upstream/main)"
echo "merge-base = $(git merge-base HEAD refs/remotes/upstream/main | cut -c1-12)"

- uses: actions/setup-python@a309ff8b426b58ec0e2a45f0f869d46889d02405 # v6.2.0
with:
python-version: "3.12"

- name: Verify fork.yaml against the real diff
run: |
python3 -m pip install --quiet pyyaml
python3 scripts/check_forkdiff.py --upstream-ref refs/remotes/upstream/main

- uses: actions/setup-go@40f1582b2485089dde7abd97c1529aa768e1baff # v5.6.0
with:
go-version: "1.24"
cache: false

- name: Render the fork-diff page
# fork.yaml names refs/heads/main; on a PR the checkout is a detached
# merge commit, so point the local branch at what we are checking.
run: |
git update-ref refs/heads/main HEAD
mkdir -p tmp/pages
go run github.com/protolambda/forkdiff@v0.1.1 \
-repo . -fork fork.yaml -out tmp/pages/index.html
echo "rendered $(wc -c < tmp/pages/index.html) bytes"

- name: Upload rendered page
uses: actions/upload-artifact@043fb46d1a93c77aae656e7c1c64a875d1fc6a0a # v7
with:
name: forkdiff-page
path: tmp/pages/index.html
retention-days: 7
99 changes: 99 additions & 0 deletions .github/workflows/forkdiff-pages.yml
Original file line number Diff line number Diff line change
@@ -0,0 +1,99 @@
name: Deploy Fork Diff

# Renders `fork.yaml` with protolambda/forkdiff and publishes the result as the
# repository's GitHub Pages site — the same setup ethereum-optimism/op-geth
# uses for its go-ethereum fork diff. Runs the analysis gate first, so a page
# that lies about the fork is never published.

on:
push:
branches: [main]
workflow_dispatch:

permissions:
contents: read
pages: write
id-token: write

concurrency:
group: "pages"
cancel-in-progress: true

jobs:
deploy:
name: Render and deploy
environment:
name: github-pages
url: ${{ steps.deployment.outputs.page_url }}
runs-on: ubuntu-latest
timeout-minutes: 20
steps:
- uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2
with:
fetch-depth: 0 # forkdiff diffs against base.hash, deep in history

- name: Fetch upstream main
run: |
git fetch --no-tags --quiet https://github.com/ml-explore/mlx-c.git \
main:refs/remotes/upstream/main

- uses: actions/setup-python@a309ff8b426b58ec0e2a45f0f869d46889d02405 # v6.2.0
with:
python-version: "3.12"

- name: Refuse to publish a stale analysis
run: |
python3 -m pip install --quiet pyyaml
python3 scripts/check_forkdiff.py --upstream-ref refs/remotes/upstream/main

- uses: actions/setup-go@40f1582b2485089dde7abd97c1529aa768e1baff # v5.6.0
with:
go-version: "1.24"
cache: false

- name: Build forkdiff
run: |
mkdir -p tmp/pages
go run github.com/protolambda/forkdiff@v0.1.1 \
-repo . -fork fork.yaml -out tmp/pages/index.html
touch tmp/pages/.nojekyll

- name: Is GitHub Pages enabled?
# Pages must be switched on once by a repo admin (Settings → Pages →
# Source: "GitHub Actions"). The workflow token cannot do that itself
# (`configure-pages` enablement needs an admin PAT), so until then the
# page is built and kept as an artifact but not deployed — a notice,
# not a red run.
id: pages
env:
GH_TOKEN: ${{ github.token }}
run: |
if gh api "repos/${GITHUB_REPOSITORY}/pages" --silent 2>/dev/null; then
echo "enabled=true" >> "$GITHUB_OUTPUT"
else
echo "enabled=false" >> "$GITHUB_OUTPUT"
echo "::notice::GitHub Pages is not enabled for ${GITHUB_REPOSITORY}; built the page but skipped the deploy. Enable it once under Settings → Pages (Source: GitHub Actions) and re-run this workflow."
fi

- name: Keep the rendered page as an artifact
if: steps.pages.outputs.enabled != 'true'
uses: actions/upload-artifact@043fb46d1a93c77aae656e7c1c64a875d1fc6a0a # v7
with:
name: forkdiff-page
path: tmp/pages/index.html
retention-days: 30

- name: Setup Pages
if: steps.pages.outputs.enabled == 'true'
uses: actions/configure-pages@983d7736d9b0ae728b81ab479565c72886d7745b # v5.0.0

- name: Upload artifact
if: steps.pages.outputs.enabled == 'true'
uses: actions/upload-pages-artifact@56afc609e74202658d3ffba0e8f6dda462b719fa # v3.0.1
with:
path: tmp/pages

- name: Deploy to GitHub Pages
if: steps.pages.outputs.enabled == 'true'
id: deployment
uses: actions/deploy-pages@d6db90164ac5ed86f2b6aed7e0febac5b3c0c03e # v4.0.5
79 changes: 79 additions & 0 deletions FORKDIFF.md
Original file line number Diff line number Diff line change
@@ -0,0 +1,79 @@
# Fork diff: what this fork changes, and keeping that page honest

This repository is a fork of [`ml-explore/mlx-c`](https://github.com/ml-explore/mlx-c).
Everything it changes relative to upstream is published as a browsable page:

**https://layr-labs.github.io/mlx-c/**

The page is rendered by [`protolambda/forkdiff`](https://github.com/protolambda/forkdiff)
from [`fork.yaml`](fork.yaml) at the repo root, in the style of
[op-geth's go-ethereum fork diff](https://op-geth.optimism.io/). `fork.yaml`
groups the changed files into sections with a paragraph each, names the exact
upstream commit the fork is rebased onto (`base.hash`), and lists files that
are not code (`ignore`). Every push to `main` re-renders and redeploys it
(`.github/workflows/forkdiff-pages.yml`).

**One-time setup.** GitHub Pages has to be switched on by a repo admin:
Settings → Pages → Source: **GitHub Actions**. The workflow token cannot do
this itself. Until it's done the deploy workflow still builds the page (kept
as a run artifact, `forkdiff-page`) and exits with a notice instead of failing.

## The gate

A fork-diff page is only useful while it is true, and two things make it go
stale silently. `scripts/check_forkdiff.py` runs on every PR
(`.github/workflows/forkdiff-check.yml`) and fails on both:

| Drift | Check | Why it matters |
|-------|-------|----------------|
| **Rebase onto newer upstream** | `base.hash` must equal `git merge-base HEAD upstream/main` and be an ancestor of upstream `main` | After a rebase the old base makes upstream's own commits look like fork changes: files that are not ours. |
| **New fork change nobody described** | every path in `git diff --name-only base.hash HEAD` must match a section glob or a global `ignore` | An undescribed file is a change the page can't explain. |
| **Section describing code we no longer carry** | every glob must match at least one changed path | Stale sections are as misleading as missing ones. |

The check also renders the page, so a `fork.yaml` that forkdiff itself rejects
cannot merge. The deploy workflow runs the same check before publishing, so a
stale analysis is never served.

## Day to day

**Adding or changing fork files in a PR.** If the check lists uncovered files,
add each to the section that explains it in `fork.yaml` (or start a new
section with a short description). Files that are not code go under the
top-level `ignore`. Keep globs specific: a `mlx/**` catch-all would swallow
upstream's changes after a bad rebase and defeat the gate.

**Rebasing onto newer upstream.** The gate will fail with the new merge-base
in its message:

```
git fetch https://github.com/ml-explore/mlx-c.git main:refs/remotes/upstream/main
git merge-base HEAD refs/remotes/upstream/main # → new base.hash
```

Set `base.hash` to that value, then run the check locally and fix what it
reports — usually files upstream absorbed (stale globs to delete) and files
that moved (globs to rename):

```
python3 -m pip install pyyaml
python3 scripts/check_forkdiff.py --upstream-ref refs/remotes/upstream/main
```

**Previewing the page locally** (Go 1.21+):

```
go run github.com/protolambda/forkdiff@v0.1.1 -repo . -fork fork.yaml -out tmp/index.html
open tmp/index.html
```

## Design notes

- `base.hash` is a full 40-hex commit id, never a branch name: a symbolic base
would move underneath the page and the gate alike.
- Section `ignore` lists count as coverage (forkdiff still lists those files,
grayed out); the top-level `ignore` is for things that aren't code at all.
- The glob semantics are forkdiff's: `*` and `?` stop at `/`, `**` spans
directories (and may match none), `[!x]` negates a class.
- The check is pure git + PyYAML so the same command runs locally and in CI.
It is shared verbatim across the Layr-Labs MLX forks (`mlx`, `mlx-c`,
`mlx-swift`, `mlx-swift-lm`).
2 changes: 2 additions & 0 deletions README.md
Original file line number Diff line number Diff line change
@@ -1,5 +1,7 @@
# MLX C

> **This is a fork.** `Layr-Labs/mlx-c` tracks [`ml-explore/mlx-c`](https://github.com/ml-explore/mlx-c) and adds the C entry points Layr-Labs' Swift inference stack needs from its forked MLX core. Everything changed relative to upstream is published as a fork diff at **https://layr-labs.github.io/mlx-c/**, described in [`fork.yaml`](fork.yaml) and kept honest by CI — see [FORKDIFF.md](FORKDIFF.md).

MLX C is a C API for [MLX](https://github.com/ml-explore/mlx).

MLX is an array framework for machine learning on Apple silicon. MLX C expands
Expand Down
124 changes: 124 additions & 0 deletions fork.yaml
Original file line number Diff line number Diff line change
@@ -0,0 +1,124 @@
title: "Layr-Labs/mlx-c - MLX C fork diff overview"
footer: |
Fork-diff overview of [`Layr-Labs/mlx-c`](https://github.com/Layr-Labs/mlx-c), a fork of
[`ml-explore/mlx-c`](https://github.com/ml-explore/mlx-c) &middot; the C bridge in Layr-Labs'
Apple-silicon inference stack ([`mlx`](https://github.com/Layr-Labs/mlx) →
[`mlx-c`](https://github.com/Layr-Labs/mlx-c) → [`mlx-swift`](https://github.com/Layr-Labs/mlx-swift) →
[`mlx-swift-lm`](https://github.com/Layr-Labs/mlx-swift-lm)) &middot; created with
[Forkdiff](https://github.com/protolambda/forkdiff)
base:
name: ml-explore/mlx-c
url: https://github.com/ml-explore/mlx-c
# The upstream commit this fork is rebased onto. CI (`scripts/check_forkdiff.py`)
# requires this to equal `git merge-base HEAD upstream/main`, so a rebase onto
# newer upstream fails the gate until the hash — and the sections below — are
# brought up to date. Upstream main as of 2026-04-23 (bindings for MLX 0.31.2).
hash: fba4470b89073180056c9ea46c443051375f7399
fork:
name: Layr-Labs/mlx-c
url: https://github.com/Layr-Labs/mlx-c
ref: refs/heads/main
def:
title: "Layr-Labs/mlx-c"
description: |
This is an overview of the changes in [`Layr-Labs/mlx-c`](https://github.com/Layr-Labs/mlx-c),
a fork of [`ml-explore/mlx-c`](https://github.com/ml-explore/mlx-c).

MLX C is the C boundary between the MLX core and every other language binding. Layr-Labs
runs an LLM/VLM serving stack on Apple silicon — continuous batching, paged KV caches,
MoE and quantized kernels — built as [`mlx-swift-lm`](https://github.com/Layr-Labs/mlx-swift-lm)
on [`mlx-swift`](https://github.com/Layr-Labs/mlx-swift), which links a forked
[`mlx`](https://github.com/Layr-Labs/mlx) core through this fork. Every change here is a
**carry-patch**: a thin C wrapper over a function the forked core added, so Swift can reach
it. The upstream API and ABI are left untouched; the fork only adds entry points, and pins
the core it builds against to the Layr-Labs revision that provides them.

Downstream consumers are named per section so a change here can be traced to the Swift
symbol that needs it.
sub:
- title: "Core pin"
description: |
`CMakeLists.txt` fetches the MLX core from `Layr-Labs/mlx` at an exact commit instead of
an upstream release tag. The C functions below wrap core functions that only exist on the
fork, so the pin is what makes them link; it is bumped together with the header changes
each time the core gains a new entry point.
globs:
- "CMakeLists.txt"
- title: "Memory: allocator observability and footprint policy"
description: |
The Swift admission code decides whether a request fits by asking the allocator, and the
upstream C API only offered independent counters. The fork adds, in order of arrival:

- `mlx_get_num_resources` / `mlx_get_resource_limit` — the live Metal buffer *count* and
its hard ceiling, the quantity behind the `[metal::malloc] Resource limit exceeded`
crash (consumed as `Memory.numResources` / `Memory.resourceLimit` in mlx-swift).
- `mlx_get_memory_snapshot` — one coherent active/cache/peak observation instead of three
separate calls that can straddle an allocation.
- `mlx_get_allocation_footprint_policy` and its checked `_bound` / `_maximum_extra`
helpers — the allocator's immutable rounding geometry (alignment, rounding threshold,
minimum allocation, power-of-two cut-off, cache page size) so callers can compute the
real footprint of a requested size without allocating; bound helpers return 1 on an
invalid policy or overflow without raising through the error callback.
- `mlx_get_allocation_size_upper_bound` — the same bound for the current policy.

The generator override in `python/mlxhooks.py` keeps `mlx_get_memory_snapshot` intact
when the bindings are regenerated from the core headers.
globs:
- "mlx/c/memory.h"
- "mlx/c/memory.cpp"
- "python/mlxhooks.py"
- title: "Array: buffer metadata and constant-cache identity"
description: |
Two non-evaluating inspections the quantized-inference path in mlx-swift relies on:

- `mlx_array_get_buffer_info` — backing metadata (allocated bytes, data offset, element
count, row-contiguity, uniqueness) for an *already evaluated* array. Unavailable backing
is reported as `available=false` rather than forcing an eval or a stream sync; views may
report shared backing and the fields confer no ownership.
- `_mlx_array_constant_cache_identity` — the backing descriptor's identity plus a
transform-safety predicate, so a Swift constant cache can reuse an unchanged constant's
exact cast across calls without mistaking a mutated array for its former contents. It
refuses compile/autodiff tracing and retained graphs, and never evaluates, hashes, or
reads data back. Marked internal: the caller must retain the source descriptor while
the identity is cached.
globs:
- "mlx/c/array.h"
- "mlx/c/array.cpp"
- title: "Fast: mutable-input Metal kernels and the SDPA signature"
description: |
- `mlx_fast_metal_kernel_new_mutable` — a second constructor for custom Metal kernels
that declares which caller-owned input buffers the kernel writes (the core's
mutable-input contract, used for in-place paged KV updates). The existing constructor
and its read-only ABI are unchanged.
- The scaled-dot-product-attention binding passes the `force_fused` argument MLX 0.32.2
added to the core call, as `false`, so the C signature and pre-0.32.2 behaviour are
preserved.
globs:
- "mlx/c/fast.h"
- "mlx/c/fast.cpp"
- title: "Stream: cross-thread default streams"
description: |
MLX 0.32 made the default stream's Metal command encoder thread-local, so a graph built
on one thread and evaluated on another — Swift `async/await` continuations, actor
executors — aborts with `There is no Stream(gpu, N) in current thread`.
`mlx_thread_unsafe_cpu_stream_new` / `mlx_thread_unsafe_gpu_stream_new` expose the
core's `new_thread_unsafe_stream`, whose encoder is registered globally, letting
mlx-swift restore a single process-wide default stream. "Thread-unsafe" means the caller
serializes submission; it is not locked for concurrent multi-thread use.
globs:
- "mlx/c/stream.h"
- "mlx/c/stream.cpp"
- title: "Fork tooling"
description: |
The gate that keeps this page honest: `check_forkdiff.py` fails CI when `base.hash`
is not the merge-base with upstream, or when a file the fork changes is not described
by a section above. `FORKDIFF.md` explains the day-to-day; the README links the page.
globs:
- "scripts/check_forkdiff.py"
- "FORKDIFF.md"
- "README.md"

# ignored globally, does not count towards line count
ignore:
- "fork.yaml"
- ".github/**"
Loading
Loading