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.

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

Respect ReadOnly Property in Nanobind Adapter #1809

Merged
merged 94 commits into from
Oct 30, 2024
Merged

Respect ReadOnly Property in Nanobind Adapter #1809

merged 94 commits into from
Oct 30, 2024

Conversation

fthaler
Copy link
Contributor

@fthaler fthaler commented Oct 29, 2024 •
edited
Loading

Create a SID with const element type if a Nanobind array is defined read-only. Requires Nanobind 2.1.0 (with merged wjakob/nanobind#491).

@fthaler
Renamed BlockSizes → ThreadBlockSizes
eb86cbd
@fthaler
Add loop blocking to fn GPU backend
f10f9ad
@fthaler
Cleanup/refactor
c616749
@fthaler
Treat compile-time block sizes as compile-time
44f13fd
@fthaler
Merge remote-tracking branch 'origin/master' into loop-blocking
5da7ed2
@fthaler
Merge remote-tracking branch 'origin/master' into loop-blocking
b1577c1
@fthaler
Minor cleanup
9a5edf1
@fthaler
Use uint3 instead of dim3
90e3a50
@fthaler
Silent warning
f70b885
@fthaler
Add __launch_bounds__
6e77d53
@fthaler
Fixed comment
508852d
@fthaler
Slightly simpler meta:: calculation
bade455
@fthaler
Maybe fix for Clang-CUDA compilation
1641040
@fthaler
Merge remote-tracking branch 'origin/master' into loop-blocking
85dbec8
@fthaler
Added another ::template
990555e
@fthaler
Make clang-14-cuda-11 happy
ea9e3c9
@fthaler
Fix HIP compilation
d3d5228
@fthaler
Some cleanup
bba6c78
@fthaler
Fix formatting
1b547d0
@fthaler
Check types of ThreadBlockSizes and LoopBlockSizes
6bc7bcf
@fthaler
Merge branch 'master' into loop-blocking
cf22abe
@fthaler
Merge branch 'master' into loop-blocking
afb3201
@fthaler
Faster loop blocking WIP state
21e27ea
@fthaler
Add const_host_view to const data store
6be319a
@fthaler
Added sid::make_unrolled_loop
ade9e8f
@fthaler
Use explicitly unrolled loops in GPU fn backend
0cfff11
@fthaler
Re-enable verification
980933c
@fthaler
Revert some irrelevant changes
64098e2
@fthaler
Revert irrelevant changes to neighbor tables
8a021dc
@fthaler
Revert more irrelevant changes
8ab05d1
fthaler and others added 20 commits October 7, 2024 09:36
@fthaler
Merge remote-tracking branch 'upstream/master' into loop-blocking
d16c1ea
@fthaler
Revert "Unconditional lookup in neighbor table"
4e115a5 
This reverts commit c896ea1.
@fthaler
Revert accidental changes to Daint envs
3e39f82
@fthaler
Use default loop block sizes in fn_select
494f8de
@fthaler
Disable loop unrolling for now
27fc972
@fthaler
Updated perftest references
05206ac
@iomaganaris
Only 8 loads for neighbors and passing tests
071992a
@iomaganaris
Reduces loads but errors with "operation not supported on global/shar…
ac48f24 
…ed address space" because of __ldg on array
@iomaganaris
Only read neighbor tables twice
d69fd86
@iomaganaris
Fix pointing to pointer of temporary array and m_index == -1
b39c679
@iomaganaris
Keep only the necessary changes (const table ptr)
7884d4d
@iomaganaris
Add const in sid in sid_neighbor_table
2d2a97f
@iomaganaris
Create as_const_sid for nanobind bindings to pass input fields as con…
0642864 
…st and avoid reading fields that are independent of the vertical level in each iteration
@iomaganaris
Add const in as_neighbor_table changes
3aede67
@iomaganaris @fthaler
Apply suggestions from code review
0a08822 
Improvements in `sid` handling as `const` in `as_neighbor_table`

Co-authored-by: Felix Thaler <felix.thaler@nummi.ch>
@iomaganaris
Fix as_sid and as_const_sid variable names
45fcc27
@iomaganaris
Use ndarray.ReadOnly to add_const if necessary in as_sid and remove a…
f341795 
…s_const_sid
@fthaler
Merge remote-tracking branch 'upstream/master' into readonly-nanobind…
d1be29b 
…-adapter
@fthaler
Exclude changes to SID neighbor table
abd61d3
@fthaler
Added test with a read-only nanobind array
b73a91b
@gridtoolsjenkins
Copy link
Collaborator

Hi there, this is jenkins continuous integration...
Do you want me to verify this patch?

@fthaler
Copy link
Contributor Author

fthaler commented Oct 29, 2024

launch jenkins

@fthaler
Copy link
Contributor Author

fthaler commented Oct 29, 2024

launch perftest

@fthaler fthaler requested a review from havogt October 29, 2024 15:02
@havogt havogt merged commit 4908e55 into GridTools:master Oct 30, 2024
70 checks passed
havogt pushed a commit that referenced this pull request Oct 30, 2024
@fthaler @iomaganaris @havogt
Respect ReadOnly Property in Nanobind Adapter (#1809)
6c600e4 
Create a SID with const element type if a Nanobind array is defined
read-only. Requires Nanobind 2.1.0 (with merged
wjakob/nanobind#491).

---------

Co-authored-by: Ioannis Magkanaris <ioannis.magkanaris@cscs.ch>
@iomaganaris iomaganaris mentioned this pull request Oct 31, 2024
Closed
3 tasks
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Reviewers

@havogt havogt havogt approved these changes

Assignees
No one assigned
Labels
None yet
Projects
None yet
Milestone
No milestone
Development

Successfully merging this pull request may close these issues.
4 participants
@fthaler @gridtoolsjenkins @havogt @iomaganaris