其他 AI 加速器

其他 AI 加速器#

vLLM 是一个 Python 库，支持以下 AI 加速器。选择您的 AI 加速器类型以查看特定于供应商的说明

张量处理单元 (TPU) 是 Google 定制开发的专用集成电路 (ASIC)，用于加速机器学习工作负载。TPU 有不同的版本，每个版本都有不同的硬件规格。有关 TPU 的更多信息，请参阅 TPU 系统架构。有关 vLLM 支持的 TPU 版本的更多信息，请参阅

这些 TPU 版本允许您配置 TPU 芯片的物理排列。这可以提高吞吐量和网络性能。有关更多信息，请参阅

为了使用 Cloud TPU，您需要为您的 Google Cloud Platform 项目授予 TPU 配额。TPU 配额指定您可以在 GPC 项目中使用多少 TPU，并根据 TPU 版本、您想要使用的 TPU 数量和配额类型进行指定。有关更多信息，请参阅 TPU 配额。

有关 TPU 价格信息，请参阅 Cloud TPU 价格。

您的 TPU VM 可能需要额外的持久存储。有关更多信息，请参阅 Cloud TPU 数据的存储选项。

注意

此设备没有预构建的 wheels，因此您必须使用预构建的 Docker 镜像或从源码构建 vLLM。

此选项卡提供有关使用 Intel Gaudi 设备运行 vLLM 的说明。

注意

此设备没有预构建的 wheels 或镜像，因此您必须从源码构建 vLLM。

vLLM 0.3.3 及更高版本支持在 AWS Trainium/Inferentia 上使用 Neuron SDK 进行模型推理和服务，并支持连续批处理。分页注意力和分块预填充目前正在开发中，即将推出。Neuron SDK 当前支持的数据类型为 FP16 和 BF16。

注意

此设备没有预构建的 wheels 或镜像，因此您必须从源码构建 vLLM。

由 OpenVINO 提供支持的 vLLM 支持来自 vLLM 支持的模型列表 的所有 LLM 模型，并且可以在所有 x86-64 CPU（至少支持 AVX2）以及集成和独立 Intel® GPU 上执行最佳模型服务（支持的 GPU 列表）。

注意

此设备没有预构建的 wheels 或镜像，因此您必须从源码构建 vLLM。

要求#

  • Google Cloud TPU VM

  • TPU 版本：v6e、v5e、v5p、v4

  • Python：3.10 或更高版本

配置 Cloud TPU

您可以使用 Cloud TPU API 或 排队资源 API 配置 Cloud TPU。本节介绍如何使用排队资源 API 创建 TPU。有关使用 Cloud TPU API 的更多信息，请参阅 使用创建节点 API 创建 Cloud TPU。排队资源使您能够以排队方式请求 Cloud TPU 资源。当您请求排队资源时，该请求将添加到 Cloud TPU 服务维护的队列中。当请求的资源可用时，它将被分配到您的 Google Cloud 项目供您立即独占使用。

注意

在以下所有命令中，将全部大写的参数名称替换为适当的值。有关更多信息，请参阅参数描述表。

使用 GKE 配置 Cloud TPU

有关将 TPU 与 GKE 一起使用的更多信息，请参阅

  • 操作系统：Ubuntu 22.04 LTS

  • Python：3.10

  • Intel Gaudi 加速器

  • Intel Gaudi 软件版本 1.18.0

请按照 Gaudi 安装指南 中提供的说明设置执行环境。为了获得最佳性能，请遵循 优化训练平台指南 中概述的方法。

  • 操作系统：Linux

  • Python：3.9 – 3.11

  • 加速器：NeuronCore_v2（在 trn1/inf2 实例中）

  • Pytorch 2.0.1/2.1.1

  • AWS Neuron SDK 2.16/2.17（在 python 3.8 上验证）

  • 操作系统：Linux

  • 指令集架构 (ISA) 要求：至少 AVX2。

配置新环境#

使用排队资源 API 配置 Cloud TPU

创建具有 4 个 TPU 芯片的 TPU v5e

gcloud alpha compute tpus queued-resources create QUEUED_RESOURCE_ID \
--node-id TPU_NAME \
--project PROJECT_ID \
--zone ZONE \
--accelerator-type ACCELERATOR_TYPE \
--runtime-version RUNTIME_VERSION \
--service-account SERVICE_ACCOUNT
参数描述#

参数名称

描述

QUEUED_RESOURCE_ID

排队资源请求的用户分配 ID。

TPU_NAME

TPU 的用户分配名称，在分配排队资源请求时创建。

PROJECT_ID

您的 Google Cloud 项目

ZONE

您要在其中创建 Cloud TPU 的 GCP 区域。您使用的值取决于您使用的 TPU 版本。有关更多信息，请参阅 TPU 区域 和 可用区 <https://cloud.google.com/tpu/docs/regions-zones>_

ACCELERATOR_TYPE

您要使用的 TPU 版本。指定 TPU 版本，例如 v5litepod-4 指定具有 4 个内核的 v5e TPU。有关更多信息，请参阅 TPU 版本 <https://cloud.devsite.corp.google.com/tpu/docs/system-architecture-tpu-vm#versions>_。

RUNTIME_VERSION

要使用的 TPU VM 运行时版本。有关更多信息，请参阅 TPU VM 镜像 <https://cloud.google.com/tpu/docs/runtimes>_。

SERVICE_ACCOUNT

您的服务帐户的电子邮件地址。您可以在 IAM Cloud Console 的服务帐户下找到它。例如：tpu-service-account@<your_project_ID>.iam.gserviceaccount.com

使用 SSH 连接到您的 TPU

gcloud compute tpus tpu-vm ssh TPU_NAME --zone ZONE

环境验证

要验证 Intel Gaudi 软件是否已正确安装，请运行

hl-smi # verify that hl-smi is in your PATH and each Gaudi accelerator is visible
apt list --installed | grep habana # verify that habanalabs-firmware-tools, habanalabs-graph, habanalabs-rdma-core, habanalabs-thunk and habanalabs-container-runtime are installed
pip list | grep habana # verify that habana-torch-plugin, habana-torch-dataloader, habana-pyhlml and habana-media-loader are installed
pip list | grep neural # verify that neural_compressor is installed

有关更多详细信息，请参阅 Intel Gaudi 软件堆栈验证。

运行 Docker 镜像

强烈建议使用 Intel Gaudi vault 中的最新 Docker 镜像。有关更多详细信息，请参阅 Intel Gaudi 文档。

使用以下命令运行 Docker 镜像

docker pull vault.habana.ai/gaudi-docker/1.18.0/ubuntu22.04/habanalabs/pytorch-installer-2.4.0:latest
docker run -it --runtime=habana -e HABANA_VISIBLE_DEVICES=all -e OMPI_MCA_btl_vader_single_copy_mechanism=none --cap-add=sys_nice --net=host --ipc=host vault.habana.ai/gaudi-docker/1.18.0/ubuntu22.04/habanalabs/pytorch-installer-2.4.0:latest

启动 Trn1/Inf2 实例

以下是在 Ubuntu 22.04 LTS 上启动 trn1/inf2 实例以安装 PyTorch Neuron (“torch-neuronx”) 设置的步骤。

  • 请按照 启动 Amazon EC2 实例 中的说明启动实例。在 EC2 控制台中选择实例类型时，请确保选择正确的实例类型。

  • 要获取有关实例大小和定价的更多信息，请参阅：Trn1 网页，Inf2 网页

  • 选择 Ubuntu Server 22.04 TLS AMI

  • 启动 Trn1/Inf2 时，请将您的主 EBS 卷大小调整为至少 512GB。

  • 启动实例后，请按照 连接到您的实例 中的说明连接到实例

安装驱动程序和工具

如果安装了 Deep Learning AMI Neuron，则无需安装驱动程序和工具。如果操作系统上未安装驱动程序和工具，请按照以下步骤操作

# Configure Linux for Neuron repository updates
. /etc/os-release
sudo tee /etc/apt/sources.list.d/neuron.list > /dev/null <<EOF
deb https://apt.repos.neuron.amazonaws.com ${VERSION_CODENAME} main
EOF
wget -qO - https://apt.repos.neuron.amazonaws.com/GPG-PUB-KEY-AMAZON-AWS-NEURON.PUB | sudo apt-key add -

# Update OS packages
sudo apt-get update -y

# Install OS headers
sudo apt-get install linux-headers-$(uname -r) -y

# Install git
sudo apt-get install git -y

# install Neuron Driver
sudo apt-get install aws-neuronx-dkms=2.* -y

# Install Neuron Runtime
sudo apt-get install aws-neuronx-collectives=2.* -y
sudo apt-get install aws-neuronx-runtime-lib=2.* -y

# Install Neuron Tools
sudo apt-get install aws-neuronx-tools=2.* -y

# Add PATH
export PATH=/opt/aws/neuron/bin:$PATH

您可以使用 conda 创建新的 Python 环境

# (Recommended) Create a new conda environment.
conda create -n vllm python=3.12 -y
conda activate vllm

注意

PyTorch 已弃用 conda 发布通道。如果您使用 conda，请仅使用它来创建 Python 环境，而不是安装软件包。

或者，您可以使用 uv（一个非常快速的 Python 环境管理器）创建一个新的 Python 环境。请按照 文档 安装 uv。安装 uv 后，您可以使用以下命令创建一个新的 Python 环境

# (Recommended) Create a new uv environment. Use `--seed` to install `pip` and `setuptools` in the environment.
uv venv vllm --python 3.12 --seed
source vllm/bin/activate

使用 Python 设置#

预构建 wheels#

目前，没有预构建的 TPU wheels。

目前，没有预构建的 Intel Gaudi wheels。

目前，没有预构建的 Neuron wheels。

目前，没有预构建的 OpenVINO wheels。

从源码构建 wheel#

安装 Miniconda

wget https://repo.anaconda.com/miniconda/Miniconda3-latest-Linux-x86_64.sh
bash Miniconda3-latest-Linux-x86_64.sh
source ~/.bashrc

为 vLLM 创建并激活 Conda 环境

conda create -n vllm python=3.10 -y
conda activate vllm

克隆 vLLM 仓库并转到 vLLM 目录

git clone https://github.com/vllm-project/vllm.git && cd vllm

卸载现有的 torch 和 torch_xla 软件包

pip uninstall torch torch-xla -y

安装构建依赖项

pip install -r requirements-tpu.txt
sudo apt-get install libopenblas-base libopenmpi-dev libomp-dev

运行 setup 脚本

VLLM_TARGET_DEVICE="tpu" python setup.py develop

要从源码构建和安装 vLLM，请运行

git clone https://github.com/vllm-project/vllm.git
cd vllm
pip install -r requirements-hpu.txt
python setup.py develop

目前，最新的特性和性能优化是在 Gaudi 的 vLLM-fork 中开发的，我们定期将其上游到 vLLM 主仓库。要安装最新的 HabanaAI/VLLM-fork，请运行以下命令

git clone https://github.com/HabanaAI/vllm-fork.git
cd vllm-fork
git checkout habana_main
pip install -r requirements-hpu.txt
python setup.py develop

注意

当前 Neuron 支持的 Pytorch 版本安装了 triton 版本 2.1.0。这与 vllm >= 0.5.3 不兼容。您可能会看到错误 cannot import name 'default_dump_dir...。要解决此问题，请在安装 vLLM wheel 后运行 pip install --upgrade triton==3.0.0。

以下说明适用于 Neuron SDK 2.16 及更高版本。

安装 transformers-neuronx 及其依赖项

transformers-neuronx 将成为支持 trn1/inf2 实例推理的后端。请按照以下步骤安装 transformer-neuronx 软件包及其依赖项。

# Install Python venv
sudo apt-get install -y python3.10-venv g++

# Create Python venv
python3.10 -m venv aws_neuron_venv_pytorch

# Activate Python venv
source aws_neuron_venv_pytorch/bin/activate

# Install Jupyter notebook kernel
pip install ipykernel
python3.10 -m ipykernel install --user --name aws_neuron_venv_pytorch --display-name "Python (torch-neuronx)"
pip install jupyter notebook
pip install environment_kernels

# Set pip repository pointing to the Neuron repository
python -m pip config set global.extra-index-url https://pip.repos.neuron.amazonaws.com

# Install wget, awscli
python -m pip install wget
python -m pip install awscli

# Update Neuron Compiler and Framework
python -m pip install --upgrade neuronx-cc==2.* --pre torch-neuronx==2.1.* torchvision transformers-neuronx

从源码安装 vLLM

一旦 neuronx-cc 和 transformers-neuronx 软件包安装完成，我们将能够按如下方式安装 vllm

git clone https://github.com/vllm-project/vllm.git
cd vllm
pip install -U -r requirements-neuron.txt
VLLM_TARGET_DEVICE="neuron" pip install .

如果在安装过程中正确检测到 neuron 软件包，则将安装 vllm-0.3.0+neuron212。

首先，安装 Python 并确保您拥有最新的 pip。例如，在 Ubuntu 22.04 上，您可以运行

sudo apt-get update  -y
sudo apt-get install python3
pip install --upgrade pip

其次，克隆 vLLM 并安装 vLLM OpenVINO 后端安装的先决条件

git clone https://github.com/vllm-project/vllm.git
cd vllm
pip install -r requirements-build.txt --extra-index-url https://download.pytorch.org/whl/cpu

最后，安装带有 OpenVINO 后端的 vLLM

PIP_EXTRA_INDEX_URL="https://download.pytorch.org/whl/cpu" VLLM_TARGET_DEVICE=openvino python -m pip install -v .

提示

要将 vLLM OpenVINO 后端与 GPU 设备一起使用，请确保您的系统已正确设置。请按照此处提供的说明进行操作：https://docs.openvino.ai/2024/get-started/configurations/configurations-intel-gpu.html。

使用 Docker 设置#

预构建镜像#

有关使用官方 Docker 镜像的说明，请参阅 使用 vLLM 官方 Docker 镜像，确保将镜像名称 vllm/vllm-openai 替换为 vllm/vllm-tpu。

目前，没有预构建的 Intel Gaudi 镜像。

目前，没有预构建的 Neuron 镜像。

目前，没有预构建的 OpenVINO 镜像。

从源码构建镜像#

您可以使用 Dockerfile.tpu 构建具有 TPU 支持的 Docker 镜像。

docker build -f Dockerfile.tpu -t vllm-tpu .

使用以下命令运行 Docker 镜像

# Make sure to add `--privileged --net host --shm-size=16G`.
docker run --privileged --net host --shm-size=16G -it vllm-tpu

注意

由于 TPU 依赖于需要静态形状的 XLA，因此 vLLM 对可能的输入形状进行分桶，并为每个形状编译一个 XLA 图。首次运行时，编译时间可能需要 20~30 分钟。但是，编译时间之后会减少到约 5 分钟，因为 XLA 图缓存在磁盘中（默认情况下在 VLLM_XLA_CACHE_PATH 或 ~/.cache/vllm/xla_cache 中）。

提示

如果您遇到以下错误

from torch._C import *  # noqa: F403
ImportError: libopenblas.so.0: cannot open shared object file: No such
file or directory

使用以下命令安装 OpenBLAS

sudo apt-get install libopenblas-base libopenmpi-dev libomp-dev

docker build -f Dockerfile.hpu -t vllm-hpu-env  .
docker run -it --runtime=habana -e HABANA_VISIBLE_DEVICES=all -e OMPI_MCA_btl_vader_single_copy_mechanism=none --cap-add=sys_nice --net=host --rm vllm-hpu-env

提示

如果您观察到以下错误：docker: Error response from daemon: Unknown runtime specified habana.，请参阅 Intel Gaudi 软件堆栈和驱动程序安装 的“使用容器安装”部分。确保您已安装 habana-container-runtime 软件包，并且已注册 habana 容器运行时。

有关构建 Docker 镜像的说明，请参阅 从源码构建 vLLM 的 Docker 镜像。

确保使用 Dockerfile.neuron 代替默认的 Dockerfile。

docker build -f Dockerfile.openvino -t vllm-openvino-env .
docker run -it --rm vllm-openvino-env

额外信息#

此设备没有额外信息。

支持的特性

  • 离线推理

  • 通过 OpenAI 兼容服务器 进行在线服务

  • HPU 自动检测 - 无需在 vLLM 中手动选择设备

  • 分页 KV 缓存，为 Intel Gaudi 加速器启用了算法

  • Paged Attention、KV 缓存操作、预填充注意力、均方根层归一化、旋转位置编码的自定义 Intel Gaudi 实现

  • 用于多卡推理的张量并行支持

  • 使用 HPU Graphs 进行推理，以加速低批次延迟和吞吐量

  • 具有线性偏差的注意力 (ALiBi)

不支持的特性

  • 束搜索

  • LoRA 适配器

  • 量化

  • 预填充分块（混合批次推理）

支持的配置

以下配置已验证可在 Gaudi2 设备上正常工作。未列出的配置可能可以工作，也可能无法工作。

性能调优

执行模式

目前在 vLLM for HPU 中，我们支持四种执行模式，具体取决于所选的 HPU PyTorch Bridge 后端（通过 PT_HPU_LAZY_MODE 环境变量）和 --enforce-eager 标志。

vLLM 执行模式#

PT_HPU_LAZY_MODE

enforce_eager

执行模式

0

0

torch.compile

0

1

PyTorch eager 模式

1

0

HPU Graphs

1

1

PyTorch lazy 模式

警告

在 1.18.0 中，所有利用 PT_HPU_LAZY_MODE=0 的模式都处于高度实验性阶段，仅应用于验证功能正确性。它们的性能将在后续版本中得到改进。为了在 1.18.0 中获得最佳性能，请使用 HPU Graphs 或 PyTorch lazy 模式。

分桶机制

Intel Gaudi 加速器在处理具有固定张量形状的模型时效果最佳。Intel Gaudi Graph Compiler 负责生成优化的二进制代码，该代码在 Gaudi 上实现给定的模型拓扑。在其默认配置中，生成的二进制代码可能严重依赖于输入和输出张量形状，并且当在同一拓扑中遇到不同形状的张量时，可能需要重新编译图。虽然生成的二进制文件有效地利用了 Gaudi，但编译本身可能会在端到端执行中引入明显的开销。在动态推理服务场景中，需要最大限度地减少图编译的次数，并降低服务器运行时发生图编译的风险。目前，这是通过在两个维度（batch_size 和 sequence_length）上对模型的正向传递进行“分桶”来实现的。

注意

分桶允许我们显着减少所需图的数量，但它不处理任何图编译和设备代码生成 - 这在预热和 HPUGraph 捕获阶段完成。

分桶范围由 3 个参数确定 - min、step 和 max。它们可以分别为 prompt 和 decode 阶段以及 batch size 和 sequence length 维度单独设置。这些参数可以在 vLLM 启动期间的日志中观察到

INFO 08-01 21:37:59 hpu_model_runner.py:493] Prompt bucket config (min, step, max_warmup) bs:[1, 32, 4], seq:[128, 128, 1024]
INFO 08-01 21:37:59 hpu_model_runner.py:499] Generated 24 prompt buckets: [(1, 128), (1, 256), (1, 384), (1, 512), (1, 640), (1, 768), (1, 896), (1, 1024), (2, 128), (2, 256), (2, 384), (2, 512), (2, 640), (2, 768), (2, 896), (2, 1024), (4, 128), (4, 256), (4, 384), (4, 512), (4, 640), (4, 768), (4, 896), (4, 1024)]
INFO 08-01 21:37:59 hpu_model_runner.py:504] Decode bucket config (min, step, max_warmup) bs:[1, 128, 4], seq:[128, 128, 2048]
INFO 08-01 21:37:59 hpu_model_runner.py:509] Generated 48 decode buckets: [(1, 128), (1, 256), (1, 384), (1, 512), (1, 640), (1, 768), (1, 896), (1, 1024), (1, 1152), (1, 1280), (1, 1408), (1, 1536), (1, 1664), (1, 1792), (1, 1920), (1, 2048), (2, 128), (2, 256), (2, 384), (2, 512), (2, 640), (2, 768), (2, 896), (2, 1024), (2, 1152), (2, 1280), (2, 1408), (2, 1536), (2, 1664), (2, 1792), (2, 1920), (2, 2048), (4, 128), (4, 256), (4, 384), (4, 512), (4, 640), (4, 768), (4, 896), (4, 1024), (4, 1152), (4, 1280), (4, 1408), (4, 1536), (4, 1664), (4, 1792), (4, 1920), (4, 2048)]

min 确定桶的最小值。step 确定桶之间的间隔，max 确定桶的上限。此外，min 和 step 之间的间隔有特殊处理 – min 乘以 2 的连续幂，直到达到 step。我们称之为 ramp-up 阶段，它用于处理较低的批次大小，并最大限度地减少浪费，同时允许在较大的批次大小上进行更大的填充。

示例（带 ramp-up）

min = 2, step = 32, max = 64
=> ramp_up = (2, 4, 8, 16)
=> stable = (32, 64)
=> buckets = ramp_up + stable => (2, 4, 8, 16, 32, 64)

示例（不带 ramp-up）

min = 128, step = 128, max = 512
=> ramp_up = ()
=> stable = (128, 256, 384, 512)
=> buckets = ramp_up + stable => (128, 256, 384, 512)

在记录的场景中，为 prompt（预填充）运行生成了 24 个桶，为 decode 运行生成了 48 个桶。每个桶对应于给定模型的单独优化的设备二进制文件，并具有指定的张量形状。每当处理一批请求时，都会在批次和序列长度维度上将其填充到尽可能小的桶。

警告

如果请求在任何维度上超过最大桶大小，则将在没有填充的情况下处理该请求，并且其处理可能需要图编译，从而可能显着增加端到端延迟。桶的边界是用户可配置的，通过环境变量，可以增加上限桶边界以避免这种情况。

例如，如果一个包含 3 个序列且最大序列长度为 412 的请求进入空闲的 vLLM 服务器，它将被填充并作为 (4, 512) 预填充桶执行，因为 batch_size（序列数）将被填充到 4（最接近的批次大小维度，大于 3），并且最大序列长度将被填充到 512（最接近的序列长度维度，大于 412）。在预填充阶段之后，它将作为 (4, 512) decode 桶执行，并将继续作为该桶，直到批次维度发生变化（由于请求完成） - 在这种情况下，它将变为 (2, 512) 桶，或者上下文长度增加到 512 个 token 以上，在这种情况下，它将变为 (4, 640) 桶。

注意

分桶对于客户端是透明的 – 序列长度维度中的填充永远不会返回给客户端，并且批次维度中的填充不会创建新请求。

预热

预热是一个可选但强烈推荐的步骤，在 vLLM 服务器开始监听之前发生。它使用虚拟数据为每个桶执行正向传递。目标是预编译所有图，并且在服务器运行时在桶边界内不产生任何图编译开销。每个预热步骤都会在 vLLM 启动期间记录

INFO 08-01 22:26:47 hpu_model_runner.py:1066] [Warmup][Prompt][1/24] batch_size:4 seq_len:1024 free_mem:79.16 GiB
INFO 08-01 22:26:47 hpu_model_runner.py:1066] [Warmup][Prompt][2/24] batch_size:4 seq_len:896 free_mem:55.43 GiB
INFO 08-01 22:26:48 hpu_model_runner.py:1066] [Warmup][Prompt][3/24] batch_size:4 seq_len:768 free_mem:55.43 GiB
...
INFO 08-01 22:26:59 hpu_model_runner.py:1066] [Warmup][Prompt][24/24] batch_size:1 seq_len:128 free_mem:55.43 GiB
INFO 08-01 22:27:00 hpu_model_runner.py:1066] [Warmup][Decode][1/48] batch_size:4 seq_len:2048 free_mem:55.43 GiB
INFO 08-01 22:27:00 hpu_model_runner.py:1066] [Warmup][Decode][2/48] batch_size:4 seq_len:1920 free_mem:55.43 GiB
INFO 08-01 22:27:01 hpu_model_runner.py:1066] [Warmup][Decode][3/48] batch_size:4 seq_len:1792 free_mem:55.43 GiB
...
INFO 08-01 22:27:16 hpu_model_runner.py:1066] [Warmup][Decode][47/48] batch_size:2 seq_len:128 free_mem:55.43 GiB
INFO 08-01 22:27:16 hpu_model_runner.py:1066] [Warmup][Decode][48/48] batch_size:1 seq_len:128 free_mem:55.43 GiB

此示例使用与 分桶机制 部分相同的桶。每行输出对应于单个桶的执行。当首次执行桶时，其图将被编译，并且可以稍后重用，从而跳过进一步的图编译。

提示

编译所有桶可能需要一些时间，并且可以使用 VLLM_SKIP_WARMUP=true 环境变量关闭。请记住，如果您这样做，则在首次执行给定桶时可能会遇到图编译。禁用预热以进行开发是可以的，但强烈建议在部署中启用它。

HPU Graph 捕获

HPU Graphs 是目前 vLLM 在 Intel Gaudi 上最有效的执行方法。当启用 HPU Graphs 时，执行图将被提前跟踪（记录）（在执行预热后），以便稍后在推理期间重放，从而显着减少主机开销。记录可能占用大量内存，在分配 KV 缓存时需要考虑这一点。启用 HPU Graphs 将影响可用 KV 缓存块的数量，但 vLLM 提供了用户可配置的变量来控制内存管理。

当使用 HPU 图形时，它们与 KV 缓存共享通用内存池（“可用内存”），这由 gpu_memory_utilization 标志（默认值为 0.9）确定。在分配 KV 缓存之前，模型权重会被加载到设备上，并使用虚拟数据执行模型的前向传递，以估计内存使用量。 之后，才会使用 gpu_memory_utilization 标志 - 在其默认值下，会将那时点 90% 的可用设备内存标记为可用。 接下来，分配 KV 缓存，预热模型，并捕获 HPU 图形。 环境变量 VLLM_GRAPH_RESERVED_MEM 定义了为 HPU 图形捕获保留的内存比例。 使用其默认值 (VLLM_GRAPH_RESERVED_MEM=0.1)，10% 的可用内存将保留用于图形捕获（以下称为“可用图形内存”），其余 90% 将用于 KV 缓存。 环境变量 VLLM_GRAPH_PROMPT_RATIO 决定了为预填充和解码图形保留的可用图形内存的比例。 默认情况下 (VLLM_GRAPH_PROMPT_RATIO=0.3)，两个阶段都具有相等的内存约束。 较低的值对应于为预填充阶段保留的较少可用图形内存，例如 VLLM_GRAPH_PROMPT_RATIO=0.2 将为预填充图形保留 20% 的可用图形内存，为解码图形保留 80% 的可用图形内存。

注意

gpu_memory_utilization 不对应于整个 HPU 的绝对内存使用量。 它指定了在加载模型和执行性能分析运行后剩余的内存余量。 如果设备总内存为 100 GiB，并且在加载模型权重和执行性能分析运行后剩余 50 GiB 的可用内存，则默认值下的 gpu_memory_utilization 将把 50 GiB 中的 90% 标记为可用，留下 5 GiB 的余量，而与设备总内存无关。

用户还可以分别配置用于 prompt 和 decode 阶段的 HPU 图形捕获策略。 策略会影响图形捕获的顺序。 实现了两种策略： - max_bs - 图形捕获队列将按批大小降序排序。 批大小相同的存储桶按序列长度升序排序（例如 (64, 128), (64, 256), (32, 128), (32, 256), (1, 128), (1,256)），解码的默认策略 - min_tokens - 图形捕获队列将按每个图形处理的令牌数量 (batch_size*sequence_length) 升序排序，prompt 的默认策略

当有大量请求等待处理时，vLLM 调度器将尝试尽快填充解码的最大批大小。 当一个请求完成时，解码批大小会减小。 当这种情况发生时，vLLM 将尝试为等待队列中的请求调度一个预填充迭代，以将解码批大小填充到之前的状态。 这意味着在满负载情况下，解码批大小通常处于最大值，这使得捕获大批大小 HPU 图形至关重要，正如 max_bs 策略所反映的那样。 另一方面，预填充将最频繁地以非常小的批大小 (1-4) 执行，这反映在 min_tokens 策略中。

注意

VLLM_GRAPH_PROMPT_RATIO 不会对每个阶段（预填充和解码）的图形所占用的内存设置硬性限制。 vLLM 将首先尝试用完所有可用的预填充图形内存（可用图形内存 * VLLM_GRAPH_PROMPT_RATIO）来捕获预填充 HPU 图形，接下来它将尝试对解码图形和可用的解码图形内存池执行相同的操作。 如果一个阶段已完全捕获，并且可用图形内存池中还剩余未使用的内存，则 vLLM 将尝试为另一个阶段进一步捕获图形，直到无法在不超出保留内存池的情况下捕获更多 HPU 图形为止。 可以在下面的示例中观察到该机制的行为。

vLLM 服务器会记录每个描述的步骤，如下所示（负值对应于正在释放的内存）

INFO 08-02 17:37:44 hpu_model_runner.py:493] Prompt bucket config (min, step, max_warmup) bs:[1, 32, 4], seq:[128, 128, 1024]
INFO 08-02 17:37:44 hpu_model_runner.py:499] Generated 24 prompt buckets: [(1, 128), (1, 256), (1, 384), (1, 512), (1, 640), (1, 768), (1, 896), (1, 1024), (2, 128), (2, 256), (2, 384), (2, 512), (2, 640), (2, 768), (2, 896), (2, 1024), (4, 128), (4, 256), (4, 384), (4, 512), (4, 640), (4, 768), (4, 896), (4, 1024)]
INFO 08-02 17:37:44 hpu_model_runner.py:504] Decode bucket config (min, step, max_warmup) bs:[1, 128, 4], seq:[128, 128, 2048]
INFO 08-02 17:37:44 hpu_model_runner.py:509] Generated 48 decode buckets: [(1, 128), (1, 256), (1, 384), (1, 512), (1, 640), (1, 768), (1, 896), (1, 1024), (1, 1152), (1, 1280), (1, 1408), (1, 1536), (1, 1664), (1, 1792), (1, 1920), (1, 2048), (2, 128), (2, 256), (2, 384), (2, 512), (2, 640), (2, 768), (2, 896), (2, 1024), (2, 1152), (2, 1280), (2, 1408), (2, 1536), (2, 1664), (2, 1792), (2, 1920), (2, 2048), (4, 128), (4, 256), (4, 384), (4, 512), (4, 640), (4, 768), (4, 896), (4, 1024), (4, 1152), (4, 1280), (4, 1408), (4, 1536), (4, 1664), (4, 1792), (4, 1920), (4, 2048)]
INFO 08-02 17:37:52 hpu_model_runner.py:430] Pre-loading model weights on hpu:0 took 14.97 GiB of device memory (14.97 GiB/94.62 GiB used) and 2.95 GiB of host memory (475.2 GiB/1007 GiB used)
INFO 08-02 17:37:52 hpu_model_runner.py:438] Wrapping in HPU Graph took 0 B of device memory (14.97 GiB/94.62 GiB used) and -252 KiB of host memory (475.2 GiB/1007 GiB used)
INFO 08-02 17:37:52 hpu_model_runner.py:442] Loading model weights took in total 14.97 GiB of device memory (14.97 GiB/94.62 GiB used) and 2.95 GiB of host memory (475.2 GiB/1007 GiB used)
INFO 08-02 17:37:54 hpu_worker.py:134] Model profiling run took 504 MiB of device memory (15.46 GiB/94.62 GiB used) and 180.9 MiB of host memory (475.4 GiB/1007 GiB used)
INFO 08-02 17:37:54 hpu_worker.py:158] Free device memory: 79.16 GiB, 39.58 GiB usable (gpu_memory_utilization=0.5), 15.83 GiB reserved for HPUGraphs (VLLM_GRAPH_RESERVED_MEM=0.4), 23.75 GiB reserved for KV cache
INFO 08-02 17:37:54 hpu_executor.py:85] # HPU blocks: 1519, # CPU blocks: 0
INFO 08-02 17:37:54 hpu_worker.py:190] Initializing cache engine took 23.73 GiB of device memory (39.2 GiB/94.62 GiB used) and -1.238 MiB of host memory (475.4 GiB/1007 GiB used)
INFO 08-02 17:37:54 hpu_model_runner.py:1066] [Warmup][Prompt][1/24] batch_size:4 seq_len:1024 free_mem:55.43 GiB
...
INFO 08-02 17:38:22 hpu_model_runner.py:1066] [Warmup][Decode][48/48] batch_size:1 seq_len:128 free_mem:55.43 GiB
INFO 08-02 17:38:22 hpu_model_runner.py:1159] Using 15.85 GiB/55.43 GiB of free device memory for HPUGraphs, 7.923 GiB for prompt and 7.923 GiB for decode (VLLM_GRAPH_PROMPT_RATIO=0.3)
INFO 08-02 17:38:22 hpu_model_runner.py:1066] [Warmup][Graph/Prompt][1/24] batch_size:1 seq_len:128 free_mem:55.43 GiB
...
INFO 08-02 17:38:26 hpu_model_runner.py:1066] [Warmup][Graph/Prompt][11/24] batch_size:1 seq_len:896 free_mem:48.77 GiB
INFO 08-02 17:38:27 hpu_model_runner.py:1066] [Warmup][Graph/Decode][1/48] batch_size:4 seq_len:128 free_mem:47.51 GiB
...
INFO 08-02 17:38:41 hpu_model_runner.py:1066] [Warmup][Graph/Decode][48/48] batch_size:1 seq_len:2048 free_mem:47.35 GiB
INFO 08-02 17:38:41 hpu_model_runner.py:1066] [Warmup][Graph/Prompt][12/24] batch_size:4 seq_len:256 free_mem:47.35 GiB
INFO 08-02 17:38:42 hpu_model_runner.py:1066] [Warmup][Graph/Prompt][13/24] batch_size:2 seq_len:512 free_mem:45.91 GiB
INFO 08-02 17:38:42 hpu_model_runner.py:1066] [Warmup][Graph/Prompt][14/24] batch_size:1 seq_len:1024 free_mem:44.48 GiB
INFO 08-02 17:38:43 hpu_model_runner.py:1066] [Warmup][Graph/Prompt][15/24] batch_size:2 seq_len:640 free_mem:43.03 GiB
INFO 08-02 17:38:43 hpu_model_runner.py:1128] Graph/Prompt captured:15 (62.5%) used_mem:14.03 GiB buckets:[(1, 128), (1, 256), (1, 384), (1, 512), (1, 640), (1, 768), (1, 896), (1, 1024), (2, 128), (2, 256), (2, 384), (2, 512), (2, 640), (4, 128), (4, 256)]
INFO 08-02 17:38:43 hpu_model_runner.py:1128] Graph/Decode captured:48 (100.0%) used_mem:161.9 MiB buckets:[(1, 128), (1, 256), (1, 384), (1, 512), (1, 640), (1, 768), (1, 896), (1, 1024), (1, 1152), (1, 1280), (1, 1408), (1, 1536), (1, 1664), (1, 1792), (1, 1920), (1, 2048), (2, 128), (2, 256), (2, 384), (2, 512), (2, 640), (2, 768), (2, 896), (2, 1024), (2, 1152), (2, 1280), (2, 1408), (2, 1536), (2, 1664), (2, 1792), (2, 1920), (2, 2048), (4, 128), (4, 256), (4, 384), (4, 512), (4, 640), (4, 768), (4, 896), (4, 1024), (4, 1152), (4, 1280), (4, 1408), (4, 1536), (4, 1664), (4, 1792), (4, 1920), (4, 2048)]
INFO 08-02 17:38:43 hpu_model_runner.py:1206] Warmup finished in 49 secs, allocated 14.19 GiB of device memory
INFO 08-02 17:38:43 hpu_executor.py:91] init_cache_engine took 37.92 GiB of device memory (53.39 GiB/94.62 GiB used) and 57.86 MiB of host memory (475.4 GiB/1007 GiB used)

  • 我们建议在 Gaudi 2 上使用 128 的 block_size 和 BF16 数据类型运行推理。 使用默认值 (16, 32) 可能会由于矩阵乘法引擎利用率不足而导致次优性能（请参阅 Gaudi 架构）。

  • 为了在 Llama 7B 上获得最大吞吐量，我们建议在启用 HPU 图形的情况下，以 128 或 256 的批大小和 2048 的最大上下文长度运行。 如果您遇到内存不足问题，请参阅故障排除部分。

环境变量

诊断和性能分析旋钮

  • VLLM_PROFILER_ENABLED：如果为 true，则将启用高级分析器。 生成的 JSON 跟踪可以在 perfetto.habana.ai 中查看。 默认禁用。

  • VLLM_HPU_LOG_STEP_GRAPH_COMPILATION：如果为 true，则将记录每个 vLLM 引擎步骤的图形编译，仅当有任何编译时才记录 - 强烈建议与 PT_HPU_METRICS_GC_DETAILS=1 一起使用。 默认禁用。

  • VLLM_HPU_LOG_STEP_GRAPH_COMPILATION_ALL：如果为 true，则将记录每个 vLLM 引擎步骤的图形编译，始终记录，即使没有编译。 默认禁用。

  • VLLM_HPU_LOG_STEP_CPU_FALLBACKS：如果为 true，则将记录每个 vLLM 引擎步骤的 CPU 回退，仅当有任何回退时才记录。 默认禁用。

  • VLLM_HPU_LOG_STEP_CPU_FALLBACKS_ALL：如果为 true，则将记录每个 vLLM 引擎步骤的 CPU 回退，始终记录，即使没有回退。 默认禁用。

性能调优旋钮

  • VLLM_SKIP_WARMUP：如果为 true，将跳过预热，默认为 false

  • VLLM_GRAPH_RESERVED_MEM：专用于 HPUGraph 捕获的内存百分比，默认为 0.1

  • VLLM_GRAPH_PROMPT_RATIO：专用于 prompt 图形的保留图形内存百分比，默认为 0.3

  • VLLM_GRAPH_PROMPT_STRATEGY：确定 prompt 图形捕获顺序的策略，min_tokens 或 max_bs，默认为 min_tokens

  • VLLM_GRAPH_DECODE_STRATEGY：确定 decode 图形捕获顺序的策略，min_tokens 或 max_bs，默认为 max_bs

  • VLLM_{phase}_{dim}_BUCKET_{param} - 配置分桶机制范围的 12 个环境变量的集合

    • {phase} 是 PROMPT 或 DECODE

    • {dim} 是 BS, SEQ 或 BLOCK

    • {param} 是 MIN, STEP 或 MAX

    • 默认值

      • Prompt

        • 批大小最小值 (VLLM_PROMPT_BS_BUCKET_MIN)：1

        • 批大小步长 (VLLM_PROMPT_BS_BUCKET_STEP)：min(max_num_seqs, 32)

        • 批大小最大值 (VLLM_PROMPT_BS_BUCKET_MAX)：min(max_num_seqs, 64)

        • 序列长度最小值 (VLLM_PROMPT_SEQ_BUCKET_MIN)：block_size

        • 序列长度步长 (VLLM_PROMPT_SEQ_BUCKET_STEP)：block_size

        • 序列长度最大值 (VLLM_PROMPT_SEQ_BUCKET_MAX)：max_model_len

      • Decode

        • 批大小最小值 (VLLM_DECODE_BS_BUCKET_MIN)：1

        • 批大小步长 (VLLM_DECODE_BS_BUCKET_STEP)：min(max_num_seqs, 32)

        • 批大小最大值 (VLLM_DECODE_BS_BUCKET_MAX)：max_num_seqs

        • 序列长度最小值 (VLLM_DECODE_BLOCK_BUCKET_MIN)：block_size

        • 序列长度步长 (VLLM_DECODE_BLOCK_BUCKET_STEP)：block_size

        • 序列长度最大值 (VLLM_DECODE_BLOCK_BUCKET_MAX)：max(128, (max_num_seqs*max_model_len)/block_size)

此外，还有影响 vLLM 执行的 HPU PyTorch Bridge 环境变量

  • PT_HPU_LAZY_MODE：如果为 0，将使用 Gaudi 的 PyTorch Eager 后端；如果为 1，将使用 Gaudi 的 PyTorch Lazy 后端，默认为 1

  • PT_HPU_ENABLE_LAZY_COLLECTIVES：对于使用 HPU 图形的张量并行推理，需要设置为 true

故障排除：调整 HPU 图形

如果您遇到设备内存不足问题，或者想要尝试以更大的批大小进行推理，请尝试通过以下方式调整 HPU 图形

  • 调整 gpu_memory_utilization 旋钮。 这将减少 KV 缓存的分配，为捕获更大批大小的图形留出一些空间。 默认情况下，gpu_memory_utilization 设置为 0.9。 它尝试为短时间性能分析运行后剩余的 HBM 分配约 90% 的空间用于 KV 缓存。 请注意，减少此值会减少您可用的 KV 缓存块数量，因此会减少您在给定时间内可以处理的有效最大令牌数量。

  • 如果此方法效率不高，您可以完全禁用 HPUGraph。 禁用 HPU 图形后，您将在较低批次时牺牲延迟和吞吐量，以换取在较高批次时可能更高的吞吐量。 您可以通过向服务器添加 --enforce-eager 标志（对于在线服务），或者将 enforce_eager=True 参数传递给 LLM 构造函数（对于离线推理）来完成此操作。

此设备没有额外信息。

支持的特性

OpenVINO vLLM 后端支持以下高级 vLLM 功能

  • 前缀缓存 (--enable-prefix-caching)

  • 分块预填充 (--enable-chunked-prefill)

性能提示

vLLM OpenVINO 后端环境变量

  • VLLM_OPENVINO_DEVICE 用于指定哪个设备用于推理。 如果系统中有多个 GPU，则可以使用其他索引来选择合适的 GPU（例如，VLLM_OPENVINO_DEVICE=GPU.1）。 如果未指定值，则默认使用 CPU 设备。

  • VLLM_OPENVINO_ENABLE_QUANTIZED_WEIGHTS=ON 用于在模型加载阶段启用 U8 权重压缩。 默认情况下，压缩处于关闭状态。 您还可以使用 optimum-cli 导出具有不同压缩技术的模型，并将导出的文件夹作为 <model_id> 传递。

CPU 性能提示

CPU 使用以下环境变量来控制行为

  • VLLM_OPENVINO_KVCACHE_SPACE 用于指定 KV 缓存大小（例如，VLLM_OPENVINO_KVCACHE_SPACE=40 表示 40 GB 的 KV 缓存空间），更大的设置将允许 vLLM 并行运行更多请求。 此参数应根据硬件配置和用户的内存管理模式进行设置。

  • VLLM_OPENVINO_CPU_KV_CACHE_PRECISION=u8 用于控制 KV 缓存精度。 默认情况下，根据平台使用 FP16 / BF16。

为了实现更好的 TPOT / TTFT 延迟，您可以使用 vLLM 的分块预填充功能 (--enable-chunked-prefill)。 根据实验，建议的批大小为 256 (--max-num-batched-tokens)

OpenVINO CPU 的最佳已知配置是

$ VLLM_OPENVINO_KVCACHE_SPACE=100 VLLM_OPENVINO_CPU_KV_CACHE_PRECISION=u8 VLLM_OPENVINO_ENABLE_QUANTIZED_WEIGHTS=ON \
    python3 vllm/benchmarks/benchmark_throughput.py --model meta-llama/Llama-2-7b-chat-hf --dataset vllm/benchmarks/ShareGPT_V3_unfiltered_cleaned_split.json --enable-chunked-prefill --max-num-batched-tokens 256

GPU 性能提示

GPU 设备实现了自动检测可用 GPU 内存的逻辑，默认情况下，它会尝试为 KV 缓存保留尽可能多的内存（考虑到 gpu_memory_utilization 选项）。 但是，可以通过使用 VLLM_OPENVINO_KVCACHE_SPACE 环境变量显式指定 KV 缓存所需的内存量来覆盖此行为（例如，VLLM_OPENVINO_KVCACHE_SPACE=8 表示 8 GB 的 KV 缓存空间）。

目前，对于具有量化权重的模型（支持 8 位和 4 位整数数据类型）和 preemption-mode=swap，使用 GPU 可以实现最佳性能，采用默认的 vLLM 执行参数。

OpenVINO GPU 的最佳已知配置是

$ VLLM_OPENVINO_DEVICE=GPU VLLM_OPENVINO_ENABLE_QUANTIZED_WEIGHTS=ON \
    python3 vllm/benchmarks/benchmark_throughput.py --model meta-llama/Llama-2-7b-chat-hf --dataset vllm/benchmarks/ShareGPT_V3_unfiltered_cleaned_split.json

局限性

  • 不支持 LoRA 服务。

  • 目前仅支持 LLM 模型。 LLaVa 和编码器-解码器模型目前在 vLLM OpenVINO 集成中未启用。

  • 张量和流水线并行目前在 vLLM 集成中未启用。