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

[Modes] DPC++ backend #494

Merged
merged 116 commits into from
Sep 9, 2021
Merged
Show file tree
Hide file tree
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 Aug 6, 2020
4ca9bdc
oneAPI mode registration
mpanoop Aug 18, 2020
937da89
changing include directory from oneapi to dpcpp
Sep 3, 2020
27ca53f
modifying oneqpi to dpcpp in src/modes
Sep 3, 2020
b25018f
Modified dpcpp/memory to match dpcpp implementation. There is a todo …
Sep 3, 2020
74cbb61
Translation of device.cpp
Sep 21, 2020
358c073
Some modifications on kernel.cpp
Sep 21, 2020
5c105ee
Other modifications on kernel
Sep 21, 2020
26e7398
Implementation of kernel.cpp. Still issue with the fact that SYCL ker…
Sep 23, 2020
ecfc203
It compiles now :) but won't work before we find a way to enqueue arg…
Sep 24, 2020
dffee51
almost done, examploe not compiling
Sep 25, 2020
0d4e35b
want to reverse
Sep 25, 2020
9e7146c
reverseRevert "almost done, examploe not compiling"
Sep 25, 2020
8e2aacd
Revert "It compiles now :) but won't work before we find a way to enq…
Sep 25, 2020
c727102
test
Sep 25, 2020
b1bc32a
test :wq
Sep 25, 2020
0afe7ea
Revert "Implementation of kernel.cpp. Still issue with the fact that …
Sep 25, 2020
50151b8
testRevert "Revert "It compiles now :) but won't work before we find …
Sep 25, 2020
a54e363
Revert "Revert "It compiles now :) but won't work before we find a wa…
Sep 25, 2020
2f95d38
Revert "want to reverse"
Sep 25, 2020
fdff55f
Fixing issues, adding a way to access parameter addresses, but the ex…
Sep 25, 2020
26c130d
Adding some headers
Sep 26, 2020
b652e62
changing the parallel architecture of dpcpp
Sep 28, 2020
8f3853d
changing example to show the problem
Sep 28, 2020
a2f9985
solving compilation issues Merge branch 'dpcpp' of github.com:mpanoop…
Sep 28, 2020
fa2b21d
Makefile fix for dpcpp mode registration
mpanoop Sep 28, 2020
599bf2c
Fix to expose all the DPCPP devices
mpanoop Sep 28, 2020
85eaaa0
able to compile addVectors example in DPCPP
Sep 28, 2020
68872c2
Merge branch 'dpcpp' of github.com:mpanoop/occa-dev into dpcpp
Sep 28, 2020
cd69112
correcting issue but still not running
Sep 28, 2020
1cd5be1
Few issues solved
Sep 28, 2020
92ad93d
testing stupid example
Sep 29, 2020
f4216f1
fixing issues and kernel is now running
Sep 29, 2020
43fa82a
Solving pointer free issue
Sep 29, 2020
ea2d3de
adding parser for dpcpp, not working but registered to the framework
Sep 29, 2020
18c2725
Updated runEnv on bin/occa.cpp to emit status of OCCA_DPCPP_ENABLED
mpanoop Sep 29, 2020
0985835
Adding modificaiton to the DPCPP parser
Oct 1, 2020
7690066
Merge branch 'dpcpp' of github.com:mpanoop/occa-dev into dpcpp
Oct 1, 2020
7d5f4c7
Modifying example to take one additional parameter
Oct 1, 2020
9806201
Commiting OKL quick and dirty
Oct 8, 2020
dad63fd
Fixing OKL, it is now working
Oct 9, 2020
7111703
Patching computation of local and global ndranges
Oct 9, 2020
e31c507
Introduced lambda_t header and updated exprNode header and cpp for la…
mpanoop Oct 9, 2020
f2b519c
Added capture mode for lambda in lambda type header
mpanoop Oct 9, 2020
ea867ec
OKL translator is working
Oct 10, 2020
61a4a29
Merge branch 'dpcpp' of github.com:mpanoop/occa-dev into dpcpp
Oct 10, 2020
deb7884
fixing indexing
Oct 11, 2020
414bee0
Added support for device functions
mpanoop Oct 11, 2020
3a85b86
Doing some cleanup in src/lang/mode
Oct 11, 2020
6ac2b59
Modifying iso3dfd OKL version to be compliant with serial mode
Oct 11, 2020
1eafc36
adding blocking to OKL
Oct 12, 2020
7e12526
Modified the device function example to reflect the latest change in …
mpanoop Oct 12, 2020
e420997
Fixed the DPC++ device selection logic to reflect the OCCA platform I…
mpanoop Oct 15, 2020
55ff2cb
Adding some correction to example 24. The stencil computation was not…
Oct 16, 2020
f8f1aea
Solving conflicts
Oct 16, 2020
846b6f0
Strating to add shared local variable support
Oct 28, 2020
468d45c
adding support for shared variables, iso is failing for so,e unknown …
Oct 28, 2020
57aaae8
Remove the static allocation statement from OKL when generating DPCPP…
mpanoop Oct 28, 2020
fe1cc14
Support for Shared variables
Oct 28, 2020
3220b1f
Merge branch 'dpcpp' of github.com:mpanoop/occa-dev into dpcpp
Oct 28, 2020
0071afa
- Add global and local ranges at top of kernel before queue submit
kris-rowe Nov 24, 2020
e57c435
Merge branch 'master' into sync-with-libocca
kris-rowe Jan 13, 2021
456b1f6
Move file for `occa::dpcpp` classes to src/occa/internal.
kris-rowe Jan 13, 2021
8982eb6
Remove build artifacts.
kris-rowe Jan 13, 2021
8625c3a
- Move classes for OKL translation to src/internal
kris-rowe Jan 14, 2021
b0ea495
Merge pull request #1 from mpanoop/kris-update
mpanoop Jan 20, 2021
84b0a64
Begin to transition dpcpp backend to use "launcher" approach.
kris-rowe Jan 20, 2021
efe74f1
Insert command queue and nd_range and beginning of argument list.
kris-rowe Jan 20, 2021
d2376d8
Native DPC++ kernels now work with "launcher" approach.
kris-rowe Jan 26, 2021
d7d502b
Initial translation logic for dpc++ parser.
kris-rowe Feb 16, 2021
d21c1e8
Completed OKL translation logic for dpc++.
kris-rowe Feb 22, 2021
a7d2721
Add exception handling to dpcpp backend.
kris-rowe Feb 23, 2021
15fa7b8
Polyfill and clean-up.
kris-rowe Feb 26, 2021
6bb47b0
Change command queue to store a/pass-by value: a pointer is not needed.
kris-rowe Mar 2, 2021
d92aa6c
Merge remote-tracking branch 'libocca/main' into kris-update
kris-rowe Mar 2, 2021
2bb405b
Remove dpcpp::memory::getPtr as this is now defined in the base class.
kris-rowe Mar 2, 2021
743a9a2
Clean-up:
kris-rowe Mar 3, 2021
ce78a35
Adds support for atomics to SYCL/DPC++ backend.
kris-rowe Mar 9, 2021
38302f0
Merge pull request #2 from mpanoop/kris-update
mpanoop Mar 10, 2021
7d1d310
Use FindCUDAToolkit CMake module to avoid include path conflicts with
kris-rowe Mar 11, 2021
b706574
- Remove the sycl::queue and nd_range from kernel metadata for inline…
kris-rowe Mar 11, 2021
c962f0e
Add interoperability example for DPC++.
kris-rowe Mar 12, 2021
567a481
Add example using hostMalloc and unifiedMalloc.
kris-rowe Mar 12, 2021
9828706
Merge pull request #3 from mpanoop/kris-update
mpanoop Mar 15, 2021
4a80a68
Check if more than just the SYCL host device is available when
kris-rowe Mar 18, 2021
dab171f
Remove outdated comments.
kris-rowe May 10, 2021
d0c1ee0
Merge remote-tracking branch 'libocca/main' into dpcpp
kris-rowe May 10, 2021
3d05701
Remove extra examples.
kris-rowe May 10, 2021
3da14b0
Clean-up unused files.
kris-rowe May 10, 2021
d7d55e9
Update pointer class to follow the _Rule of Zero_
kris-rowe May 21, 2021
c1264ce
Update the translation of `@shared` variables so they can be passed to
kris-rowe May 21, 2021
234c463
Update `@shared` translation to use `group_local_memory` extension
kris-rowe Aug 5, 2021
c275be3
Change `dpcpp::stream::memcpy` integer parameter to match other
kris-rowe Aug 6, 2021
0aa1413
[dpcpp] Check the device id is less than the total number of devices not
kris-rowe Aug 6, 2021
736c346
[DPCPP] Fix indexing issue in memory::copyTo/copyFrom.
kris-rowe Aug 25, 2021
4522b85
Return 0 from `dpcpp::streamTag` timing routines for now.
kris-rowe Aug 27, 2021
4daaee1
Update name of native DPC++ example kernel to match other cases.
kris-rowe Sep 7, 2021
3c467fa
Update device registration and counting to include the SYCL host device.
kris-rowe Sep 7, 2021
2c81b0e
Format argument list to be consistent with other examples.
kris-rowe Sep 7, 2021
ec404ee
Move DPC++ related primitives to the header for `dpcppParser`
kris-rowe Sep 7, 2021
85d7d1b
Add missing `device::is_host()` function to dpcpp/polyfill.
kris-rowe Sep 7, 2021
f6927fc
Resolve @dmed256 comments on PR for the DPC++ backend.
kris-rowe Sep 8, 2021
c654233
Update CMake build to handle case where DPC++ components aren't found.
kris-rowe Sep 8, 2021
35196fe
Merge branch 'main' into main
kris-rowe Sep 8, 2021
cffe1c3
Get Makefile build working for DPC++ backend.
kris-rowe Sep 8, 2021
238c469
Merge remote-tracking branch 'origin/main'
kris-rowe Sep 8, 2021
fd269c1
Update DPC++ backend to use lockless caching like libocca/main.
kris-rowe Sep 8, 2021
dd54605
Update standard flags to C++17 to allow for inline static members in
kris-rowe Sep 8, 2021
944c53e
Update Makefile for example 20_native_dpcpp_kernel.
kris-rowe Sep 8, 2021
c274892
Remove swp file.
kris-rowe Sep 8, 2021
eeaebaa
Update examples test script to match shuffled directories.
kris-rowe Sep 8, 2021
121e4c1
Remove opencl tests causing CI pipeline failures.
kris-rowe Sep 8, 2021
bc72d8d
Update src/occa/internal/bin/occa.cpp
kris-rowe Sep 8, 2021
7103b19
Update code coverage script.
kris-rowe Sep 8, 2021
0aa9183
Undo source auto formatting.
kris-rowe Sep 8, 2021
e9bdde7
[DPCPP] Increase test coverage for the DPC++ backend.
kris-rowe Sep 9, 2021
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
Prev Previous commit
Next Next commit
Update DPC++ backend to use lockless caching like libocca/main.
  • Loading branch information
kris-rowe committed Sep 8, 2021
commit fd269c18cbcdd48a984febbeb45cf697540a4985
31 changes: 13 additions & 18 deletions src/occa/internal/modes/dpcpp/device.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -126,28 +126,27 @@ namespace occa
const bool usingOkl,
lang::sourceMetadata_t &launcherMetadata,
lang::sourceMetadata_t &deviceMetadata,
const occa::json &kernelProps,
io::lock_t lock)
const occa::json &kernelProps)
{
compileKernel(hashDir,
kernelName,
kernelProps,
lock);
kernelProps);

if (usingOkl)
{
return buildOKLKernelFromBinary(kernelHash,
hashDir,
kernelName,
sourceFilename,
binaryFilename,
launcherMetadata,
deviceMetadata,
kernelProps,
lock);
kernelProps);
}
else
{
void *kernel_dlhandle = sys::dlopen(binaryFilename, lock);
occa::functionPtr_t kernel_function = sys::dlsym(kernel_dlhandle, kernelName, lock);
void *kernel_dlhandle = sys::dlopen(binaryFilename);
occa::functionPtr_t kernel_function = sys::dlsym(kernel_dlhandle, kernelName);

return new dpcpp::kernel(this,
kernelName,
Expand All @@ -164,8 +163,7 @@ namespace occa

void device::compileKernel(const std::string &hashDir,
const std::string &kernelName,
const occa::json &kernelProps,
io::lock_t &lock)
const occa::json &kernelProps)
{
occa::json allProps = kernelProps;
const bool verbose = allProps.get("verbose", false);
Expand Down Expand Up @@ -215,7 +213,6 @@ namespace occa

const int compileError = system(sCommand.c_str());

lock.release();
if (compileError)
{
OCCA_FORCE_ERROR("Error compiling [" << kernelName << "],"
Expand All @@ -226,14 +223,12 @@ namespace occa
modeKernel_t *device::buildOKLKernelFromBinary(const hash_t kernelHash,
const std::string &hashDir,
const std::string &kernelName,
const std::string &sourceFilename,
const std::string &binaryFilename,
lang::sourceMetadata_t &launcherMetadata,
lang::sourceMetadata_t &deviceMetadata,
const occa::json &kernelProps,
io::lock_t lock)
const occa::json &kernelProps)
{
const std::string sourceFilename = hashDir + kc::sourceFile;
const std::string binaryFilename = hashDir + kc::binaryFile;

dpcpp::kernel &k = *(new dpcpp::kernel(this,
kernelName,
sourceFilename,
Expand All @@ -248,7 +243,7 @@ namespace occa
kernelName,
deviceMetadata);

void *dl_handle = sys::dlopen(binaryFilename,lock);
void *dl_handle = sys::dlopen(binaryFilename);

const int launchedKernelsCount = (int)launchedKernelsMetadata.size();
for (int i = 0; i < launchedKernelsCount; ++i)
Expand All @@ -262,7 +257,7 @@ namespace occa
arguments.erase(arguments.begin());
arguments.erase(arguments.begin());

occa::functionPtr_t kernel_function = sys::dlsym(dl_handle, metadata.name,lock);
occa::functionPtr_t kernel_function = sys::dlsym(dl_handle, metadata.name);

kernel *dpcppKernel = new dpcpp::kernel(this,
metadata.name,
Expand Down
11 changes: 5 additions & 6 deletions src/occa/internal/modes/dpcpp/device.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -53,23 +53,22 @@ namespace occa
const bool usingOkl,
lang::sourceMetadata_t &launcherMetadata,
lang::sourceMetadata_t &deviceMetadata,
const occa::json &kernelProps,
io::lock_t lock) override;
const occa::json &kernelProps) override;

void setArchCompilerFlags(occa::json &kernelProps);

void compileKernel(const std::string &hashDir,
const std::string &kernelName,
const occa::json &kernelProps,
io::lock_t &lock);
const occa::json &kernelProps);

virtual modeKernel_t *buildOKLKernelFromBinary(const hash_t kernelHash,
const std::string &hashDir,
const std::string &kernelName,
const std::string &sourceFilename,
const std::string &binaryFilename,
lang::sourceMetadata_t &launcherMetadata,
lang::sourceMetadata_t &deviceMetadata,
const occa::json &kernelProps,
io::lock_t lock) override;
const occa::json &kernelProps) override;

virtual modeKernel_t *buildKernelFromBinary(const std::string &filename,
const std::string &kernelName,
Expand Down
1 change: 0 additions & 1 deletion src/occa/internal/modes/dpcpp/utils.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -5,7 +5,6 @@

#include <occa/internal/modes/dpcpp/polyfill.hpp>
#include <occa/internal/core/device.hpp>
#include <occa/internal/io/lock.hpp>

namespace occa {

Expand Down