Closed upsj closed 2 months ago
Another part of code that fails with CUDA 12.4 is https://github.com/ginkgo-project/ginkgo/blob/develop/common/cuda_hip/multigrid/pgm_kernels.hpp.inc#L32-L48
When I applied the const_cast
workarounds, I still can't build the full Ginkgo 1.7.0 as the Kokkos example now fails to build. I don't know if this is caused by CUDA 12.4 or Kokkos 4.1.00, though.
[1/3] Building CXX object examples/kokkos_assembly/CMakeFiles/kokkos-assembly.dir/kokkos_assembly.cpp.o
FAILED: examples/kokkos_assembly/CMakeFiles/kokkos-assembly.dir/kokkos_assembly.cpp.o
/usr/bin/c++ -DKOKKOS_DEPENDENCE -I/home/lahwaacz/ginkgo/build-cuda/include -I/home/lahwaacz/ginkgo/include -I/home/lahwaacz/ginkgo -MD -MT examples/kokkos_assembly/CMakeFiles/kokkos-assembly.dir/kokkos_assembly.cpp.o -MF examples/kokkos_assembly/CMakeFiles/kokkos-assembly.dir/kokkos_assembly.cpp.o.d -o examples/kokkos_assembly/CMakeFiles/kokkos-assembly.dir/kokkos_assembly.cpp.o -c /home/lahwaacz/ginkgo/examples/kokkos_assembly/kokkos_assembly.cpp
/home/lahwaacz/ginkgo/examples/kokkos_assembly/kokkos_assembly.cpp: In lambda function:
/home/lahwaacz/ginkgo/examples/kokkos_assembly/kokkos_assembly.cpp:127:43: error: ‘abs’ is not a member of ‘Kokkos::Experimental’
127 | lsum += Kokkos::Experimental::abs(
| ^~~
/home/lahwaacz/ginkgo/examples/kokkos_assembly/kokkos_assembly.cpp:127:43: note: suggested alternatives:
In file included from /usr/include/c++/13.2.1/cstdlib:81,
from /usr/include/c++/13.2.1/ext/string_conversions.h:43,
from /usr/include/c++/13.2.1/bits/basic_string.h:4097,
from /usr/include/c++/13.2.1/string:54,
from /usr/include/c++/13.2.1/bits/locale_classes.h:40,
from /usr/include/c++/13.2.1/bits/ios_base.h:41,
from /usr/include/c++/13.2.1/ios:44,
from /usr/include/c++/13.2.1/ostream:40,
from /usr/include/c++/13.2.1/iostream:41,
from /home/lahwaacz/ginkgo/examples/kokkos_assembly/kokkos_assembly.cpp:33:
/usr/include/c++/13.2.1/bits/std_abs.h:56:3: note: ‘abs’
56 | abs(long __i) { return __builtin_labs(__i); }
| ^~~
In file included from /usr/include/Kokkos_Complex.hpp:28,
from /usr/include/Kokkos_MathematicalSpecialFunctions.hpp:31,
from /usr/include/Kokkos_Core.hpp:54,
from /home/lahwaacz/ginkgo/examples/kokkos_assembly/kokkos_assembly.cpp:39:
/usr/include/c++/13.2.1/complex:887:5: note: ‘std::abs’
887 | abs(const complex<_Tp>& __z) { return __complex_abs(__z.__rep()); }
| ^~~
/usr/include/Kokkos_Complex.hpp:677:33: note: ‘Kokkos::abs’
677 | KOKKOS_INLINE_FUNCTION RealType abs(const complex<RealType>& x) {
| ^~~
In file included from /home/lahwaacz/ginkgo/include/ginkgo/core/base/matrix_data.hpp:45,
from /home/lahwaacz/ginkgo/include/ginkgo/core/base/mtx_io.hpp:40,
from /home/lahwaacz/ginkgo/include/ginkgo/core/base/batch_multi_vector.hpp:45,
from /home/lahwaacz/ginkgo/include/ginkgo/core/base/batch_lin_op.hpp:43,
from /home/lahwaacz/ginkgo/include/ginkgo/ginkgo.hpp:43,
from /home/lahwaacz/ginkgo/examples/kokkos_assembly/kokkos_assembly.cpp:40:
/home/lahwaacz/ginkgo/include/ginkgo/core/base/range.hpp:727:49: note: ‘gko::abs’
727 | GKO_ENABLE_UNARY_RANGE_OPERATION(abs_operation, abs,
| ^~~
/home/lahwaacz/ginkgo/include/ginkgo/core/base/range.hpp:648:9: note: in definition of macro ‘GKO_BIND_UNARY_RANGE_OPERATION_TO_OPERATOR’
648 | _operator_name(const range<Accessor>& operand) \
| ^~~~~~~~~~~~~~
/home/lahwaacz/ginkgo/include/ginkgo/core/base/range.hpp:727:1: note: in expansion of macro ‘GKO_ENABLE_UNARY_RANGE_OPERATION’
727 | GKO_ENABLE_UNARY_RANGE_OPERATION(abs_operation, abs,
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
In file included from /usr/include/c++/13.2.1/mutex:43,
from /usr/include/Serial/Kokkos_Serial.hpp:33,
from /usr/include/decl/Kokkos_Declare_SERIAL.hpp:21,
from /usr/include/KokkosCore_Config_DeclareBackend.hpp:22,
from /usr/include/Kokkos_Core.hpp:45:
/usr/include/c++/13.2.1/bits/chrono.h:457:7: note: ‘std::chrono::abs’
457 | abs(duration<_Rep, _Period> __d)
| ^~~
In file included from /usr/include/c++/13.2.1/cstdlib:79:
/usr/include/stdlib.h:980:12: note: ‘std::abs’
980 | extern int abs (int __x) __THROW __attribute__ ((__const__)) __wur;
| ^~~
/usr/include/stdlib.h:980:12: note: ‘std::abs’
/usr/include/stdlib.h:980:12: note: ‘std::abs’
/usr/include/stdlib.h:980:12: note: ‘std::abs’
/usr/include/stdlib.h:980:12: note: ‘std::abs’
/home/lahwaacz/ginkgo/examples/kokkos_assembly/kokkos_assembly.cpp:129:39: error: ‘abs’ is not a member of ‘Kokkos::Experimental’
129 | Kokkos::Experimental::abs(correct_u(xi)));
| ^~~
/home/lahwaacz/ginkgo/examples/kokkos_assembly/kokkos_assembly.cpp:129:39: note: suggested alternatives:
/usr/include/c++/13.2.1/bits/std_abs.h:56:3: note: ‘abs’
56 | abs(long __i) { return __builtin_labs(__i); }
| ^~~
/usr/include/c++/13.2.1/complex:887:5: note: ‘std::abs’
887 | abs(const complex<_Tp>& __z) { return __complex_abs(__z.__rep()); }
| ^~~
/usr/include/Kokkos_Complex.hpp:677:33: note: ‘Kokkos::abs’
677 | KOKKOS_INLINE_FUNCTION RealType abs(const complex<RealType>& x) {
| ^~~
/home/lahwaacz/ginkgo/include/ginkgo/core/base/range.hpp:727:49: note: ‘gko::abs’
727 | GKO_ENABLE_UNARY_RANGE_OPERATION(abs_operation, abs,
| ^~~
/home/lahwaacz/ginkgo/include/ginkgo/core/base/range.hpp:648:9: note: in definition of macro ‘GKO_BIND_UNARY_RANGE_OPERATION_TO_OPERATOR’
648 | _operator_name(const range<Accessor>& operand) \
| ^~~~~~~~~~~~~~
/home/lahwaacz/ginkgo/include/ginkgo/core/base/range.hpp:727:1: note: in expansion of macro ‘GKO_ENABLE_UNARY_RANGE_OPERATION’
727 | GKO_ENABLE_UNARY_RANGE_OPERATION(abs_operation, abs,
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/usr/include/c++/13.2.1/bits/chrono.h:457:7: note: ‘std::chrono::abs’
457 | abs(duration<_Rep, _Period> __d)
| ^~~
/usr/include/stdlib.h:980:12: note: ‘std::abs’
980 | extern int abs (int __x) __THROW __attribute__ ((__const__)) __wur;
| ^~~
/usr/include/stdlib.h:980:12: note: ‘std::abs’
/usr/include/stdlib.h:980:12: note: ‘std::abs’
/usr/include/stdlib.h:980:12: note: ‘std::abs’
/usr/include/stdlib.h:980:12: note: ‘std::abs’
@lahwaacz Regarding the Kokkos example, yes that is due to changes in Kokkos 4.x. We are reworking the Kokkos example anway ATM, which will also fix this issue.
For info, this also came up in spack: https://github.com/spack/spack/pull/42748#issuecomment-2015383713
Sorry, I need to hold this up. From the comments, something like pgm also fails. I am compiling it with cuda 12.4 to ensure everything work
Yes, I didn't get around to fixing that one, should be an easy one as well, though.
Yes, it requires const_cast though. Do you have any update from nvidia?
void compute_coarse_coo(std::shared_ptr<const DefaultExecutor> exec, ...
{
auto vals_it = as_device_type(vals);
auto key_it =
- thrust::make_zip_iterator(thrust::make_tuple(row_idxs, col_idxs));
+ thrust::make_zip_iterator(thrust::make_tuple(const_cast<IndexType*>(row_idxs), const_cast<IndexType*>(col_idxs)));
It's the last piece for compiling ginkgo with CUDA 12.4
The fix should be in the next release of CUDA https://github.com/NVIDIA/cccl/issues/1527
Nice! then probably guard this const_cast only with CUDA 12.4.
Fixes #1564 #1568