Skip to content

[SYCL][DOC] Added group sorting algorithms #3514

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.

Sign up for GitHub

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

Jump to bottom
Merged
merged 7 commits into from
Apr 28, 2021
Merged

[SYCL][DOC] Added group sorting algorithms #3514

merged 7 commits into from
Apr 28, 2021

Conversation

andreyfe1
Copy link
Contributor

@andreyfe1 andreyfe1 commented Apr 8, 2021 •
edited
Loading

This extension add the following:

  • joint_sort, sort_over_group algorithms
  • sorter that is a special type that let SYCL backends choose a method for sorting (e. g. radix sort, ...).
  • 2 predefined sorters: default_sorter, radix_sorter
  • Some issues that are mentioned in the Issue section.

For future:

  • Introduce interfaces for sorting (and for other Group algorithms) basing on static arrays allocated in private memory for better performance.
  • Introduce interfaces for key-value sorting (see the 2nd issue in the Issue section )

Signed-off-by: Fedorov, Andrey andrey.fedorov@intel.com

@andreyfe1 andreyfe1 requested a review from a team as a code owner April 8, 2021 16:45
@bader bader added the spec extension All issues/PRs related to extensions specifications label Apr 14, 2021
@andreyfe1
Copy link
Contributor Author

is_sorter is not needed since it can be expressed by users using pure C++. Currently we can not imagine use-cases when users may need it

@andreyfe1
Copy link
Contributor Author

I think the PR is ready for merging.
@intel/dpcpp-specification-reviewers, could you please take a look?

gmlueck
gmlueck reviewed Apr 21, 2021
View reviewed changes
@andreyfe1
Copy link
Contributor Author

By the way, key-value interface (keys are compared, but keys and values both are moved) can be expressed using default_sorter (or using interface without sorters) and oneapi::dpl::zip_iterator.
Usage:

auto first_zip = oneapi::dpl::make_zip_iterator( keys_first, values_first );
joint_sort( g, first_zip, first_zip + n, [ ]( auto x, auto y ){ return std::get<0>(x) < std::get<0>(y); } );

@andreyfe1
Copy link
Contributor Author

Does anyone have any more comments or questions?

gmlueck
gmlueck previously approved these changes Apr 26, 2021
View reviewed changes
Copy link
Contributor

@gmlueck gmlueck left a comment

Choose a reason for hiding this comment

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

LGTM

bader
bader reviewed Apr 26, 2021
View reviewed changes
Copy link
Contributor

@bader bader left a comment

Choose a reason for hiding this comment

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

Does anyone have any more comments or questions?

Please, resolve merge conflicts in sycl/doc/extensions/README.md.

@andreyfe1
added initial version for group sort
f543940 
Signed-off-by: Fedorov, Andrey <andrey.fedorov@intel.com>
@andreyfe1
update radix_sorter and remove is_sorter
d0d9ead 
Signed-off-by: Fedorov, Andrey <andrey.fedorov@intel.com>
@andreyfe1
reworked issues
85cc411 
Signed-off-by: Fedorov, Andrey <andrey.fedorov@intel.com>
@andreyfe1
apply comments
17efe73 
Signed-off-by: Fedorov, Andrey <andrey.fedorov@intel.com>
@andreyfe1
add O(N*log_2(N))
7c75ab3 
Signed-off-by: Fedorov, Andrey <andrey.fedorov@intel.com>
@andreyfe1
fixed a wording
0869b05
@andreyfe1
update copyright
1420c39 
Signed-off-by: Fedorov, Andrey <andrey.fedorov@intel.com>
@andreyfe1
Copy link
Contributor Author

Does anyone have any more comments or questions?

Please, resolve merge conflicts in sycl/doc/extensions/README.md.

Done

@bader bader merged commit edaee9b into intel:sycl Apr 28, 2021
alexbatashev pushed a commit to alexbatashev/llvm that referenced this pull request Apr 28, 2021
Merge remote-tracking branch 'upstream/sycl' into l0_make_kernel_bundle
9c45cf7 
* upstream/sycl:
  [SYCL][Doc] Add group sorting algorithms extension specification (intel#3514)
  [Buildbot] Update Windows GPU driver to 27.20.100.9466 (intel#3594)
  [SYCL][NFC] Update tests for FPGA attributes (intel#3632)
  [CODEOWNERS] Add @kbobrovs back to few projects (intel#3638)
  [NFC] Update codeowners (intel#3619)
  [SYCL] Support 3-, 16-elements vectors in SG load/store (intel#3617)
  [SYCL-PTX] Fix libclc dependencies (intel#3624)
  [SYCL] Add sycl::span for SYCL2020 support (intel#3569)
  [NFC][SYCL] Avoid -Wreorder warning about order of initialization (intel#3620)
  [SYCL][DOC] Initial Draft of Extension for querying free device memory on Level Zero (intel#3468)
  [SYCL][PI][L0] Submit open command batch on event status query (intel#3612)
  [NFC] Fix the comment (intel#3613)
  Rename misleading attribute flag (intel#3610)
  [SYCL] Generate an opt report of kernel arguments.  (intel#3492)
  [SYCL] Support extra environment variables in LIT (intel#3598)
  [SYCL][Matrix] Make joint_matrix_mad return A*B+C's result instead of C=A*B+C (intel#3586)
@andreyfe1 andreyfe1 mentioned this pull request Jun 11, 2021
Closed
@andreyfe1 andreyfe1 mentioned this pull request Jun 24, 2021
Merged
bader pushed a commit that referenced this pull request Aug 27, 2021
@andreyfe1
[SYCL][DOC] Extend group sort algorithms with interfaces to scratchpa…
f57091d 
…d memory (#3989)

Previously it was found that we need an additional local memory for more performant implementation. The proposal fixes it.
Previous version of proposal was discussed here: #3514
Design review: #3754

Signed-off-by: Fedorov, Andrey <andrey.fedorov@intel.com>
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Reviewers

@bader bader bader approved these changes

@gmlueck gmlueck gmlueck left review comments

Assignees
No one assigned
Labels
spec extension All issues/PRs related to extensions specifications
Projects
None yet
Milestone
No milestone
Development

Successfully merging this pull request may close these issues.
3 participants
@andreyfe1 @bader @gmlueck