diff --git a/sycl/source/detail/scheduler/commands.cpp b/sycl/source/detail/scheduler/commands.cpp index 5c42709930436..63eb048212776 100644 --- a/sycl/source/detail/scheduler/commands.cpp +++ b/sycl/source/detail/scheduler/commands.cpp @@ -3592,21 +3592,10 @@ ur_result_t ExecCGCommand::enqueueImpQueue() { return UR_RESULT_SUCCESS; } - case CGType::None: { - if (RawEvents.empty()) { - // urEnqueueEventsWait with zero events acts like a barrier which is NOT - // what we want here. On the other hand, there is nothing to wait for, so - // we don't need to enqueue anything. - return UR_RESULT_SUCCESS; - } - const detail::AdapterPtr &Adapter = MQueue->getAdapter(); - ur_event_handle_t Event; - ur_result_t Result = Adapter->call_nocheck( - MQueue->getHandleRef(), RawEvents.size(), - RawEvents.size() ? &RawEvents[0] : nullptr, &Event); - MEvent->setHandle(Event); - return Result; - } + case CGType::None: + throw sycl::exception(sycl::make_error_code(sycl::errc::runtime), + "CG type not implemented. " + + codeToString(UR_RESULT_ERROR_INVALID_OPERATION)); } return UR_RESULT_ERROR_INVALID_OPERATION; } diff --git a/sycl/source/handler.cpp b/sycl/source/handler.cpp index 1daa436b1ee87..a7ac73f9e4c34 100644 --- a/sycl/source/handler.cpp +++ b/sycl/source/handler.cpp @@ -496,8 +496,21 @@ event handler::finalize() { MCodeLoc)); break; case detail::CGType::None: - CommandGroup.reset(new detail::CG(detail::CGType::None, - std::move(impl->CGData), MCodeLoc)); + if (detail::ur::trace(detail::ur::TraceLevel::TRACE_ALL)) { + std::cout << "WARNING: An empty command group is submitted." << std::endl; + } + + // Empty nodes are handled by Graph like standard nodes + // For Standard mode (non-graph), + // empty nodes are not sent to the scheduler to save time + if (impl->MGraph || (MQueue && MQueue->getCommandGraph())) { + CommandGroup.reset(new detail::CG(detail::CGType::None, + std::move(impl->CGData), MCodeLoc)); + } else { + detail::EventImplPtr Event = std::make_shared(); + MLastEvent = detail::createSyclObjFromImpl(Event); + return MLastEvent; + } break; } diff --git a/sycl/test-e2e/Basic/empty_command.cpp b/sycl/test-e2e/Basic/empty_command.cpp index dac5865ae8d72..313ca81743c36 100644 --- a/sycl/test-e2e/Basic/empty_command.cpp +++ b/sycl/test-e2e/Basic/empty_command.cpp @@ -26,11 +26,18 @@ void test_host_task_dep() { auto empty_cg_event = q.submit([&](handler &cgh) { cgh.depends_on(host_event); }); + // FIXME: This should deadlock, but the dependency is ignored currently. + empty_cg_event.wait(); + assert(x == 0); start_execution.count_down(); empty_cg_event.wait(); - assert(x == 42); + // FIXME: uncomment once the bug mentioned above is fixed. + // assert(x == 42); + + // I'm seeing some weird hang without this: + host_event.wait(); } void test_device_event_dep() { @@ -46,12 +53,17 @@ void test_device_event_dep() { auto empty_cg_event = q.submit([&](handler &cgh) { cgh.depends_on(device_event); }); + // FIXME: This should deadlock, but the dependency is ignored currently. + empty_cg_event.wait(); + assert(*p == 0); start_execution.count_down(); empty_cg_event.wait(); - assert(*p == 42); + // FIXME: uncomment once the bug mentioned above is fixed. + // assert(*p == 42); + q.wait(); sycl::free(p, q); } @@ -78,12 +90,17 @@ void test_accessor_dep() { auto empty_cg_event = q.submit([&](handler &cgh) { sycl::accessor a{b, cgh}; }); + // FIXME: This should deadlock, but the dependency is ignored currently. + empty_cg_event.wait(); + assert(*p == 0); start_execution.count_down(); empty_cg_event.wait(); - assert(*p == 42); + // FIXME: uncomment once the bug mentioned above is fixed. + // assert(*p == 42); + q.wait(); sycl::free(p, q); }