From 566fe113c782522de30c2f98b89bf78c7c2fc85f Mon Sep 17 00:00:00 2001
From: =?UTF-8?q?Nicol=C3=B2=20Scipione?= <nicolo.scipione@codeplay.com>
Date: Thu, 6 Jul 2023 11:50:35 +0200
Subject: [PATCH] Fix compilation error with hipSYCL (#447)

This PR resolves a compilation error that occurs when targeting NVIDIA GPUs and using hipSYCL as the SYCL compiler. Additionally, it re-enables benchmarking with hipSYCL.

It's important to note that the fix for this compiler issue with all targets will be complete once PR #434 is merged. This is because PR #434 includes essential safeguards to prevent compilation failures, which are necessary due to a missing feature in hipSYCL.
---
 benchmark/syclblas/utils.hpp    |  4 ++--
 cmake/CmakeFunctionHelper.cmake | 14 ++++++++------
 cmake/Modules/SYCL.cmake        |  8 ++------
 3 files changed, 12 insertions(+), 14 deletions(-)

diff --git a/benchmark/syclblas/utils.hpp b/benchmark/syclblas/utils.hpp
index a73ae7574..0c326fefb 100644
--- a/benchmark/syclblas/utils.hpp
+++ b/benchmark/syclblas/utils.hpp
@@ -51,10 +51,10 @@ namespace utils {
 template <>
 inline double time_event<cl::sycl::event>(cl::sycl::event& e) {
   // get start and end times
-  cl_ulong start_time = e.template get_profiling_info<
+  auto start_time = e.template get_profiling_info<
       cl::sycl::info::event_profiling::command_start>();
 
-  cl_ulong end_time = e.template get_profiling_info<
+  auto end_time = e.template get_profiling_info<
       cl::sycl::info::event_profiling::command_end>();
 
   // return the delta
diff --git a/cmake/CmakeFunctionHelper.cmake b/cmake/CmakeFunctionHelper.cmake
index 59341efc8..0624b6d30 100644
--- a/cmake/CmakeFunctionHelper.cmake
+++ b/cmake/CmakeFunctionHelper.cmake
@@ -712,14 +712,16 @@ elseif(${TUNING_TARGET} STREQUAL "AMD_GPU")  # need investigation
       64 4 4 4 4 1 1 1 1 4 4 1 1 1 float float "no_local" "standard" "full" 4 "interleaved" "false")
   endforeach()
 elseif(${TUNING_TARGET} STREQUAL "NVIDIA_GPU")
- set(supported_types
+  set(supported_types
     "float"
     "double"
-  )
-  string(FIND ${DPCPP_SYCL_ARCH} "_" start_idx)
-  if(start_idx)
-    MATH(EXPR start_idx "${start_idx} + 1")
-    string(SUBSTRING ${DPCPP_SYCL_ARCH} ${start_idx} "2" sm_val)
+    )
+  if(is_dpcpp AND DEFINED DPCPP_SYCL_ARCH)
+    string(FIND ${DPCPP_SYCL_ARCH} "_" start_idx)
+    if(start_idx)
+      MATH(EXPR start_idx "${start_idx} + 1")
+      string(SUBSTRING ${DPCPP_SYCL_ARCH} ${start_idx} "2" sm_val)
+    endif()
   endif()
   foreach(data ${supported_types})
     # Joint Matrix specific GEMM configurations (only for float)
diff --git a/cmake/Modules/SYCL.cmake b/cmake/Modules/SYCL.cmake
index dd5d4306a..0d6fbbdb6 100644
--- a/cmake/Modules/SYCL.cmake
+++ b/cmake/Modules/SYCL.cmake
@@ -26,7 +26,7 @@ include(CheckCXXCompilerFlag)
 include(ConfigureSYCLBLAS)
 
 # find_package(hipSYCL) requires HIPSYCL_TARGETS to be set, so set it to a default value before find_package(hipSYCL)
-if(SYCL_COMPILER MATCHES "hipsycl" AND NOT HISPYCL_TARGETS AND NOT ENV{HIPSYCL_TARGETS})
+if(SYCL_COMPILER MATCHES "hipsycl" AND NOT HIPSYCL_TARGETS AND NOT ENV{HIPSYCL_TARGETS})
   message(STATUS "Using `omp` as HIPSYCL_TARGETS")
   set(HIPSYCL_TARGETS "omp")
 endif()
@@ -53,7 +53,7 @@ else()
       message(WARNING "Selected DPC++ as backend, but -fsycl not supported")
     endif()
   elseif(SYCL_COMPILER MATCHES "hipsycl")
-    find_package(hipSYCL CONFIG)
+    find_package(hipSYCL REQUIRED CONFIG)
     set(is_hipsycl ON)
   elseif(SYCL_COMPILER MATCHES "computecpp")
     set(is_computecpp ON)
@@ -92,8 +92,4 @@ elseif(is_hipsycl)
   set(CMAKE_CXX_STANDARD 17)
   set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -O3")
   get_target_property(SYCL_INCLUDE_DIRS hipSYCL::hipSYCL-rt INTERFACE_INCLUDE_DIRECTORIES)
-  if(NOT ${BLAS_ENABLE_BENCHMARK})
-    # hipSYCL currently does not support queue profiling. Thus disable benchmarks by default.
-    set(BLAS_ENABLE_BENCHMARK Off)
-  endif()
 endif()