From e7dfc5554d2d3f2c9b1c965c1a30553d4954e3f1 Mon Sep 17 00:00:00 2001 From: Chris Perkins Date: Mon, 8 Jul 2024 17:44:14 -0700 Subject: [PATCH 1/2] kernel_compiler include_file support shouldn't have files that might collide. --- .../kernel_compiler/kernel_compiler_sycl.cpp | 1 + .../KernelCompiler/kernel_compiler_sycl.cpp | 15 ++++++++++++--- 2 files changed, 13 insertions(+), 3 deletions(-) diff --git a/sycl/source/detail/kernel_compiler/kernel_compiler_sycl.cpp b/sycl/source/detail/kernel_compiler/kernel_compiler_sycl.cpp index 2967786673a5..00e6f6502df1 100644 --- a/sycl/source/detail/kernel_compiler/kernel_compiler_sycl.cpp +++ b/sycl/source/detail/kernel_compiler/kernel_compiler_sycl.cpp @@ -138,6 +138,7 @@ void outputIncludeFiles(const std::filesystem::path &Dirpath, using pairStrings = std::pair; for (pairStrings p : IncludePairs) { std::filesystem::path FilePath = Dirpath / p.first; + std::filesystem::create_directories(FilePath.parent_path()); std::ofstream outfile(FilePath, std::ios::out | std::ios::trunc); if (outfile.is_open()) { outfile << p.second << std::endl; diff --git a/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp b/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp index fc294f49fad4..6a9e06b452f3 100644 --- a/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp +++ b/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp @@ -21,10 +21,17 @@ auto constexpr AddEmH = R"===( } )==="; +auto constexpr PlusEmH = R"===( + int PlusEm(int a, int b){ + return a + b + 6; + } +)==="; + // TODO: remove SYCL_EXTERNAL once it is no longer needed. auto constexpr SYCLSource = R"===( #include -#include "AddEm.h" +#include "intermediate/AddEm.h" +#include "intermediate/PlusEm.h" // use extern "C" to avoid name mangling extern "C" SYCL_EXTERNAL SYCL_EXT_ONEAPI_FUNCTION_PROPERTY((sycl::ext::oneapi::experimental::nd_range_kernel<1>)) @@ -45,7 +52,7 @@ void ff_templated(T *ptr) { sycl::nd_item<1> Item = sycl::ext::oneapi::this_work_item::get_nd_item<1>(); sycl::id<1> GId = Item.get_global_id(); - ptr[GId.get(0)] = GId.get(0) + 39; + ptr[GId.get(0)] = PlusEm(GId.get(0), 38); } )==="; @@ -127,9 +134,11 @@ void test_build_and_run() { } // Create from source. + syclex::include_files incFiles{"intermediate/AddEm.h", AddEmH}; + incFiles.add("intermediate/PlusEm.h", PlusEmH); source_kb kbSrc = syclex::create_kernel_bundle_from_source( ctx, syclex::source_language::sycl, SYCLSource, - syclex::properties{syclex::include_files{"AddEm.h", AddEmH}}); + syclex::properties{incFiles}); // Double check kernel_bundle.get_source() / get_backend(). sycl::context ctxRes = kbSrc.get_context(); From 674008430cb668b539093889879ea4db076fb668 Mon Sep 17 00:00:00 2001 From: Chris Perkins Date: Wed, 10 Jul 2024 12:19:27 -0700 Subject: [PATCH 2/2] fix test issue --- sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp b/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp index 6a9e06b452f3..304a1322aa4e 100644 --- a/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp +++ b/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp @@ -107,8 +107,8 @@ void test_1(sycl::queue &Queue, sycl::kernel &Kernel, int seed) { Queue.wait(); for (int i = 0; i < Range; i++) { - std::cout << usmPtr[i] << " "; - assert(usmPtr[i] = i + seed); + std::cout << usmPtr[i] << "=" << (i + seed) << " "; + assert(usmPtr[i] == i + seed); } std::cout << std::endl; @@ -177,8 +177,8 @@ void test_build_and_run() { // clang-format on // Test the kernels. - test_1(q, k, 37 + 5); // AddEm will add 5 more. - test_1(q, k2, 39); + test_1(q, k, 37 + 5); // ff_cp seeds 37. AddEm will add 5 more. + test_1(q, k2, 38 + 6); // ff_templated seeds 38. PlusEm adds 6 more. } void test_error() {