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

Compatibility issues between Eigen and GPUs (OpenACC/CUDA) #728

Merged
merged 8 commits into from
Sep 6, 2021

Conversation

kotsaloscv
Copy link
Contributor

This PR solves the compatibility issues between Eigen and GPUs (OpenACC/CUDA).

Two factors contribute to the above solution:

  1. New Eigen branch (version 3.5 and above). Currently we are using a mirrored version of Eigen (https://github.com/BlueBrain/eigen/releases/tag/v3.5-alpha.1).
  2. An API that makes possible to call any Eigen __device__ function from within OpenACC regions. In more details, Eigen-3.5+ provides better GPU support; however, some functions cannot be called directly from within OpenACC regions. Therefore, we need to wrap them in a special API (decorate them with __device__ & acc routine tokens), which allows us to eventually call them from OpenACC. Calling these functions from CUDA kernels presents no issue ...

This PR closes/replaces the following PRs:

  1. Eigen openacc compatibility #690
  2. Fix the issue with Eigen+OpenACC while executing on GPU #726

This PR works in combination with CoreNEURON's PR # 624 (minor changes to include the new Eigen API).

This PR closes the following issues:

  1. Eigen compatibility with OpenACC #311
  2. Issue with newton solver when used with OpenACC backend #135

@bbpbuildbot
Copy link
Collaborator

Logfiles from GitLab pipeline #15814 (:white_check_mark:) have been uploaded here!

Status and direct links:

Copy link
Contributor

@iomaganaris iomaganaris left a comment

Choose a reason for hiding this comment

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

LGTM. Just a few questions

src/codegen/codegen_acc_visitor.cpp Show resolved Hide resolved
src/codegen/codegen_c_visitor.hpp Outdated Show resolved Hide resolved
src/codegen/codegen_c_visitor.cpp Show resolved Hide resolved
@bbpbuildbot
Copy link
Collaborator

Logfiles from GitLab pipeline #15838 (:white_check_mark:) have been uploaded here!

Status and direct links:

@codecov-commenter
Copy link

Codecov Report

Merging #728 (20ed7d9) into master (1321c7b) will decrease coverage by 0.04%.
The diff coverage is 24.32%.

Impacted file tree graph

@@            Coverage Diff             @@
##           master     #728      +/-   ##
==========================================
- Coverage   60.71%   60.67%   -0.05%     
==========================================
  Files         200      200              
  Lines       33362    33385      +23     
==========================================
  Hits        20255    20255              
- Misses      13107    13130      +23     
Impacted Files Coverage Δ
src/codegen/codegen_acc_visitor.cpp 0.00% <0.00%> (ø)
src/codegen/codegen_acc_visitor.hpp 0.00% <ø> (ø)
src/codegen/codegen_c_visitor.hpp 91.66% <ø> (ø)
src/solver/newton/newton.hpp 94.33% <ø> (ø)
src/codegen/codegen_c_visitor.cpp 63.09% <32.14%> (-0.35%) ⬇️

Continue to review full report at Codecov.

Legend - Click here to learn more
Δ = absolute <relative> (impact), ø = not affected, ? = missing data
Powered by Codecov. Last update 1321c7b...20ed7d9. Read the comment docs.

Copy link
Contributor

@pramodk pramodk left a comment

Choose a reason for hiding this comment

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

Apart from minor comment, LGTM

src/solver/newton/newton.hpp Show resolved Hide resolved
src/solver/newton/newton.hpp Show resolved Hide resolved
src/solver/newton/newton.hpp Show resolved Hide resolved
@pramodk pramodk merged commit 5f4462a into master Sep 6, 2021
@pramodk pramodk deleted the kotsalos/solve_eigen_GPU_issues branch September 6, 2021 15:27
pramodk pushed a commit to BlueBrain/CoreNeuron that referenced this pull request Sep 7, 2021
* Resolve Eigen-OpenACC compatibility issues
   - See the changes done in NMODL: BlueBrain/nmodl/pull/728
* As part of this PR we now build `partial_piv_lu.cu` from NMODL
* Some CUDA warnings are suppressed because they are assumed
   safe and can be ignored by end users.
* Update README to fix code examples
pramodk pushed a commit to neuronsimulator/nrn that referenced this pull request Nov 2, 2022
…ueBrain/CoreNeuron#624)

* Resolve Eigen-OpenACC compatibility issues
   - See the changes done in NMODL: BlueBrain/nmodl/pull/728
* As part of this PR we now build `partial_piv_lu.cu` from NMODL
* Some CUDA warnings are suppressed because they are assumed
   safe and can be ignored by end users.
* Update README to fix code examples

CoreNEURON Repo SHA: BlueBrain/CoreNeuron@2b3e746
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.

Eigen compatibility with OpenACC Issue with newton solver when used with OpenACC backend
5 participants