add the default MPI broadcast routine

CMakeLists.txt

@@ -34,6 +34,7 @@ set( CMAKE_CXX_STANDARD 14 )
 set(CMAKE_CXX_FLAGS_DEBUG "${CMAKE_CXX_FLAGS_DEBUG} -O0")
 set(CMAKE_CXX_FLAGS_Release "${CMAKE_C_FLAGS_Release} -O3")
 ADD_DEFINITIONS(-DROCM=1)
+ADD_DEFINITIONS(-DHPL_COPY_L=1)
 file(COPY ${CMAKE_CURRENT_SOURCE_DIR}/HPL.dat
      DESTINATION ${CMAKE_CURRENT_BINARY_DIR})
 file(COPY ${CMAKE_CURRENT_SOURCE_DIR}/scripts/runHPL_singleNode.sh
@@ -116,7 +117,7 @@ src/comm/HPL_blonM.cpp src/comm/HPL_1ring.cpp src/comm/HPL_2ring.cpp
 src/comm/HPL_1rinM.cpp src/comm/HPL_2rinM.cpp src/comm/HPL_packL.cpp
 src/comm/HPL_sdrv.cpp src/comm/HPL_send.cpp src/pgesv/HPL_pdlaswp00N.cpp
 src/comm/HPL_recv.cpp src/grid/HPL_reduce.cpp src/comm/HPL_binit.cpp
-src/comm/HPL_bwait.cpp
+src/comm/HPL_bwait.cpp src/comm/HPL_copyL.cpp
 src/pgesv/HPL_pdlaswp00T.cpp src/pgesv/HPL_pdlaswp01N.cpp
 src/pgesv/HPL_pdlaswp01T.cpp src/pgesv/HPL_pdupdateNT.cpp
 src/pgesv/HPL_pdupdateTN.cpp src/pgesv/HPL_pdupdateTT.cpp

HPL.dat

@@ -20,7 +20,7 @@ HPL.out      output file name (if any)
 1            # of recursive panel fact.
 2            RFACTs (0=left, 1=Crout, 2=Right)
 1            # of broadcast
-3            BCASTs (0=1rg,1=1rM,2=2rg,3=2rM,4=Lng,5=LnM,6=ibcast)
+6            BCASTs (0=1rg,1=1rM,2=2rg,3=2rM,4=Lng,5=LnM,6=ibcast)
 1            # of lookahead depth
 1            DEPTHs (>=0)
 1            SWAP (0=bin-exch,1=long,2=mix)

include/backend/hpl_backendHIP.h

@@ -93,6 +93,10 @@ namespace HIP {
     void stream_synchronize(enum HPL_STREAM);
     void stream_wait_event(enum HPL_STREAM, enum HPL_EVENT);
     void device_sync();
+
+    int binit_ibcst(HPL_T_panel*);
+    int bcast_ibcst(HPL_T_panel*, int*);
+    int bwait_ibcst(HPL_T_panel*);
 /*
 *  ----------------------------------------------------------------------
 *  - BLAS ---------------------------------------------------------------

include/backend/hpl_backendWrapper.h

@@ -52,6 +52,13 @@ void HPL_BE_stream_synchronize(enum HPL_STREAM, enum HPL_TARGET);
 void HPL_BE_stream_wait_event(enum HPL_STREAM, enum HPL_EVENT, enum HPL_TARGET);
 
 void HPL_BE_stream_sync(enum HPL_STREAM, enum HPL_TARGET);
+
+/*
+    Broadcast routine
+*/
+void HPL_BE_binit_ibcast(HPL_T_panel*);
+void HPL_BE_bcast_ibcast(HPL_T_panel*, int*);
+int HPL_BE_bwait_ibcast(HPL_T_panel*);
 /*
 *  ----------------------------------------------------------------------
 *  - BLAS ---------------------------------------------------------------

include/hpl_comm.h

@@ -64,7 +64,8 @@ typedef enum
    HPL_2RING         = 403,                      /* Increasing 2-ring */
    HPL_2RING_M       = 404,           /* Increasing 2-ring (modified) */
    HPL_BLONG         = 405,                         /* long broadcast */
-   HPL_BLONG_M       = 406               /* long broadcast (modified) */
+   HPL_BLONG_M       = 406,               /* long broadcast (modified) */
+   HPL_IBCAST       = 407               /* default bcast in MPI */
 } HPL_T_TOP;
 /*
  * ---------------------------------------------------------------------

scripts/runHPL_singleNode.sh

@@ -9,5 +9,5 @@ export OMP_NUM_THREADS=${num_cpu_cores}
 export LD_LIBRARY_PATH=openblas:$LD_LIBRARY_PATH
 
 # ./xhpl
-HSA_ENABLE_SDMA=1 ${MPI_DIR}/mpirun --allow-run-as-root -np ${num_process} --map-by node:PE=${num_cpu_cores} --bind-to core:overload-allowed --report-bindings ./xhplhip
+HSA_ENABLE_SDMA=1 ${MPI_DIR}/mpirun  -mca btl '^openib' --mca pml ucx -x UCX_RNDV_PIPELINE_SEND_THRESH=256k -x UCX_RNDV_FRAG_SIZE=rocm:4m --allow-run-as-root -np ${num_process} --map-by node:PE=${num_cpu_cores} --bind-to core:overload-allowed --report-bindings ./xhplhip
 grep --color "e+" HPL.out

src/comm/HPL_bcast.cpp

@@ -108,6 +108,7 @@ int HPL_bcast
       case HPL_2RING   : ierr = HPL_bcast_2ring( PANEL, IFLAG ); break;
       case HPL_BLONG_M : ierr = HPL_bcast_blonM( PANEL, IFLAG ); break;
       case HPL_BLONG   : ierr = HPL_bcast_blong( PANEL, IFLAG ); break;
+      case HPL_IBCAST  : ierr = HPL_BE_bcast_ibcast( PANEL, IFLAG ); break;
       default          : ierr = HPL_SUCCESS;
    }
 

src/comm/HPL_binit.cpp

@@ -98,6 +98,7 @@ int HPL_binit
       case HPL_2RING   : ierr = HPL_binit_2ring( PANEL ); break;
       case HPL_BLONG_M : ierr = HPL_binit_blonM( PANEL ); break;
       case HPL_BLONG   : ierr = HPL_binit_blong( PANEL ); break;
+      case HPL_IBCAST  : ierr = HPL_BE_binit_ibcast( PANEL ); break;
       default          : ierr = HPL_SUCCESS;
    }
 

src/comm/HPL_bwait.cpp

@@ -99,6 +99,7 @@ int HPL_bwait
       case HPL_2RING   : ierr = HPL_bwait_2ring( PANEL ); break;
       case HPL_BLONG_M : ierr = HPL_bwait_blonM( PANEL ); break;
       case HPL_BLONG   : ierr = HPL_bwait_blong( PANEL ); break;
+      case HPL_IBCAST  : ierr = HPL_BE_bwait_ibcast( PANEL ); break;
       default          : ierr = HPL_SUCCESS;
    }
 

src/comm/HPL_packL.cpp

@@ -117,7 +117,7 @@ int HPL_packL
  * Panel + L1 + DPIV  have been copied into a contiguous buffer - Create
  * and commit a contiguous data type
  */
-   PANEL->buffers[IBUF] = (void *)(PANEL->L2 + INDEX);
+   PANEL->buffers[IBUF] = (void ***)(PANEL->L2 + INDEX);
    PANEL->counts [IBUF] = 1;
 
    ierr =      MPI_Type_contiguous( LEN, MPI_DOUBLE, &PANEL->dtypes[IBUF] );

testing/backend/HPL_backendHIP.cpp

@@ -939,3 +939,59 @@ void HIP::pdlaswp(HPL_T_panel *PANEL, const int NN){
     hipLaunchKernelGGL(_dlaswp00N, dim3(grid_size), dim3(block_size), 0, pdlaswpStream,
                                       nn, jb, Aptr, lda, ipiv);
 }
+
+int HIP::binit_ibcst(HPL_T_panel* PANEL) {
+
+    return (HPL_SUCCESS);
+}
+
+#define _M_BUFF (void*)(PANEL->dL2)
+#define _M_COUNT PANEL->len
+#define _M_TYPE MPI_DOUBLE
+
+static MPI_Request request  = MPI_REQUEST_NULL;
+static MPI_Request request2 = MPI_REQUEST_NULL;
+
+int HIP::bcast_ibcst(HPL_T_panel* PANEL, int* IFLAG) {
+  MPI_Comm comm;
+  int      ierr, ierr2, go, next, msgid, prev, rank, root, size;
+
+  if(PANEL == NULL) {
+    *IFLAG = HPL_SUCCESS;
+    return (HPL_SUCCESS);
+  }
+  if((size = PANEL->grid->npcol) <= 1) {
+    *IFLAG = HPL_SUCCESS;
+    return (HPL_SUCCESS);
+  }
+
+  rank  = PANEL->grid->mycol;
+  comm  = PANEL->grid->row_comm;
+  root  = PANEL->pcol;
+  msgid = PANEL->msgid;
+
+  ierr  = MPI_Ibcast(_M_BUFF, _M_COUNT, _M_TYPE, root, comm, &request);
+  ierr2 = MPI_Ibcast(PANEL->dIWORK, PANEL->jb * 2, MPI_INT, root, comm, &request2);
+  /*
+   * If the message was received and being forwarded,  return HPL_SUCCESS.
+   * If an error occured in an MPI call, return HPL_FAILURE.
+   */
+  *IFLAG = (ierr == MPI_SUCCESS ? HPL_SUCCESS : HPL_FAILURE);
+  *IFLAG = (ierr2 == MPI_SUCCESS ? *IFLAG : HPL_FAILURE);
+
+  return (*IFLAG);
+}
+
+int HIP::bwait_ibcst(HPL_T_panel* PANEL) {
+  int ierr1, ierr2;
+
+  if(PANEL == NULL) { return (HPL_SUCCESS); }
+  if(PANEL->grid->npcol <= 1) { return (HPL_SUCCESS); }
+
+  ierr1 = MPI_Wait(&request, MPI_STATUS_IGNORE);
+  ierr2 = MPI_Wait(&request2, MPI_STATUS_IGNORE);
+
+  return ((ierr1 == MPI_SUCCESS
+               ? (ierr2 == MPI_SUCCESS ? HPL_SUCCESS : HPL_FAILURE)
+               : HPL_FAILURE));
+}

testing/backend/HPL_backendWrapper.cpp

@@ -236,6 +236,51 @@ extern "C" {
       }
    }
 
+   /*
+   * Broadcast routine
+   */
+   void HPL_BE_binit_ibcast(HPL_T_panel* PANEL)
+   {
+      switch(TR) {
+         case T_CPU :
+            DO_NOTHING();
+            break;
+         case T_HIP:
+            HPL::dispatch(HIP::binit_ibcst, PANEL);
+            break;
+         default:
+            DO_NOTHING();
+      }
+   }
+
+   void HPL_BE_bcast_ibcast(HPL_T_panel* PANEL, int* FLAG)
+   {
+      switch(TR) {
+         case T_CPU :
+            DO_NOTHING();
+            break;
+         case T_HIP:
+            HPL::dispatch(HIP::bcast_ibcst, PANEL, FLAG);
+            break;
+         default:
+            DO_NOTHING();
+      }
+   }
+
+   int HPL_BE_bwait_ibcast(HPL_T_panel* PANEL)
+   {
+      switch(TR) {
+         case T_CPU :
+            DO_NOTHING();
+            break;
+         case T_HIP:
+            HPL::dispatch(HIP::bwait_ibcst, PANEL);
+            break;
+         default:
+            DO_NOTHING();
+      }
+   }
+
 /*
 *  ----------------------------------------------------------------------
 *  - BLAS ---------------------------------------------------------------

testing/ptest/HPL_pdinfo.cpp

@@ -538,6 +538,7 @@ void HPL_pdinfo
          else if( j == 3 ) TP[ i ] = HPL_2RING_M;
          else if( j == 4 ) TP[ i ] = HPL_BLONG;
          else if( j == 5 ) TP[ i ] = HPL_BLONG_M;
+         else if( j == 6 ) TP[ i ] = HPL_IBCAST;
          else              TP[ i ] = HPL_1RING_M;
       }
 /*
@@ -702,6 +703,7 @@ void HPL_pdinfo
          else if( TP[i] == HPL_2RING_M ) iwork[j] = 3;
          else if( TP[i] == HPL_BLONG   ) iwork[j] = 4;
          else if( TP[i] == HPL_BLONG_M ) iwork[j] = 5;
+         else if( TP[i] == HPL_IBCAST )  iwork[j] = 6;
          j++;
       }
       for( i = 0; i < *NDHS; i++ ) { iwork[j] = DH[i]; j++; }
@@ -745,6 +747,7 @@ void HPL_pdinfo
          else if( iwork[j] == 3 ) TP[i] = HPL_2RING_M;
          else if( iwork[j] == 4 ) TP[i] = HPL_BLONG;
          else if( iwork[j] == 5 ) TP[i] = HPL_BLONG_M;
+         else if( iwork[j] == 6 ) TP[i] = HPL_IBCAST;
          j++;
       }
       for( i = 0; i < *NDHS; i++ ) { DH[i] = iwork[j]; j++; }
@@ -1008,6 +1011,8 @@ void HPL_pdinfo
             HPL_fprintf( TEST->outfp,       "   Blong " );
          else if( TP[i] == HPL_BLONG_M )
             HPL_fprintf( TEST->outfp,       "  BlongM " );
+         else if( TP[i] == HPL_IBCAST )
+            HPL_fprintf( TEST->outfp,       "  IBCAST " );
       }
       if( *NTPS > 8 )
       {
@@ -1026,6 +1031,8 @@ void HPL_pdinfo
                HPL_fprintf( TEST->outfp,       "   Blong " );
             else if( TP[i] == HPL_BLONG_M )
                HPL_fprintf( TEST->outfp,       "  BlongM " );
+            else if( TP[i] == HPL_IBCAST )
+               HPL_fprintf( TEST->outfp,       "  IBCAST " );
          }
          if( *NTPS > 16 )
          {
@@ -1044,6 +1051,8 @@ void HPL_pdinfo
                   HPL_fprintf( TEST->outfp,       "   Blong " );
                else if( TP[i] == HPL_BLONG_M )
                   HPL_fprintf( TEST->outfp,       "  BlongM " );
+               else if( TP[i] == HPL_IBCAST )
+                  HPL_fprintf( TEST->outfp,       "  IBCAST " );
             }
          }
       }

testing/ptest/HPL_pdtest.cpp

@@ -271,6 +271,7 @@ void HPL_pdtest
       else if( ALGO->btopo == HPL_2RING   ) ctop = '2';
       else if( ALGO->btopo == HPL_2RING_M ) ctop = '3';
       else if( ALGO->btopo == HPL_BLONG   ) ctop = '4';
+      else if( ALGO->btopo == HPL_IBCAST  ) ctop = '6';
       else /* if( ALGO->btopo == HPL_BLONG_M ) */ ctop = '5';
 
       if( wtime[0] > HPL_rzero ) {