Felipe97/llama-cpp-compiled
01.1k
1# llama.cpp for SYCL2 3- [Background](#background)4- [Recommended Release](#recommended-release)5- [News](#news)6- [OS](#os)7- [Hardware](#hardware)8- [Performance Reference](#performance-reference)9- [Docker](#docker)10- [Linux](#linux)11- [Windows](#windows-1)12- [Environment Variable](#environment-variable)13- [Design Rule](#design-rule)14- [Known Issue](#known-issues)15- [Q&A](#qa)16- [TODO](#todo)17 18## Background19 20**SYCL** is a high-level parallel programming model designed to improve developers productivity writing code across various hardware accelerators such as CPUs, GPUs, and FPGAs. It is a single-source language designed for heterogeneous computing and based on standard C++17.21 22**oneAPI** is an open ecosystem and a standard-based specification, supporting multiple architectures including but not limited to Intel CPUs, GPUs and FPGAs. The key components of the oneAPI ecosystem include:23 24- **DPCPP** *(Data Parallel C++)*: The primary oneAPI SYCL implementation, which includes the icpx/icx Compilers.25- **oneAPI Libraries**: A set of highly optimized libraries targeting multiple domains *(e.g. Intel oneMKL, oneMath and oneDNN)*.26- **oneAPI LevelZero**: A high performance low level interface for fine-grained control over Intel iGPUs and dGPUs.27 28### Llama.cpp + SYCL29 30The llama.cpp SYCL backend is primarily designed for **Intel GPUs**.31SYCL cross-platform capabilities enable support for other vendor GPUs as well.32 33## Recommended Release34 35### Windows36 37The following releases are verified and recommended:38 39|Commit ID|Tag|Release|Verified Platform| Update date|40|-|-|-|-|-|41|24e86cae7219b0f3ede1d5abdf5bf3ad515cccb8|b5377 |[llama-b5377-bin-win-sycl-x64.zip](https://github.com/ggml-org/llama.cpp/releases/download/b5377/llama-b5377-bin-win-sycl-x64.zip) |Arc B580/Linux/oneAPI 2025.1<br>LNL Arc GPU/Windows 11/oneAPI 2025.1.1|2025-05-15|42|3bcd40b3c593d14261fb2abfabad3c0fb5b9e318|b4040 |[llama-b4040-bin-win-sycl-x64.zip](https://github.com/ggml-org/llama.cpp/releases/download/b4040/llama-b4040-bin-win-sycl-x64.zip) |Arc A770/Linux/oneAPI 2024.1<br>MTL Arc GPU/Windows 11/oneAPI 2024.1| 2024-11-19|43|fb76ec31a9914b7761c1727303ab30380fd4f05c|b3038 |[llama-b3038-bin-win-sycl-x64.zip](https://github.com/ggml-org/llama.cpp/releases/download/b3038/llama-b3038-bin-win-sycl-x64.zip) |Arc A770/Linux/oneAPI 2024.1<br>MTL Arc GPU/Windows 11/oneAPI 2024.1||44 45### Ubuntu 24.0446 47The release packages for Ubuntu 24.04 x64 (FP32/FP16) only include the binary files of the llama.cpp SYCL backend. They require the target machine to have pre-installed Intel GPU drivers and oneAPI packages that are the same version as the build package. To get the version and installation info, refer to [.github/workflows/release.yml#L713](../../.github/workflows/release.yml#L713): ubuntu-24-sycl -> Download & Install oneAPI.48 49It is recommended to use them with [Intel Docker](https://hub.docker.com/r/intel/deep-learning-essentials).50 51The packages for FP32 and FP16 would have different accuracy and performance on LLMs. Please choose it according to the test result.52 53## News54 55- 2026.04-0556 - Optimize mul_mat by reorder feature for data type: Q4_K, Q5_K, Q6_K, Q8_0.57 - Fused MoE.58 - Upgrate CI and built package for oneAPI 2025.3.3, support Ubuntu 24.04 built package.59 60- 2026.0361 - Support Flash-Attention: less memory usage, performance impact depends on LLM.62 63- 2026.0264 - Remove support for Nvidia & AMD GPU, because the oneAPI plugin for Nvidia & AMD GPU is unavailable: download/installation channels are out of work. User can't build up the software for Nvidia & AMD GPU.65 66- 2025.1167 - Support malloc memory on device more than 4GB.68 69- 2025.270 - Optimize MUL_MAT Q4_0 on Intel GPU for all dGPUs and built-in GPUs since MTL. Increase the performance of LLM (llama-2-7b.Q4_0.gguf) 21%-87% on Intel GPUs (MTL, ARL-H, Arc, Flex, PVC).71 |GPU|Base tokens/s|Increased tokens/s|Percent|72 |-|-|-|-|73 |PVC 1550|39|73|+87%|74 |Flex 170|39|50|+28%|75 |Arc A770|42|55|+30%|76 |MTL|13|16|+23%|77 |ARL-H|14|17|+21%|78 79- 2024.1180 - Use syclcompat to improve the performance on some platforms. This requires to use oneAPI 2025.0 or newer.81 82- 2024.883 - Use oneDNN as the default GEMM library, improve the compatibility for new Intel GPUs.84 85- 2024.586 - Performance is increased: 34 -> 37 tokens/s of llama-2-7b.Q4_0 on Arc A770.87 - Arch Linux is verified successfully.88 89- 2024.490 - Support data types: GGML_TYPE_IQ4_NL, GGML_TYPE_IQ4_XS, GGML_TYPE_IQ3_XXS, GGML_TYPE_IQ3_S, GGML_TYPE_IQ2_XXS, GGML_TYPE_IQ2_XS, GGML_TYPE_IQ2_S, GGML_TYPE_IQ1_S, GGML_TYPE_IQ1_M.91 92- 2024.393 - Release binary files of Windows.94 - A blog is published: **Run LLM on all Intel GPUs Using llama.cpp**: [intel.com](https://www.intel.com/content/www/us/en/developer/articles/technical/run-llm-on-all-gpus-using-llama-cpp-artical.html) or [medium.com](https://medium.com/@jianyu_neo/run-llm-on-all-intel-gpus-using-llama-cpp-fd2e2dcbd9bd).95 - New base line is ready: [tag b2437](https://github.com/ggml-org/llama.cpp/tree/b2437).96 - Support multiple cards: **--split-mode**: [none|layer]; not support [row], it's on developing.97 - Support to assign main GPU by **--main-gpu**, replace $GGML_SYCL_DEVICE.98 - Support detecting all GPUs with level-zero and same top **Max compute units**.99 - Support OPs100 - hardsigmoid101 - hardswish102 - pool2d103 104- 2024.1105 - Create SYCL backend for Intel GPU.106 - Support Windows build107 108## OS109 110| OS | Status | Verified |111|---------|---------|------------------------------------------------|112| Linux | Support | Ubuntu 22.04, Fedora Silverblue 39, Arch Linux |113| Windows | Support | Windows 11 |114 115 116## Hardware117 118### Intel GPU119 120SYCL backend supports Intel GPU Family:121 122- Intel Data Center Max Series123- Intel Flex Series, Arc Series124- Intel Built-in Arc GPU125- Intel iGPU in Core CPU (11th Generation Core CPU and newer, refer to [oneAPI supported GPU](https://www.intel.com/content/www/us/en/developer/articles/system-requirements/intel-oneapi-base-toolkit-system-requirements.html#inpage-nav-1-1)).126 127On older Intel GPUs, you may try [OpenCL](/docs/backend/OPENCL.md) although the performance is not optimal, and some GPUs may not support OpenCL nor have any GPGPU capabilities.128 129#### Verified devices130 131| Intel GPU | Status | Verified Model |132|-------------------------------|---------|---------------------------------------|133| Intel Data Center Max Series | Support | Max 1550, 1100 |134| Intel Data Center Flex Series | Support | Flex 170 |135| Intel Arc A-Series | Support | Arc A770, Arc A730M, Arc A750 |136| Intel Arc B-Series | Support | Arc B580 |137| Intel built-in Arc GPU | Support | built-in Arc GPU in Meteor Lake, Arrow Lake, Lunar Lake |138| Intel iGPU | Support | iGPU in 13700k, 13400, i5-1250P, i7-1260P, i7-1165G7 |139 140*Notes:*141 142- **Memory**143 - The device memory is a limitation when running a large model. The loaded model size, *`llm_load_tensors: buffer_size`*, is displayed in the log when running `./bin/llama-completion`.144 - Please make sure the GPU shared memory from the host is large enough to account for the model's size. For e.g. the *llama-2-7b.Q4_0* requires at least 8.0GB for integrated GPU and 4.0GB for discrete GPU.145 146- **Execution Unit (EU)**147 - If the iGPU has less than 80 EUs, the inference speed will likely be too slow for practical use.148 149### Other Vendor GPU150 151NA152 153## Performance Reference154 155 156To get the supported LLMs, GPUs, and performance reference, please check [Performance of llama.cpp on Intel GPU with SYCL backend](https://github.com/ggml-org/llama.cpp/discussions/23313).157 158You could update your test result in it directly.159 160## Docker161 162Please refer to [Docker with SYCL](../docker.md#docker-with-sycl) for details.163 164## Quick Development WOW165 166This chapter is for quick development & try with SYCL backend on Intel GPU.167 168You need to install following sofeware before development:169 - Intel GPU driver170 - oneAPI package171 - other development tools.172 173Please refer to [Linux](#linux) or [Windows](#windows-1) for above installation and resolve the trouble in usage. There are the detailed guide.174 175- Linux176 177```178## build from source code179./examples/sycl/build.sh180 181## run CONV_2D_DW unit test cases182./build/bin/test-backend-ops -b SYCL0 -o CONV_2D_DW183 184## run all unit test cases185./build/bin/test-backend-ops -b SYCL0186 187## run with LLM on the first GPU188./examples/sycl/test.sh -mg 0 -m xxxx.gguf189 190## run service with LLM on the first GPU191export ONEAPI_DEVICE_SELECTOR="level_zero:0"192./examples/sycl/start-svr.sh -m xxxx.gguf193 194## update the docs/ops.md for new/update OPs195./examples/sycl/update-ops-doc.sh196```197 198- Windows199 200```201## build from source code202examples\sycl\win-build-sycl.bat203 204## run CONV_2D_DW unit test cases205build\bin\test-backend-ops.exe -b SYCL0 -o CONV_2D_DW206 207## run all unit test cases208build\bin\test-backend-ops.exe -b SYCL0209 210## run LLM on the first GPU211examples\sycl\win-test.bat -mg 0 -m xxxx.gguf212 213## run service with LLM on the first GPU214set ONEAPI_DEVICE_SELECTOR="level_zero:0"215examples\sycl\win-start-svr.bat -m xxxx.gguf216 217## update the docs/ops.md for new/update OPs218examples\sycl\win-update-ops-doc.bat219```220 221 222## Linux223 224### I. Setup Environment225 2261. **Install GPU drivers**227 228 - **Intel GPU**229 230Intel data center GPUs drivers installation guide and download page can be found here: [Get Intel dGPU Drivers](https://dgpu-docs.intel.com/driver/installation.html#ubuntu-install-steps).231 232*Note*: for client GPUs *(iGPU & Arc A-Series)*, please refer to the [client iGPU driver installation](https://dgpu-docs.intel.com/driver/client/overview.html).233 234Once installed, add the user(s) to the `video` and `render` groups.235 236```sh237sudo usermod -aG render $USER238sudo usermod -aG video $USER239```240 241*Note*: logout/re-login for the changes to take effect.242 243Verify installation through `clinfo`:244 245```sh246sudo apt install clinfo247sudo clinfo -l248```249 250Sample output:251 252```sh253Platform #0: Intel(R) OpenCL Graphics254 `-- Device #0: Intel(R) Arc(TM) A770 Graphics255 256Platform #0: Intel(R) OpenCL HD Graphics257 `-- Device #0: Intel(R) Iris(R) Xe Graphics [0x9a49]258```259 2602. **Install Intel® oneAPI Base toolkit**261 262SYCL backend depends on:263 - Intel® oneAPI DPC++/C++ compiler/running-time.264 - Intel® oneAPI DPC++/C++ library (oneDPL).265 - Intel® oneAPI Deep Neural Network Library (oneDNN).266 - Intel® oneAPI Math Kernel Library (oneMKL).267 268- **For Intel GPU**269 270All above are included in both **Intel® oneAPI Base toolkit** and **Intel® Deep Learning Essentials** packages.271 272It's recommended to install **Intel® Deep Learning Essentials** which only provides the necessary libraries with less size.273 274The **Intel® oneAPI Base toolkit** and **Intel® Deep Learning Essentials** can be obtained from the official [Intel® oneAPI Base Toolkit](https://www.intel.com/content/www/us/en/developer/tools/oneapi/base-toolkit.html) page.275 276Please follow the instructions for downloading and installing the Toolkit for Linux, and preferably keep the default installation values unchanged, notably the installation path *(`/opt/intel/oneapi` by default)*.277 278Following guidelines/code snippets assume the default installation values. Otherwise, please make sure the necessary changes are reflected where applicable.279 280Upon a successful installation, SYCL is enabled for the available Intel devices, along with relevant libraries such as oneAPI oneDNN for Intel GPUs.281 282|Verified release|283|-|284|2025.3.3 |285|2025.2.1|286|2025.1|287|2024.1|288 2893. **Verify installation and environment**290 291In order to check the available SYCL devices on the machine, please use the `sycl-ls` command.292```sh293source /opt/intel/oneapi/setvars.sh294sycl-ls295```296 297- **Intel GPU**298 299When targeting an intel GPU, the user should expect one or more devices among the available SYCL devices. Please make sure that at least one GPU is present via `sycl-ls`, for instance `[level_zero:gpu]` in the sample output below:300 301```302[level_zero:gpu][level_zero:0] Intel(R) oneAPI Unified Runtime over Level-Zero, Intel(R) Arc(TM) A770 Graphics 12.55.8 [1.3.29735+27]303[level_zero:gpu][level_zero:1] Intel(R) oneAPI Unified Runtime over Level-Zero, Intel(R) UHD Graphics 730 12.2.0 [1.3.29735+27]304[opencl:cpu][opencl:0] Intel(R) OpenCL, 13th Gen Intel(R) Core(TM) i5-13400 OpenCL 3.0 (Build 0) [2025.20.8.0.06_160000]305[opencl:gpu][opencl:1] Intel(R) OpenCL Graphics, Intel(R) Arc(TM) A770 Graphics OpenCL 3.0 NEO [24.39.31294]306[opencl:gpu][opencl:2] Intel(R) OpenCL Graphics, Intel(R) UHD Graphics 730 OpenCL 3.0 NEO [24.39.31294]307```308 309### II. Build llama.cpp310 311#### Intel GPU312 313```sh314# Uses FP32, consider using FP16 for better performance in most cases315./examples/sycl/build.sh316```317 318or319 320```sh321# Export relevant ENV variables322source /opt/intel/oneapi/setvars.sh323 324# Option 1: Use FP16 (recommended for better performance in most cases)325cmake -B build -DGGML_SYCL=ON -DCMAKE_C_COMPILER=icx -DCMAKE_CXX_COMPILER=icpx -DGGML_SYCL_F16=ON326 327# Option 2: Use FP32328cmake -B build -DGGML_SYCL=ON -DCMAKE_C_COMPILER=icx -DCMAKE_CXX_COMPILER=icpx329 330# build all binary331cmake --build build --config Release -j -v332```333 334It is possible to come across some precision issues when running tests that stem from using faster335instructions, which can be circumvented by setting the environment variable `SYCL_PROGRAM_COMPILE_OPTIONS`336as `-cl-fp32-correctly-rounded-divide-sqrt`337 338### III. Run the inference339 340#### Retrieve and prepare model341 342You can refer to the general [*Obtaining and quantizing models*](../../README.md#obtaining-and-quantizing-models) guide for model preparation, or download an already quantized model like [llama-2-7b.Q4_0.gguf](https://huggingface.co/TheBloke/Llama-2-7B-GGUF/resolve/main/llama-2-7b.Q4_0.gguf?download=true) or [Meta-Llama-3-8B-Instruct-Q4_0.gguf](https://huggingface.co/aptha/Meta-Llama-3-8B-Instruct-Q4_0-GGUF/resolve/main/Meta-Llama-3-8B-Instruct-Q4_0.gguf).343 344##### Check device345 3461. Enable oneAPI running environment347 348```sh349source /opt/intel/oneapi/setvars.sh350```351 3522. List devices information353 354Similar to the native `sycl-ls`, available SYCL devices can be queried as follow:355 356```sh357./build/bin/llama-ls-sycl-device358```359 360This command will only display the selected backend that is supported by SYCL. The default backend is level_zero. For example, in a system with 2 *Intel GPU* it would look like the following:361```362found 2 SYCL devices:363 364| | | |Compute |Max compute|Max work|Max sub| |365|ID| Device Type| Name|capability|units |group |group |Global mem size|366|--|------------------|---------------------------------------------|----------|-----------|--------|-------|---------------|367| 0|[level_zero:gpu:0]| Intel(R) Arc(TM) A770 Graphics| 1.3| 512| 1024| 32| 16225243136|368| 1|[level_zero:gpu:1]| Intel(R) UHD Graphics 770| 1.3| 32| 512| 32| 53651849216|369```370 371#### Choose level-zero devices372 373|Chosen Device ID|Setting|374|-|-|375|0|`export ONEAPI_DEVICE_SELECTOR="level_zero:0"` or no action|376|1|`export ONEAPI_DEVICE_SELECTOR="level_zero:1"`|377|0 & 1|`export ONEAPI_DEVICE_SELECTOR="level_zero:0;level_zero:1"`|378 379#### Execute380 381Choose one of following methods to run.382 3831. Script384 385- Use device 0:386 387```sh388./examples/sycl/test.sh -mg 0389```390- Use multiple devices:391 392```sh393./examples/sycl/test.sh394```395 396- Run llama-server:397 398```sh399./examples/sycl/start-svr.sh -m PATH/MODEL_FILE400```401 4022. Command line403Launch inference404 405There are two device selection modes:406 407- Single device: Use one device assigned by user. Default device id is 0.408- Multiple devices: Automatically choose the devices with the same backend.409 410In two device selection modes, the default SYCL backend is level_zero, you can choose other backend supported by SYCL by setting environment variable ONEAPI_DEVICE_SELECTOR.411 412| Device selection | Parameter |413|------------------|----------------------------------------|414| Single device | --split-mode none --main-gpu DEVICE_ID |415| Multiple devices | --split-mode layer (default) |416| Multiple devices | --split-mode tensor (tensor parallelism) |417 418`--split-mode tensor` (tensor parallelism) shards each layer across the selected419GPUs. It requires flash attention, which is auto-enabled when `--flash-attn` is420left at its default `auto`, so `--split-mode tensor` works out of the box.421Passing `--flash-attn off` together with `--split-mode tensor` is rejected at422context creation. The default `f16` KV cache is recommended. Tensor parallelism423is currently optimized for 2 GPUs; other device counts fall back to a generic424all-reduce.425 426Examples:427 428- Use device 0:429 430```sh431ZES_ENABLE_SYSMAN=1 ./build/bin/llama-completion -no-cnv -m models/llama-2-7b.Q4_0.gguf -p "Building a website can be done in 10 simple steps:" -n 400 -e -ngl 99 -sm none -mg 0 --load-mode auto432```433 434- Use multiple devices:435 436```sh437ZES_ENABLE_SYSMAN=1 ./build/bin/llama-completion -no-cnv -m models/llama-2-7b.Q4_0.gguf -p "Building a website can be done in 10 simple steps:" -n 400 -e -ngl 99 -sm layer --load-mode auto438```439 440*Notes:*441 442- Upon execution, verify the selected device(s) ID(s) in the output log, which can for instance be displayed as follow:443 444```sh445detect 1 SYCL GPUs: [0] with top Max compute units:512446```447Or448```sh449use 1 SYCL GPUs: [0] with Max compute units:512450```451 452User can use the device management in [docs/multi-gpu.md](https://github.com/ggml-org/llama.cpp/blob/master/docs/multi-gpu.md), like parameter `--device SYCL0,SYCL1` to assign one or more devices.453 454## Windows455 456### Install GPU driver457 458Intel GPU drivers instructions guide and download page can be found here: [Get Intel GPU Drivers](https://www.intel.com/content/www/us/en/products/docs/discrete-gpus/arc/software/drivers.html).459 460### Option 1: download the binary package directly461 462Download the binary package for Windows from: https://github.com/ggml-org/llama.cpp/releases.463 464Extract the package to local folder, run the llama tools directly. Refer to [Run the inference](#iii-run-the-inference-1).465 466Note, the package includes the SYCL running time and all depended dll files, no need to install oneAPI package and activte them.467 468### Option 2: build locally from the source code.469 470#### I. Setup environment471 4721. Install Visual Studio473 474If you already have a recent version of Microsoft Visual Studio, you can skip this step. Otherwise, please refer to the official download page for [Microsoft Visual Studio](https://visualstudio.microsoft.com/).475 4762. Install Intel® oneAPI Base toolkit477 478SYCL backend depends on:479 - Intel® oneAPI DPC++/C++ compiler/running-time.480 - Intel® oneAPI DPC++/C++ library (oneDPL).481 - Intel® oneAPI Deep Neural Network Library (oneDNN).482 - Intel® oneAPI Math Kernel Library (oneMKL).483 484All above are included in both **Intel® oneAPI Base toolkit** and **Intel® Deep Learning Essentials** packages.485 486It's recommended to install **Intel® Deep Learning Essentials** which only provides the necessary libraries with less size.487 488The **Intel® oneAPI Base toolkit** and **Intel® Deep Learning Essentials** can be obtained from the official [Intel® oneAPI Base Toolkit](https://www.intel.com/content/www/us/en/developer/tools/oneapi/base-toolkit.html) page.489 490Please follow the instructions for downloading and installing the Toolkit for Windows, and preferably keep the default installation values unchanged, notably the installation path *(`C:\Program Files (x86)\Intel\oneAPI` by default)*.491 492Following guidelines/code snippets assume the default installation values. Otherwise, please make sure the necessary changes are reflected where applicable.493 494b. Enable oneAPI running environment:495 496- Type "oneAPI" in the search bar, then open the `Intel oneAPI command prompt for Intel 64 for Visual Studio 2022` App.497 498- On the command prompt, enable the runtime environment with the following:499```500"C:\Program Files (x86)\Intel\oneAPI\setvars.bat" intel64501```502 503- if you are using Powershell, enable the runtime environment with the following:504 505```506cmd.exe "/K" '"C:\Program Files (x86)\Intel\oneAPI\setvars.bat" && powershell'507```508 509c. Verify installation510 511In the oneAPI command line, run the following to print the available SYCL devices:512 513```514sycl-ls.exe515```516 517There should be one or more *level-zero* GPU devices displayed as **[ext_oneapi_level_zero:gpu]**. Below is example of such output detecting an *Intel Iris Xe* GPU as a Level-zero SYCL device:518 519Output (example):520```521[opencl:acc:0] Intel(R) FPGA Emulation Platform for OpenCL(TM), Intel(R) FPGA Emulation Device OpenCL 1.2 [2023.16.10.0.17_160000]522[opencl:cpu:1] Intel(R) OpenCL, 11th Gen Intel(R) Core(TM) i7-1185G7 @ 3.00GHz OpenCL 3.0 (Build 0) [2023.16.10.0.17_160000]523[opencl:gpu:2] Intel(R) OpenCL Graphics, Intel(R) Iris(R) Xe Graphics OpenCL 3.0 NEO [31.0.101.5186]524[ext_oneapi_level_zero:gpu:0] Intel(R) Level-Zero, Intel(R) Iris(R) Xe Graphics 1.3 [1.3.28044]525```526 5273. Install build tools528 529a. Download & install cmake for Windows: https://cmake.org/download/ (CMake can also be installed from Visual Studio Installer)530b. The new Visual Studio will install Ninja as default. (If not, please install it manually: https://ninja-build.org/)531 532 533#### II. Build llama.cpp534 535You could download the release package for Windows directly, which including binary files and depended oneAPI dll files.536 537Choose one of following methods to build from source code.538 539##### Option 1: Script540 541```sh542# Uses FP32, consider using FP16 for better performance in most cases543.\examples\sycl\win-build-sycl.bat544```545 546##### Option 2: CMake547 548On the oneAPI command line window, step into the llama.cpp main directory and run the following:549 550```551@call "C:\Program Files (x86)\Intel\oneAPI\setvars.bat" intel64 --force552 553# Option 1: Use FP16 (recommended for better performance in most cases)554cmake -B build -G "Ninja" -DGGML_SYCL=ON -DCMAKE_C_COMPILER=cl -DCMAKE_CXX_COMPILER=icx -DCMAKE_BUILD_TYPE=Release -DGGML_SYCL_F16=ON555 556# Option 2: Or FP32557cmake -B build -G "Ninja" -DGGML_SYCL=ON -DCMAKE_C_COMPILER=cl -DCMAKE_CXX_COMPILER=icx -DCMAKE_BUILD_TYPE=Release558 559cmake --build build --config Release -j560```561 562Or, use CMake presets to build:563 564```sh565cmake -DGGML_SYCL_F16=ON --preset x64-windows-sycl-release566cmake --build build-x64-windows-sycl-release -j --target llama-completion567 568cmake --preset x64-windows-sycl-release569cmake --build build-x64-windows-sycl-release -j --target llama-completion570 571cmake --preset x64-windows-sycl-debug572cmake --build build-x64-windows-sycl-debug -j --target llama-completion573```574 575##### Option 3: Visual Studio576 577You have two options to use Visual Studio to build llama.cpp:578- As CMake Project using CMake presets.579- Creating a Visual Studio solution to handle the project.580 581**Note**:582 583All following commands are executed in PowerShell.584 585###### - Open as a CMake Project586 587You can use Visual Studio to open the `llama.cpp` folder directly as a CMake project. Before compiling, select one of the SYCL CMake presets:588 589- `x64-windows-sycl-release`590 591- `x64-windows-sycl-debug`592 593*Notes:*594- For a minimal experimental setup, you can build only the inference executable using:595 596 ```Powershell597 cmake --build build --config Release -j --target llama-completion598 ```599 600###### - Generating a Visual Studio Solution601 602You can use Visual Studio solution to build and work on llama.cpp on Windows. You need to convert the CMake Project into a `.sln` file.603 604If you want to use the Intel C++ Compiler for the entire `llama.cpp` project, run the following command:605 606```Powershell607cmake -B build -G "Visual Studio 17 2022" -T "Intel C++ Compiler 2025" -A x64 -DGGML_SYCL=ON -DCMAKE_BUILD_TYPE=Release608```609 610If you prefer to use the Intel C++ Compiler only for `ggml-sycl`, ensure that `ggml` and its backend libraries are built as shared libraries ( i.e. `-DBUILD_SHARED_LIBRARIES=ON`, this is default behaviour):611 612```Powershell613cmake -B build -G "Visual Studio 17 2022" -A x64 -DGGML_SYCL=ON -DCMAKE_BUILD_TYPE=Release \614 -DSYCL_INCLUDE_DIR="C:\Program Files (x86)\Intel\oneAPI\compiler\latest\include" \615 -DSYCL_LIBRARY_DIR="C:\Program Files (x86)\Intel\oneAPI\compiler\latest\lib"616```617 618If successful the build files have been written to: *path/to/llama.cpp/build*619Open the project file **build/llama.cpp.sln** with Visual Studio.620 621Once the Visual Studio solution is created, follow these steps:622 6231. Open the solution in Visual Studio.624 6252. Right-click on `ggml-sycl` and select **Properties**.626 6273. In the left column, expand **C/C++** and select **DPC++**.628 6294. In the right panel, find **Enable SYCL Offload** and set it to `Yes`.630 6315. Apply the changes and save.632 633 634*Navigation Path:*635 636```637Properties -> C/C++ -> DPC++ -> Enable SYCL Offload (Yes)638```639 640Now, you can build `llama.cpp` with the SYCL backend as a Visual Studio project.641To do it from menu: `Build -> Build Solution`.642Once it is completed, final results will be in **build/Release/bin**643 644*Additional Note*645 646- You can avoid specifying `SYCL_INCLUDE_DIR` and `SYCL_LIBRARY_DIR` in the CMake command by setting the environment variables:647 648 - `SYCL_INCLUDE_DIR_HINT`649 650 - `SYCL_LIBRARY_DIR_HINT`651 652- Above instruction has been tested with Visual Studio 17 Community edition and oneAPI 2025.0. We expect them to work also with future version if the instructions are adapted accordingly.653 654### III. Run the inference655 656#### Retrieve and prepare model657 658You can refer to the general [*Obtaining and quantizing models*](../../README.md#obtaining-and-quantizing-models) guide for model preparation, or download an already quantized model like [llama-2-7b.Q4_0.gguf](https://huggingface.co/TheBloke/Llama-2-7B-GGUF/blob/main/llama-2-7b.Q4_0.gguf) or [Meta-Llama-3-8B-Instruct-Q4_0.gguf](https://huggingface.co/aptha/Meta-Llama-3-8B-Instruct-Q4_0-GGUF/resolve/main/Meta-Llama-3-8B-Instruct-Q4_0.gguf).659 660##### Check device661 6621. Enable oneAPI running environment663 664On the oneAPI command line window, run the following and step into the llama.cpp directory:665```666"C:\Program Files (x86)\Intel\oneAPI\setvars.bat" intel64667```668 6692. List devices information670 671Similar to the native `sycl-ls`, available SYCL devices can be queried as follow:672 673```674build\bin\llama-ls-sycl-device.exe675```676 677This command will only display the selected backend that is supported by SYCL. The default backend is level_zero. For example, in a system with 2 *Intel GPU* it would look like the following:678```679found 2 SYCL devices:680| | | |Compute |Max compute|Max work|Max sub| |681|ID| Device Type| Name|capability|units |group |group |Global mem size|682|--|------------------|---------------------------------------------|----------|-----------|--------|-------|---------------|683| 0|[level_zero:gpu:0]| Intel(R) Arc(TM) A770 Graphics| 1.3| 512| 1024| 32| 16225243136|684| 1|[level_zero:gpu:1]| Intel(R) UHD Graphics 770| 1.3| 32| 512| 32| 53651849216|685 686```687 688##### Choose level-zero devices689 690|Chosen Device ID|Setting|691|-|-|692|0|Default option. You may also want to `set ONEAPI_DEVICE_SELECTOR="level_zero:0"`|693|1|`set ONEAPI_DEVICE_SELECTOR="level_zero:1"`|694|0 & 1|`set ONEAPI_DEVICE_SELECTOR="level_zero:0;level_zero:1"` or `set ONEAPI_DEVICE_SELECTOR="level_zero:*"`|695 696##### Execute697 698Choose one of following methods to run.699 7001. Script701 702- Run test:703 704```705examples\sycl\win-test.bat706```707 708- Run llama-server:709 710```711examples\sycl\win-start-svr.bat -m PATH\MODEL_FILE712```713 7142. Command line715 716Launch inference717 718There are two device selection modes:719 720- Single device: Use one device assigned by user. Default device id is 0.721- Multiple devices: Automatically choose the devices with the same backend.722 723In two device selection modes, the default SYCL backend is level_zero, you can choose other backend supported by SYCL by setting environment variable ONEAPI_DEVICE_SELECTOR.724 725| Device selection | Parameter |726|------------------|----------------------------------------|727| Single device | --split-mode none --main-gpu DEVICE_ID |728| Multiple devices | --split-mode layer (default) |729| Multiple devices | --split-mode tensor (tensor parallelism) |730 731`--split-mode tensor` (tensor parallelism) shards each layer across the selected732GPUs. It requires flash attention, which is auto-enabled when `--flash-attn` is733left at its default `auto`, so `--split-mode tensor` works out of the box.734Passing `--flash-attn off` together with `--split-mode tensor` is rejected at735context creation. The default `f16` KV cache is recommended. Tensor parallelism736is currently optimized for 2 GPUs; other device counts fall back to a generic737all-reduce.738 739Examples:740 741- Use device 0:742 743```744build\bin\llama-completion.exe -no-cnv -m models\llama-2-7b.Q4_0.gguf -p "Building a website can be done in 10 simple steps:\nStep 1:" -n 400 -e -ngl 99 -sm none -mg 0 --load-mode auto745```746 747- Use multiple devices:748 749```750build\bin\llama-completion.exe -no-cnv -m models\llama-2-7b.Q4_0.gguf -p "Building a website can be done in 10 simple steps:\nStep 1:" -n 400 -e -ngl 99 -sm layer --load-mode auto751```752 753 754Note:755 756- Upon execution, verify the selected device(s) ID(s) in the output log, which can for instance be displayed as follow:757 758```sh759detect 1 SYCL GPUs: [0] with top Max compute units:512760```761 762Or763 764```sh765use 1 SYCL GPUs: [0] with Max compute units:512766```767 768User can use the device management in [docs/multi-gpu.md](https://github.com/ggml-org/llama.cpp/blob/master/docs/multi-gpu.md), like parameter `--device SYCL0,SYCL1` to assign one or more devices.769 770## Environment Variable771 772### Build773 774| Name | Value | Function |775|--------------------|---------------------------------------|---------------------------------------------|776| GGML_SYCL | ON (mandatory) | Enable build with SYCL code path. |777| GGML_SYCL_TARGET | INTEL *(default)* | Set the SYCL target device type. |778| GGML_SYCL_DEVICE_ARCH | Optional | Set the SYCL device architecture. Setting the device architecture can improve the performance. See the table [--offload-arch](https://github.com/intel/llvm/blob/sycl/sycl/doc/design/OffloadDesign.md#--offload-arch) for a list of valid architectures. |779| GGML_SYCL_F16 | OFF *(default)* \|ON *(optional)* | Enable FP16 build with SYCL code path. (1.) |780| GGML_SYCL_GRAPH | ON *(default)* \|OFF *(Optional)* | Enable build with [SYCL Graph extension](https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/experimental/sycl_ext_oneapi_graph.asciidoc). |781| GGML_SYCL_DNN | ON *(default)* \|OFF *(Optional)* | Enable build with oneDNN. |782| GGML_SYCL_HOST_MEM_FALLBACK | ON *(default)* \|OFF *(Optional)* | Allow host memory fallback when device memory is full during quantized weight reorder. Enables inference to continue at reduced speed (reading over PCIe) instead of failing. Requires Linux kernel 6.8+. |783| GGML_SYCL_SUPPORT_LEVEL_ZERO_API | ON *(default)* \|OFF *(Optional)* | Support to use Level Zero API for device memory allocation. Requires Level Zero headers/library at build time and Intel GPU driver (Level Zero runtime) at run time. Reduces system RAM usage during multi-GPU inference. SYCL backend always runs on Level Zero running time even if it's set as OFF (The SYCL api will be usage for memory allocation).|784| CMAKE_C_COMPILER | `icx` *(Linux)*, `icx/cl` *(Windows)* | Set `icx` compiler for SYCL code path. |785| CMAKE_CXX_COMPILER | `icpx` *(Linux)*, `icx` *(Windows)* | Set `icpx/icx` compiler for SYCL code path. |786 7871. FP32 or FP16 have different performance impact to LLM. Recommended to test them for better prompt processing performance on your models. You need to rebuild the code after change `GGML_SYCL_F16=OFF/ON`.788 789### Runtime790 791| Name | Value | Function |792|-------------------|------------------|---------------------------------------------------------------------------------------------------------------------------|793| GGML_SYCL_DEBUG | 0 (default) or 1 | Enable log function: GGML_SYCL_DEBUG() for common debug. |794| GGML_SYCL_DEV_DEBUG | 0 (default) or 1 | Enable log function: GGML_SYCL_DEV_DEBUG() for developmental purposes by replacing GGML_SYCL_DEBUG() in special codes. Restore to GGML_SYCL_DEBUG() before committing code.|795| GGML_SYCL_DEV2DEV_MEMCPY | 0 (default), 1, 2 | Choose the method of dev2dev memory copy.<br>Value: <br>* 0: SYCL API (default), only support dGPUs.<br>* 1: L0 API -- Better performance, only support dGPUs, found to lead to abnormal crash in some case. <br>* 2: Host Forward -- Most stable method for all cases (including iGPU + dGPU*N), but with lower performance (-2% to -5%).<br>SYCL & L0 API are easy to be impacted by Intel GPU driver issue. When you meet the garbled output or crash issues in multiple GPUs case, try with this debug flag to work around or check the issue.|796| GGML_SYCL_ENABLE_FLASH_ATTN | 1 (default) or 0| Enable Flash-Attention. It can reduce memory usage. The performance impact depends on the LLM.|797| GGML_SYCL_ENABLE_OPT | 0 or 1 (default)| Enable optimize features for Intel GPUs. (Recommended to 0 for Intel devices older than Gen 10) |798| GGML_SYCL_ENABLE_GRAPH | 0 (default) or 1 | Enable running computations through SYCL Graphs feature. Disabled by default because SYCL Graph is still on development, no better performance. |799| GGML_SYCL_ENABLE_HOST_PINNED_MEM | 0 or 1 (default) | Enable host pinned memory to speed up copy data from host to device. When disable it, host memory will common malloc() on CPU. Disable it when use `--load-model mlock`.|800| GGML_SYCL_HOST_PINNED_MEM_2G | 0 (default) or 1 | Limit the max memory allocation to be no more than 2GB when enable host pinned memory. USM allocations above 2 GiB take the relaxed/large-allocation path, which serializes H2D copies with compute and prevents copy/compute overlap. It will impact the startup time. Need more test. Depend on `GGML_SYCL_ENABLE_HOST_PINNED_MEM=1`.|801| GGML_SYCL_GET_MEM_API | 0 (default) or 1 | Set to get memory info (free, total) by Level Zero or SYCL API:<br>0 - Level Zero API: support more GPUs, only run on Level Zero running time. When there is an error, fallback to call SYCL API. Depend on GGML_SYCL_SUPPORT_LEVEL_ZERO_API.<br>1 - SYCL API: legacy, support more running time, it can't get the free size of some GPUs (like Arc770). In such case, return the free size as value of total size.|802| GGML_SYCL_USE_LEVEL_ZERO_API | 1 (default) or 0 | Use Level Zero API for device memory allocation instead of SYCL. Reduces system RAM usage on Intel dGPUs by avoiding DMA-buf/TTM host memory staging. Requires GGML_SYCL_SUPPORT_LEVEL_ZERO_API=ON at build time. SYCL backend always runs on Level Zero running time even if it's set as OFF (The SYCL api will be usage for memory allocation).|803| GGML_SYCL_ENABLE_DNN | 0 or 1 (default)| Enable running computations through oneDNN and always use oneMKL. |804| GGML_SYCL_FA_ONEDNN | 1 (default) or 0 | Enable the oneDNN fused SDPA (flash-attention) path on supported GPUs. Set to 0 to always use the native SYCL flash-attention kernel. |805| GGML_SYCL_FA_ONEDNN_MAX_KV | 0 (default, disabled) or positive integer | By default (0), all sequences are handled by the oneDNN fused SDPA path, regardless of KV length; a positive value caps that length, past which sequences fall back to the native kernel. If GPU driver watchdog resets (DEVICE_LOST) occur during long-context inference, set this near the context depth where they start, e.g. 24576. |806| GGML_SYCL_ENABLE_VMM | 0 or 1 (default) | Enable the virtual-memory device pool. |807| GGML_SYCL_ENABLE_MKL_FA | 1 (default) or 0 | Enable oneMKL GEMM flash attention for XMX-accelerated prompt processing with quantized KV cache. Automatically activates during prefill (prompt processing) when all conditions are met: (1) flash-attn enabled (`-fa` or `--flash-attn on`), (2) KV cache quantized (`--cache-type-k q8_0 --cache-type-v q8_0` or other `*_0/*_1` types), (3) batch size ≥ 1024 (`--batch-size 1024`), (4) prompt length ≥ 1024 tokens. Set to 0 to force the TILE kernel for A/B testing. Example minimum command: `llama-cli -m model.gguf -fa -ngl 99 --cache-type-k q8_0 --cache-type-v q8_0 --batch-size 1024 -p "your prompt"` |808| GGML_SYCL_MKL_FA_DEBUG | 0 (default) or 1 | Enable per-call diagnostic logging for MKL flash attention: GEMM/softmax timings, interleaved-head detection, and buffer memory usage. |809| GGML_SYCL_MEMTRACE | 0 (default), 1, 2 | Enable record and output memory allocation diagnostics. Requires `-lv 4`. <br>0 - Disable<br>1 - Basic memory info, including current and peak allocations, as well allocations from other sources, around 50 lines per model load.<br>2 - More verbose, logging around 900 specific allocations and deallocations. |810| GGML_SYCL_MEMTRACE_STEP | 64 (default) or positive integer | With GGML_SYCL_MEMTRACE=1, the minimum growth in memory usage to trigger another log record. |811| GGML_SYCL_MKL_FA_DIAG | 0 (default) or 1 | Enable output fingerprinting for MKL flash attention. Dumps the first 64 float output values for the first 6 FA calls with n_kv ≥ 1024, labeled with kernel type (MKL/TILE/VEC) for cross-kernel comparison. |812| GGML_SYCL_ENABLE_FUSION | 0 or 1 (default) | Enable fused-kernel dispatch in graph compute. Unsupported types and layouts fall back to the standalone op kernels. See `ggml_sycl_can_fuse()`. |813| GGML_SYCL_ENABLE_ESIMD | 0 or 1 (default)| Enable ESIMD kernels when available. |814| ZES_ENABLE_SYSMAN | 0 (default) or 1 | Support to get free memory of GPU by sycl::aspect::ext_intel_free_memory.<br>Recommended to use when --split-mode = layer |815| UR_L0_ENABLE_RELAXED_ALLOCATION_LIMITS | 0 (default) or 1 | Allow SYCL/Unified Runtime Level Zero device allocations larger than 4 GiB. llama.cpp's direct Level Zero allocation path requests the relaxed maximum-size limit itself when GGML_SYCL_ENABLE_LEVEL_ZERO=1. |816| GGML_SYCL_USM_SYSTEM | 0 (default) or 1 | Enable experimental support for [USM system allocations](https://github.khronos.org/SYCL_Reference/iface/usm_basic_concept.html#system-allocations) for large GPU buffers. This requires enough host memory for model weights and caches, an Intel Xe2+ GPU such as BMG or newer and supported on Linux only, with CONFIG_DRM_XE_GPUSVM enabled. |817 818## Compile-time Flags819 820Pass these via `CXXFLAGS` or add a one-off `#define` to enable a flag on the spot.821 822| Name | Function |823|-----------------|----------------------------------------------------------------------------------|824| DEBUG_SYCL_POOL | Enable device memory pool logging on teardown. Useful for profiling allocations. |825| DEBUG_SYCL_MALLOC | Enable verbose per-call logging of device pool alloc/free operations. |826| GGML_SYCL_SUPPORT_VMM | Support to building with VMM code. Default is Yes. |827 828## Design Rule829 830- Open to all contributors.831 832- All code change should be useful to user:833 - Fix bug.834 - Add new function.835 - Improve the performance/usage.836 - Make code be easy to maintain.837 - ...838 839- Don't accept the codes of following cases:840 - Break legacy function.841 - Reduce the performance of legacy case in default.842 - Not completed work/the functionality cannot be demonstrated.843 844- Encourage to use environment variable to control features to be opened/closed.845 - User can evaluate the feature without rebuild the code.846 - Recommend the best features to user by setting them be opened as default.847 848- Design the code based on the published official releases of oneAPI packages: compiler, library, driver, OS kernel.849 850- Developers need to maintain the code they submit.851 852## Known Issues853 854- `Split-mode:[row]` is not supported.855 856- Missed the AOT (Ahead-of-Time) in building.857 - Good: Builds quickly, smaller size of binary file.858 - Bad: The startup is slow (JIT) in first time, but subsequent performance is unaffected.859 860## Q&A861 862- Error: `error while loading shared libraries: libsycl.so: cannot open shared object file: No such file or directory`.863 864 - Potential cause: Unavailable oneAPI installation or not set ENV variables.865 - Solution: Install *oneAPI base toolkit* and enable its ENV through: `source /opt/intel/oneapi/setvars.sh`.866 867- General compiler error:868 869 - Remove **build** folder or try a clean-build.870 871- I can **not** see `[ext_oneapi_level_zero:gpu]` after installing the GPU driver on Linux.872 873 Please double-check with `sudo sycl-ls`.874 875 If it's present in the list, please add video/render group to your user then **logout/login** or restart your system:876 877 ```878 sudo usermod -aG render $USER879 sudo usermod -aG video $USER880 ```881 Otherwise, please double-check the GPU driver installation steps.882 883- Can I report Ollama issue on Intel GPU to llama.cpp SYCL backend?884 885 No. We can't support Ollama issue directly, because we aren't familiar with Ollama.886 887 Suggest reproducing on llama.cpp and report similar issue to llama.cpp. We will support it.888 889 It's same for other projects including llama.cpp SYCL backend.890 891- `Native API failed. Native API returns: 39 (UR_RESULT_ERROR_OUT_OF_DEVICE_MEMORY)`, `ggml_backend_sycl_buffer_type_alloc_buffer: can't allocate 3503030272 Bytes of memory on device`, or `failed to allocate SYCL0 buffer`892 893 You are running out of Device Memory.894 895 |Reason|Solution|896 |-|-|897 | The default context is too big. It leads to excessive memory usage.|Set `-c 8192` or a smaller value.|898 | The model is too big and requires more memory than what is available.|Choose a smaller model or change to a smaller quantization, like Q5 -> Q4;<br>Alternatively, use more than one device to load model.|899 900- `ggml_backend_sycl_buffer_type_alloc_buffer: can't allocate 5000000000 Bytes of memory on device`901 902 With the default `GGML_SYCL_ENABLE_LEVEL_ZERO=1`, llama.cpp requests Level Zero's relaxed maximum-size allocation limit directly. If Level Zero support is disabled at build time or runtime and the allocation goes through SYCL/Unified Runtime instead, enable support for allocations larger than 4 GiB by:903 ```904 export UR_L0_ENABLE_RELAXED_ALLOCATION_LIMITS=1905 set UR_L0_ENABLE_RELAXED_ALLOCATION_LIMITS=1906 ```907 908- When I set `SYCL_CACHE_PERSISTENT=1` in running time, I meet crash.909 910 `SYCL_CACHE_PERSISTENT=1` is not recommended by llama.cpp SYCL backend.911 When cache is enabled, SYCL runtime will try to cache and reuse JIT-compiled binaries.912 913 We find some AI will tell user this cmd to speed up SYCL backend. It only speeds up the startup to skip the JIT process, instead of running speed.914 915 It will bring negative impact when the SYCL binary file is changed frequently in your running environment. The new & old codes mix will lead to crash.916 917 Compare to the benefit, it has brought more failed cases.918 If you are not familiar with the SYCL compiler principle of JIT and AOT, please don't use it.919 920 To restore, you need to remove the local cache: `~/.cache/libsycl_cache/` and execute `unset SYCL_CACHE_PERSISTENT` in running time.921 922- How to use iGPU and dGPU in same time?923 924 1. Detect the devices in your running time.925 ```926 source /opt/intel/oneapi/setvars.sh927 ./build/bin/llama-server --list-devices928 929 or930 ./build/bin/llama-cli --list-devices931 ./build/bin/llama-bench --list-devices932 ./build/bin/llama-completion --list-devices933 934 Available devices:935 SYCL0: Intel(R) Arc(TM) A770 Graphics (15473 MiB, 15473 MiB free)936 SYCL1: Intel(R) UHD Graphics 770 (59675 MiB, 44986 MiB free)937 ```938 939 The dGPU will be in the head of this list and iGPU will be the end.940 If not all GPUs are listed, please check the env var: ONEAPI_DEVICE_SELECTOR and unset it.941 942 2. Set the iGPU and dGPU943 944 Set the iGPU and dGPU by `./build/bin/llama-server --device SYCL0,SYCL1,SYCLxxx`.945 946 947### **GitHub contribution**:948Please add the `[SYCL]` prefix/tag in issues/PRs titles to help the SYCL contributors to check/address them without delay.949 950## TODO951 952- Review ZES_ENABLE_SYSMAN: https://github.com/intel/compute-runtime/blob/master/programmers-guide/SYSMAN.md#support-and-limitations953