Allow using GPU-aware MPI via env variable

packages/adelus/CMakeLists.txt

@@ -54,10 +54,10 @@ TRIBITS_ADD_OPTION_AND_DEFINE(${PACKAGE_NAME}_ENABLE_TIMING
   "Enable timing."
   OFF NONCACHE)
 
-TRIBITS_ADD_OPTION_AND_DEFINE(${PACKAGE_NAME}_ENABLE_HOSTPINNEDMEM
-  ADELUS_HOST_PINNED_MEM_MPI
-  "Use Cuda/HIP Host Pinned memory for MPI."
-  ON )
+TRIBITS_ADD_OPTION_AND_DEFINE(${PACKAGE_NAME}_USE_GPU_AWARE_MPI
+  ADELUS_USE_GPU_AWARE_MPI
+  "Build with GPU-aware MPI."
+  OFF )
 
 TRIBITS_ADD_OPTION_AND_DEFINE(${PACKAGE_NAME}_ENABLE_USEDEEPCOPY
   USE_DEEPCOPY

packages/adelus/README.md

@@ -161,8 +161,8 @@ options can be found below.
 * ```Adelus_ENABLE_TIMING```
   * Whether to enable internal solver timing
   * ```BOOL``` Default: OFF
-* ```Adelus_ENABLE_HOSTPINNEDMEM```
-  * Whether to use Cuda/HIP Host Pinned memory for MPI
+* ```Adelus_USE_GPU_AWARE_MPI```
+  * Whether to use GPU-aware MPI
   * ```BOOL``` Default: OFF
 * ```Adelus_ENABLE_USEDEEPCOPY```
   * Whether to Use Kokkos::deep_copy for BLAS copy
@@ -171,6 +171,12 @@ options can be found below.
   * Whether to enable status prints
   * ```BOOL``` Default: OFF
 
+ Users can tell Adelus whether it uses GPU-aware MPI via the environment variable
+ ```ADELUS_USE_GPU_AWARE_MPI``` (```export ADELUS_USE_GPU_AWARE_MPI=1``` to enable or
+```export ADELUS_USE_GPU_AWARE_MPI=0``` to disable). The CMake build flag
+```-D Adelus_USE_GPU_AWARE_MPI:BOOL=ON/OFF``` sets up the default behavior for Adelus
+at the compile time, while the environment variable acts as a runtime override. 
+
  We refer readers to Trilinos', Kokkos', and Kokkos Kernels' documentations for
 further details of building Trilinos, Kokkos, and Kokkos Kernels.
 
@@ -236,7 +242,7 @@ cmake \
 \
 -D Adelus_ENABLE_ZCPLX:BOOL=ON \
 -D Adelus_ENABLE_TIMING:BOOL=ON \
--D Adelus_ENABLE_HOSTPINNEDMEM:BOOL=OFF \
+-D Adelus_USE_GPU_AWARE_MPI:BOOL=ON \
 -D Adelus_ENABLE_PRINTSTATUS:BOOL=ON \
 \
 ${TRILINOS_HOME}/Trilinos

packages/adelus/cmake/Adelus_config.h.in

@@ -55,8 +55,8 @@
 /* Whether to enable timing */
 #cmakedefine GET_TIMING
 
-/* Whether to use Cuda/HIP Host Pinned memory for MPI */
-#cmakedefine ADELUS_HOST_PINNED_MEM_MPI
+/* Whether to enable building with GPU-aware MPI */
+#cmakedefine ADELUS_USE_GPU_AWARE_MPI
 
 /* Whether to use Kokkos::deep_copy for BLAS copy */
 #cmakedefine USE_DEEPCOPY

packages/adelus/cmake/Dependencies.cmake

@@ -1,7 +1,7 @@
-SET(LIB_REQUIRED_DEP_PACKAGES Kokkos KokkosKernels)
+SET(LIB_REQUIRED_DEP_PACKAGES Kokkos KokkosKernels Teuchos)
 SET(LIB_OPTIONAL_DEP_PACKAGES)
 SET(TEST_REQUIRED_DEP_PACKAGES)
-SET(TEST_OPTIONAL_DEP_PACKAGES Teuchos)
+SET(TEST_OPTIONAL_DEP_PACKAGES)
 SET(LIB_REQUIRED_DEP_TPLS BLAS)
 SET(LIB_OPTIONAL_DEP_TPLS)
 SET(TEST_REQUIRED_DEP_TPLS)

packages/adelus/src/Adelus.hpp

@@ -18,6 +18,7 @@
 #include <Adelus_global_comm.hpp>
 #include <Adelus_defines.h>
 #include <Adelus_distribute.hpp>
+#include <Adelus_mpi_behavior.hpp>
 #include <Adelus_xlu_solve.hpp>
 #include <Adelus_x_factor.hpp>
 #include <Adelus_x_solve.hpp>

packages/adelus/src/Adelus_factor.hpp

@@ -65,12 +65,21 @@ void factor(HandleType& ahandle,           // handle containg metadata
   using execution_space = typename ZDView::device_type::execution_space;
   using memory_space    = typename ZDView::device_type::memory_space;
 #endif
-#ifdef ADELUS_HOST_PINNED_MEM_MPI
 #if defined(KOKKOS_ENABLE_CUDA)
   using View1DHostPinnType = Kokkos::View<value_type*, Kokkos::LayoutLeft, Kokkos::CudaHostPinnedSpace>;//CudaHostPinnedSpace
 #elif defined(KOKKOS_ENABLE_HIP)
   using View1DHostPinnType = Kokkos::View<value_type*, Kokkos::LayoutLeft, Kokkos::HIPHostPinnedSpace>;//HIPHostPinnedSpace
+#else
+  using View1DHostPinnType = Kokkos::View<value_type*, Kokkos::LayoutLeft>;//fallback placeholder
 #endif
+
+  constexpr bool isOnDeviceSpace =
+#if defined( KOKKOS_ENABLE_CUDA )
+    std::is_same_v<memory_space, Kokkos::CudaSpace>;
+#elif defined( KOKKOS_ENABLE_HIP )
+    std::is_same_v<memory_space, Kokkos::HIPSpace>;
+#else
+    false;
 #endif
 
   MPI_Comm comm     = ahandle.get_comm();
@@ -137,9 +146,7 @@ void factor(HandleType& ahandle,           // handle containg metadata
   double t1,t2;
   double msgtime,copytime,dgemmtime,totalfactortime;
   double iamaxtime,getlocalpivtime,localpivtime;
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined(KOKKOS_ENABLE_HIP))
   double copyhostpinnedtime;
-#endif
 #endif
 
   MPI_Request msgrequest;
@@ -194,9 +201,7 @@ void factor(HandleType& ahandle,           // handle containg metadata
   copycoltime=copyrowtime=copyrow1time=copypivrowtime=copypivrow1time=pivotswaptime=0.0;
   updatetime=colupdtime=rowupdtime=scaltime=0.0;
   iamaxtime=getlocalpivtime=localpivtime=0.0;
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
-  copyhostpinnedtime=0.0;
-#endif
+  if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace) copyhostpinnedtime=0.0;
 #endif
 
 #ifdef PRINT_STATUS
@@ -210,11 +215,12 @@ void factor(HandleType& ahandle,           // handle containg metadata
   }
 #endif
 
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
-  View1DHostPinnType h_coltmp( "h_coltmp", my_rows );
-  View1DHostPinnType h_row2  ( "h_row2",   my_cols + blksz + nrhs );
-  View1DHostPinnType h_row3  ( "h_row3",   my_cols + blksz + nrhs );
-#endif
+  View1DHostPinnType h_coltmp, h_row2, h_row3;
+  if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace) {
+    h_coltmp = View1DHostPinnType( Kokkos::view_alloc(Kokkos::WithoutInitializing, "h_coltmp"), my_rows );
+    h_row2   = View1DHostPinnType( Kokkos::view_alloc(Kokkos::WithoutInitializing, "h_row2"), my_cols + blksz + nrhs );
+    h_row3   = View1DHostPinnType( Kokkos::view_alloc(Kokkos::WithoutInitializing, "h_row3"), my_cols + blksz + nrhs );
+  }
 
   //Kokkos::fence();
   for (j=0; j<ncols_matrix; j++) {
@@ -391,27 +397,26 @@ void factor(HandleType& ahandle,           // handle containg metadata
       bcastpivstime += (MPI_Wtime()-t1);
 #endif
 
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
+      if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace) {
 #ifdef GET_TIMING
-      t1 = MPI_Wtime();
+        t1 = MPI_Wtime();
 #endif
-      Kokkos::deep_copy(subview(h_coltmp, Kokkos::make_pair(0, col_len)), subview(col1_view, Kokkos::make_pair(sav_col_i, sav_col_i+col_len), sav_col_j));
+        Kokkos::deep_copy(subview(h_coltmp, Kokkos::make_pair(0, col_len)), subview(col1_view, Kokkos::make_pair(sav_col_i, sav_col_i+col_len), sav_col_j));
 #ifdef GET_TIMING
-      copyhostpinnedtime += (MPI_Wtime()-t1);
-#endif
+        copyhostpinnedtime += (MPI_Wtime()-t1);
 #endif
+      }
 
 #ifdef GET_TIMING
       t1 = MPI_Wtime();
 #endif
       for (rdist = 1;rdist <= MAXDIST;rdist++) {
         if (rowplus(rdist) == c_owner) break;
         bytes=sizeof(ADELUS_DATA_TYPE)*col_len;
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
-        MPI_Send(h_coltmp.data(),bytes,MPI_BYTE,rowplus(rdist),LUROWTYPE+j,comm);
-#else //GPU-aware MPI
-        MPI_Send(col1_view.data()+sav_col_j*col1_view.stride(1)+sav_col_i,bytes,MPI_BYTE,rowplus(rdist),LUROWTYPE+j,comm);
-#endif
+        if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace)
+          MPI_Send(h_coltmp.data(),bytes,MPI_BYTE,rowplus(rdist),LUROWTYPE+j,comm);
+        else //GPU-aware MPI
+          MPI_Send(col1_view.data()+sav_col_j*col1_view.stride(1)+sav_col_i,bytes,MPI_BYTE,rowplus(rdist),LUROWTYPE+j,comm);
       }
 #ifdef GET_TIMING
       bcastcolstime += (MPI_Wtime()-t1);
@@ -433,12 +438,11 @@ void factor(HandleType& ahandle,           // handle containg metadata
       // recv column and pivot
 
       bytes=col_len*sizeof(ADELUS_DATA_TYPE);
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
-      MPI_Irecv(h_coltmp.data(),bytes,MPI_BYTE,MPI_ANY_SOURCE,LUROWTYPE+j,comm,&msgrequest);
-#else //GPU-aware MPI
-      MPI_Irecv(col1_view.data()+sav_col_j*col1_view.stride(1)+sav_col_i,bytes,MPI_BYTE,
-                MPI_ANY_SOURCE,LUROWTYPE+j,comm,&msgrequest);
-#endif
+      if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace)
+        MPI_Irecv(h_coltmp.data(),bytes,MPI_BYTE,MPI_ANY_SOURCE,LUROWTYPE+j,comm,&msgrequest);
+      else //GPU-aware MPI
+        MPI_Irecv(col1_view.data()+sav_col_j*col1_view.stride(1)+sav_col_i,bytes,MPI_BYTE,
+                  MPI_ANY_SOURCE,LUROWTYPE+j,comm,&msgrequest);
 
 #ifdef GET_TIMING
       t1 = MPI_Wtime();
@@ -474,28 +478,27 @@ void factor(HandleType& ahandle,           // handle containg metadata
         bcastcolrtime += (MPI_Wtime()-t1);
 #endif
 
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
+        if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace) {
 #ifdef GET_TIMING
-        t1 = MPI_Wtime();
+          t1 = MPI_Wtime();
 #endif
-        bytes=col_len*sizeof(ADELUS_DATA_TYPE);
-        Kokkos::deep_copy(subview(col1_view, Kokkos::make_pair(sav_col_i, sav_col_i+col_len), sav_col_j), subview(h_coltmp, Kokkos::make_pair(0, col_len)));
+          bytes=col_len*sizeof(ADELUS_DATA_TYPE);
+          Kokkos::deep_copy(subview(col1_view, Kokkos::make_pair(sav_col_i, sav_col_i+col_len), sav_col_j), subview(h_coltmp, Kokkos::make_pair(0, col_len)));
 #ifdef GET_TIMING
-        copyhostpinnedtime += (MPI_Wtime()-t1);
-#endif
+          copyhostpinnedtime += (MPI_Wtime()-t1);
 #endif
+        }
 
 #ifdef GET_TIMING
         t1 = MPI_Wtime();
 #endif
         for (rdist = 1;rdist <= MAXDIST;rdist++) {
           if (rowplus(rdist) == c_owner) break;
           bytes=col_len*sizeof(ADELUS_DATA_TYPE);
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
-          MPI_Send(h_coltmp.data(),bytes,MPI_BYTE,rowplus(rdist),LUROWTYPE+j,comm);
-#else //GPU-aware MPI
-          MPI_Send(col1_view.data()+sav_col_j*col1_view.stride(1)+sav_col_i,bytes,MPI_BYTE,rowplus(rdist),LUROWTYPE+j,comm);
-#endif
+          if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace)
+            MPI_Send(h_coltmp.data(),bytes,MPI_BYTE,rowplus(rdist),LUROWTYPE+j,comm);
+          else //GPU-aware MPI
+            MPI_Send(col1_view.data()+sav_col_j*col1_view.stride(1)+sav_col_i,bytes,MPI_BYTE,rowplus(rdist),LUROWTYPE+j,comm);
         }
 #ifdef GET_TIMING
         bcastcolstime += (MPI_Wtime()-t1);
@@ -629,40 +632,38 @@ void factor(HandleType& ahandle,           // handle containg metadata
 
     // broadcast pivot row
 
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
+    if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace) {
 #ifdef GET_TIMING
-    t1 = MPI_Wtime();
+      t1 = MPI_Wtime();
 #endif
-    Kokkos::deep_copy(subview(h_row3, Kokkos::make_pair(0, row_len + colcnt)), subview(row3_view, Kokkos::make_pair(0, row_len + colcnt)));
+      Kokkos::deep_copy(subview(h_row3, Kokkos::make_pair(0, row_len + colcnt)), subview(row3_view, Kokkos::make_pair(0, row_len + colcnt)));
 #ifdef GET_TIMING
-    copyhostpinnedtime += (MPI_Wtime()-t1);
-#endif
+      copyhostpinnedtime += (MPI_Wtime()-t1);
 #endif
+    }
 
 #ifdef GET_TIMING
     t1 = MPI_Wtime();
 #endif
     bytes=sizeof(ADELUS_DATA_TYPE)*row_size ;
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
-    MPI_Bcast(reinterpret_cast<char *>(h_row3.data()), row_size, MPI_CHAR, mesh_row(pivot_owner), col_comm);
-    MPI_Barrier(col_comm);
-#else //GPU-aware MPI -- Note: Looks like MPI_Bcast is still working well with device (cuda) pointers (and faster than using cuda host pinned memory) for 2 nodes
-    MPI_Bcast(reinterpret_cast<char *>(row3_view.data()), row_size, MPI_CHAR, mesh_row(pivot_owner), col_comm);
+    if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace)
+      MPI_Bcast(reinterpret_cast<char *>(h_row3.data()), row_size, MPI_CHAR, mesh_row(pivot_owner), col_comm);
+    else //GPU-aware MPI -- Note: Looks like MPI_Bcast is still working well with device (cuda) pointers (and faster than using cuda host pinned memory) for 2 nodes
+      MPI_Bcast(reinterpret_cast<char *>(row3_view.data()), row_size, MPI_CHAR, mesh_row(pivot_owner), col_comm);
     MPI_Barrier(col_comm);
-#endif
 #ifdef GET_TIMING
     bcastrowtime += (MPI_Wtime()-t1);
 #endif
 
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
+    if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace) {
 #ifdef GET_TIMING
-    t1 = MPI_Wtime();
+      t1 = MPI_Wtime();
 #endif
-    Kokkos::deep_copy(subview(row3_view, Kokkos::make_pair(0, row_len + colcnt)), subview(h_row3, Kokkos::make_pair(0, row_len + colcnt)));
+      Kokkos::deep_copy(subview(row3_view, Kokkos::make_pair(0, row_len + colcnt)), subview(h_row3, Kokkos::make_pair(0, row_len + colcnt)));
 #ifdef GET_TIMING
-    copyhostpinnedtime += (MPI_Wtime()-t1);
-#endif
+      copyhostpinnedtime += (MPI_Wtime()-t1);
 #endif
+    }
 
 #ifdef GET_TIMING
     t1 = MPI_Wtime();
@@ -685,25 +686,24 @@ void factor(HandleType& ahandle,           // handle containg metadata
 
     if (gpivot_row != j) {
       if (me != pivot_owner && me == r_owner) {
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
+        if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace) {
 #ifdef GET_TIMING
-        t1 = MPI_Wtime();
+          t1 = MPI_Wtime();
 #endif
-        Kokkos::deep_copy(subview(h_row2, Kokkos::make_pair(0, row_len + colcnt)), subview(row2_view, Kokkos::make_pair(0, row_len + colcnt)));
+          Kokkos::deep_copy(subview(h_row2, Kokkos::make_pair(0, row_len + colcnt)), subview(row2_view, Kokkos::make_pair(0, row_len + colcnt)));
 #ifdef GET_TIMING
-        copyhostpinnedtime += (MPI_Wtime()-t1);
-#endif
+          copyhostpinnedtime += (MPI_Wtime()-t1);
 #endif
+        }
 
 #ifdef GET_TIMING
         t1 = MPI_Wtime();
 #endif
         bytes=(row_len+colcnt)*sizeof(ADELUS_DATA_TYPE);
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
-        MPI_Send(h_row2.data(),bytes,MPI_BYTE,pivot_owner,LUSENDTYPE+j,comm);
-#else //GPU-aware MPI
-        MPI_Send(row2_view.data(),bytes,MPI_BYTE,pivot_owner,LUSENDTYPE+j,comm);
-#endif
+        if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace)
+          MPI_Send(h_row2.data(),bytes,MPI_BYTE,pivot_owner,LUSENDTYPE+j,comm);
+        else //GPU-aware MPI
+          MPI_Send(row2_view.data(),bytes,MPI_BYTE,pivot_owner,LUSENDTYPE+j,comm);
 #ifdef GET_TIMING
         sendrowtime += (MPI_Wtime()-t1);
 #endif
@@ -716,26 +716,25 @@ void factor(HandleType& ahandle,           // handle containg metadata
 #endif
         if (me != r_owner) {
           bytes=(row_len+colcnt)*sizeof(ADELUS_DATA_TYPE);
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
-          MPI_Recv(h_row2.data(),bytes,MPI_BYTE,r_owner,LUSENDTYPE+j,comm,&msgstatus);
-#else //GPU-aware MPI
-          MPI_Recv(row2_view.data(),bytes,MPI_BYTE,r_owner,LUSENDTYPE+j,comm,&msgstatus);
-#endif
+          if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace)
+            MPI_Recv(h_row2.data(),bytes,MPI_BYTE,r_owner,LUSENDTYPE+j,comm,&msgstatus);
+          else //GPU-aware MPI
+            MPI_Recv(row2_view.data(),bytes,MPI_BYTE,r_owner,LUSENDTYPE+j,comm,&msgstatus);
         }
 #ifdef GET_TIMING
         recvrowtime += (MPI_Wtime()-t1);
 #endif
 
         if (me != r_owner) {
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
+          if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace) {
 #ifdef GET_TIMING
-          t1 = MPI_Wtime();
+            t1 = MPI_Wtime();
 #endif
-          Kokkos::deep_copy(subview(row2_view, Kokkos::make_pair(0, row_len + colcnt)), subview(h_row2, Kokkos::make_pair(0, row_len + colcnt)));
+            Kokkos::deep_copy(subview(row2_view, Kokkos::make_pair(0, row_len + colcnt)), subview(h_row2, Kokkos::make_pair(0, row_len + colcnt)));
 #ifdef GET_TIMING
-          copyhostpinnedtime += (MPI_Wtime()-t1);
-#endif
+            copyhostpinnedtime += (MPI_Wtime()-t1);
 #endif
+          }
         }
 
 #ifdef GET_TIMING
@@ -805,15 +804,15 @@ void factor(HandleType& ahandle,           // handle containg metadata
       bcastcolrtime += (MPI_Wtime()-t1);
 #endif
 
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
+      if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace) {
 #ifdef GET_TIMING
-      t1 = MPI_Wtime();
+        t1 = MPI_Wtime();
 #endif
-      Kokkos::deep_copy(subview(col1_view, Kokkos::make_pair(sav_col_i, sav_col_i+col_len), sav_col_j), subview(h_coltmp, Kokkos::make_pair(0, col_len)));
+        Kokkos::deep_copy(subview(col1_view, Kokkos::make_pair(sav_col_i, sav_col_i+col_len), sav_col_j), subview(h_coltmp, Kokkos::make_pair(0, col_len)));
 #ifdef GET_TIMING
-      copyhostpinnedtime += (MPI_Wtime()-t1);
-#endif
+        copyhostpinnedtime += (MPI_Wtime()-t1);
 #endif
+      }
     }
     // saved this active row and column so get ready for next ones
 
@@ -926,11 +925,11 @@ void factor(HandleType& ahandle,           // handle containg metadata
   showtime(ahandle.get_comm_id(),comm,me,numprocs,"Total msg passing time",&msgtime);
   tmp = 100*msgtime/totalfactortime;
   showtime(ahandle.get_comm_id(),comm,me,numprocs,"Percent msg passing time",&tmp);
-#if defined(ADELUS_HOST_PINNED_MEM_MPI) && (defined(KOKKOS_ENABLE_CUDA) || defined (KOKKOS_ENABLE_HIP))
-  showtime(ahandle.get_comm_id(),comm,me,numprocs,"Total copy between host pinned mem and dev mem time",&copyhostpinnedtime);
-  tmp = 100*copyhostpinnedtime/totalfactortime;
-  showtime(ahandle.get_comm_id(),comm,me,numprocs,"Percent copy between host pinned mem and dev mem time",&tmp);
-#endif
+  if (!(ahandle.get_gpuawarempi_behavior()) && isOnDeviceSpace) {
+    showtime(ahandle.get_comm_id(),comm,me,numprocs,"Total copy between host pinned mem and dev mem time",&copyhostpinnedtime);
+    tmp = 100*copyhostpinnedtime/totalfactortime;
+    showtime(ahandle.get_comm_id(),comm,me,numprocs,"Percent copy between host pinned mem and dev mem time",&tmp);
+  }
 #ifdef ADELUS_SHOW_TIMING_DETAILS
   showtime(ahandle.get_comm_id(),comm,me,numprocs,"Time to swap pivot",&pivotswaptime);
   showtime(ahandle.get_comm_id(),comm,me,numprocs,"Time to copy cur col",&copycoltime);