Skip to content

Commit 91a9796

Browse files
authored
[SYCL] Test for fix of linked alloca's deps (intel#1470)
The tests provided here are for the fix introduced with intel#1471. Use case fixed by the fix. Imagine us having two device queues (`Q1` and `Q2`) and a single buffer `B`. We initialize the buffer with host accessor. Also lets have two kernels: kernel `K1` executed at `Q1`; and kernel `K2` which executes at `Q2`. `K1` writes to `B`, `K2` reads from `B`. After submitting `K1` to `Q1` there will also be an `AllocaCommand` `A1` which allocates buffer on device. `K1` depends on `A1` via memory object `B`. After submitting `K2` to `Q2` there will be another `AllocaCommand` `A2` (for queue `Q2` and its device). `K2` will depend on `A2`. `A2`, however, should depend on both `A1` and `K1`. `A2->K1` dependency eliminates data race. Signed-off-by: Sergey Kanaev <[email protected]>
1 parent fa8953d commit 91a9796

File tree

4 files changed

+109
-0
lines changed

4 files changed

+109
-0
lines changed

sycl/source/detail/scheduler/scheduler.hpp

Lines changed: 5 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -166,6 +166,9 @@
166166
/// clReleaseContext(ContextCPU);
167167
/// \endcode
168168

169+
// For testing purposes
170+
class MockScheduler;
171+
169172
__SYCL_INLINE_NAMESPACE(cl) {
170173
namespace sycl {
171174
namespace detail {
@@ -576,6 +579,8 @@ class Scheduler {
576579
friend class Command;
577580

578581
private:
582+
friend class ::MockScheduler;
583+
579584
/// Searches for suitable alloca in memory record.
580585
///
581586
/// If none found, creates new one.

sycl/unittests/scheduler/CMakeLists.txt

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -6,5 +6,6 @@ add_sycl_unittest(SchedulerTests OBJECT
66
MemObjCommandCleanup.cpp
77
CommandsWaitForEvents.cpp
88
WaitAfterCleanup.cpp
9+
LinkedAllocaDependencies.cpp
910
utils.cpp
1011
)
Lines changed: 96 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,96 @@
1+
//==------ LinkedAllocaDependencies.cpp --- Scheduler unit tests -----------==//
2+
//
3+
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4+
// See https://llvm.org/LICENSE.txt for license information.
5+
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6+
//
7+
//===----------------------------------------------------------------------===//
8+
9+
#include "SchedulerTest.hpp"
10+
#include "SchedulerTestUtils.hpp"
11+
12+
using namespace cl::sycl;
13+
14+
class MemObjMock : public cl::sycl::detail::SYCLMemObjI {
15+
public:
16+
using ContextImplPtr = std::shared_ptr<cl::sycl::detail::context_impl>;
17+
18+
MemObjMock(const std::shared_ptr<cl::sycl::detail::MemObjRecord> &Record)
19+
: SYCLMemObjI() {
20+
MRecord = Record;
21+
}
22+
23+
~MemObjMock() = default;
24+
25+
MemObjType getType() const override { return MemObjType::BUFFER; }
26+
27+
void *allocateMem(ContextImplPtr, bool, void *,
28+
cl::sycl::detail::pi::PiEvent &) {
29+
return nullptr;
30+
}
31+
32+
void *allocateHostMem() { return nullptr; }
33+
void releaseMem(ContextImplPtr, void *) {}
34+
void releaseHostMem(void *) {}
35+
size_t getSize() const override { return 10; }
36+
};
37+
38+
TEST_F(SchedulerTest, LinkedAllocaDependencies) {
39+
default_selector Selector{};
40+
if (Selector.select_device().is_host()) {
41+
std::cerr << "Not run due to host-only environment\n";
42+
return;
43+
}
44+
45+
// 1. create two commands: alloca + alloca and link them
46+
// 2. call Scheduler::GraphBuilder::getOrCreateAllocaForReq
47+
detail::Requirement Req = getMockRequirement();
48+
49+
cl::sycl::queue Queue1;
50+
cl::sycl::detail::QueueImplPtr Q1 = cl::sycl::detail::getSyclObjImpl(Queue1);
51+
52+
sycl::device HostDevice;
53+
std::shared_ptr<detail::queue_impl> DefaultHostQueue(new detail::queue_impl(
54+
detail::getSyclObjImpl(HostDevice), /*AsyncHandler=*/{},
55+
/*PropList=*/{}));
56+
57+
std::shared_ptr<cl::sycl::detail::MemObjRecord> Record{
58+
new cl::sycl::detail::MemObjRecord(DefaultHostQueue->getContextImplPtr(),
59+
10)};
60+
61+
MemObjMock MemObj(Record);
62+
Req.MSYCLMemObj = &MemObj;
63+
64+
cl::sycl::detail::AllocaCommand AllocaCmd1(DefaultHostQueue, Req, false);
65+
Record->MAllocaCommands.push_back(&AllocaCmd1);
66+
67+
MockCommand DepCmd(DefaultHostQueue, Req);
68+
MockCommand DepDepCmd(DefaultHostQueue, Req);
69+
DepCmd.MDeps.push_back({&DepDepCmd, DepDepCmd.getRequirement(), &AllocaCmd1});
70+
DepDepCmd.MUsers.insert(&DepCmd);
71+
Record->MWriteLeaves.push_back(&DepCmd);
72+
73+
MockScheduler MS;
74+
cl::sycl::detail::Command *AllocaCmd2 =
75+
MS.getOrCreateAllocaForReq(Record.get(), &Req, Q1);
76+
77+
ASSERT_TRUE(!!AllocaCmd1.MLinkedAllocaCmd)
78+
<< "No link appeared in existing command";
79+
ASSERT_EQ(AllocaCmd1.MLinkedAllocaCmd, AllocaCmd2) << "Invalid link appeared";
80+
ASSERT_GT(AllocaCmd1.MUsers.count(AllocaCmd2), 0u)
81+
<< "New alloca isn't in users of the old one";
82+
ASSERT_GT(AllocaCmd2->MDeps.size(), 1u)
83+
<< "No deps appeared in the new alloca";
84+
ASSERT_GT(DepCmd.MUsers.count(AllocaCmd2), 0u)
85+
<< "No deps appeared for leaves of record (i.e. deps of existing alloca)";
86+
ASSERT_TRUE(std::find_if(AllocaCmd2->MDeps.begin(), AllocaCmd2->MDeps.end(),
87+
[&](const cl::sycl::detail::DepDesc &Dep) -> bool {
88+
return Dep.MDepCommand == &AllocaCmd1;
89+
}) != AllocaCmd2->MDeps.end())
90+
<< "No deps for existing alloca appeared in new alloca";
91+
ASSERT_TRUE(std::find_if(AllocaCmd2->MDeps.begin(), AllocaCmd2->MDeps.end(),
92+
[&](const cl::sycl::detail::DepDesc &Dep) -> bool {
93+
return Dep.MDepCommand == &DepCmd;
94+
}) != AllocaCmd2->MDeps.end())
95+
<< "No deps for leaves (deps of existing alloca) appeared in new alloca";
96+
}

sycl/unittests/scheduler/SchedulerTestUtils.hpp

Lines changed: 7 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -91,6 +91,13 @@ class MockScheduler : public cl::sycl::detail::Scheduler {
9191
cl::sycl::detail::BlockingT Blocking) {
9292
return GraphProcessor::enqueueCommand(Cmd, EnqueueResult, Blocking);
9393
}
94+
95+
cl::sycl::detail::AllocaCommandBase *
96+
getOrCreateAllocaForReq(cl::sycl::detail::MemObjRecord *Record,
97+
const cl::sycl::detail::Requirement *Req,
98+
cl::sycl::detail::QueueImplPtr Queue) {
99+
return MGraphBuilder.getOrCreateAllocaForReq(Record, Req, Queue);
100+
}
94101
};
95102

96103
void addEdge(cl::sycl::detail::Command *User, cl::sycl::detail::Command *Dep,

0 commit comments

Comments
 (0)