Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

[SYCL][Docs] Implement compile and link for source-based kernel bundles #17442

Open
wants to merge 14 commits into
base: sycl
Choose a base branch
from

Conversation

steffenlarsen
Copy link
Contributor

@steffenlarsen steffenlarsen commented Mar 13, 2025

Implements a new compile variant for source-based kernel_bundle and the corresponding linking functionality.

@steffenlarsen
Copy link
Contributor Author

Built on #17380.

@steffenlarsen steffenlarsen changed the title [SYCL] Move kernel_compiler related information [SYCL] Implement compile and link for source-based kernel bundles Mar 13, 2025
@steffenlarsen steffenlarsen force-pushed the steffen/kernel_bundle_compile_link_2 branch from 4ca96ad to 81b3f39 Compare March 13, 2025 09:52
@steffenlarsen steffenlarsen changed the title [SYCL] Implement compile and link for source-based kernel bundles [SYCL][Docs] Implement compile and link for source-based kernel bundles Mar 13, 2025
@steffenlarsen steffenlarsen force-pushed the steffen/kernel_bundle_compile_link_2 branch from 81b3f39 to 668f2b0 Compare March 13, 2025 10:05
@steffenlarsen steffenlarsen force-pushed the steffen/kernel_bundle_compile_link_2 branch from 668f2b0 to 8e76b38 Compare March 13, 2025 10:49
@steffenlarsen steffenlarsen force-pushed the steffen/kernel_bundle_compile_link_2 branch from 8e76b38 to 814ad25 Compare March 13, 2025 10:50
@steffenlarsen steffenlarsen force-pushed the steffen/kernel_bundle_compile_link_2 branch from 814ad25 to cf2840e Compare March 13, 2025 11:29
@steffenlarsen steffenlarsen force-pushed the steffen/kernel_bundle_compile_link_2 branch from cf2840e to 811f9e9 Compare March 13, 2025 13:37
@steffenlarsen steffenlarsen force-pushed the steffen/kernel_bundle_compile_link_2 branch from 811f9e9 to 3772cc2 Compare March 13, 2025 15:25
@steffenlarsen steffenlarsen force-pushed the steffen/kernel_bundle_compile_link_2 branch from 3772cc2 to 148fddb Compare March 14, 2025 06:28
@steffenlarsen steffenlarsen force-pushed the steffen/kernel_bundle_compile_link_2 branch from 148fddb to 7fb1fce Compare March 14, 2025 11:48
Copy link
Contributor

@jopperm jopperm left a comment

Choose a reason for hiding this comment

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

Two questions pertaining to the potential non-uniqueness of symbol names in the context of RTC (e.g., application compiling multiple source strings with the same kernel/function names).

`sourceBundle`.

* An `exception` with the `errc::invalid` error code if `props` contains an
`options` property that specifies an invalid option.
Copy link
Contributor

Choose a reason for hiding this comment

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

The corresponding error for build is:

An exception with the errc::invalid error code if the source language lang does not support one of the properties in PropertyListT or if props contains a build_options property that contains an option that is not supported by lang.

Is there a reason the wording here isn't the same? Did you intend to rename build_options to options, or was that just a typo?

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 was contemplating adding a separate option property or renaming the existing one, but I worried it would just grow the current addition too much for a single change. What do we think is best here? Are we happy with it being build_options? Do we rename? Should we have a separate compile_options to make it easier to separate the valid input and do any potential checks?


* An `exception` with the `errc::invalid` error code if `source` was not created
with `source_language::sycl` or was the result of `sycl::join` taking one or
more `kernel_bundle` objects not created with `source_language::sycl`.
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 similar restriction for build. Is there a reason we need this restriction for compile? Can we not compile other languages like OpenCL or SPIR-V?

It seems weird to list the sycl language specifically here. Should we instead split the query ext_oneapi_can_compile into two queries (for compile and build)? Or, do you think the limitation with sycl is only temporary, and that's why you didn't add a query?

The whole business about join seems like a can of worms. As you allude, the application could joint two source bundles from different languages! I think it's better to avoid the whole can of worms by adding a constraint to join, preventing it from joining bundles of source state.

Copy link
Contributor Author

Choose a reason for hiding this comment

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

There is no similar restriction for build. Is there a reason we need this restriction for compile? Can we not compile other languages like OpenCL or SPIR-V?

I believe we could relax this restriction to allow compiling OpenCL, SPIR-V, etc, but by allowing these source languages in object state complicates the logic for linking a little, as we wouldn't know their imported/exported symbols. There are ways we could solve that, but since my focus was on SYCL RTC kernel bundles in this initial variant, I elected to disallow them for now.

It seems weird to list the sycl language specifically here. Should we instead split the query ext_oneapi_can_compile into two queries (for compile and build)? Or, do you think the limitation with sycl is only temporary, and that's why you didn't add a query?

I did not consider splitting ext_oneapi_can_compile. I think it would make sense (unless we know we want to expand the above and don't think there will be cases where a target can build but not compile,) but since it would be ext_oneapi_can_compile for compile we have to repurpose the one we already have and probably change the uses of it around existing code. Would it make sense to do this in a follow-up for the sake of change granularity?

The whole business about join seems like a can of worms. As you allude, the application could joint two source bundles from different languages! I think it's better to avoid the whole can of worms by adding a constraint to join, preventing it from joining bundles of source state.

I'm still not sure it's necessarily that big a can of worms. I agree, the wording here is a little awkward because of there not currently being any restrictions on join and I think it might make sense to at least specify that joining kernel bundles in source state must be of the same state. Logically that makes sense in my head, as I think the source language of a source state bundle is intrinsic to it. I would be happy to sketch up some wording for a section describing the behavior of join with source state bundles. 😄

Signed-off-by: Larsen, Steffen <[email protected]>
Signed-off-by: Larsen, Steffen <[email protected]>
Signed-off-by: Larsen, Steffen <[email protected]>
Signed-off-by: Larsen, Steffen <[email protected]>
Signed-off-by: Larsen, Steffen <[email protected]>
Signed-off-by: Larsen, Steffen <[email protected]>
Signed-off-by: Larsen, Steffen <[email protected]>
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.

3 participants