Skip to content

[SYCL] Improve testing of host-task #2540

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

Merged
merged 6 commits into from
Sep 28, 2020
Merged

[SYCL] Improve testing of host-task #2540

Changes from all commits
Commits
Show all changes
6 commits
Select commit Hold shift + click to select a range
97e690c
[SYCL] Improve testing of host-task
Sep 25, 2020
14e9517
[SYCL] Disable test-case as it hangs now
Sep 25, 2020
299797b
Address style issues
Sep 25, 2020
e7797f4
[SYCL] Improve testing of host-task
Sep 25, 2020
f0fce8f
[SYCL] Remove unwanted include
Sep 25, 2020
c94a245
[SYCL] Use default C++ std
Sep 25, 2020
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
195 changes: 186 additions & 9 deletions sycl/test/host-interop-task/host-task.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,34 +1,58 @@
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out %threads_lib -lOpenCL
// RUN: %CPU_RUN_PLACEHOLDER %t.out
// RUN: %GPU_RUN_PLACEHOLDER %t.out
// RUN: %ACC_RUN_PLACEHOLDER %t.out
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
// RUN: %CPU_RUN_PLACEHOLDER %t.out 1
// RUN: %GPU_RUN_PLACEHOLDER %t.out 1
// RUN: %ACC_RUN_PLACEHOLDER %t.out 1

// RUN: %CPU_RUN_PLACEHOLDER %t.out 2
// RUN: %GPU_RUN_PLACEHOLDER %t.out 2
// RUN: %ACC_RUN_PLACEHOLDER %t.out 2

// RUN: %CPU_RUN_PLACEHOLDER %t.out 3
// RUN: %GPU_RUN_PLACEHOLDER %t.out 3
// RUN: %ACC_RUN_PLACEHOLDER %t.out 3

// RUNx: %CPU_RUN_PLACEHOLDER %t.out 4
// RUNx: %GPU_RUN_PLACEHOLDER %t.out 4
// RUNx: %ACC_RUN_PLACEHOLDER %t.out 4
Comment on lines +14 to +16
Copy link
Contributor Author

Choose a reason for hiding this comment

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

These are going to be enabled after the issue is fixed.


#include <CL/sycl.hpp>
#include <chrono>
#include <iostream>
#include <vector>

using namespace cl::sycl;
using namespace cl::sycl::access;

static constexpr size_t BUFFER_SIZE = 1024;

static auto EH = [](exception_list EL) {
for (const std::exception_ptr &E : EL) {
throw E;
}
};

// Check that a single host-task with a buffer will work
void test1() {
buffer<int, 1> Buffer{BUFFER_SIZE};

queue Q;
queue Q(EH);

Q.submit([&](handler &CGH) {
auto Acc = Buffer.get_access<mode::write>(CGH);
CGH.codeplay_host_task([=] {
// A no-op
});
});

Q.wait_and_throw();
}

// Check that a host task after the kernel (deps via buffer) will work
void test2() {
buffer<int, 1> Buffer1{BUFFER_SIZE};
buffer<int, 1> Buffer2{BUFFER_SIZE};

queue Q;
queue Q(EH);

Q.submit([&](handler &CGH) {
auto Acc = Buffer1.template get_access<mode::write>(CGH);
Expand All @@ -55,10 +79,163 @@ void test2() {
assert(Acc[Idx] == 123);
}
}

Q.wait_and_throw();
}

// Host-task depending on another host-task via both buffers and
// handler::depends_on() should not hang
void test3() {
queue Q(EH);

static constexpr size_t BufferSize = 10 * 1024;

buffer<int, 1> B0{range<1>{BufferSize}};
buffer<int, 1> B1{range<1>{BufferSize}};
buffer<int, 1> B2{range<1>{BufferSize}};
buffer<int, 1> B3{range<1>{BufferSize}};
buffer<int, 1> B4{range<1>{BufferSize}};
buffer<int, 1> B5{range<1>{BufferSize}};
buffer<int, 1> B6{range<1>{BufferSize}};
buffer<int, 1> B7{range<1>{BufferSize}};
buffer<int, 1> B8{range<1>{BufferSize}};
buffer<int, 1> B9{range<1>{BufferSize}};

std::vector<event> Deps;

using namespace std::chrono_literals;
static constexpr size_t Count = 10;

auto Start = std::chrono::steady_clock::now();
for (size_t Idx = 0; Idx < Count; ++Idx) {
event E = Q.submit([&](handler &CGH) {
CGH.depends_on(Deps);

std::cout << "Submit: " << Idx << std::endl;

auto Acc0 = B0.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc1 = B1.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc2 = B2.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc3 = B3.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc4 = B4.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc5 = B5.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc6 = B6.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc7 = B7.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc8 = B8.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc9 = B9.get_access<mode::read_write, target::host_buffer>(CGH);

CGH.codeplay_host_task([=] {
uint64_t X = 0;

X ^= reinterpret_cast<uint64_t>(&Acc0[Idx + 0]);
X ^= reinterpret_cast<uint64_t>(&Acc1[Idx + 1]);
X ^= reinterpret_cast<uint64_t>(&Acc2[Idx + 2]);
X ^= reinterpret_cast<uint64_t>(&Acc3[Idx + 3]);
X ^= reinterpret_cast<uint64_t>(&Acc4[Idx + 4]);
X ^= reinterpret_cast<uint64_t>(&Acc5[Idx + 5]);
X ^= reinterpret_cast<uint64_t>(&Acc6[Idx + 6]);
X ^= reinterpret_cast<uint64_t>(&Acc7[Idx + 7]);
X ^= reinterpret_cast<uint64_t>(&Acc8[Idx + 8]);
X ^= reinterpret_cast<uint64_t>(&Acc9[Idx + 9]);
});
});

Deps = {E};
}

Q.wait_and_throw();
auto End = std::chrono::steady_clock::now();

constexpr auto Threshold = 2s;

assert(End - Start < Threshold && "Host tasks were waiting for too long");
}

// Host-task depending on another host-task via handler::depends_on() only
// should not hang
void test4() {
queue Q(EH);

static constexpr size_t BufferSize = 10 * 1024;

buffer<int, 1> B0{range<1>{BufferSize}};
buffer<int, 1> B1{range<1>{BufferSize}};
buffer<int, 1> B2{range<1>{BufferSize}};
buffer<int, 1> B3{range<1>{BufferSize}};
buffer<int, 1> B4{range<1>{BufferSize}};
buffer<int, 1> B5{range<1>{BufferSize}};

// This host task should be submitted without hesitation
event E1 = Q.submit([&](handler &CGH) {
std::cout << "Submit 1" << std::endl;

auto Acc0 = B0.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc1 = B1.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc2 = B2.get_access<mode::read_write, target::host_buffer>(CGH);

CGH.codeplay_host_task([=] {
Acc0[0] = 1;
Acc1[0] = 2;
Acc2[0] = 3;
});
});

// This host task is going to depend on blocked empty node of the first
// host-task (via buffer #2). Still this one should be enqueued.
event E2 = Q.submit([&](handler &CGH) {
std::cout << "Submit 2" << std::endl;

auto Acc2 = B2.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc3 = B3.get_access<mode::read_write, target::host_buffer>(CGH);

CGH.codeplay_host_task([=] {
Acc2[1] = 1;
Acc3[1] = 2;
});
});

// This host-task only depends on the second host-task via
// handler::depends_on(). This one should not hang and should be enqueued
// after host-task #2.
event E3 = Q.submit([&](handler &CGH) {
CGH.depends_on(E2);

std::cout << "Submit 3" << std::endl;

auto Acc4 = B4.get_access<mode::read_write, target::host_buffer>(CGH);
auto Acc5 = B5.get_access<mode::read_write, target::host_buffer>(CGH);

CGH.codeplay_host_task([=] {
Acc4[2] = 1;
Acc5[2] = 2;
});
});

Q.wait_and_throw();
}

int main() {
test1();
test2();
int main(int Argc, const char *Argv[]) {
if (Argc < 2)
return 1;

int TestIdx = std::stoi(Argv[1]);

switch (TestIdx) {
case 1:
test1();
break;
case 2:
test2();
break;
case 3:
test3();
break;
case 4:
test4();
break;
default:
return 1;
}

return 0;
}