-
Notifications
You must be signed in to change notification settings - Fork 86
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
[Modes] DPC++ backend #494
Merged
Merged
Changes from 1 commit
Commits
Show all changes
116 commits
Select commit
Hold shift + click to select a range
e71cdd9
oneAPI headers changes phase 1
subarnar 4ca9bdc
oneAPI mode registration
mpanoop 937da89
changing include directory from oneapi to dpcpp
27ca53f
modifying oneqpi to dpcpp in src/modes
b25018f
Modified dpcpp/memory to match dpcpp implementation. There is a todo …
74cbb61
Translation of device.cpp
358c073
Some modifications on kernel.cpp
5c105ee
Other modifications on kernel
26e7398
Implementation of kernel.cpp. Still issue with the fact that SYCL ker…
ecfc203
It compiles now :) but won't work before we find a way to enqueue arg…
dffee51
almost done, examploe not compiling
0d4e35b
want to reverse
9e7146c
reverseRevert "almost done, examploe not compiling"
8e2aacd
Revert "It compiles now :) but won't work before we find a way to enq…
c727102
test
b1bc32a
test :wq
0afe7ea
Revert "Implementation of kernel.cpp. Still issue with the fact that …
50151b8
testRevert "Revert "It compiles now :) but won't work before we find …
a54e363
Revert "Revert "It compiles now :) but won't work before we find a wa…
2f95d38
Revert "want to reverse"
fdff55f
Fixing issues, adding a way to access parameter addresses, but the ex…
26c130d
Adding some headers
b652e62
changing the parallel architecture of dpcpp
8f3853d
changing example to show the problem
a2f9985
solving compilation issues Merge branch 'dpcpp' of github.com:mpanoop…
fa2b21d
Makefile fix for dpcpp mode registration
mpanoop 599bf2c
Fix to expose all the DPCPP devices
mpanoop 85eaaa0
able to compile addVectors example in DPCPP
68872c2
Merge branch 'dpcpp' of github.com:mpanoop/occa-dev into dpcpp
cd69112
correcting issue but still not running
1cd5be1
Few issues solved
92ad93d
testing stupid example
f4216f1
fixing issues and kernel is now running
43fa82a
Solving pointer free issue
ea2d3de
adding parser for dpcpp, not working but registered to the framework
18c2725
Updated runEnv on bin/occa.cpp to emit status of OCCA_DPCPP_ENABLED
mpanoop 0985835
Adding modificaiton to the DPCPP parser
7690066
Merge branch 'dpcpp' of github.com:mpanoop/occa-dev into dpcpp
7d5f4c7
Modifying example to take one additional parameter
9806201
Commiting OKL quick and dirty
dad63fd
Fixing OKL, it is now working
7111703
Patching computation of local and global ndranges
e31c507
Introduced lambda_t header and updated exprNode header and cpp for la…
mpanoop f2b519c
Added capture mode for lambda in lambda type header
mpanoop ea867ec
OKL translator is working
61a4a29
Merge branch 'dpcpp' of github.com:mpanoop/occa-dev into dpcpp
deb7884
fixing indexing
414bee0
Added support for device functions
mpanoop 3a85b86
Doing some cleanup in src/lang/mode
6ac2b59
Modifying iso3dfd OKL version to be compliant with serial mode
1eafc36
adding blocking to OKL
7e12526
Modified the device function example to reflect the latest change in …
mpanoop e420997
Fixed the DPC++ device selection logic to reflect the OCCA platform I…
mpanoop 55ff2cb
Adding some correction to example 24. The stencil computation was not…
f8f1aea
Solving conflicts
846b6f0
Strating to add shared local variable support
468d45c
adding support for shared variables, iso is failing for so,e unknown …
57aaae8
Remove the static allocation statement from OKL when generating DPCPP…
mpanoop fe1cc14
Support for Shared variables
3220b1f
Merge branch 'dpcpp' of github.com:mpanoop/occa-dev into dpcpp
0071afa
- Add global and local ranges at top of kernel before queue submit
kris-rowe e57c435
Merge branch 'master' into sync-with-libocca
kris-rowe 456b1f6
Move file for `occa::dpcpp` classes to src/occa/internal.
kris-rowe 8982eb6
Remove build artifacts.
kris-rowe 8625c3a
- Move classes for OKL translation to src/internal
kris-rowe b0ea495
Merge pull request #1 from mpanoop/kris-update
mpanoop 84b0a64
Begin to transition dpcpp backend to use "launcher" approach.
kris-rowe efe74f1
Insert command queue and nd_range and beginning of argument list.
kris-rowe d2376d8
Native DPC++ kernels now work with "launcher" approach.
kris-rowe d7d502b
Initial translation logic for dpc++ parser.
kris-rowe d21c1e8
Completed OKL translation logic for dpc++.
kris-rowe a7d2721
Add exception handling to dpcpp backend.
kris-rowe 15fa7b8
Polyfill and clean-up.
kris-rowe 6bb47b0
Change command queue to store a/pass-by value: a pointer is not needed.
kris-rowe d92aa6c
Merge remote-tracking branch 'libocca/main' into kris-update
kris-rowe 2bb405b
Remove dpcpp::memory::getPtr as this is now defined in the base class.
kris-rowe 743a9a2
Clean-up:
kris-rowe ce78a35
Adds support for atomics to SYCL/DPC++ backend.
kris-rowe 38302f0
Merge pull request #2 from mpanoop/kris-update
mpanoop 7d1d310
Use FindCUDAToolkit CMake module to avoid include path conflicts with
kris-rowe b706574
- Remove the sycl::queue and nd_range from kernel metadata for inline…
kris-rowe c962f0e
Add interoperability example for DPC++.
kris-rowe 567a481
Add example using hostMalloc and unifiedMalloc.
kris-rowe 9828706
Merge pull request #3 from mpanoop/kris-update
mpanoop 4a80a68
Check if more than just the SYCL host device is available when
kris-rowe dab171f
Remove outdated comments.
kris-rowe d0c1ee0
Merge remote-tracking branch 'libocca/main' into dpcpp
kris-rowe 3d05701
Remove extra examples.
kris-rowe 3da14b0
Clean-up unused files.
kris-rowe d7d55e9
Update pointer class to follow the _Rule of Zero_
kris-rowe c1264ce
Update the translation of `@shared` variables so they can be passed to
kris-rowe 234c463
Update `@shared` translation to use `group_local_memory` extension
kris-rowe c275be3
Change `dpcpp::stream::memcpy` integer parameter to match other
kris-rowe 0aa1413
[dpcpp] Check the device id is less than the total number of devices not
kris-rowe 736c346
[DPCPP] Fix indexing issue in memory::copyTo/copyFrom.
kris-rowe 4522b85
Return 0 from `dpcpp::streamTag` timing routines for now.
kris-rowe 4daaee1
Update name of native DPC++ example kernel to match other cases.
kris-rowe 3c467fa
Update device registration and counting to include the SYCL host device.
kris-rowe 2c81b0e
Format argument list to be consistent with other examples.
kris-rowe ec404ee
Move DPC++ related primitives to the header for `dpcppParser`
kris-rowe 85d7d1b
Add missing `device::is_host()` function to dpcpp/polyfill.
kris-rowe f6927fc
Resolve @dmed256 comments on PR for the DPC++ backend.
kris-rowe c654233
Update CMake build to handle case where DPC++ components aren't found.
kris-rowe 35196fe
Merge branch 'main' into main
kris-rowe cffe1c3
Get Makefile build working for DPC++ backend.
kris-rowe 238c469
Merge remote-tracking branch 'origin/main'
kris-rowe fd269c1
Update DPC++ backend to use lockless caching like libocca/main.
kris-rowe dd54605
Update standard flags to C++17 to allow for inline static members in
kris-rowe 944c53e
Update Makefile for example 20_native_dpcpp_kernel.
kris-rowe c274892
Remove swp file.
kris-rowe eeaebaa
Update examples test script to match shuffled directories.
kris-rowe 121e4c1
Remove opencl tests causing CI pipeline failures.
kris-rowe bc72d8d
Update src/occa/internal/bin/occa.cpp
kris-rowe 7103b19
Update code coverage script.
kris-rowe 0aa9183
Undo source auto formatting.
kris-rowe e9bdde7
[DPCPP] Increase test coverage for the DPC++ backend.
kris-rowe File filter
Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
Completed OKL translation logic for dpc++.
- Loading branch information
commit d21c1e830b1b3bd2368f1ed713c1830d5d4dc34b
There are no files selected for viewing
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,65 @@ | ||
#include <occa/internal/lang/expr/dpcppAccessorNode.hpp> | ||
#include <occa/internal/utils/string.hpp> | ||
|
||
namespace occa | ||
{ | ||
namespace lang | ||
{ | ||
|
||
dpcppAccessorNode::dpcppAccessorNode(token_t *token_, | ||
const vartype_t &shared_type_, | ||
const std::string &handler_name_) | ||
: exprNode{token_}, shared_type{shared_type_}, handler_name{handler_name_} | ||
{ | ||
} | ||
|
||
dpcppAccessorNode::dpcppAccessorNode(const dpcppAccessorNode &node) | ||
: exprNode{node.token}, shared_type{node.shared_type}, handler_name{node.handler_name} | ||
{} | ||
|
||
exprNode *dpcppAccessorNode::clone() const | ||
{ | ||
return new dpcppAccessorNode(*this); | ||
} | ||
|
||
void dpcppAccessorNode::print(printer &pout) const | ||
{ | ||
const arrayVector &var_dims = shared_type.arrays; | ||
const std::size_t var_rank = var_dims.size(); | ||
|
||
pout << "sycl::accessor<"; | ||
pout << shared_type.name(); | ||
pout << ","; | ||
pout << occa::toString(var_rank); | ||
pout << ","; | ||
pout << "sycl::access::mode::read_write,"; | ||
pout << "sycl::access::target::local>("; | ||
|
||
if (var_rank > 0) | ||
{ | ||
pout << "{" << *(var_dims[0].size); | ||
|
||
if(var_rank > 1) | ||
pout << "," << *(var_dims[1].size); | ||
|
||
if(var_rank > 2) | ||
pout << "," << *(var_dims[1].size); | ||
|
||
pout<< "},"; | ||
} | ||
|
||
pout << handler_name; | ||
pout << ")"; | ||
} | ||
|
||
void dpcppAccessorNode::debugPrint(const std::string &prefix) const | ||
{ | ||
printer pout(io::stderr); | ||
io::stderr << prefix << "|\n" | ||
<< prefix << "|---["; | ||
pout << (*this); | ||
io::stderr << "] (dpcppAccessor)\n"; | ||
} | ||
|
||
} // namespace lang | ||
} // namespace occa |
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,32 @@ | ||
#ifndef OCCA_INTERNAL_LANG_EXPR_DPCPPACCESSORNODE_HEADER | ||
#define OCCA_INTERNAL_LANG_EXPR_DPCPPACCESSORNODE_HEADER | ||
|
||
#include <occa/internal/lang/expr/exprNode.hpp> | ||
|
||
namespace occa { | ||
namespace lang { | ||
class dpcppAccessorNode : public exprNode { | ||
public: | ||
vartype_t shared_type; | ||
std::string handler_name; | ||
|
||
dpcppAccessorNode(token_t *token_, | ||
const vartype_t& shared_type_, | ||
const std::string& handler_name_); | ||
|
||
dpcppAccessorNode(const dpcppAccessorNode& node); | ||
|
||
~dpcppAccessorNode() = default; | ||
|
||
inline udim_t type() const { return exprNodeType::dpcppAccessor; } | ||
|
||
virtual exprNode* clone() const; | ||
|
||
virtual void print(printer &pout) const; | ||
|
||
virtual void debugPrint(const std::string &prefix) const; | ||
}; | ||
} | ||
} | ||
|
||
#endif |
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Oops, something went wrong.
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Nit: We aren't using the
{}
intializers, but it doesn't seem to impact anything either so might be ok to keep