[SYCL] Deprecate discard_events queue prop, make impl no-op

[JackAKirk]

Deprecates sycl_ext_oneapi_discard_queue_events in favour of the sycl_ext_oneapi_enqueue_functions extension.
In order to simplify/allow further simplifications to the handler/scheduler code the implementation of the discard_events queue property is also made a no-op immediately.

The main consequence of this is that users of discard_events will have to ensure they use the new submit_without_event / nd_launch etc APIs from https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/experimental/sycl_ext_oneapi_enqueue_functions.asciidoc
sycl::event is very lightweight already in the level_zero v2 adapter and discard_events already doesn't have a very big performance advantage for this case. Therefore the biggest consequence of this PR from a performance point of view is in the cuda/hip backends. Users of these backends need to be aware of the performance implications of making the discard_events no-op, and should ensure that they switch to using the sycl_ext_oneapi_enqueue_functions.

This PR leaves all other functionality working correctly.

[jbrodman]

So I think we'd like to still have the property exist to avoid a breaking change, but when it's seen, it's just ignored.

[JackAKirk]

OK sure, I'll fix this. Thanks

[JackAKirk]

I've now kept the property but made it a no-op. I've updated the PR description accordingly.
I've kept all discard events tests other than those which test that the underlying sycl::event is nullptr, this this PR makes those tests obsolete.
I've also updated to docs to mark sycl_ext_oneapi_discard_queue_events deprecated.

[igchor]

@JackAKirk can you also cleanup event_impl? I still see some references to discard events there. Or is this still needed for some paths?

[JackAKirk]

@JackAKirk can you also cleanup event_impl? I still see some references to discard events there. Or is this still needed for some paths?

When we use the submit_without_event path the handler_impl::MEventNeeded is still set false which leads to event_impl::isDiscarded being called which is currently used in queue::ext_oneapi_empty. So I think most of the "discard" in event_impl is still needed until events are fully removed from the submit_without_event path, which sounds like it is going to happen since apparently there is still a 1-2% overhead due to that.

But I think you might be right that I can still remove some of the event_impl "discard" stuff already here, since maybe the "discard" in event_impl::wait is only used with the discard_events property.
I'll check this. Thanks for the idea.

[JackAKirk]

@JackAKirk can you also cleanup event_impl? I still see some references to discard events there. Or is this still needed for some paths?

I've removed one case that is definitely safe to remove. waitInternal could also potentially be cleaned up, although this depends on the way forward with host_task; in general I think that all of the internal discard event checks could be removed following on from #18018 . Apart from host_task the only other reason for keeping LastEvent is to keep this little potential optimization: https://github.com/intel/llvm/pull/18018/files#diff-f412476f167a41ca486ea41ab76bc7e861df86e8cf0542d4931b93947f41e6edL687

However as discussed this doesn't seem worth it.
You do still have a LastHostTaskEvent in the #18018 draft, so maybe it makes sense to wait until this PR is finalized/merged to finish the internal discard_event cleanup.
Could you get rid of LastHostTaskEvent completely, and instead submit a empty barrier to the native queue after the host_task is completed? This would guarantee that the host_task is synced with the in-order queue, and if users want to use multiple in order queues they can submit_with_event the host_task, and sync manually?
This would mean that sycl::queue::host_task matches exactly the semantics of cudaLaunchHostFunc(cudaStream_t, ...), see
https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#host-functions-callbacks

[igchor]

@JackAKirk can you also cleanup event_impl? I still see some references to discard events there. Or is this still needed for some paths?

I've removed one case that is definitely safe to remove. waitInternal could also potentially be cleaned up, although this depends on the way forward with host_task; in general I think that all of the internal discard event checks could be removed following on from #18018 . Apart from host_task the only other reason for keeping LastEvent is to keep this little potential optimization: https://github.com/intel/llvm/pull/18018/files#diff-f412476f167a41ca486ea41ab76bc7e861df86e8cf0542d4931b93947f41e6edL687

However as discussed this doesn't seem worth it. You do still have a LastHostTaskEvent in the #18018 draft, so maybe it makes sense to wait until this PR is finalized/merged to finish the internal discard_event cleanup. Could you get rid of LastHostTaskEvent completely, and instead submit a empty barrier to the native queue after the host_task is completed? This would guarantee that the host_task is synced with the in-order queue, and if users want to use multiple in order queues they can submit_with_event the host_task, and sync manually? This would mean that sycl::queue::host_task matches exactly the semantics of cudaLaunchHostFunc(cudaStream_t, ...), see https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#host-functions-callbacks

I believe @KseniyaTikhomirova is working on a similar approach (having host task signal an event that we can pass to next enqueue function). However, there are some issues related to memory residence with this approach so this will probably happen as a next step.

[gmlueck]

The title of the PR says that it deprecates the discard_events property, but I don't see anything in this PR that adds a deprecation warning. Was that done in a previous PR? The deprecation warning should direct people to use the sycl_ext_oneapi_enqueue_functions extension instead.

[gmlueck]

You need to define dpcpp as an asciidoc attribute at the top of the file.

[gmlueck]

I think you could add some words here noting that this extension no longer provides any benefit, and direct people to the enqueue function extension as an alternative.