Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Jump to bottom

[SYCL] kernel_compiler include file paths collision fix #14490

Merged
merged 2 commits into from
Jul 16, 2024
Merged

[SYCL] kernel_compiler include file paths collision fix #14490

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
e7dfc55
kernel_compiler include_file support shouldn't have files that might …
cperkinsintel Jul 9, 2024
6740084
fix test issue
cperkinsintel Jul 10, 2024
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
1 change: 1 addition & 0 deletions sycl/source/detail/kernel_compiler/kernel_compiler_sycl.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -138,6 +138,7 @@ void outputIncludeFiles(const std::filesystem::path &Dirpath,
using pairStrings = std::pair<std::string, std::string>;
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;
Expand Down
23 changes: 16 additions & 7 deletions sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -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 <sycl/sycl.hpp>
#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>))
Expand All @@ -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);
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Why 39 replaced with 38?

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Thanks @dm-vodopyanov ! There was a mistake in the test assertions and your question made me realize it existed. Fixed.

}
)===";

Expand Down Expand Up @@ -100,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;

Expand All @@ -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();
Expand Down Expand Up @@ -168,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() {
Expand Down
Loading