[GSD-13259] Arc B390 (Panther Lake, 8086:b082) — Level Zero kernel dispatch fails with UR_RESULT_ERROR_UNSUPPORTED_FEATURE

[ApexDevelopment]

Environment

• CPU: Intel Core Ultra X7 358H (Panther Lake)
• iGPU PCI ID: 8086:b082 (Panther Lake [Arc B390], Xe3 architecture, rev 04)
• Board: Framework Computer Inc. laptop
• Host OS: Bazzite (Fedora Atomic/rpm-ostree based), kernel 7.1.5-ogc5.1.fc44.x86_64, xe kernel driver
• Test environment: Ubuntu 24.04.4 LTS in a rootless Podman/distrobox container sharing the host kernel and /dev/dri
• oneAPI DPC++ compiler: 2026.1.1 (2026.1.1.20260724)

Version matrix tested (both fail identically)

compute-runtime	libze-intel-gpu1	IGC	Result
26.27.39122.11	26.27.39122.11	2.38.2	Fails
26.18.38308.1	26.18.38308.1	2.34.4	Fails

The 26.18.38308.1 + IGC 2.34.4 pair is the exact combination reported as a working fix for the related Panther Lake Xe3 segfault in #918 (PCI ID 8086:b0a0). It does not fix this device.

clinfo -l / sycl-ls both enumerate the device correctly in every combination tested:

$ sycl-ls
[level_zero:gpu][level_zero:0] Intel(R) oneAPI Unified Runtime over Level-Zero V2, Intel(R) Arc(TM) B390 GPU 30.0.4 [1.15.39122+11]
[opencl:cpu][opencl:0] Intel(R) OpenCL, Intel(R) Core(TM) Ultra X7 358H OpenCL 3.0 (Build 0) [2026.21.7.0.24_160000]
[opencl:gpu][opencl:1] Intel(R) OpenCL Graphics, Intel(R) Arc(TM) B390 GPU OpenCL 3.0 NEO  [26.27.39122.11]

Problem

Enumeration and device query both work fine, but any Level Zero kernel dispatch fails with UR_RESULT_ERROR_UNSUPPORTED_FEATURE (error 44), even for a trivial kernel with no special requirements. This is not a segfault (unlike #918/#920) — it's a clean exception thrown from the Level Zero adapter at kernel submission.

Confirmed at three levels of isolation:

1. llama.cpp SYCL backend (upstream ggml-org/llama.cpp, built from source against the above stack) — fails loading any model on first RMS_NORM kernel dispatch.
2. Same failure with the RMS_NORM+MUL kernel fusion path disabled and with GGML_SYCL_WARP_SIZE forced to 32 instead of the Intel-target default of 16 — rules out a reqd_sub_group_size mismatch.
3. Minimal standalone reproduction with zero llama.cpp/ggml involvement — a bare parallel_for over a sycl::range<1>, no reqd_sub_group_size, no local memory, nothing beyond basic USM allocation and kernel submission:

#include <sycl/sycl.hpp>
#include <iostream>
int main() {
    sycl::queue q(sycl::gpu_selector_v);
    std::cout << "Device: " << q.get_device().get_info<sycl::info::device::name>() << std::endl;
    int *data = sycl::malloc_shared<int>(16, q);
    q.submit([&](sycl::handler &h) {
        h.parallel_for(sycl::range<1>(16), [=](sycl::id<1> i) {
            data[i] = i * 2;
        });
    }).wait();
    for (int i = 0; i < 16; i++) std::cout << data[i] << " ";
    std::cout << std::endl << "SUCCESS" << std::endl;
    sycl::free(data, q);
    return 0;
}

Built with icpx -fsycl test.cpp -o test, run with ONEAPI_DEVICE_SELECTOR=level_zero:0 ./test:

Device: Intel(R) Arc(TM) B390 GPU
terminate called after throwing an instance of 'sycl::_V1::exception'
  what():  level_zero backend failed with error: 44 (UR_RESULT_ERROR_UNSUPPORTED_FEATURE)
Aborted (core dumped)

Adding [[sycl::reqd_sub_group_size(16)]] or [[sycl::reqd_sub_group_size(32)]] to the kernel (both are advertised as supported — clinfo reports Sub-group sizes (Intel): 16, 32) makes no difference; the unattributed kernel above fails identically. So this does not appear to be a subgroup-size issue.

OpenCL-only paths (clinfo, device enumeration) are unaffected — the failure is specific to Level Zero kernel dispatch via the SYCL/UR adapter.

Relation to #918 / #920 / #891

• [GSD-12664] Panther Lake Xe3 iGPU (8086:b0a0) — Level Zero GPU driver segfaults on initialization #918 and [GSD-12691] Panther Lake Xe3 iGPU (8086:B03E) — Level Zero GPU driver segfaults on initialization #920 covered PCI ID 8086:b0a0 (segfault in libze_intel_gpu.so on Level Zero init, fixed by upgrading IGC 2.30.3 → 2.34.4).
• feature: brand string update for Arc B390/B370 on PTL #891 covered 8086:b082/8086:b0a3 (Arc B390/B370) but was only a device brand-string cosmetic fix, unrelated to compute.
• This is, as far as I can tell, an unreported issue specific to 8086:b082 (Arc B390): enumeration and init succeed, but kernel dispatch fails, and the IGC upgrade that fixed the sibling chip's issue in [GSD-12664] Panther Lake Xe3 iGPU (8086:b0a0) — Level Zero GPU driver segfaults on initialization #918 does not fix this one.

Request

Is 8086:b082 (Arc B390 on Panther Lake) missing something in the Level Zero UR adapter's supported-feature set that the sibling b0a0/b03e chips have, or is this a distinct regression? Happy to run further diagnostics (env var traces, ZE_ENABLE_VALIDATION_LAYER, etc.) — I can reproduce this on demand.

Diagnostics available on request

• lspci -vnn -s 00:02.0 output
• Full llama.cpp SYCL backtrace (crashes in ggml_sycl_op_rms_norm_fused / ggml_sycl_op_rms_norm)
• Build logs for the 26.18.38308.1 + IGC 2.34.4 downgrade test

[kgibala]

Hello @ApexDevelopment,

Thank you for your contribution.

Could you please follow the Issue Submission Guide and create an issue based on it? You can edit this one.

Thank you.