Import next changes SHA[beb928035, 8aa440482]

Author: stgeke

3rd_party/hypre/src/IJ_mv/IJMatrix_parcsr_device.c

@@ -214,9 +214,9 @@ hypre_IJMatrixSetAddValuesParCSRDevice( hypre_IJMatrix       *matrix,
    return hypre_error_flag;
 }
 
-#if defined(HYPRE_USING_SYCL)
 template<typename T1, typename T2>
 struct hypre_IJMatrixAssembleFunctor
+#if defined(HYPRE_USING_SYCL)
 {
    typedef std::tuple<T1, T2> Tuple;
 
@@ -227,9 +227,6 @@ struct hypre_IJMatrixAssembleFunctor
    }
 };
 #else
-template<typename T1, typename T2>
-struct hypre_IJMatrixAssembleFunctor : public
-   thrust::binary_function< thrust::tuple<T1, T2>, thrust::tuple<T1, T2>, thrust::tuple<T1, T2> >
 {
    typedef thrust::tuple<T1, T2> Tuple;
 
@@ -315,15 +312,16 @@ hypre_IJMatrixAssembleSortAndReduce1(HYPRE_Int  N0, HYPRE_BigInt  *I0, HYPRE_Big
 
    *N1 = std::get<0>(new_end.first.base()) - I;
 #else
-   HYPRE_THRUST_CALL(
-      exclusive_scan_by_key,
-      make_reverse_iterator(thrust::make_zip_iterator(thrust::make_tuple(I0 + N0, J0 + N0))),
-      make_reverse_iterator(thrust::make_zip_iterator(thrust::make_tuple(I0,    J0))),
-      make_reverse_iterator(thrust::device_pointer_cast<char>(X0) + N0),
-      make_reverse_iterator(thrust::device_pointer_cast<char>(X) + N0),
-      char(0),
-      thrust::equal_to< thrust::tuple<HYPRE_BigInt, HYPRE_BigInt> >(),
-      thrust::maximum<char>() );
+  HYPRE_THRUST_CALL(
+    exclusive_scan_by_key,
+    thrust::make_reverse_iterator(thrust::make_zip_iterator(thrust::make_tuple(I0 + N0, J0 + N0))),
+    thrust::make_reverse_iterator(thrust::make_zip_iterator(thrust::make_tuple(I0,    J0))),
+    thrust::make_reverse_iterator(thrust::device_pointer_cast<char>(X0) + N0),
+    thrust::make_reverse_iterator(thrust::device_pointer_cast<char>(X) + N0),
+    char(0),
+    thrust::equal_to<thrust::tuple<HYPRE_BigInt, HYPRE_BigInt>>(),
+    thrust::maximum<char>()
+  );
 
    HYPRE_THRUST_CALL(replace_if, A0, A0 + N0, X, thrust::identity<char>(), 0.0);
 
@@ -347,9 +345,9 @@ hypre_IJMatrixAssembleSortAndReduce1(HYPRE_Int  N0, HYPRE_BigInt  *I0, HYPRE_Big
    return hypre_error_flag;
 }
 
-#if defined(HYPRE_USING_SYCL)
 template<typename T1, typename T2>
 struct hypre_IJMatrixAssembleFunctor2
+#if defined(HYPRE_USING_SYCL)
 {
    typedef std::tuple<T1, T2> Tuple;
 
@@ -364,9 +362,6 @@ struct hypre_IJMatrixAssembleFunctor2
    }
 };
 #else
-template<typename T1, typename T2>
-struct hypre_IJMatrixAssembleFunctor2 : public
-   thrust::binary_function< thrust::tuple<T1, T2>, thrust::tuple<T1, T2>, thrust::tuple<T1, T2> >
 {
    typedef thrust::tuple<T1, T2> Tuple;
 
@@ -484,13 +479,14 @@ hypre_IJMatrixAssembleSortAndReduce3(HYPRE_Int  N0, HYPRE_BigInt  *I0, HYPRE_Big
 #else
    /* output in X0: 0: keep, 1: zero-out */
    HYPRE_THRUST_CALL(
-      inclusive_scan_by_key,
-      make_reverse_iterator(thrust::make_zip_iterator(thrust::make_tuple(I0 + N0, J0 + N0))),
-      make_reverse_iterator(thrust::make_zip_iterator(thrust::make_tuple(I0,    J0))),
-      make_reverse_iterator(thrust::device_pointer_cast<char>(X0) + N0),
-      make_reverse_iterator(thrust::device_pointer_cast<char>(X0) + N0),
-      thrust::equal_to< thrust::tuple<HYPRE_BigInt, HYPRE_BigInt> >(),
-      thrust::maximum<char>() );
+     inclusive_scan_by_key,
+     thrust::make_reverse_iterator(thrust::make_zip_iterator(thrust::make_tuple(I0 + N0, J0 + N0))),
+     thrust::make_reverse_iterator(thrust::make_zip_iterator(thrust::make_tuple(I0,    J0))),
+     thrust::make_reverse_iterator(thrust::device_pointer_cast<char>(X0) + N0),
+     thrust::make_reverse_iterator(thrust::device_pointer_cast<char>(X0) + N0),
+     thrust::equal_to<thrust::tuple<HYPRE_BigInt, HYPRE_BigInt>>(),
+     thrust::maximum<char>()
+   );
 
    HYPRE_THRUST_CALL(replace_if, A0, A0 + N0, X0, thrust::identity<char>(), 0.0);
 

3rd_party/hypre/src/IJ_mv/IJVector_parcsr_device.c

@@ -21,9 +21,9 @@
  * hypre_IJVectorAssembleFunctor
  *--------------------------------------------------------------------*/
 
-#if defined(HYPRE_USING_SYCL)
 template<typename T1, typename T2>
 struct hypre_IJVectorAssembleFunctor
+#if defined(HYPRE_USING_SYCL)
 {
    typedef std::tuple<T1, T2> Tuple;
 
@@ -34,9 +34,6 @@ struct hypre_IJVectorAssembleFunctor
    }
 };
 #else
-template<typename T1, typename T2>
-struct hypre_IJVectorAssembleFunctor : public
-   thrust::binary_function< thrust::tuple<T1, T2>, thrust::tuple<T1, T2>, thrust::tuple<T1, T2> >
 {
    typedef thrust::tuple<T1, T2> Tuple;
 

3rd_party/hypre/src/configure

@@ -3369,7 +3369,7 @@ hypre_using_memory_tracker=no
 
 hypre_test_using_host=no
 
-hypre_cxxstd=11
+hypre_cxxstd=17
 
 hypre_using_magma=no
 hypre_user_gave_magma_inc=no
@@ -10828,7 +10828,7 @@ printf "%s\n" "#define HYPRE_USING_HIP 1" >>confdefs.h
            HYPRE_CUDA_GENCODE="`echo ${HYPRE_CUDA_GENCODE}|sed 's/,$//'`"
         fi
 
-                                        HIPCXXFLAGS="-x hip -std=c++14 ${HIPCXXFLAGS}"
+                                        HIPCXXFLAGS="-x hip -std=c++${hypre_cxxstd} ${HIPCXXFLAGS}"
 
                                 if test x"$hypre_using_debug" == x"yes"
 then :

3rd_party/hypre/src/parcsr_ls/ams.c

@@ -495,11 +495,7 @@ HYPRE_Int hypre_ParCSRMatrixFixZeroRows(hypre_ParCSRMatrix *A)
  *--------------------------------------------------------------------------*/
 
 #if defined(HYPRE_USING_GPU)
-#if defined(HYPRE_USING_SYCL)
 struct l1_norm_op1
-#else
-struct l1_norm_op1 : public thrust::binary_function<HYPRE_Complex, HYPRE_Complex, HYPRE_Complex>
-#endif
 {
    __host__ __device__
    HYPRE_Complex operator()(const HYPRE_Complex &x, const HYPRE_Complex &y) const
@@ -510,11 +506,7 @@ struct l1_norm_op1 : public thrust::binary_function<HYPRE_Complex, HYPRE_Complex
 #endif
 
 #if defined(HYPRE_USING_GPU)
-#if defined(HYPRE_USING_SYCL)
 struct l1_norm_op6
-#else
-struct l1_norm_op6 : public thrust::binary_function<HYPRE_Complex, HYPRE_Complex, HYPRE_Complex>
-#endif
 {
    __host__ __device__
    HYPRE_Complex operator()(const HYPRE_Complex &d, const HYPRE_Complex &l) const

3rd_party/hypre/src/parcsr_ls/par_lr_restr_device.c

@@ -455,7 +455,7 @@ hypre_BoomerAMGBuildRestrNeumannAIR_assembleRdiag( hypre_DeviceItem    &item,
 }
 
 #if !defined(HYPRE_USING_SYCL)
-struct setTo1minus1 : public thrust::unary_function<HYPRE_Int, HYPRE_Int>
+struct setTo1minus1
 {
    __host__ __device__ HYPRE_Int operator()(const HYPRE_Int &x) const
    {

3rd_party/hypre/src/parcsr_ls/par_mgr_device.c

@@ -19,11 +19,7 @@
 #if defined (HYPRE_USING_GPU)
 
 template<typename T>
-#if defined(HYPRE_USING_SYCL)
 struct functor
-#else
-struct functor : public thrust::binary_function<T, T, T>
-#endif
 {
    T scale;
 

3rd_party/hypre/src/parcsr_ls/par_mod_multi_interp_device.c

@@ -51,8 +51,7 @@ struct globalC_functor
 };
 #else
 template<typename T>
-struct tuple_plus : public
-   thrust::binary_function<thrust::tuple<T, T>, thrust::tuple<T, T>, thrust::tuple<T, T> >
+struct tuple_plus
 {
    __host__ __device__
    thrust::tuple<T, T> operator()( const thrust::tuple<T, T> & x1, const thrust::tuple<T, T> & x2)
@@ -63,8 +62,7 @@ struct tuple_plus : public
 };
 
 template<typename T>
-struct tuple_minus : public
-   thrust::binary_function<thrust::tuple<T, T>, thrust::tuple<T, T>, thrust::tuple<T, T> >
+struct tuple_minus
 {
    __host__ __device__
    thrust::tuple<T, T> operator()( const thrust::tuple<T, T> & x1, const thrust::tuple<T, T> & x2)
@@ -74,8 +72,7 @@ struct tuple_minus : public
    }
 };
 
-struct local_equal_plus_constant : public
-   thrust::binary_function<HYPRE_BigInt, HYPRE_BigInt, HYPRE_BigInt>
+struct local_equal_plus_constant
 {
    HYPRE_BigInt _value;
 
@@ -86,7 +83,7 @@ struct local_equal_plus_constant : public
 };
 
 /* transform from local C index to global C index */
-struct globalC_functor : public thrust::unary_function<HYPRE_Int, HYPRE_BigInt>
+struct globalC_functor
 {
    HYPRE_BigInt C_first;
 

3rd_party/hypre/src/parcsr_mv/par_csr_fffc_device.c

@@ -21,11 +21,7 @@ typedef thrust::tuple<HYPRE_Int, HYPRE_Int> Tuple;
 /* transform from local F/C index to global F/C index,
  * where F index "x" are saved as "-x-1"
  */
-#if defined(HYPRE_USING_SYCL)
 struct FFFC_functor
-#else
-struct FFFC_functor : public thrust::unary_function<Tuple, HYPRE_BigInt>
-#endif
 {
    HYPRE_BigInt CF_first[2];
 
@@ -48,11 +44,7 @@ struct FFFC_functor : public thrust::unary_function<Tuple, HYPRE_BigInt>
 
 /* this predicate selects A^s_{FF} */
 template<typename T>
-#if defined(HYPRE_USING_SYCL)
 struct FF_pred
-#else
-struct FF_pred : public thrust::unary_function<Tuple, bool>
-#endif
 {
    HYPRE_Int  option;
    HYPRE_Int *row_CF_marker;
@@ -86,11 +78,7 @@ struct FF_pred : public thrust::unary_function<Tuple, bool>
 
 /* this predicate selects A^s_{FC} */
 template<typename T>
-#if defined(HYPRE_USING_SYCL)
 struct FC_pred
-#else
-struct FC_pred : public thrust::unary_function<Tuple, bool>
-#endif
 {
    HYPRE_Int *row_CF_marker;
    T         *col_CF_marker;
@@ -113,11 +101,7 @@ struct FC_pred : public thrust::unary_function<Tuple, bool>
 
 /* this predicate selects A^s_{CF} */
 template<typename T>
-#if defined(HYPRE_USING_SYCL)
 struct CF_pred
-#else
-struct CF_pred : public thrust::unary_function<Tuple, bool>
-#endif
 {
    HYPRE_Int *row_CF_marker;
    T         *col_CF_marker;
@@ -140,11 +124,7 @@ struct CF_pred : public thrust::unary_function<Tuple, bool>
 
 /* this predicate selects A^s_{CC} */
 template<typename T>
-#if defined(HYPRE_USING_SYCL)
 struct CC_pred
-#else
-struct CC_pred : public thrust::unary_function<Tuple, bool>
-#endif
 {
    HYPRE_Int *row_CF_marker;
    T         *col_CF_marker;
@@ -166,11 +146,7 @@ struct CC_pred : public thrust::unary_function<Tuple, bool>
 };
 
 /* this predicate selects A^s_{C,:} */
-#if defined(HYPRE_USING_SYCL)
 struct CX_pred
-#else
-struct CX_pred : public thrust::unary_function<Tuple, bool>
-#endif
 {
    HYPRE_Int *row_CF_marker;
 
@@ -191,11 +167,7 @@ struct CX_pred : public thrust::unary_function<Tuple, bool>
 
 /* this predicate selects A^s_{:,C} */
 template<typename T>
-#if defined(HYPRE_USING_SYCL)
 struct XC_pred
-#else
-struct XC_pred : public thrust::unary_function<Tuple, bool>
-#endif
 {
    T         *col_CF_marker;
 

3rd_party/hypre/src/parcsr_mv/par_csr_triplemat_device.c

@@ -17,11 +17,7 @@
  * option == 2, T = HYPRE_Int,
  */
 template<HYPRE_Int option, typename T>
-#if defined(HYPRE_USING_SYCL)
 struct RAP_functor
-#else
-struct RAP_functor : public thrust::unary_function<HYPRE_Int, T>
-#endif
 {
    HYPRE_Int num_col;
    T         first_col;

3rd_party/hypre/src/seq_mv/csr_matop_device.c

@@ -698,6 +698,7 @@ hypre_CSRMatrixSplitDevice_core( HYPRE_Int      job,
                         std::not_fn(pred1) );
       hypre_assert( std::get<0>(new_end.base()) == B_ext_offd_ii + B_ext_offd_nnz );
 #else
+
       auto new_end = HYPRE_THRUST_CALL(
                         copy_if,
                         thrust::make_zip_iterator(thrust::make_tuple(B_ext_ii,      B_ext_bigj,
@@ -1865,7 +1866,7 @@ struct Int2Unequal
 };
 #else
 typedef thrust::tuple<HYPRE_Int, HYPRE_Int> Int2;
-struct Int2Unequal : public thrust::unary_function<Int2, bool>
+struct Int2Unequal
 {
    __host__ __device__
    bool operator()(const Int2& t) const
@@ -2086,8 +2087,7 @@ struct cabsfirst_greaterthan_second_pred
    }
 };
 #else
-struct cabsfirst_greaterthan_second_pred : public
-   thrust::unary_function<thrust::tuple<HYPRE_Complex, HYPRE_Real>, bool>
+struct cabsfirst_greaterthan_second_pred
 {
    __host__ __device__
    bool operator()(const thrust::tuple<HYPRE_Complex, HYPRE_Real>& t) const
@@ -2346,11 +2346,7 @@ hypre_CSRMatrixDiagMatrixFromMatrixDevice(hypre_CSRMatrix *A, HYPRE_Int type)
  * adj_functor (Used in hypre_CSRMatrixPermuteDevice)
  *--------------------------------------------------------------------------*/
 
-#if defined(HYPRE_USING_SYCL)
 struct adj_functor
-#else
-struct adj_functor : public thrust::unary_function<HYPRE_Int, HYPRE_Int>
-#endif
 {
    HYPRE_Int *ia_;