Skip to content

[SYCL][FPGA] Refactor [[intel::max_work_group_size()]] attribute implementation #5392

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

Merged
merged 19 commits into from
Feb 8, 2022

Conversation

smanna12
Copy link
Contributor

@smanna12 smanna12 commented Jan 25, 2022

This patch

  1. refactors FPGA function attribute [[intel::max_work_group_size()]]
    to better fit for community standards and separates the attribute from
    [[sycl::reqd_work_group_size()]] attribute implementation.

  2. refactors the way we handled duplicate attributes and
    mutually exclusive attributes logic with when present on a given declaration.

  3. handles redeclarations or template instantiations properly.

  4. adds tests

  5. Before the refactoring patch, we silently accepted this test case below:

struct TRIFuncObjBad {
[[intel::max_work_group_size(4, 4, 4)]] void
operator()() const;
};
[[intel::max_global_work_dim(0)]]
void TRIFuncObjBad::operator()() const {}

This PR fixes the bug and closes #5449

Signed-off-by: Soumi Manna soumi.manna@intel.com

Signed-off-by: Soumi Manna <soumi.manna@intel.com>
@smanna12 smanna12 changed the title [SYCL[FPGA] Refactor of max-work-group-size attribute implementation [SYCL[FPGA] Refactor of [[intel::max_work_group_size]] attribute implementation Jan 25, 2022
@bader bader changed the title [SYCL[FPGA] Refactor of [[intel::max_work_group_size]] attribute implementation [SYCL][FPGA] Refactor [[intel::max_work_group_size]] attribute implementation Jan 25, 2022
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
@smanna12
Copy link
Contributor Author

I will refactor "reqd_work_group_size" in a separate patch to avoid big patch here for better code-reviews.

@AaronBallman, could you please provide some feedback on the implementation. I have separated support for "max_work_group_size" that was previously handled together with "reqd_work_group_size" attribute. Thanks

@smanna12 smanna12 requested a review from AaronBallman January 25, 2022 15:46
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
@smanna12 smanna12 changed the title [SYCL][FPGA] Refactor [[intel::max_work_group_size]] attribute implementation [SYCL][FPGA] Refactor [[intel::max_work_group_size()]] attribute implementation Jan 25, 2022
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
@smanna12 smanna12 marked this pull request as ready for review February 1, 2022 19:28
@smanna12 smanna12 requested a review from a team as a code owner February 1, 2022 19:28
@smanna12 smanna12 requested a review from AaronBallman February 1, 2022 19:28
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
@smanna12 smanna12 requested a review from premanandrao February 1, 2022 21:21
Copy link
Contributor

@premanandrao premanandrao left a comment

Choose a reason for hiding this comment

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

This is just a suggestion; you can decide if this reads better or if you want to retain the enum with three states better.

I am fine with it, but feel that this simplifies it at least for this case. Maybe keeping three states makes sense if this routine will generalized?

Signed-off-by: Soumi Manna <soumi.manna@intel.com>
@smanna12
Copy link
Contributor Author

smanna12 commented Feb 2, 2022

This is just a suggestion; you can decide if this reads better or if you want to retain the enum with three states better.

I am fine with it, but feel that this simplifies it at least for this case. Maybe keeping three states makes sense if this routine will generalized?

Thanks @premanandrao for the suggestion. I agree with you that this simplifies it at least for this case and reads better. I do not think we need to have enum with three states here. One state is kind of unused. So, returning true/false from the static bool function would be sufficient for this case. I have named the static bool function "InvalidWorkGroupSizeAttrs" so that i can reuse this later for reqd_work_group_size attribute refactorization work that i plan to do it after this PR.

Signed-off-by: Soumi Manna <soumi.manna@intel.com>
@smanna12 smanna12 requested a review from premanandrao February 2, 2022 23:29
Copy link
Contributor

@premanandrao premanandrao left a comment

Choose a reason for hiding this comment

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

A fix to the comment, but otherwise looks good.

Signed-off-by: Soumi Manna <soumi.manna@intel.com>
premanandrao
premanandrao previously approved these changes Feb 3, 2022
@smanna12
Copy link
Contributor Author

smanna12 commented Feb 3, 2022

Thank you for the reviews everyone!

@smanna12
Copy link
Contributor Author

smanna12 commented Feb 6, 2022

@bader, can this PR be merged? I have an approval and all checks have passed. Thanks

@bader
Copy link
Contributor

bader commented Feb 6, 2022

@bader, can this PR be merged? I have an approval and all checks have passed. Thanks

Shouldn't we get approval from @AaronBallman?

@smanna12
Copy link
Contributor Author

smanna12 commented Feb 6, 2022

@bader, can this PR be merged? I have an approval and all checks have passed. Thanks

Shouldn't we get approval from @AaronBallman?

Sorry I missed that. Yes, Agreed

Ping @AaronBallman

Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Signed-off-by: Soumi Manna <soumi.manna@intel.com>
Copy link
Contributor

@AaronBallman AaronBallman left a comment

Choose a reason for hiding this comment

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

Assuming CI passes, LGTM

@smanna12
Copy link
Contributor Author

smanna12 commented Feb 7, 2022

Assuming CI passes, LGTM

Thanks Aaron for the reviews.

@smanna12 smanna12 requested a review from premanandrao February 7, 2022 14:43
@smanna12
Copy link
Contributor Author

smanna12 commented Feb 7, 2022

Thank you for the reviews!

@bader bader merged commit 5deccd2 into intel:sycl Feb 8, 2022
@smanna12 smanna12 deleted the Refactor_Max_Work_Group_size_Attr branch February 9, 2022 21:31
smanna12 added a commit to smanna12/llvm that referenced this pull request Feb 16, 2022
…te arguments check

If the [[intel::max_work_group_size(X, Y, Z)]] attribute is specified on
a declaration along with [[sycl::reqd_work_group_size(X1, Y1, Z1)]]
attribute, this patch checks if values of reqd_work_group_size arguments are
equal or less than values of max_work_group_size attribute arguments.

Some of the test cases were missed during refactoring work with PGA function
attribute [[intel::max_work_group_size()]] on intel#5392

This patch fixes them.

Signed-off-by: Soumi Manna <soumi.manna@intel.com>
bader pushed a commit that referenced this pull request Mar 3, 2022
…te arguments check (#5592)

If the [[intel::max_work_group_size(X, Y, Z)]] attribute is specified on
a declaration along with [[sycl::reqd_work_group_size(X1, Y1, Z1)]]
attribute, this patch checks if values of reqd_work_group_size arguments are
equal or less than values of max_work_group_size attribute arguments.

Some of the test cases were missed during refactoring work with PGA function
attribute [[intel::max_work_group_size()]] on #5392

This patch adds the missing cases and fixes pulldown bug.

Signed-off-by: Soumi Manna <soumi.manna@intel.com>
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.

No diagnostic for checking max_work_group_size and max_global_work_dim attributes when merging
4 participants