Skip to content

[KHR] Prefetch host#960

Open
TApplencourt wants to merge 6 commits intomainfrom
prefetch_host
Open

[KHR] Prefetch host#960
TApplencourt wants to merge 6 commits intomainfrom
prefetch_host

Conversation

@TApplencourt
Copy link
Contributor

Replace old #280

Add prefetch_host with the new format. The text and layout are based of the prefetch code to keep consistency.

Copy link
Contributor

@illuhad illuhad left a comment

Choose a reason for hiding this comment

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

LGTM!

Sidenote: AdaptiveCpp already implements this exact API; the only difference is that we don't have yet the khr_ prefix and the right feature test macro.
https://github.com/AdaptiveCpp/AdaptiveCpp/blob/develop/doc/extensions.md#acpp_ext_prefetch_host

@illuhad
Copy link
Contributor

illuhad commented Feb 13, 2026

AdaptiveCpp implementation PR with interface as proposed here: AdaptiveCpp/AdaptiveCpp#1972

As a sidenote, I noticed that prefetch[_host] and mem_advise all take void* instead of const void* in the SYCL spec. It's not really clear to me why they don't operate on const void*. This of course limits the code that is legal when invoking these APIs. Is there a reason for this restriction?

@TApplencourt
Copy link
Contributor Author

TApplencourt commented Feb 13, 2026

Yeah, I noticed that too. I didn't wanted to modify for this KHR but as a side note, in cuda they take a const ptr (https://docs.nvidia.com/cuda/cuda-runtime-api/group__CUDART__MEMORY.html#group__CUDART__MEMORY_1g856fa41c8c0d28655e37b778cb9ffc65). Same for L0 (https://oneapi-src.github.io/level-zero-spec/level-zero/latest/core/api.html#zecommandlistappendmemoryprefetch).

I guess oversight? (also should the function be const?)

@illuhad
Copy link
Contributor

illuhad commented Feb 13, 2026

Yeah, I noticed that too. I didn't wanted to modify for this KHR but as a side note, in cuda they take a const ptr

That makes more sense to me. Agree that for this PR it's probably out of scope.

(also should the function be const?)

I think none of the command submission functions are const. This I can understand, since from a C++ point of view, they modify the state of handler/queue objects.


int main() {
sycl::queue Q;
int *A = Q.malloc_shared<int>(1,Q);
Copy link
Contributor

Choose a reason for hiding this comment

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

There is no such member function to queue.

Copy link
Contributor Author

Choose a reason for hiding this comment

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

I'm ashamed : ( Good catch!

== Extensions to the handler class

This extension adds the following function to the [code]#sycl::handler# class.

Copy link
Contributor

Choose a reason for hiding this comment

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

I think we should show a high level synopsis here of the additions to the handler class like:

namespace sycl {
class handler {
  void khr_prefetch_host(void* ptr, size_t numBytes);
  // ...
};
}

This is consistent with the other KHR's that add member functions:

Copy link
Contributor Author

Choose a reason for hiding this comment

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

I did it but then remove it due to

Can put it back if you prefer

Copy link
Contributor

Choose a reason for hiding this comment

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

I think sycl_khr_queue_empty_query is the outlier. We should add the high level synopsis there too to make them all consistent.

== Extensions to the queue class

This extension adds the following function to the [code]#sycl::queue# class.

Copy link
Contributor

Choose a reason for hiding this comment

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

Also here, add a high-level synopsis of the member functions added to queue.

@TApplencourt
Copy link
Contributor Author

2 Options:

  • A/ Merge this KHR
  • B/ Add prefetch_host to launch_enqueue PR

We will ping some user so they can comment

@tdavidcl
Copy link

I would lean toward more granularity in extension and thus merge it. Which also will speedup its adoption (since its already available almost as is in AdaptiveCpp)

@nscottnichols
Copy link

As a SYCL user, I'm in favor of accepting this sooner than other larger extensions (that are in some sense more disruptive to the specification) since it is small and self contained. As others have mentioned, adding this to the specification will speed up adoption.

@gmlueck
Copy link
Contributor

gmlueck commented Feb 25, 2026

For those who are asking for this KHR to move forward in order to speed up adoptions, which SYCL implementation(s) are you using?

If this includes DPC++, then I wanted to explain my hesitancy. DPC++ currently provides a way to prefetch to host, but it has no effect when using the Level Zero backend because there is not yet support in Level Zero to do this prefetch. Therefore, I don't see a lot of value for DPC++ to implement a standalone "prefetch host" KHR in the short term because it would just be a no-op anyway on Level Zero. By the time we have Level Zero support, I expect that the WG will have approved #922, so I'd just as soon wait until then to have a prefetch-host KHR.

@TApplencourt
Copy link
Contributor Author

@nscottnichols , @tdavidcl if you don't mind chiming in : ) Thanks!

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

None yet

Development

Successfully merging this pull request may close these issues.

5 participants