Skip to content
Open
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
191 changes: 185 additions & 6 deletions .github/workflows/unsloth-sd-prebuilt.yml
Original file line number Diff line number Diff line change
@@ -1,26 +1,28 @@
# SPDX-License-Identifier: MIT
# Copyright 2026-present the Unsloth AI Inc. team.

name: Unsloth SD prebuilt (CPU/Apple/CUDA/Vulkan)
name: Unsloth SD prebuilt (CPU/Apple/CUDA/Vulkan/ROCm)

# Build and publish our stable-diffusion.cpp (sd-cli + sd-server) prebuilts for
# CPU, Apple, CUDA, and Vulkan. MiniMax-H3 uses the native engine on GPU hosts too.
# CPU, Apple, CUDA, Vulkan, and ROCm. MiniMax-H3 uses the native engine on GPU hosts too.
# Vulkan must carry the same H3 and graph-segment fixes as the other backends,
# rather than sending AMD hosts to an older upstream fallback binary.
#
# Shape mirrors unslothai/llama.cpp's prebuilt pipeline, simplified for a 7-way
# CPU/Apple/Vulkan matrix plus one best-effort CUDA leg:
# CPU/Apple/Vulkan matrix plus best-effort CUDA and ROCm legs:
# resolve -- take THIS repository at the requested ref, merge the pinned PR set
# in scripts/unsloth/pr-set.json, stamp a source tarball, name it
# after the upstream release the tree descends from (supply-chain
# aged), decide if our release already exists.
# build-unix -- macOS arm64 (Metal) + x64, Linux x64 + arm64, Linux Vulkan.
# build-windows-- Windows x64 CPU + Vulkan (MSVC + Ninja).
# build-linux-cuda -- Linux x64, CUDA 12.8, best effort (continue-on-error).
# build-linux-rocm -- Linux x64, ROCm 7.14.0 with its userspace bundled, best effort
# like CUDA. Without it AMD Linux hosts fall back to upstream or CPU (unsloth#8814).
# assemble -- fingerprint gate, sha256 + manifest, coverage gate, atomic
# draft->publish. If any CPU/Apple/Vulkan leg fails, assemble
# is skipped and nothing is published (the Studio needs the full
# asset set). The CUDA leg is exempt: it ships when it built and is
# asset set). The CUDA and ROCm legs are exempt: each ships when it built and is
# absent when it did not, so it can never hold back the fallbacks.
#
# Asset names match the Studio installer's resolve_release_asset.
Expand Down Expand Up @@ -669,9 +671,185 @@ jobs:
if-no-files-found: error
retention-days: 7

# Linux ROCm from TheRock wheels like upstream's build.yml, but the linked ROCm userspace ships
# in the bundle with rpath $ORIGIN: upstream's zip dies on a host whose ROCm differs or is
# absent (unsloth#8814). Only the kernel driver comes from the host.
build-linux-rocm:
name: Linux-Ubuntu-24.04-x86_64-rocm-7.14.0
needs: resolve
continue-on-error: true
if: ${{ needs.resolve.outputs.exists != 'true' || github.event_name == 'workflow_dispatch' }}
runs-on: ubuntu-24.04
env:
ROCM_VERSION: "7.14.0"
# Consumer parts, upstream's Windows list. CDNA hosts run diffusers and never reach this.
GPU_TARGETS: "gfx1010;gfx1011;gfx1012;gfx1030;gfx1031;gfx1032;gfx1033;gfx1034;gfx1035;gfx1036;gfx1100;gfx1101;gfx1102;gfx1103;gfx1150;gfx1151;gfx1152;gfx1153;gfx1200;gfx1201"
# rocm-sdk re-expands clang from the wheel every run, so its mtime always differs and
# the default compiler_check=mtime misses every object.
CCACHE_COMPILERCHECK: content
steps:
- name: Checkout mirror (tooling)
uses: actions/checkout@11d5960a326750d5838078e36cf38b85af677262 # v4.4.0
with:
path: tooling
fetch-depth: 1

- name: Free disk for the ROCm wheels
run: |
set -eux
sudo rm -rf /usr/local/lib/android /usr/share/dotnet /opt/ghc /usr/local/.ghcup \
/usr/local/share/boost /usr/share/swift /opt/hostedtoolcache
docker system prune -af || true
df -h /

- name: Download source @ ${{ needs.resolve.outputs.tag }}
uses: actions/download-artifact@d3f86a106a0bac45b974a628896c90dbdf5c8093 # v4.3.0
with:
name: ${{ needs.resolve.outputs.source_artifact }}
path: srcpkg
- name: Extract source
run: |
set -eux
mkdir -p src
tar -xzf "srcpkg/sd-source-${{ needs.resolve.outputs.tag }}.tar.gz" -C src

- name: Install ROCm (TheRock wheels)
run: |
set -euo pipefail
sudo apt-get update -qq && sudo apt-get install -y -qq ninja-build patchelf
python3 -m venv .rocm
. .rocm/bin/activate
python -m pip install -q --upgrade pip
# libraries/devel carry no GPU code: rocBLAS/hipBLASLt kernels and kpack archives come
# from one rocm-sdk-device-<gfx> wheel per target, linked into the devel tree by
# `rocm-sdk path`.
DEVICE="$(echo "$GPU_TARGETS" | sed 's/;/,device-/g; s/^/device-/')"
python -m pip install -q --no-cache-dir --index-url https://repo.amd.com/rocm/whl-multi-arch/ \
"rocm[libraries,devel,${DEVICE}]==${ROCM_VERSION}"
ROCM_PATH="$(rocm-sdk path --root)"
{
echo "ROCM_PATH=$ROCM_PATH"
echo "HIP_PATH=$ROCM_PATH"
echo "CMAKE_PREFIX_PATH=$(rocm-sdk path --cmake)"
echo "LD_LIBRARY_PATH=$ROCM_PATH/lib:${LD_LIBRARY_PATH:-}"
} >> "$GITHUB_ENV"
echo "$(rocm-sdk path --bin)" >> "$GITHUB_PATH"
echo "$(pwd)/.rocm/bin" >> "$GITHUB_PATH"

- name: ccache key
id: cckey
run: echo "archs=$(echo "$GPU_TARGETS" | sha256sum | cut -c1-8)" >> "$GITHUB_OUTPUT"

- name: ccache
uses: hendrikmuhs/ccache-action@d62db5f07c26379fc4b4e0916f098a92573c3b03 # v1.2.23
with:
key: sd-rocm-${{ env.ROCM_VERSION }}-${{ steps.cckey.outputs.archs }}-${{ needs.resolve.outputs.tag }}
restore-keys: |
sd-rocm-${{ env.ROCM_VERSION }}-${{ steps.cckey.outputs.archs }}
append-timestamp: false
variant: ccache
max-size: 2G
save: false

- name: Build sd-cli + sd-server (ROCm)
working-directory: src
run: |
set -euo pipefail
cmake -B build -G Ninja \
-DCMAKE_BUILD_TYPE=Release \
-DCMAKE_HIP_COMPILER="$(hipconfig -l)/clang" \
-DCMAKE_HIP_FLAGS="-mllvm --amdgpu-unroll-threshold-local=600" \
-DSD_BUILD_EXAMPLES=ON \
-DSD_SERVER_BUILD_FRONTEND=OFF \
-DSD_WEBP=OFF -DSD_WEBM=OFF \
-DGGML_NATIVE=OFF \
-DSD_HIPBLAS=ON \
-DHIP_PLATFORM=amd \
-DGPU_TARGETS="$GPU_TARGETS" \
-DCMAKE_C_COMPILER_LAUNCHER=ccache \
-DCMAKE_CXX_COMPILER_LAUNCHER=ccache \
-DCMAKE_HIP_COMPILER_LAUNCHER=ccache
cmake --build build --config Release -j "$(nproc)" --target sd-cli sd-server
ccache --show-stats || true

- name: Stage the bundle with the ROCm runtime
run: |
set -euo pipefail
BIN="${GITHUB_WORKSPACE}/src/build/bin"
STAGE="${GITHUB_WORKSPACE}/rocm-bundle"
rm -rf "$STAGE" && mkdir -p "$STAGE/lib" "$STAGE/.kpack"
cp "$BIN/sd-cli" "$BIN/sd-server" "$STAGE/"
# ldd walks the closure and ggml is static here, so the two binaries are the whole list.
ldd "$STAGE/sd-cli" "$STAGE/sd-server" \
| awk -v r="$ROCM_PATH/" 'index($3, r) == 1 { print $3 }' | sort -u > "$RUNNER_TEMP/rocm-libs.txt"
if [ ! -s "$RUNNER_TEMP/rocm-libs.txt" ]; then
echo "ERROR: nothing resolved from $ROCM_PATH; is this a HIP build?" >&2
exit 1
fi
while read -r so; do cp -L "$so" "$STAGE/lib/"; done < "$RUNNER_TEMP/rocm-libs.txt"
# rocBLAS and hipBLASLt load kernels from <dir of the library>/<name>/library.
for t in rocblas hipblaslt; do
compgen -G "$STAGE/lib/lib$t.so*" > /dev/null || continue
mkdir -p "$STAGE/lib/$t"
cp -rL "$ROCM_PATH/lib/$t/library" "$STAGE/lib/$t/"
done
# kpack-split libraries name their GPU code in .rocm_kpack_ref as
# ../.kpack/<name>_@GFXARCH@.kpack, relative to the library, so lib/ and .kpack/ must
# stay siblings in the bundle.
for so in "$STAGE"/lib/*.so*; do
objcopy -O binary --only-section=.rocm_kpack_ref "$so" "$RUNNER_TEMP/kpack_ref.bin"
for name in $(strings "$RUNNER_TEMP/kpack_ref.bin" | grep -o '\.kpack/[A-Za-z0-9_]*_@GFXARCH@' \
| sed 's#^\.kpack/##; s#_@GFXARCH@$##' | sort -u); do
for gfx in ${GPU_TARGETS//;/ }; do
cp -L "$ROCM_PATH/.kpack/${name}_${gfx}.kpack" "$STAGE/.kpack/"
done
done
done
patchelf --set-rpath '$ORIGIN/lib' "$STAGE/sd-cli" "$STAGE/sd-server"
for f in "$STAGE"/lib/*.so*; do patchelf --set-rpath '$ORIGIN' "$f"; done
echo "bundled $(wc -l < "$RUNNER_TEMP/rocm-libs.txt") ROCm libraries, $(ls "$STAGE/.kpack" | wc -l) kpack archives"
du -sh "$STAGE"
# The proof the bundle stands alone: resolve with the wheel tree hidden.
if LD_LIBRARY_PATH= ldd "$STAGE/sd-cli" "$STAGE/sd-server" | grep -i 'not found'; then
echo "ERROR: unresolved libraries after bundling" >&2
exit 1
fi

- name: Package bundle
env:
BIN_DIR: ${{ github.workspace }}/rocm-bundle
KEEP_LAYOUT: "1"
OUT_DIR: ${{ github.workspace }}/dist
TAG: ${{ needs.resolve.outputs.tag }}
LABEL: Linux-Ubuntu-24.04-x86_64-rocm-7.14.0
COMMIT: ${{ needs.resolve.outputs.commit }}
SOURCE_REPO: ${{ github.repository }}
LICENSE_FILE: ${{ github.workspace }}/src/LICENSE
run: python3 tooling/scripts/unsloth/package_bundle.py

- name: Upload bundle
uses: actions/upload-artifact@ea165f8d65b6e75b540449e92b4886f43607fa02 # v4.6.2
with:
name: sd-${{ needs.resolve.outputs.tag }}-bin-Linux-Ubuntu-24.04-x86_64-rocm-7.14.0
path: dist/sd-${{ needs.resolve.outputs.tag }}-bin-Linux-Ubuntu-24.04-x86_64-rocm-7.14.0.zip
if-no-files-found: error
retention-days: 7

- name: Evict stale ccache files
if: ${{ !cancelled() }}
continue-on-error: true
run: ccache --evict-older-than 14d

- name: Save ccache
if: ${{ always() }}
uses: actions/cache/save@55cc8345863c7cc4c66a329aec7e433d2d1c52a9 # v6.1.0
with:
path: ${{ github.workspace }}/.ccache
key: ccache-sd-rocm-${{ env.ROCM_VERSION }}-${{ steps.cckey.outputs.archs }}-${{ needs.resolve.outputs.tag }}-

assemble:
name: Assemble + publish
needs: [resolve, build-unix, build-windows, build-linux-cuda]
needs: [resolve, build-unix, build-windows, build-linux-cuda, build-linux-rocm]
# Consumed by `reclaim` to tell "these bundles are now release assets" from
# "nothing will ever read these". Set only after draft=false lands.
outputs:
Expand Down Expand Up @@ -727,6 +905,7 @@ jobs:

# Vulkan is required: the pinned upstream fallback lacks our H3 fixes.
# Blocking publication on a Vulkan failure prevents promoting an incomplete fix.
# CUDA and ROCm are best effort and are not listed here.
- name: Coverage gate (all 7 CPU/Apple/Vulkan assets present)
run: |
set -eu
Expand Down Expand Up @@ -776,7 +955,7 @@ jobs:
AHEAD='${{ needs.resolve.outputs.ahead }}'
COMMIT='${{ needs.resolve.outputs.commit }}'
REPO="$GITHUB_REPOSITORY"
NOTES="Automated Unsloth stable-diffusion.cpp prebuild (sd-cli + sd-server), built from [\`${COMMIT}\`](https://github.com/${REPO}/commit/${COMMIT}) in this repository. Includes CPU, Apple Metal, and Linux/Windows Vulkan builds, plus Linux CUDA when its build succeeds."
NOTES="Automated Unsloth stable-diffusion.cpp prebuild (sd-cli + sd-server), built from [\`${COMMIT}\`](https://github.com/${REPO}/commit/${COMMIT}) in this repository. Includes CPU, Apple Metal, and Linux/Windows Vulkan builds, plus Linux CUDA and Linux ROCm when their builds succeed."
PINS='${{ needs.resolve.outputs.pins }}'
if [ "${AHEAD:-0}" -gt 0 ]; then
NOTES="${NOTES}"$'\n\n'"Not a stock build: the tree is ${AHEAD} commits past the ${UPSTREAM_TAG} release it descends from, which is what the \`-u\` suffix on the tag marks. \`sd-prebuilt-manifest.json\` records the commit."
Expand Down
6 changes: 4 additions & 2 deletions docs/backend.md
Original file line number Diff line number Diff line change
Expand Up @@ -156,8 +156,10 @@ the runner's graph-cut capacity checks.

Runtime capacity checks also leave 512 MiB of currently free device memory for
backend scratch buffers and pipelines, including with explicit backend assignments.
They cap stale free-memory reports by the device's total memory minus tracked
resident allocations and reject reports that exceed the device's total memory.
They cap free-memory reports by the device's total memory minus tracked
resident allocations. Vulkan reports exceeding total memory are rejected because
its heap-budget subtraction can underflow. Other backends use the cap instead of
treating such reports as zero free memory.

Components are considered in `diffusion`, `te`, `vae` order so that repeatedly
used diffusion weights have priority. Each component's weights use the first
Expand Down
30 changes: 17 additions & 13 deletions scripts/unsloth/package_bundle.py
Original file line number Diff line number Diff line change
Expand Up @@ -16,6 +16,8 @@
COMMIT source commit SHA (provenance)
SOURCE_REPO e.g. leejet/stable-diffusion.cpp
LICENSE_FILE path to the LICENSE to include (optional)
KEEP_LAYOUT "1" ships every file under BIN_DIR at its relative path instead of
flattening (the ROCm leg stages lib/ and .kpack/ that must stay siblings)

The zip unpacks into a single named dir ``sd-<TAG>-bin-<LABEL>/`` containing the
binaries, their sibling runtime libs, LICENSE, and an UNSLOTH_BUILD.txt provenance
Expand Down Expand Up @@ -57,15 +59,18 @@ def _is_runtime_lib(name: str) -> bool:
return ".so." in lowered


def _collect(bin_dir: Path) -> list[Path]:
"""The binaries + sibling runtime libs to ship. Recurse so a nested bin/ layout
(some generators emit build/bin/, some build/bin/Release/) is still captured."""
found: list[Path] = []
def _collect(bin_dir: Path, keep_layout: bool) -> list[tuple[Path, str]]:
"""(file, archive-relative name) for the binaries + sibling runtime libs to ship. Recurse
so a nested bin/ layout (some generators emit build/bin/, some build/bin/Release/) is still
captured and flattened. With keep_layout, every file ships at its path under bin_dir."""
found: list[tuple[Path, str]] = []
for p in sorted(bin_dir.rglob("*")):
if not p.is_file():
continue
if p.name in _BINARIES or _is_runtime_lib(p.name):
found.append(p)
if keep_layout:
found.append((p, p.relative_to(bin_dir).as_posix()))
elif p.name in _BINARIES or _is_runtime_lib(p.name):
found.append((p, p.name))
return found


Expand All @@ -82,12 +87,12 @@ def main() -> int:
print(f"package_bundle: BIN_DIR {bin_dir} is not a directory", file = sys.stderr)
return 2

files = _collect(bin_dir)
have_cli = any(f.name in ("sd-cli", "sd-cli.exe") for f in files)
files = _collect(bin_dir, os.environ.get("KEEP_LAYOUT", "").strip() == "1")
have_cli = any(f.name in ("sd-cli", "sd-cli.exe") for f, _ in files)
if not have_cli:
print(f"package_bundle: no sd-cli under {bin_dir}; refusing to package", file = sys.stderr)
return 1
have_server = any(f.name in ("sd-server", "sd-server.exe") for f in files)
have_server = any(f.name in ("sd-server", "sd-server.exe") for f, _ in files)

out_dir.mkdir(parents = True, exist_ok = True)
stem = f"sd-{tag}-bin-{label}"
Expand All @@ -108,10 +113,9 @@ def main() -> int:
# Deterministic-ish: sort members; drop the archive if it already exists.
zip_path.unlink(missing_ok = True)
with zipfile.ZipFile(zip_path, "w", compression = zipfile.ZIP_DEFLATED) as zf:
for f in files:
# Flatten under the named top-level dir; keep just the basename so the
# binaries sit at sd-<tag>-bin-<label>/<name> regardless of build layout.
zf.write(f, arcname = f"{stem}/{f.name}")
for f, name in files:
# Under sd-<tag>-bin-<label>/, flat unless KEEP_LAYOUT.
zf.write(f, arcname = f"{stem}/{name}")
zf.writestr(f"{stem}/UNSLOTH_BUILD.txt", provenance)
if license_file and Path(license_file).is_file():
zf.write(license_file, arcname = f"{stem}/LICENSE")
Expand Down
5 changes: 4 additions & 1 deletion src/conditioning/conditioner.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -2219,7 +2219,10 @@ struct LLMEmbedder : public Conditioner {
false,
deepstack_image_embeds,
image_grids);
GGML_ASSERT(!hidden_states.empty());
if (hidden_states.empty()) {
LOG_ERROR("LLM prompt encoding failed");
return {};
}
hidden_states = apply_token_weights(std::move(hidden_states), weights);
GGML_ASSERT(hidden_states.shape()[1] > prompt_template_encode_start_idx);

Expand Down
2 changes: 1 addition & 1 deletion src/model_manager.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -1635,7 +1635,7 @@ ModelManager::CapacityCheck ModelManager::check_capacity(
return SIZE_MAX;
}
// Vulkan's heap budget subtraction can underflow when usage exceeds the budget.
if (total_bytes > 0 && free_bytes > total_bytes) {
if (total_bytes > 0 && free_bytes > total_bytes && sd_backend_is(backend, "Vulkan")) {
return size_t{0};
}
const size_t resident = add(compute_backend_resident_bytes(backend),
Expand Down