-
Notifications
You must be signed in to change notification settings - Fork 16
Resolve compiler warnings when building with CUDA #251
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
base: develop
Are you sure you want to change the base?
Conversation
test/test_field_communication.cpp
Outdated
| Kokkos::parallel_for( | ||
| "id 0", Kokkos::RangePolicy<pcms::HostMemorySpace::execution_space>(0, n), | ||
| KOKKOS_LAMBDA(int i) { ids[i] = 0; }); | ||
| for (size_t i = 0; i < static_cast<size_t>(n); ++i) { |
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.
since we know this is on the host use std::fill
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.
This would use OpenMP threads for host parallel if available. What is the specific warning that is being resolved with this? Basicly I want to decide if it makes sense to swich this to a host for loop or stay with kokkos host parallel_for.
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.
The warning is:
warning #20011-D: calling a __host__ function("Omega_h::HostWrite<double> ::operator [](int) const") from a __host__ __device__ function("server(int, ::Omega_h::Mesh &, ::std::__cxx11::basic_string<char, ::std::char_traits<char> , ::std::allocator<char> > , int, const ::std::map< ::std::__cxx11::basic_string<char, ::std::char_traits<char> , ::std::allocator<char> > , ::std::__cxx11::basic_string<char, ::std::char_traits<char> , ::std::allocator<char> > , ::std::less< ::std::__cxx11::basic_string<char, ::std::char_traits<char> , ::std::allocator<char> > > , ::std::allocator< ::std::pair<const ::std::__cxx11::basic_string<char, ::std::char_traits<char> , ::std::allocator<char> > , ::std::__cxx11::basic_string<char, ::std::char_traits<char> , ::std::allocator<char> > > > > &, const ::std::__cxx11::basic_string<char, ::std::char_traits<char> , ::std::allocator<char> > &)::[lambda(int) (instance 1)]::operator () const") is not allowed
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.
I see...since it's in the HostSpace using KOKKOS_LAMBDA is marking that as __host__ __device__. Probably just using [=](int i){} for the lambda will resolve this. Alternatively there may be a KOKKOS_HOST_LAMBDA?
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.
I see. I'll update accordingly.
test/test_field_communication.cpp
Outdated
| Kokkos::parallel_for( | ||
| "id gid", Kokkos::RangePolicy<pcms::HostMemorySpace::execution_space>(0, n), | ||
| KOKKOS_LAMBDA(int i) { ids[i] = gids[i]; }); | ||
| for (size_t i = 0; i < static_cast<size_t>(n); ++i) { |
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.
since we know this is on the host, use std::copy
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.
This would use OpenMP threads for host parallel if available. What is the specific warning that is being resolved with this? Basicly I want to decide if it makes sense to swich this to a host for loop or stay with kokkos host parallel_for.
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.
Similar warnings about calling host functions from device functions.
| "data_d", data.size()); | ||
| Kokkos::deep_copy(data_d, Kokkos::View<const Real*, HostMemorySpace>( | ||
| data.data_handle(), data.size())); | ||
| Omega_h::parallel_for( |
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.
would it make sense to swap out the Omega_h::parallel_for with Kokkos::parallel_for if we are using the Kokkos macros?
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.
That makes sense. I think we should be consistent and use only one of them if possible.
| // element should be inside the domain (positive) | ||
| PCMS_ALWAYS_ASSERT(elem_idx >= 0 && elem_idx < mesh_.nelems()); | ||
| // PCMS_ALWAYS_ASSERT(elem_idx >= 0 && elem_idx < mesh_.nelems()); | ||
| LO count = Kokkos::atomic_sub_fetch(&elem_counts_(elem_idx), 1); |
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.
Should we define a Host macro that only evaluates on host?
test/test_field_copy.cpp
Outdated
| KOKKOS_LAMBDA(int i) { ids[i] = i; }); | ||
| for (size_t i = 0; i < static_cast<size_t>(ndata); ++i) { | ||
| ids[i] = i; | ||
| } |
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.
use std::iota
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.
This would use OpenMP threads for host parallel if available. What is the specific warning that is being resolved with this? Basicly I want to decide if it makes sense to swich this to a host for loop or stay with kokkos host parallel_for.
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.
The warning is:
warning #20011-D: calling a __host__ function("Omega_h::HostWrite<double> ::operator [](int) const") from a __host__ __device__ function("test_copy( ::std::shared_ptr< ::Omega_h::Comm> , int, int, int)::[lambda(int) (instance 1)]::operator () const") is not allowed
test/test_field_copy.cpp
Outdated
| if (std::abs(ids[i] - copied_array[i]) < 1e-12) { | ||
| sum++; | ||
| } | ||
| } |
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.
This would use OpenMP threads for host parallel if available. What is the specific warning that is being resolved with this?
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.
Similar warnings about calling host functions from device functions.
|
|
||
| // ------------------ Verification ------------------ // | ||
| double tol = 10.0 / 100.0; // 10 percent tolerance | ||
| // double tol = 10.0 / 100.0; // 10 percent tolerance |
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.
If this is unused, just delete it.
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.
I think this is intended to be used with the commented-out verification code. Would you like me to remove that entire block?
// ------------------ Verification ------------------ //
double tol = 10.0 / 100.0; // 10 percent tolerance
// XGC Originated Values
/*
for (int i = 0; i < xgc_num_nodes; ++i) {
CHECK_THAT(interpolated_back_density_at_xgc_nodes[i],
Catch::Matchers::WithinRel(density_at_xgc_nodes[i], tol) ||
Catch::Matchers::WithinAbs(density_at_xgc_nodes[i], tol));
CHECK_THAT(interpolated_back_temp_at_xgc_nodes[i],
Catch::Matchers::WithinRel(temp_at_xgc_nodes[i], tol) ||
Catch::Matchers::WithinAbs(temp_at_xgc_nodes[i], tol));
}
*/
No description provided.