diff --git a/common/src/KokkosKernels_default_types.hpp b/common/src/KokkosKernels_default_types.hpp index 672bdf3fbb..30ca52e300 100644 --- a/common/src/KokkosKernels_default_types.hpp +++ b/common/src/KokkosKernels_default_types.hpp @@ -25,7 +25,8 @@ using default_lno_t = int; #elif defined(KOKKOSKERNELS_INST_ORDINAL_INT64_T) using default_lno_t = int64_t; #else -using default_lno_t = int; +// Non-ETI build: default to int +using default_lno_t = int; #endif // Prefer int as the default offset type, because cuSPARSE doesn't support // size_t for rowptrs. @@ -34,6 +35,7 @@ using default_size_type = int; #elif defined(KOKKOSKERNELS_INST_OFFSET_SIZE_T) using default_size_type = size_t; #else +// Non-ETI build: default to int using default_size_type = int; #endif diff --git a/perf_test/sparse/KokkosSparse_spadd.cpp b/perf_test/sparse/KokkosSparse_spadd.cpp index f27d7d93db..3b347eb903 100644 --- a/perf_test/sparse/KokkosSparse_spadd.cpp +++ b/perf_test/sparse/KokkosSparse_spadd.cpp @@ -185,6 +185,31 @@ void run_experiment(int argc, char** argv, CommonInputParams) { "If running MKL, can't output the result to file"); } + // Check that offset/ordinal types are compatible with any requested TPLs +#ifdef KOKKOSKERNELS_ENABLE_TPL_MKL + if (params.use_mkl) { + if constexpr (!std::is_same_v) { + throw std::runtime_error( + "MKL configured with long long int not supported in Kokkos Kernels"); + } + if constexpr (!std::is_same_v || + !std::is_same_v) { + throw std::runtime_error( + "Must enable int as both ordinal and offset type in KokkosKernels to " + "call MKL SpAdd"); + } + } +#endif + + if (params.use_cusparse) { + if constexpr (!std::is_same_v || + !std::is_same_v) { + throw std::runtime_error( + "Must enable int as both ordinal and offset type in KokkosKernels to " + "call cuSPARSE SpAdd"); + } + } + std::cout << "************************************* \n"; crsMat_t A; crsMat_t B; @@ -319,9 +344,11 @@ void run_experiment(int argc, char** argv, CommonInputParams) { } #endif #ifdef KOKKOSKERNELS_ENABLE_TPL_MKL - sparse_matrix_t Amkl, Bmkl, Cmkl; + sparse_matrix_t Amkl = sparse_matrix_t(), Bmkl = sparse_matrix_t(), + Cmkl = sparse_matrix_t(); if (params.use_mkl) { - if constexpr (std::is_same_v) { + if constexpr (std::is_same_v && + std::is_same_v) { KOKKOSKERNELS_MKL_SAFE_CALL(mkl_sparse_d_create_csr( &Amkl, SPARSE_INDEX_BASE_ZERO, m, n, (int*)A.graph.row_map.data(), (int*)A.graph.row_map.data() + 1, A.graph.entries.data(), @@ -330,9 +357,6 @@ void run_experiment(int argc, char** argv, CommonInputParams) { &Bmkl, SPARSE_INDEX_BASE_ZERO, m, n, (int*)B.graph.row_map.data(), (int*)B.graph.row_map.data() + 1, B.graph.entries.data(), B.values.data())); - } else { - throw std::runtime_error( - "MKL configured with long long int not supported in Kokkos Kernels"); } } #endif @@ -347,22 +371,30 @@ void run_experiment(int argc, char** argv, CommonInputParams) { c_nnz = addHandle->get_c_nnz(); } else if (params.use_cusparse) { #ifdef KOKKOSKERNELS_ENABLE_TPL_CUSPARSE - // Symbolic phase: compute buffer size, then compute nnz - size_t bufferSize; - KOKKOS_CUSPARSE_SAFE_CALL(cusparseDcsrgeam2_bufferSizeExt( - cusparseHandle, A.numRows(), A.numCols(), &alphabeta, A_cusparse, - A.nnz(), A.values.data(), A.graph.row_map.data(), - A.graph.entries.data(), &alphabeta, B_cusparse, B.nnz(), - B.values.data(), B.graph.row_map.data(), B.graph.entries.data(), - C_cusparse, NULL, row_mapC.data(), NULL, &bufferSize)); - // Allocate work buffer - KOKKOS_IMPL_CUDA_SAFE_CALL( - cudaMalloc((void**)&cusparseBuffer, bufferSize)); - KOKKOS_CUSPARSE_SAFE_CALL(cusparseXcsrgeam2Nnz( - cusparseHandle, m, n, A_cusparse, A.nnz(), A.graph.row_map.data(), - A.graph.entries.data(), B_cusparse, B.nnz(), B.graph.row_map.data(), - B.graph.entries.data(), C_cusparse, row_mapC.data(), &c_nnz, - cusparseBuffer)); + if constexpr (std::is_same_v && + std::is_same_v) { + // Symbolic phase: compute buffer size, then compute nnz + size_t bufferSize; + KOKKOS_CUSPARSE_SAFE_CALL(cusparseDcsrgeam2_bufferSizeExt( + cusparseHandle, A.numRows(), A.numCols(), &alphabeta, A_cusparse, + A.nnz(), A.values.data(), A.graph.row_map.data(), + A.graph.entries.data(), &alphabeta, B_cusparse, B.nnz(), + B.values.data(), B.graph.row_map.data(), B.graph.entries.data(), + C_cusparse, NULL, row_mapC.data(), NULL, &bufferSize)); + // Allocate work buffer + KOKKOS_IMPL_CUDA_SAFE_CALL( + cudaMalloc((void**)&cusparseBuffer, bufferSize)); + KOKKOS_CUSPARSE_SAFE_CALL(cusparseXcsrgeam2Nnz( + cusparseHandle, m, n, A_cusparse, A.nnz(), A.graph.row_map.data(), + A.graph.entries.data(), B_cusparse, B.nnz(), B.graph.row_map.data(), + B.graph.entries.data(), C_cusparse, row_mapC.data(), &c_nnz, + cusparseBuffer)); + } else { + throw std::runtime_error( + "Must enable int as both ordinal and offset type in KokkosKernels " + "to " + "call cuSPARSE"); + } #endif } if (!params.use_mkl) { @@ -381,24 +413,32 @@ void run_experiment(int argc, char** argv, CommonInputParams) { for (int numericRep = 0; numericRep < params.numericRepeat; numericRep++) { if (params.use_cusparse) { #ifdef KOKKOSKERNELS_ENABLE_TPL_CUSPARSE - KOKKOS_CUSPARSE_SAFE_CALL(cusparseDcsrgeam2( - cusparseHandle, m, n, &alphabeta, A_cusparse, A.nnz(), - A.values.data(), A.graph.row_map.data(), A.graph.entries.data(), - &alphabeta, B_cusparse, B.nnz(), B.values.data(), - B.graph.row_map.data(), B.graph.entries.data(), C_cusparse, - valuesC.data(), row_mapC.data(), entriesC.data(), cusparseBuffer)); + if constexpr (std::is_same_v && + std::is_same_v) { + KOKKOS_CUSPARSE_SAFE_CALL(cusparseDcsrgeam2( + cusparseHandle, m, n, &alphabeta, A_cusparse, A.nnz(), + A.values.data(), A.graph.row_map.data(), A.graph.entries.data(), + &alphabeta, B_cusparse, B.nnz(), B.values.data(), + B.graph.row_map.data(), B.graph.entries.data(), C_cusparse, + valuesC.data(), row_mapC.data(), entriesC.data(), + cusparseBuffer)); + } #endif } else if (params.use_mkl) { #ifdef KOKKOSKERNELS_ENABLE_TPL_MKL - KOKKOSKERNELS_MKL_SAFE_CALL(mkl_sparse_d_add( - SPARSE_OPERATION_NON_TRANSPOSE, Amkl, 1.0, Bmkl, &Cmkl)); - KOKKOSKERNELS_MKL_SAFE_CALL(mkl_sparse_destroy(Cmkl)); + if constexpr (std::is_same_v && + std::is_same_v) { + KOKKOSKERNELS_MKL_SAFE_CALL(mkl_sparse_d_add( + SPARSE_OPERATION_NON_TRANSPOSE, Amkl, 1.0, Bmkl, &Cmkl)); + KOKKOSKERNELS_MKL_SAFE_CALL(mkl_sparse_destroy(Cmkl)); + } #endif } else { - spadd_numeric( - &kh, A.graph.row_map, A.graph.entries, A.values, 1.0, // A, alpha - B.graph.row_map, B.graph.entries, B.values, 1.0, // B, beta - row_mapC, entriesC, valuesC); // C + spadd_numeric(&kh, A.graph.row_map, A.graph.entries, A.values, + 1.0, // A, alpha + B.graph.row_map, B.graph.entries, B.values, + 1.0, // B, beta + row_mapC, entriesC, valuesC); // C } } numericTime += timer.seconds();