[Linux][BMG/B60] currUsableMemSize (and SYCL free_memory since oneAPI 2026.2) is capped at 95% of VRAM: ~1.4 GiB of allocatable memory reported as not free

Start with Device::getUsableMemorySize() in shared/source/device/device.cpp, then trace getGlobalMemorySize() and DrmMemoryManager::getPercentOfGlobalMemoryAvailable() to understand the 95% scaling and used-memory calculation. Compare the reported value with kernel free memory and the provided allocation reproducer. Done means currUsableMemSize reflects allocatable VRAM, or the intentional headroom is documented; first confirm what the 5% headroom is meant to cover.

Pre-submission Checklist

• I am using the latest GPU driver version (releases)
• I have searched for similar issues and found none

GPU Hardware

Intel Arc Pro B60 (one dual-GPU card: two 8086:e211 devices, 24 GB each)

DRI Devices Information

pci-0000:06:00.0-card -> ../card1
pci-0000:72:00.0-card -> ../card0
pci-0000:72:00.0-render -> ../renderD128
pci-0000:76:00.0-card -> ../card2
pci-0000:76:00.0-render -> ../renderD129

GPU Detailed Information (lspci output)

06:00.0 VGA compatible controller [0300]: ASPEED Technology, Inc. ASPEED Graphics Family [1a03:2000] (rev 52)
72:00.0 VGA compatible controller [0300]: Intel Corporation Device [8086:e211]
76:00.0 VGA compatible controller [0300]: Intel Corporation Device [8086:e211]

Driver Version

26.40.40196.8

Installed GPU Driver Packages

intel-igc-core-2 2.42.6
intel-igc-opencl-2 2.42.6
intel-ocloc 26.40.40196.8-0
intel-opencl-icd 26.40.40196.8-0
libigdgmm12 22.10.2
libze-intel-gpu1 26.40.40196.8-0
libze1 1.35.0

Driver Installation Details

• Installation method: release .deb packages from https://github.com/intel/compute-runtime/releases/tag/26.40.40196.8, SHA256 checked against the published digests, installed with dpkg -i
• Custom kernel parameters: none

Linux Distribution

Ubuntu 24.04 LTS

Other Linux Distribution

Ubuntu 24.04.5 LTS

Kernel Version & Boot Parameters

7.3.1-070301-generic (Ubuntu mainline kernel build), xe driver, GuC 70.72.1
BOOT_IMAGE=/boot/vmlinuz-7.3.1-070301-generic root=UUID=... ro quiet splash vt.handoff=7

The device runs in SR-IOV PF mode (sriov_info: mode: SR-IOV PF), no VFs enabled.

Actual Behavior

On an idle B60, ze_device_usablemem_size_ext_properties_t::currUsableMemSize is ~1.4 GiB lower than the free VRAM reported by the kernel, although that memory can be allocated and written.

Since oneAPI 2026.2, sycl::ext::intel::info::device::free_memory returns this value: the Unified Runtime Level Zero adapter now prefers currUsableMemSize over Sysman for UR_DEVICE_INFO_GLOBAL_MEM_FREE. Every SYCL application that sizes its allocations from free_memory is therefore affected.

The measurements below were taken on the same machine with the same driver and kernel, on GPU 76:00.0 with nothing else running on it:

Quantity	Value
Level Zero global_mem_size	25,669,140,480 B (23.91 GiB)
Kernel free, DRM_XE_DEVICE_QUERY_MEM_REGIONS (total_size - used)	25,625,300,992 B (23.87 GiB)
SYCL free_memory, oneAPI 2026.1.1 runtime (UR -> Sysman)	23.62-23.67 GiB
SYCL free_memory, oneAPI 2026.2.0 runtime (UR -> currUsableMemSize)	22.43-22.48 GiB
OpenCL CL_DEVICE_GLOBAL_MEM_SIZE / max single allocation	24,385,683,456 B (22.71 GiB) = exactly 0.95 x 25,669,140,480
Really allocated and filled by one process (256 MiB malloc_device + memset until nullptr)	25,501,368,320 B (23.75 GiB)

The reported value is 0.95 x total - used. The process is told ~22.4 GiB are free, and it can then allocate 23.75 GiB, about 1.3 GiB more.

Our reading of the 26.40.40196.8 sources:

• Device::getUsableMemorySize() (shared/source/device/device.cpp) returns getDeviceInfo().globalMemSize - usedMemorySize.
• deviceInfo.globalMemSize comes from Device::getGlobalMemorySize(), which multiplies the local memory size by getPercentOfGlobalMemoryAvailable(). DrmMemoryManager::getPercentOfGlobalMemoryAvailable() returns 0.95 when local memory is supported.
• usedMemorySize comes from DrmMemoryManager::getCurrentUsedLocalMemorySize(), i.e. the kernel's device-wide used figure.

So the 5% headroom of the OpenCL CL_DEVICE_GLOBAL_MEM_SIZE report is subtracted a second time from a "current usable memory" figure whose other term is the real, device-wide usage.

Expected Behavior

On an idle discrete GPU, currUsableMemSize should be close to the memory an application can actually allocate: here ~23.75 GiB, with 23.87 GiB free according to the kernel. That would be the case if the value were computed from the unscaled local memory size minus the used memory (see Additional Notes).

If the 5% headroom is intentional, it would help to document that currUsableMemSize is not the free memory and that allocations beyond it succeed.

The impact is not small on multi-GPU nodes: an application that sizes buffers from free_memory leaves ~1.4 GiB unused per 24 GB card, ~11 GiB on an 8-GPU node.

Reproduction Rate

Always reproduces - 100%

Steps to Reproduce

1. Build the reproducer below with oneAPI 2026.2: icpx -O2 -fsycl free_vs_fill.cpp -o free_vs_fill
2. Run it on one idle GPU: ONEAPI_DEVICE_SELECTOR=level_zero:0 ./free_vs_fill
3. Compare free_memory (first line) with the memory actually allocated and filled (second line).
4. Optionally, run the same binary with the oneAPI 2026.1.1 runtime libraries in LD_LIBRARY_PATH: free_memory is then ~1.2 GiB higher (Sysman path), while the allocated amount does not change.

The result is the same on both GPUs, on every run, with or without a desktop session (GNOME accounts for ~44 MiB).

Is this a regression?

• Yes, this is a regression - functionality that previously worked is now broken

Not for the driver: in 26.22.38646.4 Device::getUsableMemorySize() already used the scaled globalMemSize. What changed is oneAPI 2026.2, whose Unified Runtime started to use currUsableMemSize for free_memory: with the same 26.40 driver, oneAPI 2026.1.1 reports 23.62-23.67 GiB and 2026.2.0 reports 22.43-22.48 GiB.

Source Code / Reproducer

// free_vs_fill.cpp
#include <sycl/sycl.hpp>
#include <cstdio>
#include <vector>
int main() {
    sycl::device d{sycl::gpu_selector_v};
    sycl::queue q{d};
    const double G = 1024.0 * 1024 * 1024;
    std::printf("global_mem_size %.2f GiB, free_memory %.2f GiB\n",
                d.get_info<sycl::info::device::global_mem_size>() / G,
                d.get_info<sycl::ext::intel::info::device::free_memory>() / G);
    const size_t chunk = size_t(256) << 20;
    std::vector<void *> v;
    for (;;) {
        void *p = sycl::malloc_device(chunk, q);
        if (!p) break;
        q.memset(p, 0xA5, chunk).wait();
        v.push_back(p);
    }
    std::printf("allocated and filled %.2f GiB\n", v.size() * chunk / G);
    for (void *p : v) sycl::free(p, q);
}

Command Line / Application Details

$ icpx -O2 -fsycl free_vs_fill.cpp -o free_vs_fill
$ ONEAPI_DEVICE_SELECTOR=level_zero:1 ./free_vs_fill        # oneAPI 2026.2.0 runtime
global_mem_size 23.91 GiB, free_memory 22.48 GiB
allocated and filled 23.75 GiB
$ ONEAPI_DEVICE_SELECTOR=level_zero:1 ./free_vs_fill        # oneAPI 2026.1.1 runtime, same binary
global_mem_size 23.91 GiB, free_memory 23.67 GiB
allocated and filled 23.75 GiB

The kernel figure can be read without any runtime through DRM_IOCTL_XE_DEVICE_QUERY / DRM_XE_DEVICE_QUERY_MEM_REGIONS on the render node (total_size - used of the VRAM region): 23.87 GiB.

oneAPI Version (if applicable)

intel-oneapi-compiler-dpcpp-cpp-runtime-2026.2 2026.2.0-863  (Intel(R) oneAPI DPC++/C++ Compiler 2026.2.0 (2026.2.0.20261001))
intel-oneapi-compiler-dpcpp-cpp-runtime-2026.1 2026.1.1-325  (used only for the comparison)

Additional Notes

A possible fix, for discussion. In Device::getUsableMemorySize(), subtract the used memory from the unscaled local memory size instead of the 95%-scaled globalMemSize, e.g.:

const auto totalMemorySize = getMemoryManager()->isLocalMemorySupported(getRootDeviceIndex())
    ? getMemoryManager()->getLocalMemorySize(getRootDeviceIndex(), static_cast<uint32_t>(getDeviceBitfield().to_ulong()))
    : getDeviceInfo().globalMemSize;

This would leave CL_DEVICE_GLOBAL_MEM_SIZE, maxMemAllocSize and the integrated-GPU path unchanged. We have not built or tested this change, and we do not know whether the 5% is meant to cover driver-internal allocations, which is why we are not opening a pull request yet.

Related but different: #1047 reports allocations accepted beyond VRAM on B70 through deferred backing. Here the reported free memory is too low. In a single process on the B60 with 26.40 we saw no allocation beyond VRAM: allocation stops cleanly with nullptr at 23.75 GiB, and deferred backing switches on at 51% (PrintDeferBackingLogs=1).