Remove flush workaround. (#429)

Author: maleadt

deps/generate_interfaces.jl

@@ -23,7 +23,6 @@ version_types_header = Dict{Char, String}('S' => "float",
 comments = ["namespace", "#", "}", "/*", "*", "//", "[[", "ONEMKL_DECLARE_", "ONEMKL_INLINE_DECLARE"]
 
 void_output = ["init_matrix_handle", "init_matmat_descr", "release_matmat_descr", "set_matmat_data", "get_matmat_data"]
-no_force_flush = ["release_matrix_handle", "optimize_gemv", "optimize_trsv", "optimize_trmv", "set_csr_data"]
 
 function generate_headers(library::String, filename::Vector{String}, output::String; pattern::String="")
   routines = Dict{String,Int}()
@@ -426,7 +425,6 @@ function generate_cpp(library::String, filename::Vector{String}, output::String;
     if occursin("scratchpad_size", name)
       write(oneapi_cpp, "   return scratchpad_size;\n")
     else
-      !(name ∈ no_force_flush ∪ void_output) && write(oneapi_cpp, "   __FORCE_MKL_FLUSH__(status);\n")
       write(oneapi_cpp, "   return 0;\n")
     end
     write(oneapi_cpp, "}")

deps/onemkl_epilogue.cpp

@@ -1,6 +1,5 @@
 extern "C" int onemklXsparse_matmat(syclQueue_t device_queue, matrix_handle_t A, matrix_handle_t B, matrix_handle_t C, onemklMatmatRequest req, matmat_descr_t descr, int64_t *sizeTempBuffer, void *tempBuffer) {
    auto status = oneapi::mkl::sparse::matmat(device_queue->val, (oneapi::mkl::sparse::matrix_handle_t) A, (oneapi::mkl::sparse::matrix_handle_t) B, (oneapi::mkl::sparse::matrix_handle_t) C, convert(req), (oneapi::mkl::sparse::matmat_descr_t) descr, sizeTempBuffer, tempBuffer, {});
-   __FORCE_MKL_FLUSH__(status);
    return 0;
 }
 

deps/onemkl_prologue.cpp

@@ -5,12 +5,6 @@
 #include <memory>
 #include <oneapi/mkl.hpp>
 
-// This is a workaround to flush MKL submissions into Level-zero queue, using
-// unspecified but guaranteed behavior of intel-sycl runtime. Once SYCL standard
-// committee approves sycl::queue::flush() we will change the macro to use that
-#define __FORCE_MKL_FLUSH__(cmd) \
-            sycl::get_native<sycl::backend::ext_oneapi_level_zero>(cmd)
-
 oneapi::mkl::transpose convert(onemklTranspose val) {
     switch (val) {
     case ONEMKL_TRANSPOSE_NONTRANS:
@@ -392,7 +386,6 @@ extern "C" int onemklHgemm_batch(syclQueue_t device_queue, onemklTranspose trans
                         reinterpret_cast<const sycl::half **>(&b[0]), ldb,
                         reinterpret_cast<sycl::half *>(beta), reinterpret_cast<sycl::half **>(&c[0]),
                         ldc, group_count, group_size, {});
-    __FORCE_MKL_FLUSH__(status);
     return 0;
 }
 
@@ -410,7 +403,6 @@ extern "C" int onemklSgemm_batch(syclQueue_t device_queue, onemklTranspose trans
                         (const float **)&b[0], ldb,
                         beta, &c[0], ldc,
                         group_count, group_size, {});
-    __FORCE_MKL_FLUSH__(status);
     return 0;
 }
 
@@ -428,7 +420,6 @@ extern "C" int onemklDgemm_batch(syclQueue_t device_queue, onemklTranspose trans
                         (const double **)&b[0], ldb,
                         beta, &c[0], ldc,
                         group_count, group_size, {});
-    __FORCE_MKL_FLUSH__(status);
     return 0;
 }
 
@@ -450,7 +441,6 @@ extern "C" int onemklCgemm_batch(syclQueue_t device_queue, onemklTranspose trans
                         reinterpret_cast<std::complex<float> *>(beta),
                         reinterpret_cast<std::complex<float> **>(&c[0]), ldc,
                         group_count, group_size, {});
-    __FORCE_MKL_FLUSH__(status);
     return 0;
 }
 
@@ -473,7 +463,6 @@ extern "C" int onemklZgemm_batch(syclQueue_t device_queue, onemklTranspose trans
                         reinterpret_cast<std::complex<double> *>(beta),
                         reinterpret_cast<std::complex<double> **>(&c[0]), ldc,
                         group_count, group_size, {});
-    __FORCE_MKL_FLUSH__(status);
     return 0;
 }
 
@@ -490,7 +479,6 @@ extern "C" int onemklStrsm_batch(syclQueue_t device_queue, onemklSide left_right
                         &trsmInfo.m_transa[0], &trsmInfo.m_unitdiag[0],
                         m, n, alpha, (const float **)&a[0], lda,
                         &b[0], ldb, group_count, group_size, {});
-    __FORCE_MKL_FLUSH__(status);
     return 0;
 }
 
@@ -508,7 +496,6 @@ extern "C" int onemklDtrsm_batch(syclQueue_t device_queue, onemklSide left_right
                         &trsmInfo.m_transa[0], &trsmInfo.m_unitdiag[0],
                         m, n, alpha, (const double **)&a[0], lda, &b[0],
                         ldb, group_count, group_size, {});
-    __FORCE_MKL_FLUSH__(status);
     return 0;
 }
 
@@ -528,7 +515,6 @@ extern "C" int onemklCtrsm_batch(syclQueue_t device_queue, onemklSide left_right
                         reinterpret_cast<const std::complex<float> **>(&a[0]),
                         lda, reinterpret_cast<std::complex<float> **>(&b[0]),
                         ldb, group_count, group_size, {});
-    __FORCE_MKL_FLUSH__(status);
     return 0;
 }
 
@@ -548,6 +534,5 @@ extern "C" int onemklZtrsm_batch(syclQueue_t device_queue, onemklSide left_right
                         reinterpret_cast<const std::complex<double> **>(&a[0]),
                         lda, reinterpret_cast<std::complex<double> **>(&b[0]),
                         ldb, group_count, group_size, {});
-    __FORCE_MKL_FLUSH__(status);
     return 0;
 }