diff --git a/sycl/source/detail/queue_impl.hpp b/sycl/source/detail/queue_impl.hpp index a2d24846ba555..1d4d66a2b8b20 100644 --- a/sycl/source/detail/queue_impl.hpp +++ b/sycl/source/detail/queue_impl.hpp @@ -439,27 +439,41 @@ class queue_impl { return MAssertHappenedBuffer; } -private: - void finalizeHandler(handler &Handler, const CG::CGTYPE &Type, +protected: + // template is needed for proper unit testing + template + void finalizeHandler(HandlerType &Handler, const CG::CGTYPE &Type, event &EventRet) { if (MIsInorder) { - bool NeedSeparateDependencyMgmt = - (Type == CG::CGTYPE::CodeplayHostTask || - Type == CG::CGTYPE::CodeplayInteropTask); + + auto IsExpDepManaged = [](const CG::CGTYPE &Type) { + return (Type == CG::CGTYPE::CodeplayHostTask || + Type == CG::CGTYPE::CodeplayInteropTask); + }; + // Accessing and changing of an event isn't atomic operation. // Hence, here is the lock for thread-safety. std::lock_guard Lock{MLastEventMtx}; + if (MLastCGType == CG::CGTYPE::None) + MLastCGType = Type; + // Also handles case when sync model changes. E.g. Last is host, new is + // kernel. + bool NeedSeparateDependencyMgmt = + IsExpDepManaged(Type) || IsExpDepManaged(MLastCGType); + if (NeedSeparateDependencyMgmt) Handler.depends_on(MLastEvent); EventRet = Handler.finalize(); MLastEvent = EventRet; + MLastCGType = Type; } else EventRet = Handler.finalize(); } +private: /// Performs command group submission to the queue. /// /// \param CGF is a function object containing command group. @@ -560,6 +574,10 @@ class queue_impl { // Access to the event should be guarded with MLastEventMtx event MLastEvent; std::mutex MLastEventMtx; + // Used for in-order queues in pair with MLastEvent + // Host tasks are explicitly synchronized in RT, pi tasks - implicitly by + // backend. Using type to setup explicit sync between host and pi tasks. + CG::CGTYPE MLastCGType = CG::CGTYPE::None; const bool MIsInorder; diff --git a/sycl/unittests/scheduler/CMakeLists.txt b/sycl/unittests/scheduler/CMakeLists.txt index 759df4076d13d..98518b2229274 100644 --- a/sycl/unittests/scheduler/CMakeLists.txt +++ b/sycl/unittests/scheduler/CMakeLists.txt @@ -20,4 +20,5 @@ add_sycl_unittest(SchedulerTests OBJECT Regression.cpp utils.cpp LeafLimitDiffContexts.cpp + InOrderQueueSyncCheck.cpp ) diff --git a/sycl/unittests/scheduler/InOrderQueueSyncCheck.cpp b/sycl/unittests/scheduler/InOrderQueueSyncCheck.cpp new file mode 100644 index 0000000000000..9d75512445bd4 --- /dev/null +++ b/sycl/unittests/scheduler/InOrderQueueSyncCheck.cpp @@ -0,0 +1,106 @@ +//==---------- InOrderQueueSyncCheck.cpp --- Scheduler unit tests ----------==// +// +// 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 +// +//===----------------------------------------------------------------------===// + +#include "SchedulerTest.hpp" +#include "SchedulerTestUtils.hpp" +#include +#include +#include + +#include + +using namespace sycl; + +// Define type with the only methods called by finalizeHandler +class LimitedHandler { +public: + virtual void depends_on(sycl::event){}; + + virtual event finalize() { + cl::sycl::detail::EventImplPtr NewEvent = + std::make_shared(); + return sycl::detail::createSyclObjFromImpl(NewEvent); + }; +}; + +// Needed to use EXPECT_CALL to verify depends_on that originally appends lst +// event as dependency to the new CG +class LimitedHandlerSimulation : public LimitedHandler { +public: + MOCK_METHOD1(depends_on, void(sycl::event)); +}; + +class MockQueueImpl : public sycl::detail::queue_impl { +public: + MockQueueImpl(const sycl::detail::DeviceImplPtr &Device, + const sycl::async_handler &AsyncHandler, + const sycl::property_list &PropList) + : sycl::detail::queue_impl(Device, AsyncHandler, PropList) {} + using sycl::detail::queue_impl::finalizeHandler; +}; + +// Only check events dependency in queue_impl::finalizeHandler +TEST_F(SchedulerTest, InOrderQueueSyncCheck) { + sycl::platform Plt{sycl::default_selector()}; + if (Plt.is_host() || Plt.get_backend() == sycl::backend::ext_oneapi_cuda || + Plt.get_backend() == sycl::backend::ext_oneapi_hip) { + std::cerr << "Test is not supported on " + << Plt.get_info() << ", skipping\n"; + GTEST_SKIP(); // test is not supported on selected platform. + } + + const sycl::device Dev = Plt.get_devices()[0]; + auto Queue = std::make_shared( + sycl::detail::getSyclObjImpl(Dev), sycl::async_handler{}, + sycl::property::queue::in_order()); + + // What we are testing here: + // Task type | Must depend on + // host | yes - always, separate sync management + // host | yes - always, separate sync management + // kernel | yes - change of sync approach + // kernel | no - sync between pi calls must be done by backend + // host | yes - always, separate sync management + + sycl::event Event; + // host task + { + LimitedHandlerSimulation MockCGH; + EXPECT_CALL(MockCGH, depends_on).Times(1); + Queue->finalizeHandler( + MockCGH, detail::CG::CGTYPE::CodeplayHostTask, Event); + } + // host task + { + LimitedHandlerSimulation MockCGH; + EXPECT_CALL(MockCGH, depends_on).Times(1); + Queue->finalizeHandler( + MockCGH, detail::CG::CGTYPE::CodeplayHostTask, Event); + } + // kernel task + { + LimitedHandlerSimulation MockCGH; + EXPECT_CALL(MockCGH, depends_on).Times(1); + Queue->finalizeHandler( + MockCGH, detail::CG::CGTYPE::Kernel, Event); + } + // kernel task + { + LimitedHandlerSimulation MockCGH; + EXPECT_CALL(MockCGH, depends_on).Times(0); + Queue->finalizeHandler( + MockCGH, detail::CG::CGTYPE::Kernel, Event); + } + // host task + { + LimitedHandlerSimulation MockCGH; + EXPECT_CALL(MockCGH, depends_on).Times(1); + Queue->finalizeHandler( + MockCGH, detail::CG::CGTYPE::CodeplayHostTask, Event); + } +} \ No newline at end of file