Hacktoberfest 2026: los issues que los mantenedores marcaron para octubre, abiertos y aptos para principiantes. Explorar issues de Hacktoberfest

oneMKL: transposed sparse gemv reads y when beta is zero

Abierto
  • ¿Algo no está bien?
#703 1 comentario 0 reacciones 0 asignados Ver en GitHub

Los mantenedores suelen responder en 1 día

Nadie ha tomado este issue todavía.

Evaluación

Dificultad
4/5
Tiempo estimado
1-2 días
Aptitud para principiantes
22/100
Tipo de issue
Error
Claridad
Bien especificado
Estado de actividad
Activo
Stack tecnológico
cpp
Área
backend

Línea de trabajo

The reproducer is the standalone repro.cpp in the issue, built with icpx -fsycl -qmkl=sequential and run on the GPU device; the calls under suspicion are sparse::gemv with trans and conjtrans and beta set to zero. The wrapper calls live in deps/src/onemkl.cpp, which is where the beta-zero workaround from #702 would sit. Done means the upstream oneMKL report is filed on the Intel forum and linked from this issue, since the fix belongs in oneMKL and not in oneAPI.jl.

Escrito por el modelo de indexación a partir del texto del issue.

Descripción

libraries upstream

On the GPU, oneapi::mkl::sparse::gemv with a CSR handle and transpose::trans or transpose::conjtrans reads y when beta is zero, so NaNs in an uninitialized output buffer end up in the result. The same call with transpose::nontrans ignores y, as do CSR gemm with any op and the CPU device. Prefilling y with 42 gives the correct result, so y is scaled by beta rather than skipped. BLAS convention is that y need not be set on input when beta is zero, and Julia's mul! relies on that: * passes uninitialized memory from similar.

In oneAPI.jl this made oneSparseMatrixCSC * x return NaNs depending on what the allocator handed out, because CSC matrices are stored as the CSR handle of their transpose and multiplied with a transposed op. That is the intermittent oneMKL.oneSparseMatrixCSC test failure in GPUArrays.jl's downstream CI. #702 works around it by zeroing the output before oneMKL calls that take a beta when beta is zero.

Reproducer

The calls mirror what our support library does (deps/src/onemkl.cpp): init_matrix_handle, set_csr_data with one-based indices, then gemv/gemm with USM pointers and no optimize_* call.

repro.cpp 
// oneMKL sparse::gemv on a CSR handle with op = trans/conjtrans reads y when beta == 0
#include <sycl/sycl.hpp>
#include <oneapi/mkl.hpp>
#include <mkl.h>
#include <algorithm>
#include <cmath>
#include <cstdint>
#include <cstdio>
#include <utility>

using oneapi::mkl::transpose;
namespace sparse = oneapi::mkl::sparse;

int main() {
    sycl::queue q{sycl::default_selector_v, sycl::property::queue::in_order()};
    char ver[256];
    mkl_get_version_string(ver, sizeof(ver));
    auto dev = q.get_device();
    printf("%s\n%s (driver %s)\n\n", ver, dev.get_info<sycl::info::device::name>().c_str(),
           dev.get_info<sycl::info::device::driver_version>().c_str());

    // 4x3 matrix in one-based CSR; row 2 and column 2 are empty
    //   [1 0 4]
    //   [0 0 0]
    //   [0 0 2]
    //   [3 0 0]
    const int64_t m = 4, n = 3, nnz = 4;
    int32_t *row_ptr = sycl::malloc_shared<int32_t>(m + 1, q);
    int32_t *col_ind = sycl::malloc_shared<int32_t>(nnz, q);
    float *val = sycl::malloc_shared<float>(nnz, q);
    const int32_t rp[] = {1, 3, 3, 4, 5}, ci[] = {1, 3, 3, 1};
    const float v[] = {1, 4, 2, 3};
    std::copy(rp, rp + m + 1, row_ptr);
    std::copy(ci, ci + nnz, col_ind);
    std::copy(v, v + nnz, val);

    sparse::matrix_handle_t A;
    sparse::init_matrix_handle(&A);
    sparse::set_csr_data(q, A, m, n, nnz, oneapi::mkl::index_base::one, row_ptr, col_ind, val).wait();

    // y = 1 * op(A) * x + 0 * y with x = ones and y prefilled; gemm uses 2 columns
    const float expect_N[] = {5, 0, 2, 3}, expect_T[] = {4, 0, 6};
    float *x = sycl::malloc_shared<float>(2 * m, q), *y = sycl::malloc_shared<float>(2 * m, q);
    const std::pair<transpose, const char *> ops[] = {
        {transpose::nontrans, "nontrans"}, {transpose::trans, "trans"}, {transpose::conjtrans, "conjtrans"}};
    for (int64_t ncols : {1, 2})
        for (auto [op, opname] : ops)
            for (float fill : {NAN, 42.0f}) {
                const bool N = op == transpose::nontrans;
                const int64_t nx = N ? n : m, ny = N ? m : n;
                std::fill(x, x + nx * ncols, 1.0f);
                std::fill(y, y + ny * ncols, fill);
                if (ncols == 1)
                    sparse::gemv(q, op, 1.0f, A, x, 0.0f, y).wait();
                else
                    sparse::gemm(q, oneapi::mkl::layout::col_major, op, transpose::nontrans, 1.0f, A,
                                 x, ncols, nx, 0.0f, y, ny).wait();
                bool ok = true;
                printf("%s %-9s y_in=%-3g y =", ncols == 1 ? "gemv" : "gemm", opname, fill);
                for (int64_t i = 0; i < ny * ncols; i++) {
                    printf(" %g", y[i]);
                    ok &= y[i] == (N ? expect_N : expect_T)[i % ny];
                }
                printf("  %s\n", ok ? "OK" : "WRONG");
            }

    sparse::release_matrix_handle(q, &A).wait();
    for (void *p : {(void *)row_ptr, (void *)col_ind, (void *)val, (void *)x, (void *)y})
        sycl::free(p, q);
}

icpx -fsycl -qmkl=sequential -O2 repro.cpp -o repro
ONEAPI_DEVICE_SELECTOR=level_zero:gpu ./repro

Intel(R) oneAPI Math Kernel Library Version 2026.1-Product Build 20260612 for Intel(R) 64 architecture applications
Intel(R) Iris(R) Xe Graphics (driver 1.15.38308)

gemv nontrans  y_in=nan y = 5 0 2 3  OK
gemv nontrans  y_in=42  y = 5 0 2 3  OK
gemv trans     y_in=nan y = nan nan nan  WRONG
gemv trans     y_in=42  y = 4 0 6  OK
gemv conjtrans y_in=nan y = nan nan nan  WRONG
gemv conjtrans y_in=42  y = 4 0 6  OK
gemm nontrans  y_in=nan y = 5 0 2 3 5 0 2 3  OK
gemm nontrans  y_in=42  y = 5 0 2 3 5 0 2 3  OK
gemm trans     y_in=nan y = 4 0 6 4 0 6  OK
gemm trans     y_in=42  y = 4 0 6 4 0 6  OK
gemm conjtrans y_in=nan y = 4 0 6 4 0 6  OK
gemm conjtrans y_in=42  y = 4 0 6 4 0 6  OK

With ONEAPI_DEVICE_SELECTOR=opencl:cpu every case is OK.

Variations:

  • Every output entry becomes NaN, not just the one for the empty column, and a matrix without empty rows or columns fails the same way.
  • Zero-based indices fail the same way.
  • Calling sparse::optimize_gemv(q, op, A) before each gemv makes all cases correct. Calling it only for nontrans and then running a trans gemv still fails.

On the older oneMKL of the Aurora LTS stack, sparse gemm with beta = 0 also propagated NaN, for every op (first CI run of #702). I haven't been able to reproduce that on 2026.1.

Versions
  • oneMKL 2026.1 (Product Build 20260612), the release oneAPI_Support_jll 0.11.0 is built from
  • Intel oneAPI DPC++/C++ Compiler 2026.1.0 (2026.1.0.20260617)
  • Intel Iris Xe Graphics (TigerLake-LP GT2), Level Zero driver 1.15.38308 (NEO 26.18.38308 from NEO_jll), Level Zero loader 1.29.0
  • Debian 13

Float64 is untested: this GPU has no FP64 support.

To do: report this to Intel on the oneMKL community forum, and link the report here.

Lenguaje dominante
Julia
Estrellas
217
Forks
37
Merge medio
1 d 7 h
PR fusionados (30 d)
45

Preparar el entorno

Este proyecto no incluye contenedor de desarrollo, Dockerfile ni guía de contribución, así que la configuración corre por tu cuenta: empieza por su README y consulta nuestra guía para la primera contribución para los pasos generales.

Primeros pasos

  1. Lee el issue completo y luego la guía de contribución del proyecto.
  2. Comenta en el issue que vas a ocuparte — evita que dos personas hagan lo mismo.
  3. Haz un fork del repositorio y trabaja en una rama.
  4. Abre un pull request que haga referencia al número del issue.

Más de JuliaGPU/oneAPI.jl

Todos los issues de JuliaGPU/oneAPI.jl

Issues similares

Más issues de Julia

Recibe los nuevos issues en tu correo

Un resumen breve de issues de GitHub para principiantes.