From 18ed40b393fa1949c0343cd125a42b0d3f9b2a3a Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Thu, 10 Feb 2022 15:57:48 -0800 Subject: [PATCH 1/5] Honor property::queue::enable_profiling Signed-off-by: Sergey V Maslov --- SYCL/Plugin/level_zero_queue_profiling.cpp | 82 ++++++++++++++++++++++ 1 file changed, 82 insertions(+) create mode 100755 SYCL/Plugin/level_zero_queue_profiling.cpp diff --git a/SYCL/Plugin/level_zero_queue_profiling.cpp b/SYCL/Plugin/level_zero_queue_profiling.cpp new file mode 100755 index 0000000000..c50206b05d --- /dev/null +++ b/SYCL/Plugin/level_zero_queue_profiling.cpp @@ -0,0 +1,82 @@ +// REQUIRES: gpu, level_zero + +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out +// RUN: env ZE_DEBUG=-1 %GPU_RUN_PLACEHOLDER %t.out 2>&1 | FileCheck --check-prefixes=WITHOUT %s +// RUN: env ZE_DEBUG=-1 %GPU_RUN_PLACEHOLDER %t.out profile 2>&1 | FileCheck --check-prefixes=WITH %s + +// Test case adapted from the SYCL version of Rodinia benchmark hotspot. + +// Check the expected output when queue::enable_profiling is not specified +// +// WITHOUT: ze_event_pool_desc_t flags set to: 1 +// WITHOUT: terminate called after throwing an instance of 'cl::sycl::runtime_error' +// WITHOUT: what(): Native API failed. Native API returns: -7 (CL_PROFILING_INFO_NOT_AVAILABLE) + +// Check the expected output when queue::enable_profiling is specified +// +// WITH: ze_event_pool_desc_t flags set to: 5 +// WITH: Kernel time: +// WITH: Device kernel time: +// WITH: Device offloading time: + +#include +using namespace cl::sycl; +#include + +// wrapper for gettimeofday +long long get_time() { + struct timeval tv; + gettimeofday(&tv, NULL); + return (tv.tv_sec * 1000000) + tv.tv_usec; +} + +int foo(queue &q, int n) { + for (int i = 0; i < n; i++) { + + long long start_time = get_time(); + sycl::event queue_event = q.submit([&](handler &cgh) { + cgh.parallel_for(range<2>(10000, 10000), + [=](item<2> item) { ; }); + }); + + q.wait(); + + // Get kernel computation time + // queue_event.wait(); + long long end_time = get_time(); + long long total_time = (end_time - start_time); + printf("\nKernel time: %.3f seconds\n", + ((float)total_time) / (1000 * 1000)); + auto startk = queue_event.template get_profiling_info< + cl::sycl::info::event_profiling::command_start>(); + auto endk = queue_event.template get_profiling_info< + cl::sycl::info::event_profiling::command_end>(); + auto kernel_time = + (float)(endk - startk) * 1e-9f; // to seconds, 1e-6f to milliseconds + printf("Device kernel time: %.12fs\n", (float)kernel_time); + } + return n; +} + +int main(int argc, char **argv) { + + bool profiling = argc > 1; + + long long start_time = get_time(); + { + gpu_selector dev_sel; + property_list propList{}; + if (profiling) + propList = cl::sycl::property::queue::enable_profiling(); + + queue q(dev_sel, propList); + // Perform the computation + foo(q, 1); + } // SYCL scope + + long long end_time = get_time(); + printf("Device offloading time: %.3f seconds\n", + ((float)(end_time - start_time)) / (1000 * 1000)); + + return 0; +} From 32c044c2e3ac1b54a88c70dcf3c39ee328f9abce Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Fri, 11 Feb 2022 07:19:39 -0800 Subject: [PATCH 2/5] update per review Signed-off-by: Sergey V Maslov --- SYCL/Plugin/level_zero_queue_profiling.cpp | 30 ++++------------------ 1 file changed, 5 insertions(+), 25 deletions(-) mode change 100755 => 100644 SYCL/Plugin/level_zero_queue_profiling.cpp diff --git a/SYCL/Plugin/level_zero_queue_profiling.cpp b/SYCL/Plugin/level_zero_queue_profiling.cpp old mode 100755 new mode 100644 index c50206b05d..47d126caa9 --- a/SYCL/Plugin/level_zero_queue_profiling.cpp +++ b/SYCL/Plugin/level_zero_queue_profiling.cpp @@ -9,44 +9,29 @@ // Check the expected output when queue::enable_profiling is not specified // // WITHOUT: ze_event_pool_desc_t flags set to: 1 -// WITHOUT: terminate called after throwing an instance of 'cl::sycl::runtime_error' -// WITHOUT: what(): Native API failed. Native API returns: -7 (CL_PROFILING_INFO_NOT_AVAILABLE) +// WITHOUT: terminate called after throwing an instance of +// 'cl::sycl::runtime_error' WITHOUT: what(): Native API failed. Native API +// returns: -7 (CL_PROFILING_INFO_NOT_AVAILABLE) // Check the expected output when queue::enable_profiling is specified // // WITH: ze_event_pool_desc_t flags set to: 5 -// WITH: Kernel time: // WITH: Device kernel time: -// WITH: Device offloading time: #include using namespace cl::sycl; -#include - -// wrapper for gettimeofday -long long get_time() { - struct timeval tv; - gettimeofday(&tv, NULL); - return (tv.tv_sec * 1000000) + tv.tv_usec; -} int foo(queue &q, int n) { for (int i = 0; i < n; i++) { - long long start_time = get_time(); sycl::event queue_event = q.submit([&](handler &cgh) { cgh.parallel_for(range<2>(10000, 10000), - [=](item<2> item) { ; }); + [=](item<2> item) {}); }); q.wait(); // Get kernel computation time - // queue_event.wait(); - long long end_time = get_time(); - long long total_time = (end_time - start_time); - printf("\nKernel time: %.3f seconds\n", - ((float)total_time) / (1000 * 1000)); auto startk = queue_event.template get_profiling_info< cl::sycl::info::event_profiling::command_start>(); auto endk = queue_event.template get_profiling_info< @@ -62,7 +47,6 @@ int main(int argc, char **argv) { bool profiling = argc > 1; - long long start_time = get_time(); { gpu_selector dev_sel; property_list propList{}; @@ -71,12 +55,8 @@ int main(int argc, char **argv) { queue q(dev_sel, propList); // Perform the computation - foo(q, 1); + foo(q, 2); } // SYCL scope - long long end_time = get_time(); - printf("Device offloading time: %.3f seconds\n", - ((float)(end_time - start_time)) / (1000 * 1000)); - return 0; } From 2e902b013af5f11b3ee0d30361a758bfd9bc214a Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Fri, 11 Feb 2022 10:56:43 -0800 Subject: [PATCH 3/5] disable clang-format Signed-off-by: Sergey V Maslov --- SYCL/Plugin/level_zero_queue_profiling.cpp | 7 ++++--- 1 file changed, 4 insertions(+), 3 deletions(-) diff --git a/SYCL/Plugin/level_zero_queue_profiling.cpp b/SYCL/Plugin/level_zero_queue_profiling.cpp index 47d126caa9..e9a5a70245 100644 --- a/SYCL/Plugin/level_zero_queue_profiling.cpp +++ b/SYCL/Plugin/level_zero_queue_profiling.cpp @@ -6,17 +6,18 @@ // Test case adapted from the SYCL version of Rodinia benchmark hotspot. +// clang-format off // Check the expected output when queue::enable_profiling is not specified // // WITHOUT: ze_event_pool_desc_t flags set to: 1 -// WITHOUT: terminate called after throwing an instance of -// 'cl::sycl::runtime_error' WITHOUT: what(): Native API failed. Native API -// returns: -7 (CL_PROFILING_INFO_NOT_AVAILABLE) +// WITHOUT: terminate called after throwing an instance of 'cl::sycl::runtime_error' +// WITHOUT: what(): Native API failed. Native API returns: -7 (CL_PROFILING_INFO_NOT_AVAILABLE) // Check the expected output when queue::enable_profiling is specified // // WITH: ze_event_pool_desc_t flags set to: 5 // WITH: Device kernel time: +// clang-format on #include using namespace cl::sycl; From 0e659442bb398b2af1f223dd49c199b60f67cb01 Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Fri, 11 Feb 2022 12:18:46 -0800 Subject: [PATCH 4/5] restart testing Signed-off-by: Sergey V Maslov --- SYCL/Plugin/level_zero_queue_profiling.cpp | 1 + 1 file changed, 1 insertion(+) diff --git a/SYCL/Plugin/level_zero_queue_profiling.cpp b/SYCL/Plugin/level_zero_queue_profiling.cpp index e9a5a70245..244feea467 100644 --- a/SYCL/Plugin/level_zero_queue_profiling.cpp +++ b/SYCL/Plugin/level_zero_queue_profiling.cpp @@ -18,6 +18,7 @@ // WITH: ze_event_pool_desc_t flags set to: 5 // WITH: Device kernel time: // clang-format on +// #include using namespace cl::sycl; From 88bb989ee278609e8db0e372568dbf240ea35404 Mon Sep 17 00:00:00 2001 From: Sergey V Maslov Date: Thu, 17 Feb 2022 10:27:00 -0800 Subject: [PATCH 5/5] fix reported regressions Signed-off-by: Sergey V Maslov --- SYCL/Plugin/level_zero_queue_profiling.cpp | 23 +++++++++++++--------- 1 file changed, 14 insertions(+), 9 deletions(-) diff --git a/SYCL/Plugin/level_zero_queue_profiling.cpp b/SYCL/Plugin/level_zero_queue_profiling.cpp index 244feea467..70600fc7b2 100644 --- a/SYCL/Plugin/level_zero_queue_profiling.cpp +++ b/SYCL/Plugin/level_zero_queue_profiling.cpp @@ -10,8 +10,7 @@ // Check the expected output when queue::enable_profiling is not specified // // WITHOUT: ze_event_pool_desc_t flags set to: 1 -// WITHOUT: terminate called after throwing an instance of 'cl::sycl::runtime_error' -// WITHOUT: what(): Native API failed. Native API returns: -7 (CL_PROFILING_INFO_NOT_AVAILABLE) +// WITHOUT: SYCL exception caught: Native API failed. Native API returns: -7 (CL_PROFILING_INFO_NOT_AVAILABLE) // Check the expected output when queue::enable_profiling is specified // @@ -34,13 +33,19 @@ int foo(queue &q, int n) { q.wait(); // Get kernel computation time - auto startk = queue_event.template get_profiling_info< - cl::sycl::info::event_profiling::command_start>(); - auto endk = queue_event.template get_profiling_info< - cl::sycl::info::event_profiling::command_end>(); - auto kernel_time = - (float)(endk - startk) * 1e-9f; // to seconds, 1e-6f to milliseconds - printf("Device kernel time: %.12fs\n", (float)kernel_time); + try { + auto startk = queue_event.template get_profiling_info< + cl::sycl::info::event_profiling::command_start>(); + auto endk = queue_event.template get_profiling_info< + cl::sycl::info::event_profiling::command_end>(); + auto kernel_time = + (float)(endk - startk) * 1e-9f; // to seconds, 1e-6f to milliseconds + printf("Device kernel time: %.12fs\n", (float)kernel_time); + + } catch (const sycl::exception &e) { + std::cout << "SYCL exception caught: " << e.what() << '\n'; + return 0; + } } return n; }