COMpile fixes

tests/debug/Test_allgather.cc

@@ -21,10 +21,10 @@ Author: Peter Boyle <pboyle@bnl.gov>
 // Regression gate for CartesianCommunicator::AllGather / AllGatherV.
 //
 // CPU build under mpirun:
-//   mpirun -n 1 ./Test_allgather --grid 8.8.8.8  --mpi 1.1.1.1
+//   mpirun -n 1 ./Test_allgather --grid 16.16.16.32 --mpi 1.1.1.1
-//   mpirun -n 2 ./Test_allgather --grid 8.8.8.8  --mpi 1.1.1.2
+//   mpirun -n 2 ./Test_allgather --grid 16.16.16.32 --mpi 1.1.1.2
-//   mpirun -n 3 ./Test_allgather --grid 8.8.8.12 --mpi 1.1.1.3
+//   mpirun -n 3 ./Test_allgather --grid 16.16.16.48 --mpi 1.1.1.3
-//   mpirun -n 4 ./Test_allgather --grid 8.8.8.8  --mpi 1.1.1.4
+//   mpirun -n 4 ./Test_allgather --grid 16.16.16.32 --mpi 1.1.1.4
 //
 //   T1 : uniform AllGather, rank-ordered concatenation.
 //   T2 : AllGatherV with non-uniform counts and displacements.

tests/debug/Test_schur2d.cc

@@ -22,10 +22,10 @@ Author: Peter Boyle <pboyle@bnl.gov>
 // Regression gate for BlockCyclicSchurInverse -- stage 3 of the 2D
 // distributed dense inverse.  CPU build under mpirun:
 //
-//   mpirun -n 1 ./Test_schur2d --grid 8.8.8.8  --mpi 1.1.1.1
+//   mpirun -n 1 ./Test_schur2d --grid 16.16.16.32 --mpi 1.1.1.1
-//   mpirun -n 2 ./Test_schur2d --grid 8.8.8.8  --mpi 1.1.1.2
+//   mpirun -n 2 ./Test_schur2d --grid 16.16.16.32 --mpi 1.1.1.2
-//   mpirun -n 3 ./Test_schur2d --grid 8.8.8.12 --mpi 1.1.1.3
+//   mpirun -n 3 ./Test_schur2d --grid 16.16.16.48 --mpi 1.1.1.3
-//   mpirun -n 4 ./Test_schur2d --grid 8.8.8.8  --mpi 1.1.1.4
+//   mpirun -n 4 ./Test_schur2d --grid 16.16.16.32 --mpi 1.1.1.4
 //
 // Sweeps all process-grid factorisations of P and a battery of (N,nb)
 // including ragged trailing blocks, a single-leaf matrix (nblocks==1),
@@ -48,6 +48,15 @@ using namespace Grid;
 
 static int failures = 0;
 
+// Portable |z|: ComplexD is std::complex on CPU builds and thrust::complex
+// under HIP, where std::abs does not resolve (same trap RecursiveSchurInverse
+// documents at FrobNorm2Local).  Member real()/imag() work on both.
+static double Cabs(const ComplexD &z)
+{
+  double re = z.real(), im = z.imag();
+  return std::sqrt(re*re + im*im);
+}
+
 static void Report(const std::string &name, bool pass, const std::string &detail="")
 {
   std::cout << GridLogMessage << "  " << name << (pass ? "  PASS" : "  ** FAIL **");
@@ -78,9 +87,9 @@ static void HostInverse(std::vector<ComplexD> A, std::vector<ComplexD> &X, int64
   X.assign((uint64_t)N*N, ComplexD(0.0,0.0));
   for(int64_t i=0;i<N;i++) X[i+i*N] = ComplexD(1.0,0.0);
   for(int64_t c=0;c<N;c++){
-    int64_t piv=c; double mx = std::abs(A[c+c*N]);
+    int64_t piv=c; double mx = Cabs(A[c+c*N]);
     for(int64_t r=c+1;r<N;r++)
-      if ( std::abs(A[r+c*N]) > mx ){ mx=std::abs(A[r+c*N]); piv=r; }
+      if ( Cabs(A[r+c*N]) > mx ){ mx=Cabs(A[r+c*N]); piv=r; }
     GRID_ASSERT( mx > 0.0 );
     if ( piv != c )
       for(int64_t j=0;j<N;j++){
@@ -152,7 +161,7 @@ int main(int argc, char **argv)
         for(int64_t j=0;j<N;j++)
           for(int64_t i=0;i<N;i++){
             ComplexD id = (i==j) ? ComplexD(1.0,0.0) : ComplexD(0.0,0.0);
-            dc = std::max(dc, std::abs(Cert[i+j*N]-id));
+            dc = std::max(dc, Cabs(Cert[i+j*N]-id));
           }
         worstC = std::max(worstC,dc);
         if ( dc > 1.0e-10 ) okC = false;
@@ -160,8 +169,8 @@ int main(int argc, char **argv)
         // reference: element-wise, scaled by the largest inverse entry
         A.ExportGlobal(Ainv);
         double mxref = 0.0, dr = 0.0;
-        for(uint64_t i=0;i<Ref.size();i++) mxref = std::max(mxref, std::abs(Ref[i]));
+        for(uint64_t i=0;i<Ref.size();i++) mxref = std::max(mxref, Cabs(Ref[i]));
-        for(uint64_t i=0;i<Ref.size();i++) dr = std::max(dr, std::abs(Ainv[i]-Ref[i]));
+        for(uint64_t i=0;i<Ref.size();i++) dr = std::max(dr, Cabs(Ainv[i]-Ref[i]));
         dr /= mxref;
         worstR = std::max(worstR,dr);
         if ( dr > 1.0e-9 ) okR = false;

tests/debug/Test_schur2d_redist.cc

@@ -46,6 +46,15 @@ using namespace Grid;
 
 static int failures = 0;
 
+// Portable |z|: ComplexD is std::complex on CPU builds and thrust::complex
+// under HIP, where std::abs does not resolve (same trap RecursiveSchurInverse
+// documents at FrobNorm2Local).  Member real()/imag() work on both.
+static double Cabs(const ComplexD &z)
+{
+  double re = z.real(), im = z.imag();
+  return std::sqrt(re*re + im*im);
+}
+
 static void Report(const std::string &name, bool pass, const std::string &detail="")
 {
   std::cout << GridLogMessage << "  " << name << (pass ? "  PASS" : "  ** FAIL **");
@@ -72,9 +81,9 @@ static void HostInverse(std::vector<ComplexD> A, std::vector<ComplexD> &X, int64
   X.assign((uint64_t)N*N, ComplexD(0.0,0.0));
   for(int64_t i=0;i<N;i++) X[i+i*N] = ComplexD(1.0,0.0);
   for(int64_t c=0;c<N;c++){
-    int64_t piv=c; double mx = std::abs(A[c+c*N]);
+    int64_t piv=c; double mx = Cabs(A[c+c*N]);
     for(int64_t r=c+1;r<N;r++)
-      if ( std::abs(A[r+c*N]) > mx ){ mx=std::abs(A[r+c*N]); piv=r; }
+      if ( Cabs(A[r+c*N]) > mx ){ mx=Cabs(A[r+c*N]); piv=r; }
     GRID_ASSERT( mx > 0.0 );
     if ( piv != c )
       for(int64_t j=0;j<N;j++){ std::swap(A[c+j*N],A[piv+j*N]); std::swap(X[c+j*N],X[piv+j*N]); }
@@ -204,7 +213,7 @@ int main(int argc, char **argv)
             h[i + j*myrows] = Ag[(rowStart[me]+i) + j*N];
 
         double mxref = 0.0;
-        for(auto &z : Ref) mxref = std::max(mxref, std::abs(z));
+        for(auto &z : Ref) mxref = std::max(mxref, Cabs(z));
 
         // ---- 2D pipeline: rows -> cyclic -> invert -> rows ----
         std::vector<ComplexD> h2d(h.size());
@@ -219,7 +228,7 @@ int main(int argc, char **argv)
         }
         for(int64_t j=0;j<N;j++)
           for(int64_t i=0;i<myrows;i++){
-            double d = std::abs(h2d[i+j*myrows]-Ref[(rowStart[me]+i)+j*N])/mxref;
+            double d = Cabs(h2d[i+j*myrows]-Ref[(rowStart[me]+i)+j*N])/mxref;
             worst3 = std::max(worst3,d);
             if ( d > 1.0e-9 ) ok3 = false;
           }
@@ -235,7 +244,7 @@ int main(int argc, char **argv)
           acceleratorCopyFromDevice(&Ar.data[0], &h1d[0], h1d.size()*sizeof(ComplexD));
           for(int64_t j=0;j<N;j++)
             for(int64_t i=0;i<myrows;i++){
-              double d = std::abs(h2d[i+j*myrows]-h1d[i+j*myrows])/mxref;
+              double d = Cabs(h2d[i+j*myrows]-h1d[i+j*myrows])/mxref;
               worst4 = std::max(worst4,d);
               if ( d > 1.0e-9 ) ok4 = false;
             }

tests/debug/Test_schur2d_scale.cc

@@ -54,6 +54,15 @@ Author: Peter Boyle <pboyle@bnl.gov>
 
 using namespace Grid;
 
+// Portable |z|: ComplexD is std::complex on CPU builds and thrust::complex
+// under HIP, where std::abs does not resolve (same trap RecursiveSchurInverse
+// documents at FrobNorm2Local).  Member real()/imag() work on both.
+static double Cabs(const ComplexD &z)
+{
+  double re = z.real(), im = z.imag();
+  return std::sqrt(re*re + im*im);
+}
+
 static ComplexD Fill(int64_t i, int64_t j, int64_t N)
 {
   double x = std::sin(0.7*i + 1.3*j);
@@ -61,7 +70,7 @@ static ComplexD Fill(int64_t i, int64_t j, int64_t N)
   if ( i==j ) return ComplexD(3.0*64 + x, 0.5);   // dominance independent of N
   // band-limit the off-diagonal so row sums stay bounded as N grows:
   // only |i-j| <= 64 entries are non-zero => sum |offdiag| <= 128*1.42 < 3*64
-  if ( std::abs((double)(i-j)) > 64.0 ) return ComplexD(0.0,0.0);
+  if ( std::fabs((double)(i-j)) > 64.0 ) return ComplexD(0.0,0.0);
   return ComplexD(x,y);
 }
 
@@ -149,7 +158,7 @@ int main(int argc, char **argv)
       for(int64_t li=0;li<L.mloc;li++){
         int64_t gi = BlockCyclicLayout::LocalToGlobal(li, nb, L.prow, Pr);
         ComplexD id = (gi==gj) ? ComplexD(1.0,0.0) : ComplexD(0.0,0.0);
-        mx = std::max(mx, std::abs(hc[li+lj*L.mloc]-id));
+        mx = std::max(mx, Cabs(hc[li+lj*L.mloc]-id));
       }
     }
     RealD gmx = mx;  grid->GlobalMax(gmx);

tests/debug/Test_summa.cc

@@ -22,10 +22,10 @@ Author: Peter Boyle <pboyle@bnl.gov>
 // Regression gate for BlockCyclicSumma -- stage 2 of the 2D distributed
 // dense inverse.  CPU build under mpirun:
 //
-//   mpirun -n 1 ./Test_summa --grid 8.8.8.8 --mpi 1.1.1.1
+//   mpirun -n 1 ./Test_summa --grid 16.16.16.32 --mpi 1.1.1.1
-//   mpirun -n 2 ./Test_summa --grid 8.8.8.8 --mpi 1.1.1.2
+//   mpirun -n 2 ./Test_summa --grid 16.16.16.32 --mpi 1.1.1.2
-//   mpirun -n 3 ./Test_summa --grid 8.8.8.12 --mpi 1.1.1.3
+//   mpirun -n 3 ./Test_summa --grid 16.16.16.48 --mpi 1.1.1.3
-//   mpirun -n 4 ./Test_summa --grid 8.8.8.8 --mpi 1.1.1.4
+//   mpirun -n 4 ./Test_summa --grid 16.16.16.32 --mpi 1.1.1.4
 //
 // Every stage sweeps all process-grid factorisations of P (including the
 // degenerate 1xP and Px1 rings) and a battery of (N,nb) with ragged
@@ -49,6 +49,15 @@ using namespace Grid;
 
 static int failures = 0;
 
+// Portable |z|: ComplexD is std::complex on CPU builds and thrust::complex
+// under HIP, where std::abs does not resolve (same trap RecursiveSchurInverse
+// documents at FrobNorm2Local).  Member real()/imag() work on both.
+static double Cabs(const ComplexD &z)
+{
+  double re = z.real(), im = z.imag();
+  return std::sqrt(re*re + im*im);
+}
+
 static void Report(const std::string &name, bool pass, const std::string &detail="")
 {
   std::cout << GridLogMessage << "  " << name << (pass ? "  PASS" : "  ** FAIL **");
@@ -83,7 +92,7 @@ static void RefGemm(ComplexD alpha, const std::vector<ComplexD> &A,
 static double MaxDiff(const std::vector<ComplexD> &X, const std::vector<ComplexD> &Y)
 {
   double m = 0.0;
-  for(uint64_t i=0;i<X.size();i++) m = std::max(m, std::abs(X[i]-Y[i]));
+  for(uint64_t i=0;i<X.size();i++) m = std::max(m, Cabs(X[i]-Y[i]));
   return m;
 }