diff --git a/.github/workflows/build-gfx11-image.yml b/.github/workflows/build-gfx11-image.yml new file mode 100644 index 000000000000..9df5eb0479ad --- /dev/null +++ b/.github/workflows/build-gfx11-image.yml @@ -0,0 +1,208 @@ +name: Docker build gfx11 vLLM image + +# Builds the runnable vLLM image for RDNA3/3.5 (docker/Dockerfile.gfx11). +# Distinct from build-gfx11-ci-image.yml, which publishes the thin *builder* +# image this repo's wheel CI runs inside. +# +# Manual only -- workflow_dispatch is the sole trigger. Unlike +# build-rocm-wheels.yml this has no sccache, so every run is a cache-cold +# compile of vLLM's HIP kernels (hours, not minutes); running it on push or on +# a schedule would burn a runner for most of a day per commit. +# +# A dispatch publishes to GHCR by default; untick `push_image` for a +# build-only dry run. The env fallback below stays false on purpose, so that +# adding a non-dispatch trigger later cannot publish by accident. + +on: + workflow_dispatch: + inputs: + runs_on: + description: 'Runner label' + default: 'ubuntu-latest' + required: false + rocm_arch: + description: 'Compile targets, ";"-separated' + default: 'gfx1151' + required: false + device_arch: + description: 'Single arch selecting the torch/vllm device wheel extras' + default: 'gfx1151' + required: false + max_jobs: + description: 'Compile parallelism (see note in the build step)' + default: '2' + required: false + vllm_ref: + description: 'vLLM commit to build. Blank = the pin in Dockerfile.gfx11' + default: '' + required: false + remote_vllm: + description: '1 = build the pinned ref, 0 = build this branch checkout' + default: '1' + required: false + push_image: + description: 'Publish to GHCR (untick for a build-only dry run)' + type: boolean + default: true + +concurrency: + group: build-gfx11-image-${{ github.ref }} + # A run is hours long; let an in-flight build finish rather than losing it. + cancel-in-progress: false + +env: + REGISTRY: ghcr.io + # github.repository may contain uppercase (e.g. ROCm/vllm); Docker requires + # lowercase tags. + IMAGE_NAME: rocm/vllm/gfx11 + +jobs: + build-image: + runs-on: ${{ inputs.runs_on || 'ubuntu-latest' }} + # A cold build compiles every HIP kernel; 6h is the GitHub-hosted ceiling. + timeout-minutes: 360 + permissions: + contents: read + packages: write + + # `inputs` is null on a push event, and an empty build arg would override + # the Dockerfile's ARG default with nothing (an empty REMOTE_VLLM breaks + # the FROM, an empty MAX_JOBS silently becomes $(nproc)). Resolve every + # knob to a concrete value once, here. + env: + IN_RUNS_ON: ${{ inputs.runs_on || 'ubuntu-latest' }} + IN_ROCM_ARCH: ${{ inputs.rocm_arch || 'gfx1151' }} + IN_DEVICE_ARCH: ${{ inputs.device_arch || 'gfx1151' }} + IN_MAX_JOBS: ${{ inputs.max_jobs || '2' }} + IN_REMOTE_VLLM: ${{ inputs.remote_vllm || '1' }} + IN_VLLM_REF: ${{ inputs.vllm_ref || '' }} + IN_PUSH_IMAGE: ${{ inputs.push_image || false }} + + steps: + - name: Checkout code + uses: actions/checkout@v4 + with: + # Only the Dockerfile is needed for a pinned build, but a + # remote_vllm=0 run compiles this checkout, and setup.py derives the + # wheel version from `git describe`. + fetch-depth: 0 + fetch-tags: true + + - name: Fetch upstream tags for versioning + if: env.IN_REMOTE_VLLM == '0' + run: | + git config --global --add safe.directory "$GITHUB_WORKSPACE" + git remote add upstream https://github.com/vllm-project/vllm.git || true + git fetch upstream --tags || echo "Warning: could not fetch upstream tags" + git describe --tags --always || echo "Warning: git describe failed" + + # Measured on run 33021450792: the runner has a 145 GB root filesystem + # with 86 GB free before any cleanup, and the whole ROCm base layer costs + # only ~19 GB. Disk is not the constraint here, so jlumbroso/free-disk-space + # is not used -- its swap-storage option deletes /mnt/swapfile, which left + # the build with 15 GB of RAM and zero swap and got the runner OOM-killed + # mid-compile twice (runs 33010389974 at MAX_JOBS=2, 33016302061 at 1). + # Spend the surplus disk on swap instead. + - name: Enlarge swap + run: | + sudo swapoff -a || true + sudo rm -f /mnt/swapfile + sudo fallocate -l 24G /mnt/swapfile + sudo chmod 600 /mnt/swapfile + sudo mkswap /mnt/swapfile + sudo swapon /mnt/swapfile + free -g + df -h / + + - name: Log in to GHCR + # Needed to read the buildcache tag, and to write it plus the image + # when push_image is on. + uses: docker/login-action@v3 + with: + registry: ${{ env.REGISTRY }} + username: ${{ github.actor }} + password: ${{ secrets.GITHUB_TOKEN }} + + - name: Set up Docker Buildx + uses: docker/setup-buildx-action@v3 + + # The image is normally built from the pin in Dockerfile.gfx11, not from + # this checkout, so github.sha would be the wrong thing to tag it with. + - name: Resolve vLLM source ref + id: src + run: | + if [ "$IN_REMOTE_VLLM" = "0" ]; then + REF="${{ github.sha }}" + else + REF="$IN_VLLM_REF" + [ -n "$REF" ] || REF=$(grep -m1 '^ARG VLLM_REF=' docker/Dockerfile.gfx11 | cut -d= -f2) + fi + echo "ref=$REF" >> "$GITHUB_OUTPUT" + + - name: Report build plan + run: | + echo "runner: $IN_RUNS_ON" + echo "max_jobs: $IN_MAX_JOBS" + echo "arch: $IN_ROCM_ARCH (device: $IN_DEVICE_ARCH)" + echo "source: ${{ steps.src.outputs.ref }} (remote_vllm=$IN_REMOTE_VLLM)" + echo "push_image: $IN_PUSH_IMAGE" + nproc; free -g; df -h / /mnt + + - name: Build and push + uses: docker/build-push-action@v5 + with: + context: . + file: docker/Dockerfile.gfx11 + target: runtime + push: ${{ env.IN_PUSH_IMAGE }} + # MAX_JOBS is memory-bound, not core-bound: the runner has 4 vCPU but + # only 15 GB of RAM. The ceiling is set by swap, not by this number -- + # runs 33010389974 (2 jobs) and 33016302061 (1 job) both died mid- + # compile with the runner OOM-killed, and the common factor was zero + # swap, not the job count. See the "Enlarge swap" step. + build-args: | + PYTORCH_ROCM_ARCH=${{ env.IN_ROCM_ARCH }} + DEVICE_ARCH=${{ env.IN_DEVICE_ARCH }} + MAX_JOBS=${{ env.IN_MAX_JOBS }} + REMOTE_VLLM=${{ env.IN_REMOTE_VLLM }} + ${{ env.IN_REMOTE_VLLM != '0' && format('VLLM_REF={0}', steps.src.outputs.ref) || '' }} + tags: | + ${{ env.REGISTRY }}/${{ env.IMAGE_NAME }}:latest + ${{ env.REGISTRY }}/${{ env.IMAGE_NAME }}:${{ steps.src.outputs.ref }} + labels: | + org.opencontainers.image.revision=${{ steps.src.outputs.ref }} + # Registry cache rather than type=gha: the GitHub Actions cache is + # capped at 10 GB per repo, which this image's layers blow past on + # the first export. Only written when publishing is on, so an + # experimental run leaves GHCR untouched. + cache-from: type=registry,ref=${{ env.REGISTRY }}/${{ env.IMAGE_NAME }}:buildcache + cache-to: >- + ${{ env.IN_PUSH_IMAGE == 'true' + && format('type=registry,ref={0}/{1}:buildcache,mode=min', env.REGISTRY, env.IMAGE_NAME) + || '' }} + + # Two runs died mid-build with the runner losing contact and its logs + # never uploading, so record what the build actually consumed while a + # step can still report it. + - name: Disk and memory after build + if: always() + run: | + df -h / /mnt || true + free -g || true + docker system df || true + + - name: Summary + if: success() + run: | + IMAGE="${{ env.REGISTRY }}/${{ env.IMAGE_NAME }}" + { + echo "## gfx11 vLLM image" + echo "" + echo "- runner: \`$IN_RUNS_ON\`, MAX_JOBS=\`$IN_MAX_JOBS\`" + echo "- vLLM source: \`${{ steps.src.outputs.ref }}\`" + if [ "$IN_PUSH_IMAGE" = "true" ]; then + echo "- **Pushed**: \`${IMAGE}:latest\`, \`${IMAGE}:${{ steps.src.outputs.ref }}\`" + else + echo "- **Not pushed** (push_image was off); image was built and discarded." + fi + } >> "$GITHUB_STEP_SUMMARY" diff --git a/docker/Dockerfile.gfx11 b/docker/Dockerfile.gfx11 new file mode 100644 index 000000000000..e4cd5600c06d --- /dev/null +++ b/docker/Dockerfile.gfx11 @@ -0,0 +1,209 @@ +# syntax=docker/dockerfile:1.7 +# +# vLLM runtime image for AMD RDNA3/3.5 (gfx11xx, e.g. Strix Halo / gfx1151). +# +# Built entirely from Python wheels -- no system-wide ROCm packages. The ROCm +# toolchain and runtime come from the `rocm[devel,libraries]` wheels, so this +# image does not inherit from rocm/vllm-dev like docker/Dockerfile.rocm does. +# +# Stages: +# rocm-base -- system deps + uv + pinned ROCm SDK / PyTorch / Triton wheels +# builder -- compiles the vLLM wheel into /dist +# runtime -- rocm-base + the built wheel (no ~15 GB of HIP objects) +# +# Build (pinned vLLM commit, the default -- reproducible, ignores your HEAD): +# docker build -f docker/Dockerfile.gfx11 --target runtime -t vllm-gfx11 . +# +# Build the working tree instead, to test local changes: +# docker build -f docker/Dockerfile.gfx11 --build-arg REMOTE_VLLM=0 \ +# --target runtime -t vllm-gfx11 . +# +# vLLM's HIP kernels are compiled from source: expect ~30-75 min cold and +# ~40 GB free under /var/lib/docker. ccache and the uv download cache are +# BuildKit cache mounts, so a rebuild after a version bump is far cheaper. + +ARG BASE_IMAGE=python:3.12-bookworm + +# REMOTE_VLLM=1: clone VLLM_REPO at the pinned VLLM_REF (tested and verified) +# REMOTE_VLLM=0: build the local source in the Docker context (ONBUILD COPY ./) +ARG REMOTE_VLLM=1 + +# --------------------------------------------------------------------------- +# Stage 1: rocm-base -- shared ROCm + PyTorch layer +# --------------------------------------------------------------------------- +FROM ${BASE_IMAGE} AS rocm-base + +ARG PYTORCH_ROCM_ARCH=gfx1151 +ARG DEVICE_ARCH=gfx1151 +ARG PYTORCH_INDEX_URL=https://rocm.nightlies.amd.com/whl-multi-arch/ + +# Pinned wheel versions, all from the same ROCm nightly date so the stack is +# self-consistent. torch and torchvision must share the +rocm build-date suffix +# or they ABI-mismatch at import. +ARG ROCM_BUILD=10.1.0a20260806 +ARG TORCH_VERSION=2.12.0 +ARG TORCHVISION_VERSION=0.27.0 +ARG TORCHAUDIO_VERSION=2.11.0 +ARG TRITON_VERSION=3.8.0+git4cff872c.rocm10.1.0a20260806 + +RUN apt-get update -qq && apt-get install -y -qq --no-install-recommends \ + git curl ca-certificates cmake ninja-build gcc g++ ccache libdrm-dev \ + && rm -rf /var/lib/apt/lists/* + +RUN groupadd -r render && groupadd -rf video + +RUN curl -LsSf https://astral.sh/uv/install.sh | env UV_INSTALL_DIR="/usr/local/bin" sh && uv --version + +# set up the ROCm SDK environment +ENV ROCM_PATH=/usr/local/lib/python3.12/site-packages/_rocm_sdk_devel +ENV PATH=/usr/local/lib/python3.12/site-packages/_rocm_sdk_devel/bin:$PATH +ENV PYTHONPATH=/usr/local/lib/python3.12/site-packages/_rocm_sdk_core/share/amd_smi + +# The pins are written to /constraints.txt as well as installed, so that the +# builder's `uv pip install -r requirements/build/rocm.txt` cannot quietly move +# torch onto a different build (that file carries its own torch== line and a +# download.pytorch.org extra index). +# +# `devel` and `libraries` extras are needed on top of `device-`: devel +# provides hipcc to compile vLLM, libraries provides the runtime math libs. +RUN --mount=type=cache,target=/root/.cache/uv \ + set -eux; \ + { \ + echo "torch==${TORCH_VERSION}+rocm${ROCM_BUILD}"; \ + echo "torchvision==${TORCHVISION_VERSION}+rocm${ROCM_BUILD}"; \ + echo "torchaudio==${TORCHAUDIO_VERSION}+rocm${ROCM_BUILD}"; \ + echo "triton==${TRITON_VERSION}"; \ + echo "rocm==${ROCM_BUILD}"; \ + } > /constraints.txt; \ + cat /constraints.txt; \ + uv pip install --system --break-system-packages --link-mode=copy \ + "torch[device-${DEVICE_ARCH}]==${TORCH_VERSION}+rocm${ROCM_BUILD}" \ + "torchvision[device-${DEVICE_ARCH}]==${TORCHVISION_VERSION}+rocm${ROCM_BUILD}" \ + "torchaudio==${TORCHAUDIO_VERSION}+rocm${ROCM_BUILD}" \ + "triton==${TRITON_VERSION}" \ + "rocm[devel,libraries,device-${DEVICE_ARCH}]==${ROCM_BUILD}" \ + --index-url "${PYTORCH_INDEX_URL}" \ + --extra-index-url https://pypi.org/simple/ \ + --index-strategy unsafe-first-match \ + --prerelease allow; \ + rocm-sdk init; \ + hipcc --version + +# --------------------------------------------------------------------------- +# Stage 2: source -- local build context, or a clone +# --------------------------------------------------------------------------- +# .git is deliberately part of the context: setup.py derives the wheel version +# from `git describe`, so stripping it yields an unversioned wheel. +FROM rocm-base AS fetch_vllm_0 +ONBUILD COPY ./ /src/vllm/ + +FROM rocm-base AS fetch_vllm_1 +ARG VLLM_REPO=https://github.com/ROCm/vllm.git +ARG VLLM_REF=ddd378cfb1e9e9232ae5af18657bc8f99587666c +ENV VLLM_REPO=${VLLM_REPO} +ENV VLLM_REF=${VLLM_REF} +ONBUILD RUN set -eux; \ + git clone "${VLLM_REPO}" /src/vllm; \ + cd /src/vllm; \ + git checkout --detach "${VLLM_REF}" \ + || { git fetch -v origin "${VLLM_REF}" && git checkout --detach FETCH_HEAD; }; \ + git log -1 --format='vLLM commit: %H %ad %s' + +# --------------------------------------------------------------------------- +# Stage 3: builder -- compile the vLLM wheel +# --------------------------------------------------------------------------- +FROM fetch_vllm_${REMOTE_VLLM} AS builder + +ARG PYTORCH_ROCM_ARCH=gfx1151 +ARG PYTORCH_INDEX_URL=https://rocm.nightlies.amd.com/whl-multi-arch/ +# Empty falls back to $(nproc) at build time; lower it if the compile OOMs. +ARG MAX_JOBS= + +WORKDIR /src/vllm + +RUN git config --global --add safe.directory /src/vllm + +RUN --mount=type=cache,target=/root/.cache/uv \ + uv pip install --system --break-system-packages --link-mode=copy \ + -c /constraints.txt \ + -r requirements/build/rocm.txt \ + --index-url "${PYTORCH_INDEX_URL}" \ + --extra-index-url https://pypi.org/simple/ \ + --index-strategy unsafe-first-match \ + --prerelease allow + +ENV VLLM_TARGET_DEVICE=rocm \ + PYTORCH_ROCM_ARCH=${PYTORCH_ROCM_ARCH} \ + CCACHE_DIR=/root/.cache/ccache \ + CMAKE_C_COMPILER_LAUNCHER=ccache \ + CMAKE_CXX_COMPILER_LAUNCHER=ccache \ + CMAKE_HIP_COMPILER_LAUNCHER=ccache + +RUN --mount=type=cache,target=/root/.cache/ccache \ + set -eux; \ + export MAX_JOBS="${MAX_JOBS:-$(nproc)}"; \ + ccache --zero-stats || true; \ + python setup.py bdist_wheel --dist-dir=/dist; \ + ccache --show-stats || true; \ + ls -la /dist + +# --------------------------------------------------------------------------- +# Stage 4: export_wheel -- wheel only, for `--output type=local,dest=...` +# --------------------------------------------------------------------------- +FROM scratch AS export_wheel +COPY --from=builder /dist / + +# --------------------------------------------------------------------------- +# Stage 5: runtime +# --------------------------------------------------------------------------- +FROM rocm-base AS runtime + +ARG DEVICE_ARCH=gfx1151 +ARG PYTORCH_INDEX_URL=https://rocm.nightlies.amd.com/whl-multi-arch/ +# transformers 5.15+ raises AmbiguousGlobalPerLayerAttributeError on Gemma-4. +ARG TRANSFORMERS_VERSION=5.14.1 + +COPY --from=builder /dist /dist + +RUN --mount=type=cache,target=/root/.cache/uv \ + set -eux; \ + uv pip install --system --break-system-packages --link-mode=copy \ + -c /constraints.txt \ + "$(echo /dist/vllm-*.whl)[device-${DEVICE_ARCH}]" \ + "huggingface_hub" \ + "transformers==${TRANSFORMERS_VERSION}" \ + --index-url "${PYTORCH_INDEX_URL}" \ + --extra-index-url https://pypi.org/simple/ \ + --index-strategy unsafe-first-match \ + --prerelease allow + +# /hf -- HuggingFace cache, bind-mount from the host to avoid re-downloads +# /cache -- Triton/MIOpen kernel caches + HOME, bind-mount so they stay warm +# across runs (a cold MIOpen cache costs seconds at every startup). +# 0777 because the container is typically run as the invoking host UID, which +# may not exist in /etc/passwd inside the image. +RUN mkdir -p /hf /cache/triton /cache/miopen /workspace \ + && chmod -R 777 /hf /cache /workspace + +ENV HOME=/cache \ + HF_HOME=/hf \ + TRITON_CACHE_DIR=/cache/triton \ + MIOPEN_USER_DB_PATH=/cache/miopen \ + MIOPEN_CUSTOM_CACHE_DIR=/cache/miopen \ + MIOPEN_FIND_DB_PATH=/cache/miopen \ + PYTHONUNBUFFERED=1 + +# TRITON_HIP_USE_IN_THREAD_TRANSPOSE=0 is the non-obvious one: Triton 3.8's +# in-thread transpose drives vLLM's TRITON_ATTN prefill kernel to the 256-VGPR +# cap on gfx1151, spilling 120 B/thread. At the ViT-encoder shape that is most +# of a ~5% VLM TTFT regression, so it matters for VLM workloads. +ENV TORCH_ROCM_AOTRITON_ENABLE_EXPERIMENTAL=1 \ + FLASH_ATTENTION_TRITON_AMD_ENABLE=TRUE \ + TORCH_BLAS_PREFER_HIPBLASLT=1 \ + TRITON_HIP_USE_IN_THREAD_TRANSPOSE=0 + +LABEL org.opencontainers.image.source="https://github.com/ROCm/vllm" +LABEL org.opencontainers.image.description="vLLM for AMD RDNA3/3.5 (gfx11xx), built from ROCm wheels" +LABEL org.opencontainers.image.licenses="Apache-2.0" + +WORKDIR /workspace