From 747cb0d5f6e70340e0a96f683a7f84bbb5959339 Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Wed, 10 Nov 2021 15:18:22 +0300 Subject: [PATCH 1/5] [SYCL] Add a test for freeing USM before kernel execution is finished --- SYCL/USM/free_during_kernel_execution.cpp | 41 +++++++++++++++++++++++ 1 file changed, 41 insertions(+) create mode 100644 SYCL/USM/free_during_kernel_execution.cpp diff --git a/SYCL/USM/free_during_kernel_execution.cpp b/SYCL/USM/free_during_kernel_execution.cpp new file mode 100644 index 000000000000..6fd079792ea0 --- /dev/null +++ b/SYCL/USM/free_during_kernel_execution.cpp @@ -0,0 +1,41 @@ +//==------------------- free_during_kernel_execution.cpp -------------------==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// +// +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t1.out +// RUN: %HOST_RUN_PLACEHOLDER %t1.out +// RUN: %CPU_RUN_PLACEHOLDER %t1.out +// RUN: %GPU_RUN_PLACEHOLDER %t1.out +// RUN: %ACC_RUN_PLACEHOLDER %t1.out + +#include + +class KernelA; + +int main() +{ + sycl::queue Queue; + sycl::buffer Buffer(64); + sycl::range<1> NumOfWorkItems{Buffer.size()}; + + auto *USM = sycl::malloc_host(1, Queue.get_context()); + + Queue.submit([&](sycl::handler &cgh) + { + auto Accessor = Buffer.get_access(cgh); + cgh.parallel_for(NumOfWorkItems, [=](sycl::id<1> WIid) + { + Accessor[WIid] = WIid.get(0); + }); + }); + + // Check that freeing USM before the kernel was finished works. + free(USM, Queue.get_context()); + + return 0; +} + From f2b5675eefc245fcb75d3ac0782bfa76cde97720 Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Mon, 29 Nov 2021 14:46:31 +0300 Subject: [PATCH 2/5] Update the test --- SYCL/USM/free_during_kernel_execution.cpp | 9 ++++++--- 1 file changed, 6 insertions(+), 3 deletions(-) diff --git a/SYCL/USM/free_during_kernel_execution.cpp b/SYCL/USM/free_during_kernel_execution.cpp index 6fd079792ea0..4b1f2553de03 100644 --- a/SYCL/USM/free_during_kernel_execution.cpp +++ b/SYCL/USM/free_during_kernel_execution.cpp @@ -18,8 +18,9 @@ class KernelA; int main() { + const int N = 128; sycl::queue Queue; - sycl::buffer Buffer(64); + sycl::buffer Buffer(N); sycl::range<1> NumOfWorkItems{Buffer.size()}; auto *USM = sycl::malloc_host(1, Queue.get_context()); @@ -27,9 +28,11 @@ int main() Queue.submit([&](sycl::handler &cgh) { auto Accessor = Buffer.get_access(cgh); - cgh.parallel_for(NumOfWorkItems, [=](sycl::id<1> WIid) + cgh.single_task([=]() { - Accessor[WIid] = WIid.get(0); + for (int I = 0; I < N; ++I) { + Accessor[I] = I; + } }); }); From ac77b62331639bbf0c739baec63ef2eeca4672ba Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Mon, 29 Nov 2021 14:50:13 +0300 Subject: [PATCH 3/5] Apply clang-format --- SYCL/USM/free_during_kernel_execution.cpp | 36 +++++++++++------------ 1 file changed, 17 insertions(+), 19 deletions(-) diff --git a/SYCL/USM/free_during_kernel_execution.cpp b/SYCL/USM/free_during_kernel_execution.cpp index 4b1f2553de03..b22a41b44ff8 100644 --- a/SYCL/USM/free_during_kernel_execution.cpp +++ b/SYCL/USM/free_during_kernel_execution.cpp @@ -18,27 +18,25 @@ class KernelA; int main() { - const int N = 128; - sycl::queue Queue; - sycl::buffer Buffer(N); - sycl::range<1> NumOfWorkItems{Buffer.size()}; - - auto *USM = sycl::malloc_host(1, Queue.get_context()); - - Queue.submit([&](sycl::handler &cgh) - { - auto Accessor = Buffer.get_access(cgh); - cgh.single_task([=]() - { - for (int I = 0; I < N; ++I) { - Accessor[I] = I; - } - }); + const int N = 128; + sycl::queue Queue; + sycl::buffer Buffer(N); + sycl::range<1> NumOfWorkItems{Buffer.size()}; + + auto *USM = sycl::malloc_host(1, Queue.get_context()); + + Queue.submit([&](sycl::handler &cgh) { + auto Accessor = Buffer.get_access(cgh); + cgh.single_task([=]() { + for (int I = 0; I < N; ++I) { + Accessor[I] = I; + } }); + }); - // Check that freeing USM before the kernel was finished works. - free(USM, Queue.get_context()); + // Check that freeing USM before the kernel was finished works. + free(USM, Queue.get_context()); - return 0; + return 0; } From 8e6c4e173866f146ce894947e612f9f8b895a723 Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Mon, 29 Nov 2021 14:58:48 +0300 Subject: [PATCH 4/5] More clang-format --- SYCL/USM/free_during_kernel_execution.cpp | 4 +--- 1 file changed, 1 insertion(+), 3 deletions(-) diff --git a/SYCL/USM/free_during_kernel_execution.cpp b/SYCL/USM/free_during_kernel_execution.cpp index b22a41b44ff8..e9fa9e3ea110 100644 --- a/SYCL/USM/free_during_kernel_execution.cpp +++ b/SYCL/USM/free_during_kernel_execution.cpp @@ -16,8 +16,7 @@ class KernelA; -int main() -{ +int main() { const int N = 128; sycl::queue Queue; sycl::buffer Buffer(N); @@ -39,4 +38,3 @@ int main() return 0; } - From 29e3a4ff0ae28ba8211d3d8e2eb34ea84fe8329c Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Mon, 29 Nov 2021 15:44:48 +0300 Subject: [PATCH 5/5] Tune test parameters to make it more reliable --- SYCL/USM/free_during_kernel_execution.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/SYCL/USM/free_during_kernel_execution.cpp b/SYCL/USM/free_during_kernel_execution.cpp index e9fa9e3ea110..ba68eaa2325a 100644 --- a/SYCL/USM/free_during_kernel_execution.cpp +++ b/SYCL/USM/free_during_kernel_execution.cpp @@ -17,7 +17,7 @@ class KernelA; int main() { - const int N = 128; + const int N = 512; sycl::queue Queue; sycl::buffer Buffer(N); sycl::range<1> NumOfWorkItems{Buffer.size()};