Niko1221/Strata
1
0 Fork 0
mirror of https://github.com/Niko1221/Strata.git synced 2026-11-07 18:35:04 -10:00
Code Issues Packages Projects Releases 1 Wiki Activity

Merge f-amd-vision (d5fda4d2f) into f-split-int: GPU picture reader for AMD (strata-vision-hip) with automatic CPU fallback; pictures on by default on AMD; packaging for Linux and Windows

Browse Source
This commit is contained in:
Niko1221 committed 2026-10-18 22:53:23 +02:00
2 parents 8667a3dbf7 d5fda4d2f3
commit d6aa6504e0
22 files changed
+825 -68
Download Patch File Download Diff File Expand all files Collapse all files

No files matched your search

apps/host/server/native-vision.mjs
+1
View File
@@ -5,5 +5,6 @@ export function runtimeVision(config,requested){
if(mode==='no')return undefined;
if(mode===undefined)return config.vision;
if(!config.vision)throw new Error('Prepare this model’s vision files before enabling picture reading.');
if(mode==='cpu'&&config.vision.cpu_exe){const {cpu_exe,...rest}=config.vision;return {...rest,exe:cpu_exe,gpu:false};} // AMD: the CPU reader program, not the HIP one with the GPU hidden
return {...config.vision,...(mode==='cpu'?{gpu:false}:mode==='gpu'?{gpu:true}:{})};
}
apps/host/tests/native-onboarding.test.mjs
+4
View File
@@ -145,6 +145,10 @@ test('runtime vision checks prepared files, validates mode, and leaves the sourc
assert.equal(runtimeVision(config,'cpu').gpu,false);
assert.equal(runtimeVision(config,'yes').mmproj,config.vision.mmproj);
assert.deepEqual(config,before);
const amd={vision:{exe:'/e/strata-vision-hip',cpu_exe:'/e/strata-vision',mmproj:'/m.gguf',gpu:true}};
assert.deepEqual(runtimeVision(amd,'cpu'),{exe:'/e/strata-vision',mmproj:'/m.gguf',gpu:false});
assert.equal(runtimeVision(amd,'gpu').exe,'/e/strata-vision-hip');assert.equal(runtimeVision(amd,'gpu').cpu_exe,'/e/strata-vision');
assert.equal(runtimeVision(amd,'yes').gpu,true);
assert.throws(()=>runtimeVision({},'yes'),/Prepare/);
assert.throws(()=>runtimeVision(config,'invalid'),/picture reading/);
});
docs/AMD_HIP.md
+8 -4
View File
@@ -51,7 +51,12 @@ the kernel's amdgpu driver (no ROCm install needed):
architectures above; the engine is compiled for each of them (cards of two families, e.g. gfx1100 + gfx1201, need
a system ROCm 7: AMD's wheels hold one family). A split pays only when no single card holds the model's experts
(see RDNA4 below).
- **Limits for now:** images only through the CPU encoder (`--vision cpu`, 0.1.32). Setup does not offer the tuning
- **Pictures:** on by default. The engine package carries two picture readers: `strata-vision-hip`, which reads a picture on the
card (llama.cpp's HIP backend; about 1.4 GiB of VRAM, measured on an R9700: 0.06 s for a 260-token picture and 0.4 s for 972
tokens, against 3.1 s and 14 s on the CPU), and `strata-vision`, the CPU reader. If the GPU reader cannot start (the card is
not one it was built for, the HIP runtime does not load, the card is busy) the server starts the CPU reader by itself and
prints one line; nothing is asked. `--vision cpu` picks the CPU reader on purpose.
- **Limits for now:** Setup does not offer the tuning
(calibration) on AMD yet: its controls are being checked on HIP one at a time (#566). Since 0.1.39 a tuning run by
hand (`./setup.sh --calibrate`) is saved for the AMD card it ran on and reused when setup runs again. The Monitor
shows the card's load, VRAM, temperature and power from Linux sysfs (0.1.32).
@@ -84,8 +89,7 @@ On a PC with no NVIDIA card Strata can use, the AMD card is chosen by itself; wi
the PATH): if the HIP runtime does not see the card, setup stops there and points to the driver. It also gives the
card's HIP number: with an integrated Radeon that is device 1, not 0 (#325). From then on setup lists the AMD cards
as HIP numbers them, so `--gpu N` and the config's `"gpu"` are HIP numbers.
- **Differences from Windows-on-NVIDIA and Linux-on-AMD:** no images yet (the CPU image encoder is Linux-only for
now), one card per model (`--gpus` is Linux-only for now), no calibration.
- **Differences from Windows-on-NVIDIA and Linux-on-AMD:** one card per model (`--gpus` is Linux-only for now), no calibration.
- Two Windows-only engine details (#247, #325): hipBLAS can return success and still leave `hipErrorInvalidValue`
set after some BF16/FP16 GEMMs (seen on gfx1201); the engine clears that one stale error after a GEMM that
succeeded, on Windows only. `hipHostGetDevicePointer` returns the host pointer itself on Windows: kernels read
@@ -260,7 +264,7 @@ sizes itself automatically and leaves 1 GiB of VRAM headroom.
GPU memory to system RAM (GTT), the OOM killer then ends KWin/plasmashell or `systemd-oomd` ends apps, or the
compositor fails with "Failed to pin framebuffer with error -12".
The installer supports this backend (see "Install with setup" above). Images run through the CPU encoder for now (`--vision cpu`).
The installer supports this backend (see "Install with setup" above). Pictures are read on the card (`strata-vision-hip`) when the engine has it, else on the CPU (`strata-vision`).
Setup installs one AMD card, or several with `--gpus` (the engine's layer split; see RDNA4 below).
**Opt-in switches for RDNA2 prompts and hangs (0.1.40, off by default):** `STRATA_DENSE_MMQ=1` (#820, HIP) runs the
docs/INSTALL.md
+1 -2
View File
@@ -80,8 +80,7 @@ On Linux setup uses a system ROCm 7 when there is one, or installs ROCm into `.v
sudo), and compiles the engine on your PC for the card (10-20 minutes, once; it needs a C++ compiler and git:
`sudo apt install build-essential git`). Several AMD cards share the model with `--gpus`, as on NVIDIA.
What differs from NVIDIA for now: pictures are read by the image encoder on the CPU (`--vision cpu`, about 3 s per
picture at the default 300 image tokens on 8 cores), `--calibrate` is NVIDIA-only, and Unsloth's 4-bit model needs an NVIDIA card. Measurements per card, the
What differs from NVIDIA for now: `--calibrate` is NVIDIA-only, and Unsloth's 4-bit model needs an NVIDIA card. Measurements per card, the
build by hand and the tuning tables: [AMD_HIP.md](AMD_HIP.md).
## Two or three cards
docs/TROUBLESHOOTING.md
+2 -1
View File
@@ -133,7 +133,8 @@ setup warns when it finds 13.2.0 or 13.2.1 for such a card, and takes an older 1
**Pictures are refused, or slow.**
"this server was started without the vision encoder": the model was set up for text only - run setup again with
`--vision gpu` (or `--vision cpu`). Pictures that take several seconds (about 3 s at 300 image tokens on 8 cores, more with more tokens) are read by the encoder on the CPU; `--vision gpu`
(NVIDIA, ~1.4 GB of VRAM) makes it 0.1-0.5 s.
(NVIDIA and AMD, ~1.4 GB of VRAM) makes it 0.1-0.5 s. On AMD the GPU reader is the default; if it cannot start, the server
prints "the GPU picture reader did not start (...); pictures are read on the CPU" and goes on with the CPU reader.
## AMD cards
packaging/ENGINE-COMPONENT.md
+12 -2
View File
@@ -95,9 +95,9 @@ the manifest file (which holds every file's hash), used as the package's digest
|---|---|---|---|
| `windows-x86_64` cuda, variant `cuda13` | the owner PC, the same cmake line as the release zips (CUDA 13.0, `75;86;89;120` real + `120` PTX, `STRATA_PORTABLE=ON`) | `strata.exe`, `strata_kq.dll`, `strata-vision.exe` (reused, unchanged since 0.1.44), `BUILD.json` | `min_driver` nvidia 580.65 |
| `windows-x86_64` cuda, variant `cuda12` | the owner PC (CUDA 12.9, `60;61;70;75;80;86;89` real + `89` PTX, `STRATA_EXPERIMENTAL_SM60=ON`) | the same four files | experimental, `min_driver` nvidia 575.51; the CUDA 13 build wins when both fit |
| `windows-x86_64` hip | the owner PC, `tools/hip/build_windows.bat` (ROCm 10.2 nightly wheels) then `tools/hip/package_windows.py` | `strata.exe`, `strata-device.exe`, the ROCm DLLs and the rocBLAS/hipBLASLt kernels of the zip (`--tree`) | no Windows AMD test machine yet |
| `windows-x86_64` hip | the owner PC, `tools/hip/build_windows.bat` (ROCm 10.2 nightly wheels) then `tools/hip/package_windows.py` | `strata.exe`, `strata-device.exe`, the ROCm DLLs and the rocBLAS/hipBLASLt kernels of the zip (`--tree`) | no Windows AMD test machine yet; with `STRATA_NO_VISION` unset it also builds the two picture readers (`strata-vision.exe` on the CPU, `strata-vision-hip.exe` on the GPU) |
| `linux-x86_64` cuda | a Linux box with the CUDA toolkit | `strata`, `libstrata_kq.so`, `libcudart/libcublas/libcublasLt` (13.x) | the GPU driver's `libcuda` stays the PC's |
| `linux-x86_64` hip | a box with ROCm (the Python wheels are enough) | `strata`, the HIP runtime, hipBLAS/hipBLASLt/rocBLAS, comgr + libLLVM, `rocblas/library` and `hipblaslt/library` for the listed gfx archs | |
| `linux-x86_64` hip | a box with ROCm (the Python wheels are enough) | `strata`, the HIP runtime, hipBLAS/hipBLASLt/rocBLAS, comgr + libLLVM, `rocblas/library` and `hipblaslt/library` for the listed gfx archs; `packaging/build_hip_linux.sh` builds it with both picture readers |
| `linux-x86_64` hip, gfx1103 (Radeon 780M / Phoenix iGPU) | `packaging/build_hip_gfx1103.sh` on the mini PC (TheRock gfx110X-all SDK) | the same set as the gfx1200/1201 package, with `rocblas/library` and `hipblaslt/library` for gfx1103 (600 MiB) | experimental, unvalidated: the iGPU resets now and then (MES REMOVE_QUEUE, notes-minipc.md); never kill the engine mid-request; run configs set `STRATA_NO_ARENA_THP=1`. `engine_component.resolve` picks a HIP package only when its `archs` list a gfx target the Linux kfd topology reports |
| `linux-x86_64` sycl | a box with oneAPI DPC++ | `strata` (JIT from SPIR-V), libsycl, the Unified Runtime adapters, MKL SYCL libraries | the Level Zero loader and compute runtime are the PC's driver |
| `macos-arm64` metal | an Apple Silicon Mac | `strata`, `metal/kernels/*.metal`, `metal/engine-msl/*.metal`, `metal/ggml-common.h`, `metal/data/*.bin` | the Metal engine compiles its kernel sources when it starts; since S1 it finds them beside the executable (`metal/src/resource_path.h`: environment variable, then `<exe dir>/metal/`, then the compiled-in source path) |
@@ -105,3 +105,13 @@ the manifest file (which holds every file's hash), used as the package's digest
`packaging/collect_libs.py` gathers the Linux libraries (an `ldd` closure with the dlopen'd extras named, kernel data filtered by
GPU architecture) into the folder `package_engine.py` packages; libraries the distribution provides are listed in
`SYSTEM-LIBS.txt` inside the package so a missing one can be named on a target PC.
## Picture readers in an AMD package
A HIP package carries two programs for pictures next to `strata` (`vision` in its BUILD.json): `strata-vision` is the CPU reader, which
needs nothing of the GPU runtime and starts on every PC, and `strata-vision-hip` is the same reader on the card (llama.cpp's HIP
backend, built for the package's `archs`; BUILD.json `vision_archs`). Setup writes the run config with the GPU reader first and the CPU
reader as `vision.cpu_exe`; when the GPU reader does not start (the card is not in its build, the HIP runtime does not load, the card is
busy) the server starts the CPU reader by itself and prints one line. Both programs have RUNPATH `$ORIGIN` on Linux; on Windows the
ROCm DLLs they import sit beside the programs or in `rocm/bin` (the engine's PATH entry, `lib_dirs`). `strata_control/vision_reader.py`
holds the choice; `tools/test_vision_reader.py` the tests.
packaging/build_hip_linux.sh
+96
View File
@@ -0,0 +1,96 @@
#!/bin/bash
# Linux AMD (HIP) engine package: strata + the CPU picture reader (strata-vision) + the GPU picture reader (strata-vision-hip),
# the ROCm libraries both load, and the rocBLAS / hipBLASLt kernel data for the listed archs, packaged as an engine component.
# Run on a Linux box that has ROCm (TheRock Python wheels are enough); the tester packages were built in an Ubuntu 22.04 chroot
# (glibc 2.35) so they run on every distribution since. All settings are environment variables:
#
# VARIANT=rx9070 rx9070 (gfx1200 gfx1201) | rx7900 (gfx1100 gfx1101 gfx1102) | halo (gfx1151) | igpu (gfx1103 gfx1150) | a name + ARCHS=
# ARCHS="gfx1200;gfx1201" the code objects (default: the variant's)
# SRC=<this checkout> the Strata source (default: the checkout this script is in)
# LLAMA=<llama.cpp checkout> at the pinned commit (setup.py LLAMA_CPP_COMMIT); mtmd and ggml come from it
# HP=<ROCm devel root> CORE=<ROCm core root> LIBW=<ROCm libraries root> (TheRock: _rocm_sdk_devel, _rocm_sdk_core, _rocm_sdk_libraries_*)
# OUT=<work folder> PKGVER=0.1.45.3 COMMIT=<source commit for the build id> JOBS=16
# VISION_ONLY=1 build only the two readers (no engine, no package)
# NO_PACKAGE=1 stop after the stage folder
# VISION_HIP_DEFS="-DGGML_CUDA_FA=OFF" extra cmake definitions for the GPU reader only
#
# Both readers get RUNPATH $ORIGIN (tools/vision/CMakeLists.txt), and every bundled library is patched to $ORIGIN, so the
# component folder runs from anywhere with nothing from the environment (packaging/binary_deps.py checks it).
set -u
V=${VARIANT:?VARIANT=rx9070|rx7900|halo|igpu|<name> with ARCHS=}
SRC=${SRC:-$(cd "$(dirname "$0")/.." && pwd)}
LLAMA=${LLAMA:?LLAMA=<llama.cpp checkout at the pinned commit>}
HP=${HP:?HP=<ROCm devel root>}
CORE=${CORE:-$HP}
LIBW=${LIBW:-$HP}
OUT=${OUT:-$PWD/work-$V}
PKGVER=${PKGVER:-0.1.45.3}
COMMIT=${COMMIT:-unknown}
JOBS=${JOBS:-16}
case "$V" in
rx9070) ARCHS=${ARCHS:-gfx1200;gfx1201} ;;
rx7900) ARCHS=${ARCHS:-gfx1100;gfx1101;gfx1102} ;;
halo) ARCHS=${ARCHS:-gfx1151} ;;
igpu) ARCHS=${ARCHS:-gfx1103;gfx1150} ;;
*) ARCHS=${ARCHS:?ARCHS=gfx...} ;;
esac
ARCHLIST=${ARCHS//;/ }
DATAARCH=${ARCHS//;/,}
mkdir -p "$OUT"; S=$OUT/build.status
st() { echo "$* $(date +%T)" >> "$S"; }
: > "$S"
export HIP_PLATFORM=amd HIP_COMPILER=clang HIP_RUNTIME=rocclr ROCM_PATH=$HP HIP_PATH=$HP
HIPDEFS=(-DCMAKE_HIP_COMPILER=$HP/llvm/bin/clang++ -DCMAKE_HIP_COMPILER_ROCM_ROOT=$HP "-DCMAKE_PREFIX_PATH=$HP;$LIBW"
"-DCMAKE_HIP_FLAGS=--rocm-path=$HP --rocm-device-lib-path=$HP/lib/llvm/amdgcn/bitcode")
st "start $V ($ARCHS) commit $COMMIT"
cd "$SRC" || exit 1
if [ -z "${VISION_ONLY:-}" ]; then
cmake -G Ninja -S . -B "$OUT/b" -DCMAKE_BUILD_TYPE=Release -DSTRATA_ENABLE_HIP=ON -DSTRATA_ENABLE_CUDA=OFF -DSTRATA_BUILD_TESTS=OFF \
-DSTRATA_PREFILL_MMQ=ON "-DCMAKE_HIP_ARCHITECTURES=$ARCHS" "${HIPDEFS[@]}" -DSTRATA_GGML_DIR="$LLAMA" -DSTRATA_PORTABLE=ON \
"-DCMAKE_EXE_LINKER_FLAGS=-Wl,-rpath,'\$ORIGIN' -Wl,--enable-new-dtags" > "$OUT/cfg.log" 2>&1
st "engine configured rc=$?"
nice -n 10 cmake --build "$OUT/b" --target strata -j "$JOBS" > "$OUT/build.log" 2>&1
st "engine built rc=$?"
[ -x "$OUT/b/strata" ] || { st "ABORT no strata"; exit 1; }
fi
# the CPU reader: portable (AVX2), the fallback that runs on any PC
cmake -G Ninja -S tools/vision -B "$OUT/bv" -DCMAKE_BUILD_TYPE=Release -DLLAMA_DIR="$LLAMA" -DSTRATA_VISION_CUDA=OFF \
-DSTRATA_PORTABLE=ON > "$OUT/cfg-vision.log" 2>&1
nice -n 10 cmake --build "$OUT/bv" --target strata-vision -j "$JOBS" > "$OUT/build-vision.log" 2>&1
st "cpu reader built rc=$?"
[ -x "$OUT/bv/bin/strata-vision" ] || { st "ABORT no strata-vision"; exit 1; }
# the GPU reader: ggml-hip for the same archs; the CPU kernels in it are portable too
cmake -G Ninja -S tools/vision -B "$OUT/bh" -DCMAKE_BUILD_TYPE=Release -DLLAMA_DIR="$LLAMA" -DSTRATA_VISION_HIP=ON \
"-DSTRATA_VISION_HIP_ARCHS=$ARCHS" -DSTRATA_PORTABLE=ON "${HIPDEFS[@]}" ${VISION_HIP_DEFS:-} > "$OUT/cfg-vision-hip.log" 2>&1
st "gpu reader configured rc=$?"
nice -n 10 cmake --build "$OUT/bh" --target strata-vision -j "$JOBS" > "$OUT/build-vision-hip.log" 2>&1
st "gpu reader built rc=$?"
[ -x "$OUT/bh/bin/strata-vision-hip" ] || { st "ABORT no strata-vision-hip"; tail -20 "$OUT/build-vision-hip.log" >> "$S"; exit 1; }
[ -n "${VISION_ONLY:-}" ] && { st "VISION-DONE"; exit 0; }
STAGE=$OUT/stage; rm -rf "$STAGE"; mkdir -p "$STAGE"
ARCHJSON=$(for a in $ARCHLIST; do printf '"%s", ' "$a"; done | sed 's/, $//')
cat > "$STAGE/BUILD.json" <<EOJ
{"version": "$PKGVER", "source": "f-amd-vision $COMMIT", "variant": "$V", "archs": [$ARCHJSON], "portable": true, "vision": "gpu", "vision_cpu": "strata-vision", "vision_gpu": "strata-vision-hip", "vision_archs": [$ARCHJSON]}
EOJ
# the libraries the engine and the GPU reader load (the CPU reader needs only the system's), collected from the build outputs
DATA=(--data "$LIBW/lib/rocblas/library=rocblas/library" --data "$LIBW/lib/hipblaslt/library=hipblaslt/library" --data-arch "$DATAARCH")
for b in "$OUT/b/strata" "$OUT/bh/bin/strata-vision-hip"; do
python3 packaging/collect_libs.py --binary "$b" --search "$LIBW/lib" --search "$HP/lib" --search "$CORE/lib" --search "$CORE/lib/llvm/lib" \
--also 'libamd_comgr.so*' --also 'libhiprtc*.so*' --out "$STAGE" \
"${DATA[@]}" >> "$OUT/collect.log" 2>&1
st "collected for $(basename "$b") rc=$?"
DATA=() # the kernel data once
done
cp "$OUT/b/strata" "$OUT/bv/bin/strata-vision" "$OUT/bh/bin/strata-vision-hip" "$STAGE/"
for f in "$STAGE"/*.so "$STAGE"/*.so.*; do [ -f "$f" ] && [ ! -L "$f" ] && patchelf --set-rpath '$ORIGIN' "$f"; done
for f in "$STAGE/strata" "$STAGE/strata-vision" "$STAGE/strata-vision-hip"; do
st "RUNPATH $(basename "$f"): $(readelf -d "$f" | grep -E 'RPATH|RUNPATH' | cut -d'[' -f2 | tr -d ']' | paste -sd' ')"
done
[ -n "${NO_PACKAGE:-}" ] && { st "STAGE-DONE"; exit 0; }
rm -rf "$OUT/components"; mkdir -p "$OUT/components/engine"
ARCHARGS=""; for a in $ARCHLIST; do ARCHARGS="$ARCHARGS --arch $a"; done
python3 packaging/package_engine.py --src "$STAGE" --backend hip --version "$PKGVER" --out "$OUT/components/engine" $ARCHARGS \
--variant "$V" --build-id "$COMMIT+rocm" --tree --notes "AMD $V build: engine, CPU picture reader, GPU picture reader" > "$OUT/package.log" 2>&1
st "packaged rc=$?"
cat "$OUT/package.log" >> "$S"
st "BUILD-DONE"
packaging/package_engine.py
+1 -1
View File
@@ -53,7 +53,7 @@ def sha256(path: Path) -> str:
def wanted(path: Path) -> bool:
name = path.name.lower()
return (name.endswith((".dll", ".so", ".dylib", ".metallib")) or ".so." in name or name == "build.json"
or name.startswith("license") or name in ("strata-vision", "strata-vision.exe")) # the picture reader beside the engine (pictures need it)
or name.startswith("license") or name in ("strata-vision", "strata-vision.exe", "strata-vision-hip", "strata-vision-hip.exe")) # the picture readers (CPU, AMD GPU) beside the engine
def git_build_id(repo: Path = ROOT) -> str:
serve/server.py
+35 -10
View File
@@ -66,6 +66,7 @@ from serve.frontend import (THINK_END, ChatTemplate, Event, OutputParser, anthro
from serve.mcp import McpCancelled, hub_from_config # noqa: E402
from serve import runconfig # noqa: E402
from strata_control import engine_component # noqa: E402
from strata_control import vision_reader # noqa: E402
from serve import ready as ready_file # noqa: E402
from serve import owner_guard # noqa: E402
from serve import launch_identity # noqa: E402
@@ -2490,18 +2491,13 @@ class Vision:
return Path(tempfile.mkdtemp(prefix="strata-vision-"))
def __init__(self, cfg: dict, log=None, env: dict | None = None, lazy: bool = False):
absolute = lambda p: os.path.abspath(p) if os.path.dirname(str(p)) else p # the encoder runs in its own dir (#480) # noqa: E731
args = [absolute(cfg["exe"]), "--mmproj", absolute(cfg["mmproj"]), "--model", absolute(cfg["model"])]
if cfg.get("gpu"):
args.append("--gpu")
if cfg.get("threads"):
args += ["--threads", str(cfg["threads"])]
if cfg.get("max_tokens"):
args += ["--max-tokens", str(cfg["max_tokens"])]
if cfg.get("min_tokens"): # #767: mtmd's image_min_tokens (a hand-edited key)
args += ["--min-tokens", str(cfg["min_tokens"])]
args = self.reader_args(cfg)
self.dir = self.work_dir()
self.spawn = (args, log, env) # to start it again after an unload
# AMD: the GPU reader (strata-vision-hip) has the CPU reader (strata-vision, same program, no GPU runtime) to
# fall back on when it does not start; _start switches to it by itself
cpu = vision_reader.cpu_fallback(cfg)
self.fallback = (self.reader_args(cpu), log, env) if cpu else None
self.proc = None
self.stopped = True
self.lock = threading.Lock()
@@ -2515,7 +2511,36 @@ class Vision:
if not lazy and not self.elastic:
self._start()
@staticmethod
def reader_args(cfg: dict) -> list:
absolute = lambda p: os.path.abspath(p) if os.path.dirname(str(p)) else p # the encoder runs in its own dir (#480) # noqa: E731
args = [absolute(cfg["exe"]), "--mmproj", absolute(cfg["mmproj"]), "--model", absolute(cfg["model"])]
if cfg.get("gpu"):
args.append("--gpu")
if cfg.get("threads"):
args += ["--threads", str(cfg["threads"])]
if cfg.get("max_tokens"):
args += ["--max-tokens", str(cfg["max_tokens"])]
if cfg.get("min_tokens"): # #767: mtmd's image_min_tokens (a hand-edited key)
args += ["--min-tokens", str(cfg["min_tokens"])]
return args
def _start(self):
"""Start the reader. When a GPU reader that has a CPU reader to fall back on does not start (the GPU runtime does not
load, the card is not in its build, the card is busy), the CPU reader is started instead, for good, with one log line
and no question."""
try:
self._start_reader()
except (RuntimeError, OSError) as e:
fallback = getattr(self, "fallback", None)
if fallback is None:
raise
why = " ".join(str(e).split())[:240]
print(f"[strata] the GPU picture reader did not start ({why}); pictures are read on the CPU", flush=True)
self.spawn, self.fallback = fallback, None
self._start_reader()
def _start_reader(self):
args, log, env = self.spawn
# cwd is the encoder's directory: the ENC lines name the files relative to it (#480: no space in the line)
self.proc = popen("the image encoder", args, role="vision", stdin=subprocess.PIPE, stdout=subprocess.PIPE,
setup.py
+89 -24
View File
@@ -574,6 +574,7 @@ VISION = {"gpu": {"max_tokens": 1024, "reserve_mib": 700},
"cpu": {"max_tokens": 300, "reserve_mib": 700}}
EXE = "strata.exe" if WIN else "strata"
VEXE = "strata-vision.exe" if WIN else "strata-vision"
HVEXE = "strata-vision-hip.exe" if WIN else "strata-vision-hip" # AMD: the picture reader on the GPU (tools/vision, STRATA_VISION_HIP)
# MODEL-UPDATE Phase 1: the K-quant prompt kernels (Unsloth's UD-Q4/Q5/Q6_K_XL experts) are a library next to the CUDA
# engine, opened only for those models; an engine without it runs them on the FP16 prompt path
KQ_LIB = "strata_kq.dll" if WIN else "libstrata_kq.so"
@@ -3340,12 +3341,13 @@ def build_engine_hip(gpu, llama, vision="none") -> Path:
floor = cpu_floor(cpu_info()[1]) # "" on an AVX2 CPU: the normal engine
engine_ok = meta.get("backend") == "hip" and (eng / EXE).exists() and meta.get("src") == src and has_archs and \
(meta.get("isa_floor") or "") == floor and not stale_binary(eng / EXE)
vision_ok = vision == "none" or ((eng / VEXE).exists() and meta.get("vision_src") == vsrc)
vision_ok = vision == "none" or ((eng / VEXE).exists() and meta.get("vision_src") == vsrc and
(vision != "gpu" or ((eng / HVEXE).exists() and set(archs) <= set(meta.get("vision_archs", [])))))
if engine_ok and vision_ok:
ok("engine already built for this PC")
return eng
if engine_ok:
return build_vision_cpu(eng, stamp, meta, llama, vsrc)
return build_vision_hip(eng, stamp, meta, llama, vsrc, archs, vision)
if not (shutil.which("c++") or shutil.which("g++")) or not shutil.which("git"):
fail("a C++ compiler and git are needed to compile the AMD engine",
"Ubuntu/Debian: sudo apt install build-essential git Fedora: sudo dnf install gcc-c++ git")
@@ -3376,23 +3378,80 @@ def build_engine_hip(gpu, llama, vision="none") -> Path:
meta = {"source": "local-hip", "backend": "hip", "version": source_version(), "archs": archs, "vision": "none",
"lib_dirs": dirs, "src": src, **({"isa_floor": floor} if floor else {})}
if vision != "none":
return build_vision_cpu(eng, stamp, meta, llama, vsrc)
return build_vision_hip(eng, stamp, meta, llama, vsrc, archs, vision)
stamp.write_text(json.dumps(meta, indent=1))
ok(f"engine compiled: {eng / EXE}")
return eng
def hip_vision(asked) -> str:
"""The image encoder with the AMD backend (--vision): the CPU one when asked for (#304); a HIP (GPU) encoder build
is a later step, so `yes`/`gpu` leave images off, as before, and say how to get them."""
if asked in ("yes", "gpu"):
warn("the AMD backend has no GPU image encoder yet: images off"
+ ("" if WIN else " (--vision cpu reads them on the CPU)"))
if asked == "cpu" and WIN:
warn("images on the CPU with an AMD card are Linux-only for now (the ready-made Windows AMD engine has no "
"image encoder): images off")
"""The picture reader wanted with the AMD backend (--vision): "gpu" for yes/gpu and when nothing was asked (pictures are on
by default), "cpu" when asked for (#304), "none" for no. "gpu" is a wish: hip_reader() settles it against what the
engine folder carries (the HIP reader, else the CPU one) once the engine is known."""
if asked in ("no", "none"):
return "none"
return "cpu" if asked == "cpu" else "none"
return "cpu" if asked == "cpu" else "gpu"
def reader_archs(eng: Path) -> list:
"""The GPU architectures the engine folder's HIP picture reader was built for: its BUILD.json (vision_archs, else archs),
else the component manifest's archs; [] when neither says (the reader then checks the card itself when it starts)."""
for name, keys in (("BUILD.json", ("vision_archs", "archs")), ("engine-component.json", ("archs",))):
try:
data = json.loads((eng / name).read_text(encoding="utf-8-sig"))
except (OSError, ValueError):
continue
for key in keys:
if isinstance(data, dict) and isinstance(data.get(key), list) and data[key]:
return [str(x) for x in data[key]]
return []
def hip_reader(eng: Path, vision: str, gpu: dict):
"""AMD: the picture reader this engine folder gets for `vision` ("gpu" or "cpu"; strata_control/vision_reader.py): the HIP
reader with the CPU reader to fall back on, else the CPU reader. Returns (vision, Reader); a Reader of mode "none" means
the folder has no reader. The one plain note (a card the HIP reader has no code for, no HIP reader in this engine) is said."""
from strata_control import vision_reader as VR
r = VR.choose(eng, vision, windows=WIN, gpu_arch=str(gpu.get("arch") or ""), archs=reader_archs(eng))
if r.note:
warn(r.note)
return r.mode, r
def build_vision_hip(eng: Path, stamp: Path, meta: dict, llama, vsrc, archs: list, vision: str) -> Path:
"""The picture readers beside the HIP engine: the CPU reader always (the fallback), and for "gpu" the HIP reader
(tools/vision, STRATA_VISION_HIP) for the same architectures. The HIP reader needs the ROCm environment the engine build
sets; a build that fails leaves the CPU reader and says so (pictures still work, on the CPU)."""
if vision == "gpu" and not (shutil.which("hipconfig") or os.environ.get("ROCM_PATH")):
root, dirs = rocm_root(archs) # the engine was built earlier: set the same environment again
os.environ.update({"HIP_PLATFORM": "amd", "HIP_COMPILER": "clang", "HIP_RUNTIME": "rocclr", "ROCM_PATH": str(root),
"HIP_PATH": str(root)})
os.environ["LD_LIBRARY_PATH"] = os.pathsep.join(dirs + [os.environ.get("LD_LIBRARY_PATH", "")]).rstrip(os.pathsep)
os.environ["PATH"] = os.pathsep.join([str(root / "bin"), str(root / "llvm" / "bin"), os.environ.get("PATH", "")])
build_vision_cpu(eng, stamp, meta, llama, vsrc)
meta = json.loads(stamp.read_text(encoding="utf-8"))
if vision != "gpu":
return eng
root, dirs = rocm_root(archs)
libs = [str(Path(d).parent) for d in dirs[1:]]
bitcode = next((p for p in (root / "lib" / "llvm" / "amdgcn" / "bitcode", root / "amdgcn" / "bitcode") if p.is_dir()),
root / "amdgcn" / "bitcode")
say(" Compiling the picture reader for your AMD GPU (a few minutes, once) ...")
try:
cmake_build(ROOT / "tools" / "vision", ROOT / "build-vision-hip", "strata-vision",
[f"-DLLAMA_DIR={llama}", "-DSTRATA_VISION_HIP=ON", "-DSTRATA_VISION_HIP_ARCHS=" + ";".join(archs),
"-DSTRATA_PORTABLE=OFF", f"-DCMAKE_HIP_COMPILER={root / 'llvm' / 'bin' / 'clang++'}",
f"-DCMAKE_HIP_COMPILER_ROCM_ROOT={root}", "-DCMAKE_PREFIX_PATH=" + ";".join([str(root), *libs]),
f"-DCMAKE_HIP_FLAGS=--rocm-path={root} --rocm-device-lib-path={bitcode}"], None, "")
shutil.copy2(ROOT / "build-vision-hip" / "bin" / HVEXE, eng / HVEXE)
except SystemExit:
raise
except Exception as e: # noqa: BLE001 - the CPU reader is there; pictures work without the GPU one
warn(f"the GPU picture reader did not compile ({e}): pictures are read on the CPU")
return eng
stamp.write_text(json.dumps({**meta, "vision": "gpu", "vision_archs": archs, "vision_gpu": HVEXE, "vision_cpu": VEXE}, indent=1))
ok(f"picture reader (GPU): {eng / HVEXE}, (CPU fallback): {eng / VEXE}")
return eng
def build_vision_cpu(eng: Path, stamp: Path, meta: dict, llama, vsrc) -> Path:
@@ -4841,7 +4900,7 @@ SETUP_ENV = frozenset({"STRATA_HIPBLASLT_TUNING", "STRATA_ROCBLAS_TUNING", "STRA
# user's own (STRATA_ROUTE_TAIL_SKIP=0 to turn the tail skip off, STRATA_MMQ_KQ=1 ...): a setup run again keeps them there
# (0.1.44 dropped them with the rest of SETUP_ENV).
SETUP_ENV_UNSLOTH = frozenset({"STRATA_MMQ_KQ", "STRATA_HC_Q8", "STRATA_ROUTE_TAIL_SKIP"})
SETUP_VISION = frozenset({"exe", "mmproj", "model", "gpu", "max_tokens", "threads"})
SETUP_VISION = frozenset({"exe", "cpu_exe", "mmproj", "model", "gpu", "max_tokens", "threads"})
def carry_over(old: dict, cfg: dict) -> list[str]:
@@ -6029,22 +6088,16 @@ def own_path_vision(a, fam: dict, gpu: dict, hip: bool) -> str:
say()
say(f" Images: {fam['title']} can also read pictures (screenshots, photos, scanned pages). This adds a "
f"{fam['sha256'][fam['mmproj']][0] / 1e9:.1f} GB download" + (
"; the encoder runs on the CPU (no VRAM), a few seconds per picture." if hip else
"; the encoder runs on the GPU when the engine has the AMD one (about 1 GB of VRAM), else on the CPU." if hip else
" and the image encoder runs on the GPU (~1.4 GB of VRAM kept free, 0.1-0.5 s per picture)." if big else
"; the encoder runs on the CPU (a few seconds per picture), because the GPU encoder would take ~1.4 GB "
"of VRAM from the experts on this card."))
default = "n" if (hip and WIN) else "y" if (big or hip) else "n"
default = "y" if (big or hip) else "n"
if ask("Do you want images?", ["y", "n"], default, a.yes) != "y":
return "none"
asked = "cpu" if (hip or not big) else "gpu"
asked = "gpu" if hip else "cpu" if not big else "gpu"
if hip:
if asked == "gpu":
say(" The AMD backend has no GPU image encoder yet: the encoder runs on the CPU.")
if WIN:
warn("images on the CPU with an AMD card are Linux-only for now (the ready-made Windows AMD engine has no "
"image encoder): images off")
return "none"
return "cpu"
return hip_vision(asked) # settled against the engine's readers in hip_reader() (GPU reader, else CPU)
if asked == "yes":
asked = "gpu" if big else "cpu"
if asked == "gpu" and not big and vram > 0:
@@ -6207,7 +6260,10 @@ def install_own_path(a, family, gpu, hip, amd, cuda_tk, multi, port, ram, adopte
step(4, "the Strata engine")
llama, eng, meta, lib_dirs, gpu, vision = engine_step(a, gpu, hip, amd, cuda_tk, vision)
if vision != "none" and not (eng / VEXE).exists():
reader = None
if hip and vision != "none":
vision, reader = hip_reader(eng, vision, gpu)
elif vision != "none" and not (eng / VEXE).exists():
warn("this engine has no image encoder: images off (run setup again with --build to compile it)")
vision = "none"
if fam["arch"] not in engine_families(eng):
@@ -6276,6 +6332,9 @@ def install_own_path(a, family, gpu, hip, amd, cuda_tk, multi, port, ram, adopte
cfg["args"].append("--vision")
cfg["vision"] = {"exe": str(eng / VEXE), "mmproj": str(mmproj), "model": str(gguf), "gpu": vision == "gpu",
"max_tokens": vt} # the text model, opened vocab-only
if reader is not None: # AMD: the HIP reader with the CPU reader behind it, or the CPU reader
from strata_control import vision_reader as VR
cfg["vision"] = VR.vision_section(reader, cfg["vision"])
if vision == "cpu":
cfg["vision"]["threads"] = vision_threads()
elif getattr(a, "vision_tokens", None) is not None:
@@ -7261,6 +7320,9 @@ def main() -> int:
# ---- 4. the engine
step(4, "the Strata engine")
llama, eng, meta, lib_dirs, gpu, vision = engine_step(a, gpu, hip, amd, cuda_tk, vision)
reader = None
if hip and vision != "none": # AMD: the engine folder's reader (GPU, else CPU), settled now the engine is known
vision, reader = hip_reader(eng, vision, gpu)
if (ready or family == "gyro") and meta.get("source") != "local" and \
tuple(int(x) for x in str(meta.get("version", "0")).split(".")[:4] if x.isdigit()) < MODELS[model]["engine"]:
# a ready-made engine from before NVFP4 / Gyro cannot read these experts: one compiled from this source can
@@ -7668,6 +7730,9 @@ def main() -> int:
cfg["vision"] = {"exe": str(eng / VEXE), "mmproj": str(mmproj), # the text model, opened vocab-only:
"model": str(ready_paths(model, models_dir)["dense"] if ready else shards[0]),
"gpu": vision == "gpu", "max_tokens": vt}
if reader is not None: # AMD: the HIP reader with the CPU reader behind it, or the CPU reader
from strata_control import vision_reader as VR
cfg["vision"] = VR.vision_section(reader, cfg["vision"])
if vision == "cpu":
cfg["vision"]["threads"] = vision_threads()
elif a.vision_tokens is not None:
strata_cli/launch/screens.py
+4 -1
View File
@@ -820,7 +820,10 @@ class PicturesScreen(Base):
def explain(self) -> Text:
vram = self.ctx.computer.main_vram_gb
extra = ("a 0.9 GB download. On a card under 24 GB the picture reader runs on the CPU (a few seconds per picture, nothing "
amd = any(str(g.get("vendor")) == "amd" for g in self.ctx.computer.usable_gpus)
extra = ("a 0.9 GB download. On your AMD card the picture reader runs on the graphics card (about 1 GB of its memory, under a second "
"per picture) and starts on the processor by itself if it cannot." if amd and self.reader() != "missing" else
"a 0.9 GB download. On a card under 24 GB the picture reader runs on the CPU (a few seconds per picture, nothing "
"taken from the graphics card). On a bigger card it runs on the GPU and keeps about 1.4 GB free, costing a few % of text speed."
if vram < 24 else "a 0.9 GB download, and about 1.4 GB of graphics memory kept free for the picture reader (text is a few % slower).")
if not self.supported:
strata_control/vision_reader.py
+101
View File
@@ -0,0 +1,101 @@
"""Which picture reader an engine folder gets: the GPU one (`strata-vision-hip`, AMD) or the CPU one (`strata-vision`).
An AMD engine package carries two programs for pictures. `strata-vision` is the CPU reader: it needs nothing from the GPU
runtime, so it starts on every PC. `strata-vision-hip` is the same reader on the card (llama.cpp's HIP backend), a
fraction of a second per picture instead of seconds. Setup writes the run config with the GPU reader first and the CPU
reader as its fallback (`vision.cpu_exe`); the server (serve/server.py, Vision) starts the CPU reader by itself, with one
log line, when the GPU reader does not start: the HIP runtime does not load, the card is not one the build has code for
(the reader says so before it runs a kernel), or the card is busy. Nothing is asked of the user.
`choose` is the whole decision; setup.py calls it, and tools/test_vision_reader.py checks every branch.
"""
from __future__ import annotations
import re
from dataclasses import dataclass
from pathlib import Path
CPU_READER = "strata-vision"
GPU_READER = "strata-vision-hip"
GFX = re.compile(r"gfx[0-9a-f]+", re.IGNORECASE)
def reader_name(base: str, windows: bool) -> str:
return base + (".exe" if windows else "")
def find(folder: Path | str, windows: bool) -> dict:
"""The readers in an engine folder: {"cpu": Path | None, "gpu": Path | None}."""
folder = Path(folder)
found = {}
for key, base in (("cpu", CPU_READER), ("gpu", GPU_READER)):
path = folder / reader_name(base, windows)
found[key] = path if path.is_file() else None
return found
def gfx_of(text: object) -> str:
""""gfx1201" out of a gcnArchName ("gfx1201:sramecc-:xnack-"), a device line or a plain arch; "" when there is none."""
m = GFX.search(str(text or ""))
return m.group(0).lower() if m else ""
@dataclass
class Reader:
mode: str # "gpu" | "cpu" | "none"
exe: Path | None = None # the program the run config starts
cpu_exe: Path | None = None # the CPU reader to fall back on when a GPU reader does not start (mode "gpu" only)
note: str = "" # one plain sentence for the setup output when it is not what was asked for
def choose(folder: Path | str, asked: str, *, windows: bool, gpu_arch: str = "", archs: list | None = None) -> Reader:
"""The reader for `asked` ("yes"/"on"/"gpu" = the best the folder has, "cpu", "no"/"none"/"off").
gpu_arch: the card's architecture when setup knows it ("gfx1201"); archs: what the GPU reader was built for
(BUILD.json `vision_archs`, else the package's `archs`). A card the build does not list gets the CPU reader here;
when the card is unknown the GPU reader is tried and falls back by itself (see the module doc)."""
asked = (asked or "").lower()
if asked in ("no", "none", "off", ""):
return Reader("none")
have = find(folder, windows)
if asked == "cpu":
if have["cpu"]:
return Reader("cpu", have["cpu"])
return Reader("none", note="this engine has no picture reader: pictures off")
# yes / on / gpu
gfx = gfx_of(gpu_arch)
built = [gfx_of(a) for a in (archs or []) if gfx_of(a)]
if have["gpu"] is None:
if have["cpu"]:
return Reader("cpu", have["cpu"], note="this engine has no GPU picture reader: pictures are read on the CPU "
"(a few seconds each)" if asked == "gpu" else "")
return Reader("none", note="this engine has no picture reader: pictures off")
if gfx and built and gfx not in built:
if have["cpu"]:
return Reader("cpu", have["cpu"], note=f"the GPU picture reader has no code for your card ({gfx}): pictures "
"are read on the CPU (a few seconds each)")
return Reader("none", note=f"the GPU picture reader has no code for your card ({gfx}): pictures off")
return Reader("gpu", have["gpu"], have["cpu"])
def vision_section(reader: Reader, base: dict) -> dict:
"""The run config's "vision" section for `reader` on top of `base` (mmproj, model, max_tokens ...): exe, gpu, and for the
GPU reader the CPU reader to fall back on."""
out = dict(base)
out["exe"] = str(reader.exe)
out["gpu"] = reader.mode == "gpu"
if reader.mode == "gpu" and reader.cpu_exe:
out["cpu_exe"] = str(reader.cpu_exe)
else:
out.pop("cpu_exe", None)
return out
def cpu_fallback(vision_cfg: dict) -> dict | None:
"""The same section for the CPU reader, when the GPU reader has one to fall back on; None otherwise."""
if not vision_cfg.get("gpu") or not vision_cfg.get("cpu_exe"):
return None
out = {k: v for k, v in vision_cfg.items() if k not in ("cpu_exe", "elastic", "elastic_mib", "elastic_idle_s")}
out["exe"], out["gpu"] = vision_cfg["cpu_exe"], False
return out
tools/hip/build_windows.bat
+32 -3
View File
@@ -16,6 +16,9 @@ rem gfx1031 = RX 6700 XT / 6800M, #915 #1078; drop it fro
rem STRATA_ROCM_VERSION 10.2.0a20260930 STRATA_ROCM_INDEX https://nightly.repo.amd.com/rocm/whl-next/
rem ROCM_VENV <repo>\.rocm-win BUILD_DIR <repo>\build-hip-win DIST_DIR <repo>\dist
rem STRATA_GGML_DIR a llama.cpp checkout at the pinned commit (default: CMake fetches it)
rem STRATA_ROCM_KEEP 1 = use the ROCM_VENV as it is (no pip install, even when its archs differ from STRATA_HIP_ARCHS)
rem STRATA_VISION_HIP_ARCHS the GPU reader's archs (default: STRATA_HIP_ARCHS; about 70 MB of the zip per arch)
rem STRATA_NO_VISION 1 = skip the two picture readers (strata-vision.exe for the CPU, strata-vision-hip.exe for the GPU)
setlocal EnableDelayedExpansion
for %%I in ("%~dp0..\..") do set "SRC=%%~fI"
if not defined STRATA_HIP_ARCHS set "STRATA_HIP_ARCHS=gfx1100;gfx1101;gfx1102;gfx1200;gfx1201;gfx1030;gfx1031;gfx1151"
@@ -39,7 +42,7 @@ set "STAMP=%ROCM_VENV%\strata-rocm.txt"
set "WANT=%STRATA_ROCM_VERSION% %EXTRAS%"
set "HAVE="
if exist "%STAMP%" set /p HAVE=<"%STAMP%"
if not "!HAVE!"=="!WANT!" (
if not "!HAVE!"=="!WANT!" if not "%STRATA_ROCM_KEEP%"=="1" (
echo Installing ROCm %STRATA_ROCM_VERSION% [%EXTRAS%] into %ROCM_VENV% ...
"%ROCM_VENV%\Scripts\python.exe" -m pip install --disable-pip-version-check --index-url "%STRATA_ROCM_INDEX%" "rocm[%EXTRAS%]==%STRATA_ROCM_VERSION%" || exit /b 1
"%ROCM_VENV%\Scripts\rocm-sdk.exe" init || exit /b 1
@@ -79,9 +82,35 @@ if "%TESTS%"=="ON" (
cmake --build "%BUILD_DIR%" --target strata strata-device || exit /b 1
)
rem ---- 4. the zip: the two programs, the ROCm DLLs they load (+ rocBLAS/hipBLASLt kernels for these archs), licenses
rem ---- 3b. the picture readers (tools/vision, the same llama.cpp as the engine): strata-vision.exe on the CPU (MSVC, portable: the
rem fallback that starts on any PC) and strata-vision-hip.exe on the GPU (ROCm's clang, ggml-hip, the archs above)
set "VISION_ARGS="
if not defined STRATA_VISION_HIP_ARCHS set "STRATA_VISION_HIP_ARCHS=%STRATA_HIP_ARCHS%"
if not "%STRATA_NO_VISION%"=="1" (
set "LLAMA=%STRATA_GGML_DIR%"
if not defined LLAMA set "LLAMA=%BUILD_DIR%\_deps\strata_llamacpp-src"
set "LLAMA=!LLAMA:\=/!"
if not exist "!LLAMA!/tools/mtmd" (echo llama.cpp with tools/mtmd not found in !LLAMA! - set STRATA_GGML_DIR & exit /b 1)
if not exist "%SRC%\build-vision-cpu-win\build.ninja" (
cmake -G Ninja -S "%SRC%\tools\vision" -B "%SRC%\build-vision-cpu-win" -DCMAKE_BUILD_TYPE=Release -DLLAMA_DIR="!LLAMA!" ^
-DSTRATA_VISION_CUDA=OFF -DSTRATA_PORTABLE=ON -DCMAKE_C_COMPILER=cl -DCMAKE_CXX_COMPILER=cl || exit /b 1
)
cmake --build "%SRC%\build-vision-cpu-win" --target strata-vision || exit /b 1
if not exist "%SRC%\build-vision-hip-win\build.ninja" (
cmake -G Ninja -S "%SRC%\tools\vision" -B "%SRC%\build-vision-hip-win" -DCMAKE_BUILD_TYPE=Release -DLLAMA_DIR="!LLAMA!" ^
-DSTRATA_VISION_HIP=ON "-DSTRATA_VISION_HIP_ARCHS=%STRATA_VISION_HIP_ARCHS%" -DSTRATA_PORTABLE=ON ^
"-DCMAKE_CXX_FLAGS=--rocm-path=%ROCM_F% --rocm-device-lib-path=%BITCODE%" ^
"-DCMAKE_C_COMPILER=%ROCM_F%/lib/llvm/bin/clang.exe" "-DCMAKE_CXX_COMPILER=%ROCM_F%/lib/llvm/bin/clang++.exe" ^
"-DCMAKE_HIP_COMPILER=%ROCM_F%/lib/llvm/bin/clang++.exe" "-DCMAKE_HIP_COMPILER_ROCM_ROOT=%ROCM_F%" ^
"-DCMAKE_PREFIX_PATH=%ROCM_F%" "-DCMAKE_HIP_FLAGS=--rocm-path=%ROCM_F% --rocm-device-lib-path=%BITCODE%" || exit /b 1
)
cmake --build "%SRC%\build-vision-hip-win" --target strata-vision || exit /b 1
set "VISION_ARGS=--vision-cpu %SRC%\build-vision-cpu-win\bin\strata-vision.exe --vision-hip %SRC%\build-vision-hip-win\bin\strata-vision-hip.exe"
)
rem ---- 4. the zip: the programs, the ROCm DLLs they load (+ rocBLAS/hipBLASLt kernels for these archs), licenses
"%ROCM_VENV%\Scripts\python.exe" "%SRC%\tools\hip\package_windows.py" --build "%BUILD_DIR%" --rocm "%ROCM%" ^
--archs "%STRATA_HIP_ARCHS%" --rocm-version "%STRATA_ROCM_VERSION%" --out "%DIST_DIR%" || exit /b 1
--archs "%STRATA_HIP_ARCHS%" --rocm-version "%STRATA_ROCM_VERSION%" --out "%DIST_DIR%" !VISION_ARGS! || exit /b 1
echo.
echo Done: %DIST_DIR%\strata-windows-x64-hip.zip
endlocal
tools/hip/package_windows.py
+13 -1
View File
@@ -7,6 +7,9 @@
tools/hip/build_windows.bat runs it after the build. The zip holds:
strata.exe, strata-device.exe, BUILD.json (backend "hip", the archs, the ROCm and hipBLASLt versions)
strata-vision.exe, strata-vision-hip.exe the picture readers when --vision-cpu / --vision-hip are given: the CPU reader
(the fallback; starts on any PC) and the GPU reader (tools/vision,
STRATA_VISION_HIP; imports the same ROCm DLLs as strata.exe)
amdhip64_7.dll, amd_comgr.dll + what they import the HIP runtime beside the exes too (#468 #461: before System32)
rocm/bin/*.dll the ROCm DLLs the two programs load (their import tables, followed
through the ROCm DLLs, + amd_comgr.dll, which the HIP runtime loads
@@ -73,6 +76,8 @@ def main() -> int:
ap.add_argument("--archs", required=True, help="the archs the engine was compiled for, ';'-separated")
ap.add_argument("--rocm-version", required=True)
ap.add_argument("--out", required=True, type=Path)
ap.add_argument("--vision-cpu", type=Path, help="strata-vision.exe built for the CPU (tools/vision, STRATA_PORTABLE=ON)")
ap.add_argument("--vision-hip", type=Path, help="strata-vision-hip.exe built for the same archs (STRATA_VISION_HIP=ON)")
a = ap.parse_args()
archs = [x for x in re.split(r"[;, ]+", a.archs) if x]
rbin = a.rocm / "bin"
@@ -83,10 +88,13 @@ def main() -> int:
for p in PROGRAMS:
shutil.copy2(a.build / p, stage / p)
readers = [(src, src.name) for src in (a.vision_cpu, a.vision_hip) if src is not None]
for src, name in readers:
shutil.copy2(src, stage / name)
# the ROCm DLLs: the programs' imports, followed through the ROCm DLLs themselves
rocm_dlls = {p.name.lower(): p for p in rbin.glob("*.dll")}
need, todo, crt = [], [stage / p for p in PROGRAMS] + [rbin / d for d in DYNAMIC], set()
need, todo, crt = [], [stage / p for p in PROGRAMS] + [stage / n for _, n in readers] + [rbin / d for d in DYNAMIC], set()
while todo:
f = todo.pop()
for name in imports(objdump, f):
@@ -157,6 +165,10 @@ def main() -> int:
meta = {"source": "prebuilt", "backend": "hip", "platform": "windows-x64", "version": version, "archs": archs,
"rocm": a.rocm_version, "hipblaslt_version": hlv[0] * 100000 + hlv[1] * 100 + hlv[2], "vision": "none",
"lib_dirs": ["rocm/bin"]}
if a.vision_cpu is not None:
meta.update({"vision": "cpu", "vision_cpu": a.vision_cpu.name})
if a.vision_hip is not None:
meta.update({"vision": "gpu", "vision_gpu": a.vision_hip.name, "vision_archs": archs})
(stage / "BUILD.json").write_text(json.dumps(meta, indent=1), encoding="utf-8")
z = a.out / ASSET
tools/test_setup_amd.py
+31 -7
View File
@@ -429,13 +429,37 @@ class CalibrationKey(unittest.TestCase):
self.assertIsNone(setup.setup_calibration(cfg, hip=True)) # Windows HIP: defaults for now
class WindowsHipVision(unittest.TestCase):
def test_no_cpu_encoder_on_windows(self):
with mock.patch.object(setup, "WIN", True), mock.patch.object(setup, "warn", lambda *a: None):
self.assertEqual(setup.hip_vision("cpu"), "none")
self.assertEqual(setup.hip_vision("yes"), "none")
with mock.patch.object(setup, "WIN", False), mock.patch.object(setup, "warn", lambda *a: None):
self.assertEqual(setup.hip_vision("cpu"), "cpu")
class HipVisionWish(unittest.TestCase):
"""AMD pictures are on by default, on Linux and Windows: the wish is "gpu" (hip_reader() settles it against the readers the
engine folder carries: the HIP one, else the CPU one), "cpu" when asked for, "none" for no."""
def test_wish(self):
for win in (True, False):
with mock.patch.object(setup, "WIN", win):
self.assertEqual(setup.hip_vision(None), "gpu")
self.assertEqual(setup.hip_vision("yes"), "gpu")
self.assertEqual(setup.hip_vision("gpu"), "gpu")
self.assertEqual(setup.hip_vision("cpu"), "cpu")
self.assertEqual(setup.hip_vision("no"), "none")
self.assertEqual(setup.hip_vision("none"), "none")
def test_reader_settled_against_the_engine_folder(self):
for win, suffix in ((False, ""), (True, ".exe")):
with tempfile.TemporaryDirectory() as d, mock.patch.object(setup, "WIN", win), mock.patch.object(setup, "warn", lambda *a: None):
eng = Path(d)
(eng / ("strata-vision" + suffix)).write_bytes(b"x")
self.assertEqual(setup.hip_reader(eng, "gpu", {"arch": "gfx1201"})[0], "cpu") # no HIP reader: the CPU one
(eng / ("strata-vision-hip" + suffix)).write_bytes(b"x")
mode, r = setup.hip_reader(eng, "gpu", {"arch": "gfx1201"})
self.assertEqual((mode, r.exe.name, r.cpu_exe.name),
("gpu", "strata-vision-hip" + suffix, "strata-vision" + suffix))
(eng / "BUILD.json").write_text('{"vision_archs": ["gfx1100"]}')
self.assertEqual(setup.hip_reader(eng, "gpu", {"arch": "gfx1201"})[0], "cpu") # the card is not in its build
self.assertEqual(setup.hip_reader(eng, "gpu", {"arch": "gfx1100"})[0], "gpu")
self.assertEqual(setup.hip_reader(eng, "cpu", {"arch": "gfx1100"})[0], "cpu")
(eng / ("strata-vision" + suffix)).unlink()
(eng / ("strata-vision-hip" + suffix)).unlink()
self.assertEqual(setup.hip_reader(eng, "gpu", {"arch": "gfx1100"})[0], "none")
class HipRuntimeBesideExe(unittest.TestCase):
tools/test_setup_choices.py
+4 -6
View File
@@ -318,15 +318,13 @@ class ExperimentalSm60(unittest.TestCase):
class HipVision(unittest.TestCase):
"""#304: --vision cpu with --backend hip builds the CPU image encoder beside the HIP engine."""
@mock.patch.object(setup, "WIN", False) # Linux: compiled here (Windows has no AMD image encoder yet)
@mock.patch.object(setup, "WIN", False)
def test_the_choice(self):
self.assertEqual(quiet(setup.hip_vision, "cpu"), ("cpu", ""))
for asked in (None, "no", "none"):
for asked in ("no", "none"):
self.assertEqual(quiet(setup.hip_vision, asked), ("none", ""))
for asked in ("yes", "gpu"):
got, out = quiet(setup.hip_vision, asked)
self.assertEqual(got, "none")
self.assertIn("--vision cpu", out)
for asked in (None, "yes", "gpu"): # pictures are on by default; the engine folder's readers settle gpu vs cpu
self.assertEqual(quiet(setup.hip_vision, asked), ("gpu", ""))
def build(self, meta, vision, vexe=False):
"""build_engine_hip with an engine that is already built: only the encoder can be missing."""
tools/test_setup_qwen36.py
+29 -6
View File
@@ -88,6 +88,7 @@ class InstallOwnPath(unittest.TestCase):
self.root = Path(self.tmp.name)
self.calls = []
self.families = {"qwen4exp", "qwen35moe"}
self.hip_reader = True
eng = self.root / "engine"
self.cfg = {}
@@ -96,6 +97,8 @@ class InstallOwnPath(unittest.TestCase):
if vision != "none": # the image encoder beside the engine
eng.mkdir(parents=True, exist_ok=True)
(eng / setup.VEXE).write_text("x")
if hip and self.hip_reader: # an AMD engine package carries the GPU reader too
(eng / setup.HVEXE).write_text("x")
return self.root / "llama", eng, {"version": "0.1.44"}, ["/libs"], gpu, vision
def write_cfg(path, cfg, source=None):
@@ -219,16 +222,36 @@ class InstallOwnPath(unittest.TestCase):
self.assertTrue(self.cfg["vision"]["gpu"])
self.assertIn("kept as you chose", out)
def test_amd_reads_pictures_on_the_cpu(self):
with mock.patch.object(setup, "WIN", False):
def test_amd_reads_pictures_on_the_gpu_with_the_cpu_reader_behind_it(self):
for win, suffix in ((False, ""), (True, ".exe")):
with mock.patch.object(setup, "WIN", win), mock.patch.object(setup, "VEXE", "strata-vision" + suffix), mock.patch.object(setup, "HVEXE", "strata-vision-hip" + suffix):
rc, out = self.install(model="UD-Q4_K_XL", vram=32.0, hip=True, vision="yes")
self.assertEqual(rc, 0, out)
v = self.cfg["vision"]
self.assertTrue(v["gpu"])
self.assertEqual(Path(v["exe"]).name, "strata-vision-hip" + suffix)
self.assertEqual(Path(v["cpu_exe"]).name, "strata-vision" + suffix)
self.assertIn("--vision", self.cfg["args"])
rc, out = self.install(model="UD-Q4_K_XL", vram=32.0, hip=True, vision="cpu")
v = self.cfg["vision"]
self.assertFalse(v["gpu"])
self.assertEqual(Path(v["exe"]).name, "strata-vision" + suffix)
self.assertNotIn("cpu_exe", v)
def test_amd_engine_without_the_gpu_reader_reads_pictures_on_the_cpu(self):
self.hip_reader = False
with mock.patch.object(setup, "WIN", False), mock.patch.object(setup, "VEXE", "strata-vision"):
rc, out = self.install(model="UD-Q4_K_XL", vram=32.0, hip=True, vision="yes")
self.assertEqual(rc, 0, out)
self.assertFalse(self.cfg["vision"]["gpu"])
self.assertNotIn("cpu_exe", self.cfg["vision"])
self.assertIn("--vision", self.cfg["args"])
with mock.patch.object(setup, "WIN", True):
rc, out = self.install(model="UD-Q4_K_XL", vram=32.0, hip=True, vision="cpu")
self.assertNotIn("vision", self.cfg)
self.assertIn("Linux-only", out)
def test_amd_pictures_are_on_when_nothing_is_asked(self):
with mock.patch.object(setup, "WIN", False), mock.patch.object(setup, "VEXE", "strata-vision"), mock.patch.object(setup, "HVEXE", "strata-vision-hip"):
rc, out = self.install(model="UD-Q4_K_XL", vram=32.0, hip=True)
self.assertEqual(rc, 0, out)
self.assertTrue(self.cfg["vision"]["gpu"])
def test_parallel_is_opt_in_and_recommended_where_the_experts_fit(self):
fam = setup.OWN_PATH_FAMILIES["qwen36"]
tools/test_vision_reader.py
+202
View File
@@ -0,0 +1,202 @@
"""The picture reader choice for an AMD engine folder (strata_control/vision_reader.py) and the server's fallback from the GPU
reader to the CPU reader (serve/server.py, Vision._start): hip present, hip missing, hip start failure, then cpu.
python -m unittest tools.test_vision_reader
"""
from __future__ import annotations
import io
import tempfile
import unittest
from pathlib import Path
from types import SimpleNamespace
from unittest import mock
from strata_control import vision_reader as VR
import serve.server as server
def folder(*names: str) -> Path:
d = Path(tempfile.mkdtemp(prefix="vr-"))
for n in names:
(d / n).write_bytes(b"x")
return d
class Choose(unittest.TestCase):
def test_hip_present_gives_the_gpu_reader_with_the_cpu_one_to_fall_back_on(self):
d = folder("strata-vision", "strata-vision-hip")
for ask in ("yes", "on", "gpu"):
r = VR.choose(d, ask, windows=False, gpu_arch="gfx1201", archs=["gfx1200", "gfx1201"])
self.assertEqual((r.mode, r.exe.name, r.cpu_exe.name), ("gpu", "strata-vision-hip", "strata-vision"), ask)
self.assertEqual(r.note, "")
def test_hip_missing_gives_the_cpu_reader(self):
d = folder("strata-vision")
r = VR.choose(d, "yes", windows=False, gpu_arch="gfx1201")
self.assertEqual((r.mode, r.exe.name, r.cpu_exe), ("cpu", "strata-vision", None))
self.assertEqual(r.note, "") # "yes" never asked for the GPU in particular
self.assertIn("no GPU picture reader", VR.choose(d, "gpu", windows=False).note) # --vision gpu: said once
def test_card_not_in_the_build_gives_the_cpu_reader(self):
d = folder("strata-vision", "strata-vision-hip")
r = VR.choose(d, "yes", windows=False, gpu_arch="gfx1100", archs=["gfx1200", "gfx1201"])
self.assertEqual((r.mode, r.exe.name), ("cpu", "strata-vision"))
self.assertIn("gfx1100", r.note)
def test_unknown_card_tries_the_gpu_reader(self):
d = folder("strata-vision", "strata-vision-hip")
self.assertEqual(VR.choose(d, "yes", windows=False, gpu_arch="", archs=["gfx1201"]).mode, "gpu")
self.assertEqual(VR.choose(d, "yes", windows=False, gpu_arch="gfx1100", archs=None).mode, "gpu")
def test_gcn_arch_names_with_features_match(self):
self.assertEqual(VR.gfx_of("gfx1201:sramecc-:xnack-"), "gfx1201")
d = folder("strata-vision", "strata-vision-hip")
self.assertEqual(VR.choose(d, "yes", windows=False, gpu_arch="gfx1151:sramecc+:xnack-", archs=["gfx1151"]).mode, "gpu")
def test_cpu_asked_and_off(self):
d = folder("strata-vision", "strata-vision-hip")
self.assertEqual(VR.choose(d, "cpu", windows=False).exe.name, "strata-vision")
for ask in ("no", "none", "off", None):
self.assertEqual(VR.choose(d, ask, windows=False).mode, "none")
def test_no_reader_at_all(self):
d = folder("strata")
for ask in ("yes", "cpu", "gpu"):
r = VR.choose(d, ask, windows=False)
self.assertEqual(r.mode, "none", ask)
self.assertIn("no picture reader", r.note)
def test_windows_names(self):
d = folder("strata-vision.exe", "strata-vision-hip.exe")
r = VR.choose(d, "yes", windows=True, gpu_arch="gfx1201", archs=["gfx1201"])
self.assertEqual((r.mode, r.exe.name, r.cpu_exe.name), ("gpu", "strata-vision-hip.exe", "strata-vision.exe"))
self.assertEqual(VR.choose(folder("strata-vision-hip"), "yes", windows=True).mode, "none") # a Linux name on Windows
def test_run_config_section(self):
d = folder("strata-vision", "strata-vision-hip")
base = {"mmproj": "m.gguf", "model": "t.gguf", "max_tokens": 1024}
sec = VR.vision_section(VR.choose(d, "yes", windows=False), base)
self.assertTrue(sec["gpu"])
self.assertEqual(Path(sec["exe"]).name, "strata-vision-hip")
self.assertEqual(Path(sec["cpu_exe"]).name, "strata-vision")
cpu = VR.cpu_fallback(sec)
self.assertFalse(cpu["gpu"])
self.assertEqual(Path(cpu["exe"]).name, "strata-vision")
self.assertNotIn("cpu_exe", cpu)
self.assertEqual(cpu["mmproj"], "m.gguf")
cpu_only = VR.vision_section(VR.choose(folder("strata-vision"), "yes", windows=False), base)
self.assertFalse(cpu_only["gpu"])
self.assertNotIn("cpu_exe", cpu_only)
self.assertIsNone(VR.cpu_fallback(cpu_only))
self.assertIsNone(VR.cpu_fallback({**sec, "cpu_exe": ""}))
self.assertIsNone(VR.cpu_fallback({"exe": "strata-vision", "gpu": True})) # an NVIDIA config has no fallback
class FakeProc:
def __init__(self, first_line: str):
self.stdin, self.stdout = io.StringIO(), io.StringIO(first_line)
self.killed = False
def kill(self):
self.killed = True
def wait(self, timeout=None):
return 0
def poll(self):
return None
def make_vision(cfg: dict) -> "server.Vision":
with mock.patch.object(server.Vision, "work_dir", staticmethod(lambda: Path(tempfile.gettempdir()))):
return server.Vision(cfg, lazy=True)
class Fallback(unittest.TestCase):
CFG = {"exe": "/e/strata-vision-hip", "cpu_exe": "/e/strata-vision", "gpu": True, "mmproj": "/m/mm.gguf",
"model": "/m/t.gguf", "max_tokens": 1024, "threads": 6}
def run_start(self, answers: dict, cfg: dict | None = None):
"""answers: exe name -> its first stdout line, or an exception to raise from popen."""
started = []
def popen(what, args, **kw):
name = Path(args[0]).name
started.append(args)
a = answers[name]
if isinstance(a, Exception):
raise a
return FakeProc(a)
v = make_vision(cfg or self.CFG)
out = io.StringIO()
with mock.patch.object(server, "popen", popen), mock.patch("builtins.print", lambda *a, **k: out.write(" ".join(map(str, a)) + "\n")):
err = None
try:
v._start()
except RuntimeError as e:
err = e
return v, started, out.getvalue(), err
def test_hip_present_and_starts(self):
v, started, log, err = self.run_start({"strata-vision-hip": "READY 2048\n"})
self.assertIsNone(err)
self.assertEqual([Path(a[0]).name for a in started], ["strata-vision-hip"])
self.assertIn("--gpu", started[0])
self.assertEqual(log, "")
self.assertFalse(v.stopped)
def test_hip_start_failure_falls_back_to_cpu_with_one_line(self):
v, started, log, err = self.run_start({"strata-vision-hip": "ERR this GPU (gfx1100) is not in the picture reader's build (gfx1200,gfx1201)\n",
"strata-vision": "READY 2048\n"})
self.assertIsNone(err)
self.assertEqual([Path(a[0]).name for a in started], ["strata-vision-hip", "strata-vision"])
self.assertNotIn("--gpu", started[1])
self.assertIn("--threads", started[1])
self.assertEqual(log.count("\n"), 1)
self.assertIn("gfx1100", log)
self.assertIn("CPU", log)
self.assertFalse(v.stopped)
self.assertIsNone(v.fallback)
def test_hip_process_that_dies_silently_falls_back(self): # a GPU runtime that aborts: stdout closes empty
v, started, log, err = self.run_start({"strata-vision-hip": "", "strata-vision": "READY 2048\n"})
self.assertIsNone(err)
self.assertEqual(len(started), 2)
self.assertEqual(log.count("\n"), 1)
def test_hip_missing_file_falls_back(self): # popen raises (exe gone, not executable)
v, started, log, err = self.run_start({"strata-vision-hip": FileNotFoundError("strata-vision-hip"),
"strata-vision": "READY 2048\n"})
self.assertIsNone(err)
self.assertEqual(len(started), 2)
def test_restart_after_a_fallback_stays_on_the_cpu_reader(self):
v, started, log, err = self.run_start({"strata-vision-hip": "ERR nope\n", "strata-vision": "READY 2048\n"})
started.clear()
with mock.patch.object(server, "popen", lambda what, args, **kw: started.append(args) or FakeProc("READY 2048\n")):
v.restart()
self.assertEqual([Path(a[0]).name for a in started], ["strata-vision"])
def test_cpu_failure_after_the_fallback_is_an_error(self):
v, started, log, err = self.run_start({"strata-vision-hip": "ERR nope\n", "strata-vision": "ERR cannot load the vision encoder\n"})
self.assertIsNotNone(err)
self.assertIn("cannot load", str(err))
def test_a_config_without_cpu_exe_has_no_fallback(self): # NVIDIA: the error is shown as before
cfg = {k: v for k, v in self.CFG.items() if k != "cpu_exe"}
v, started, log, err = self.run_start({"strata-vision-hip": "ERR nope\n"}, cfg)
self.assertIsNotNone(err)
self.assertEqual(len(started), 1)
self.assertEqual(log, "")
def test_cpu_config_is_unchanged(self):
cfg = {"exe": "/e/strata-vision", "gpu": False, "mmproj": "/m/mm.gguf", "model": "/m/t.gguf"}
v, started, log, err = self.run_start({"strata-vision": "READY 2048\n"}, cfg)
self.assertIsNone(err)
self.assertNotIn("--gpu", started[0])
if __name__ == "__main__":
unittest.main()
tools/vision/CMakeLists.txt
+39
View File
@@ -3,6 +3,7 @@
#
# cmake -S tools/vision -B build-vision -G Ninja -DCMAKE_BUILD_TYPE=Release -DLLAMA_DIR=third_party/llama.cpp
# [-DSTRATA_VISION_CUDA=ON -DCMAKE_CUDA_ARCHITECTURES=native] [-DSTRATA_PORTABLE=OFF]
# [-DSTRATA_VISION_HIP=ON -DSTRATA_VISION_HIP_ARCHS="gfx1200;gfx1201" + the HIP compiler the engine build uses]
# cmake --build build-vision --target strata-vision
#
# Portable (AVX2 baseline) is the default, so an encoder built without extra flags runs on any AVX2 PC: the 0.1.32
@@ -14,6 +15,14 @@ project(strata_vision C CXX)
set(LLAMA_DIR "${CMAKE_CURRENT_SOURCE_DIR}/../../third_party/llama.cpp" CACHE PATH "llama.cpp checkout")
option(STRATA_VISION_CUDA "run the vision encoder on the GPU (builds ggml-cuda)" OFF)
option(STRATA_VISION_HIP "run the vision encoder on an AMD GPU (builds ggml-hip); the program is named strata-vision-hip" OFF)
set(STRATA_VISION_HIP_ARCHS "" CACHE STRING "the gfx architectures strata-vision-hip is compiled for (';'-separated)")
if(STRATA_VISION_HIP AND STRATA_VISION_CUDA)
message(FATAL_ERROR "strata-vision: STRATA_VISION_CUDA and STRATA_VISION_HIP are two different programs")
endif()
if(STRATA_VISION_HIP AND NOT STRATA_VISION_HIP_ARCHS)
message(FATAL_ERROR "strata-vision: STRATA_VISION_HIP needs STRATA_VISION_HIP_ARCHS, for example \"gfx1200;gfx1201\"")
endif()
option(STRATA_PORTABLE "build for other PCs: AVX2 baseline, static MSVC runtime" ON)
if(STRATA_PORTABLE AND MSVC)
set(CMAKE_MSVC_RUNTIME_LIBRARY "MultiThreaded")
@@ -52,6 +61,13 @@ else()
endif()
set(GGML_BACKEND_DL OFF CACHE BOOL "" FORCE)
set(GGML_CUDA ${STRATA_VISION_CUDA} CACHE BOOL "" FORCE)
set(GGML_HIP ${STRATA_VISION_HIP} CACHE BOOL "" FORCE)
if(STRATA_VISION_HIP)
# the engine's HIP build says the same (cmake/hip_backend.cmake); ggml-hip reads AMDGPU_TARGETS / CMAKE_HIP_ARCHITECTURES
set(AMDGPU_TARGETS "${STRATA_VISION_HIP_ARCHS}" CACHE STRING "" FORCE)
set(GPU_TARGETS "${STRATA_VISION_HIP_ARCHS}" CACHE STRING "" FORCE)
set(CMAKE_HIP_ARCHITECTURES "${STRATA_VISION_HIP_ARCHS}" CACHE STRING "" FORCE)
endif()
add_subdirectory(${LLAMA_DIR} llama EXCLUDE_FROM_ALL)
@@ -59,9 +75,32 @@ add_executable(strata-vision strata_vision.cpp)
target_link_libraries(strata-vision PRIVATE mtmd llama)
target_compile_features(strata-vision PRIVATE cxx_std_17)
set_target_properties(strata-vision PROPERTIES RUNTIME_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR}/bin)
if(STRATA_VISION_HIP)
# beside strata (and, in a component, the CPU reader "strata-vision"): the GPU reader is its own program, so a PC where
# the HIP runtime does not load still has the CPU one to fall back on
set_target_properties(strata-vision PROPERTIES OUTPUT_NAME strata-vision-hip)
string(REPLACE ";" "," _hip_archs "${STRATA_VISION_HIP_ARCHS}")
target_compile_definitions(strata-vision PRIVATE STRATA_VISION_HIP STRATA_VISION_HIP_ARCHS="${_hip_archs}")
endif()
if(WIN32 AND STRATA_VISION_HIP)
# hipblas.dll (and rocblas.dll behind it) live in rocm\bin of the engine package: loaded on first use, after main() has put that folder on
# the DLL search path, so the reader starts from any PATH and says "no GPU" itself where there is none
target_link_libraries(strata-vision PRIVATE delayimp)
target_link_options(strata-vision PRIVATE -Wl,/delayload:hipblas.dll)
endif()
if(UNIX AND NOT APPLE)
# the libraries it needs are the ones in its own folder (a component folder is relocatable): RUNPATH $ORIGIN
set_target_properties(strata-vision PROPERTIES BUILD_WITH_INSTALL_RPATH ON INSTALL_RPATH "$ORIGIN")
target_link_options(strata-vision PRIVATE -Wl,--enable-new-dtags)
endif()
# opens a GGUF that has tensor types llama.cpp does not know (a Gyro quantization's TQK6 / TQK7) for its vocabulary only
add_executable(strata-vision-gguf-test EXCLUDE_FROM_ALL gguf_meta_test.cpp)
target_link_libraries(strata-vision-gguf-test PRIVATE llama)
target_compile_features(strata-vision-gguf-test PRIVATE cxx_std_17)
set_target_properties(strata-vision-gguf-test PROPERTIES RUNTIME_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR}/bin)
# the HIP reader's architecture check (gpu_arch.h), without a GPU
add_executable(strata-vision-arch-test EXCLUDE_FROM_ALL gpu_arch_test.cpp)
target_compile_features(strata-vision-arch-test PRIVATE cxx_std_17)
set_target_properties(strata-vision-arch-test PROPERTIES RUNTIME_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR}/bin)
tools/vision/gpu_arch.h
+33
View File
@@ -0,0 +1,33 @@
// tools/vision/gpu_arch.h - which GPU the HIP picture reader (strata-vision-hip) was compiled for, and whether the GPU it
// found is one of them. ggml-hip prints " Device 0: AMD Radeon RX 9070 XT, gfx1201:sramecc-:xnack- (0x1201), ..." when it
// starts; a card whose architecture is not in the build would fail at its first kernel ("no kernel image", an abort), so
// the reader reads that line and refuses before any kernel runs. The server then starts the CPU reader (see
// serve/server.py, Vision._start). Header only, so gpu_arch_test.cpp can check it without a GPU.
#pragma once
#include <cctype>
#include <string>
#include <vector>
// "gfx1201" out of ggml's device line (or a gcnArchName: "gfx1201:sramecc-:xnack-"); "" when there is none
inline std::string strata_gfx_in(const std::string& text) {
for (size_t at = text.find("gfx"); at != std::string::npos; at = text.find("gfx", at + 1)) {
size_t end = at + 3;
while (end < text.size() && std::isxdigit((unsigned char) text[end])) ++end;
if (end > at + 3) return text.substr(at, end - at);
}
return "";
}
// "gfx1200,gfx1201" (the build's list; ',' or ';' between) holds `arch`
inline bool strata_arch_listed(const std::string& list, const std::string& arch) {
if (arch.empty()) return false;
size_t start = 0;
while (start <= list.size()) {
size_t end = list.find_first_of(",; ", start);
if (end == std::string::npos) end = list.size();
if (list.compare(start, end - start, arch) == 0 && end - start == arch.size()) return true;
start = end + 1;
}
return false;
}
tools/vision/gpu_arch_test.cpp
+24
View File
@@ -0,0 +1,24 @@
// tools/vision/gpu_arch_test.cpp - the HIP picture reader's architecture check (gpu_arch.h) without a GPU. Exit 0 = passed.
#include "gpu_arch.h"
#include <cstdio>
static int bad = 0;
#define CHECK(c) do { if (!(c)) { std::printf("FAIL line %d: %s\n", __LINE__, #c); ++bad; } } while (0)
int main() {
CHECK(strata_gfx_in(" Device 0: AMD Radeon RX 9070 XT, gfx1201:sramecc-:xnack- (0x1201), VMM: no, Wave Size: 32") == "gfx1201");
CHECK(strata_gfx_in(" Device 0: AMD Radeon Graphics, gfx1151 (0x1151)") == "gfx1151");
CHECK(strata_gfx_in(" Device 0: gfx, nothing after it") == "");
CHECK(strata_gfx_in("no architecture here") == "");
CHECK(strata_arch_listed("gfx1200,gfx1201", "gfx1201"));
CHECK(strata_arch_listed("gfx1200;gfx1201", "gfx1200"));
CHECK(strata_arch_listed("gfx1100", "gfx1100"));
CHECK(!strata_arch_listed("gfx1200,gfx1201", "gfx1100"));
CHECK(!strata_arch_listed("gfx1100", "gfx110")); // a prefix is not an architecture
CHECK(!strata_arch_listed("gfx1100", "gfx11000"));
CHECK(!strata_arch_listed("", "gfx1100"));
CHECK(!strata_arch_listed("gfx1100", ""));
std::printf(bad ? "gpu_arch_test: %d FAILED\n" : "gpu_arch_test: ok\n", bad);
return bad ? 1 : 0;
}
tools/vision/strata_vision.cpp
+64
View File
@@ -13,9 +13,11 @@
// QUIT
// The output file is int32 {0x31455653 'SVE1', n_tokens, nx, ny, n_embd} then float32 [n_tokens][n_embd],
// row i at grid position (x = i % nx, y = i / nx). The text model is opened vocab-only (no weights).
#include "ggml-backend.h"
#include "ggml-cpu.h"
#include "gguf.h"
#include "gguf_meta.h"
#include "gpu_arch.h"
#include "llama.h"
#include "mtmd.h"
#include "mtmd-helper.h"
@@ -24,6 +26,10 @@
#if !defined(_WIN32)
#include <fcntl.h>
#include <unistd.h>
#elif defined(STRATA_VISION_HIP)
#define WIN32_LEAN_AND_MEAN
#define NOMINMAX
#include <windows.h>
#endif
#include <chrono>
#include <cstdio>
@@ -37,7 +43,14 @@
namespace {
// the GPU architectures ggml-hip printed while it started (" Device 0: ..., gfx1201:sramecc-:xnack- (0x1201), ...")
std::vector<std::string> g_gfx;
void quiet_log(ggml_log_level level, const char* text, void*) {
if (level == GGML_LOG_LEVEL_INFO && std::strstr(text, "Device ") != nullptr) {
const std::string gfx = strata_gfx_in(text);
if (!gfx.empty()) g_gfx.push_back(gfx);
}
if (level >= GGML_LOG_LEVEL_WARN) std::fputs(text, stderr);
}
@@ -56,6 +69,22 @@ bool parse_enc(const std::string& line, std::string& img, std::string& out) {
} // namespace
int main(int argc, char** argv) {
#if defined(_WIN32) && defined(STRATA_VISION_HIP)
// hipblas.dll is delay-loaded (tools/vision/CMakeLists.txt): it and the ROCm DLLs behind it (rocblas.dll finds its kernel
// library and ../.kpack beside itself) stay in rocm\bin of the engine package, which is not on the DLL search path of a
// program started with nothing but System32 in PATH. Put that folder on it before the first call.
{
wchar_t exe[MAX_PATH];
const DWORD n = GetModuleFileNameW(nullptr, exe, MAX_PATH);
std::wstring dir(exe, n);
const size_t cut = dir.find_last_of(L"\\/");
if (cut != std::wstring::npos) {
dir.resize(cut);
const std::wstring rocm = dir + L"\\rocm\\bin";
if (GetFileAttributesW(rocm.c_str()) != INVALID_FILE_ATTRIBUTES) SetDllDirectoryW(rocm.c_str());
}
}
#endif
std::string mmproj, model;
bool gpu = false;
int threads = 0, max_tokens = 0, min_tokens = 0;
@@ -70,6 +99,14 @@ int main(int argc, char** argv) {
if (a == "--mmproj") mmproj = next();
else if (a == "--model") model = next();
else if (a == "--gpu") gpu = true;
else if (a == "--version") {
#ifdef STRATA_VISION_HIP
std::printf("strata-vision hip (GPU architectures: %s)\n", STRATA_VISION_HIP_ARCHS);
#else
std::printf("strata-vision\n");
#endif
return 0;
}
else if (a == "--threads") threads = std::atoi(next().c_str());
else if (a == "--max-tokens") max_tokens = std::atoi(next().c_str());
else if (a == "--min-tokens") min_tokens = std::atoi(next().c_str()); // mtmd image_min_tokens (#767)
@@ -91,13 +128,40 @@ int main(int argc, char** argv) {
if (!gpu) {
#ifdef _WIN32
_putenv_s("CUDA_VISIBLE_DEVICES", "-1");
_putenv_s("HIP_VISIBLE_DEVICES", "-1");
#else
setenv("CUDA_VISIBLE_DEVICES", "-1", 1);
setenv("HIP_VISIBLE_DEVICES", "-1", 1);
#endif
}
llama_log_set(quiet_log, nullptr);
mtmd_helper_log_set(quiet_log, nullptr);
llama_backend_init();
#ifdef STRATA_VISION_HIP
if (gpu) {
// The AMD GPU reader: starting the backends prints the GPU line. With --gpu and no GPU, mtmd would quietly use the CPU
// (a second, slower CPU reader inside the HIP program), and a card the build has no code for would abort at its first
// kernel. Both end here with a line the server reads: it then starts the CPU reader (serve/server.py, Vision._start).
bool have_gpu = false;
for (size_t i = 0; i < ggml_backend_dev_count(); ++i) {
const auto type = ggml_backend_dev_type(ggml_backend_dev_get(i));
have_gpu = have_gpu || type == GGML_BACKEND_DEVICE_TYPE_GPU || type == GGML_BACKEND_DEVICE_TYPE_IGPU;
}
if (!have_gpu) {
std::printf("ERR this picture reader found no GPU (it was built for: %s)\n", STRATA_VISION_HIP_ARCHS);
std::fflush(stdout);
return 3;
}
const char* check = std::getenv("STRATA_VISION_ARCH_CHECK");
if (!g_gfx.empty() && (check == nullptr || std::atoi(check) != 0) &&
!strata_arch_listed(STRATA_VISION_HIP_ARCHS, g_gfx.front())) {
std::printf("ERR this GPU (%s) is not in the picture reader's build (%s)\n", g_gfx.front().c_str(),
STRATA_VISION_HIP_ARCHS);
std::fflush(stdout);
return 3;
}
}
#endif
const auto load_t0 = std::chrono::steady_clock::now();
#if !defined(_WIN32) && !defined(__APPLE__) // macOS has no posix_fadvise