From 1dddd17e3c73c37bf769ad31532e44cbb0cdc585 Mon Sep 17 00:00:00 2001 From: Dr Peter Boyle Date: Fri, 27 Apr 2018 11:44:46 +0100 Subject: [PATCH 01/15] Benchmark improvements from tesseract --- benchmarks/Benchmark_ITT.cc | 13 ++++++++++--- benchmarks/Benchmark_comms.cc | 13 +++++++++---- 2 files changed, 19 insertions(+), 7 deletions(-) diff --git a/benchmarks/Benchmark_ITT.cc b/benchmarks/Benchmark_ITT.cc index 666e4830..f811ac32 100644 --- a/benchmarks/Benchmark_ITT.cc +++ b/benchmarks/Benchmark_ITT.cc @@ -158,8 +158,10 @@ public: dbytes=0; ncomm=0; - - parallel_for(int dir=0;dir<8;dir++){ +#ifdef GRID_OMP +#pragma omp parallel for num_threads(Grid::CartesianCommunicator::nCommThreads) +#endif + for(int dir=0;dir<8;dir++){ double tbytes; int mu =dir % 4; @@ -175,9 +177,14 @@ public: int comm_proc = mpi_layout[mu]-1; Grid.ShiftedRanks(mu,comm_proc,xmit_to_rank,recv_from_rank); } +#ifdef GRID_OMP + int tid = omp_get_thread_num(); +#else + int tid = dir; +#endif tbytes= Grid.StencilSendToRecvFrom((void *)&xbuf[dir][0], xmit_to_rank, (void *)&rbuf[dir][0], recv_from_rank, - bytes,dir); + bytes,tid); #ifdef GRID_OMP #pragma omp atomic diff --git a/benchmarks/Benchmark_comms.cc b/benchmarks/Benchmark_comms.cc index 29ccf96c..304a09fc 100644 --- a/benchmarks/Benchmark_comms.cc +++ b/benchmarks/Benchmark_comms.cc @@ -169,7 +169,11 @@ int main (int argc, char ** argv) for(int lat=4;lat<=maxlat;lat+=4){ for(int Ls=8;Ls<=8;Ls*=2){ - std::vector latt_size ({lat,lat,lat,lat}); + std::vector latt_size ({lat*mpi_layout[0], + lat*mpi_layout[1], + lat*mpi_layout[2], + lat*mpi_layout[3]}); + GridCartesian Grid(latt_size,simd_layout,mpi_layout); RealD Nrank = Grid._Nprocessors; @@ -485,7 +489,8 @@ int main (int argc, char ** argv) dbytes=0; ncomm=0; - parallel_for(int dir=0;dir<8;dir++){ +#pragma omp parallel for num_threads(Grid::CartesianCommunicator::nCommThreads) + for(int dir=0;dir<8;dir++){ double tbytes; int mu =dir % 4; @@ -502,9 +507,9 @@ int main (int argc, char ** argv) int comm_proc = mpi_layout[mu]-1; Grid.ShiftedRanks(mu,comm_proc,xmit_to_rank,recv_from_rank); } - + int tid = omp_get_thread_num(); tbytes= Grid.StencilSendToRecvFrom((void *)&xbuf[dir][0], xmit_to_rank, - (void *)&rbuf[dir][0], recv_from_rank, bytes,dir); + (void *)&rbuf[dir][0], recv_from_rank, bytes,tid); #pragma omp atomic dbytes+=tbytes; From c45f24a1b51e9b6bc8a664130a04791746c183ea Mon Sep 17 00:00:00 2001 From: Peter Boyle Date: Mon, 30 Apr 2018 21:50:00 +0100 Subject: [PATCH 02/15] Improvements for tesseract --- configure.ac | 6 +- lib/communicator/SharedMemoryMPI.cc | 229 +++++++++++++++++++++++++++- lib/stencil/Stencil.h | 2 + lib/threads/Threads.h | 2 +- 4 files changed, 230 insertions(+), 9 deletions(-) diff --git a/configure.ac b/configure.ac index aced6a9c..722b1678 100644 --- a/configure.ac +++ b/configure.ac @@ -340,7 +340,7 @@ case ${ac_PRECISION} in esac ###################### Shared memory allocation technique under MPI3 -AC_ARG_ENABLE([shm],[AC_HELP_STRING([--enable-shm=shmopen|hugetlbfs|shmnone], +AC_ARG_ENABLE([shm],[AC_HELP_STRING([--enable-shm=shmopen|shmget|hugetlbfs|shmnone], [Select SHM allocation technique])],[ac_SHM=${enable_shm}],[ac_SHM=shmopen]) case ${ac_SHM} in @@ -349,6 +349,10 @@ case ${ac_SHM} in AC_DEFINE([GRID_MPI3_SHMOPEN],[1],[GRID_MPI3_SHMOPEN] ) ;; + shmget) + AC_DEFINE([GRID_MPI3_SHMGET],[1],[GRID_MPI3_SHMGET] ) + ;; + shmnone) AC_DEFINE([GRID_MPI3_SHM_NONE],[1],[GRID_MPI3_SHM_NONE] ) ;; diff --git a/lib/communicator/SharedMemoryMPI.cc b/lib/communicator/SharedMemoryMPI.cc index d534a6d9..205d61a9 100644 --- a/lib/communicator/SharedMemoryMPI.cc +++ b/lib/communicator/SharedMemoryMPI.cc @@ -114,19 +114,169 @@ void GlobalSharedMemory::Init(Grid_MPI_Comm comm) assert(WorldNode!=-1); _ShmSetup=1; } - -void GlobalSharedMemory::OptimalCommunicator(const std::vector &processors,Grid_MPI_Comm & optimal_comm) +// Gray encode support +int BinaryToGray (int binary) { + int gray = (binary>>1)^binary; + return gray; +} +int Log2Size(int TwoToPower,int MAXLOG2) { - //////////////////////////////////////////////////////////////// - // Assert power of two shm_size. - //////////////////////////////////////////////////////////////// int log2size = -1; - for(int i=0;i<=MAXLOG2RANKSPERNODE;i++){ - if ( (0x1< &processors,Grid_MPI_Comm & optimal_comm) +{ +#undef HYPERCUBE +#ifdef HYPERCUBE + //////////////////////////////////////////////////////////////// + // Assert power of two shm_size. + //////////////////////////////////////////////////////////////// + int log2size = Log2Size(WorldShmSize,MAXLOG2RANKSPERNODE); + assert(log2size != -1); + + //////////////////////////////////////////////////////////////// + // Identify the hypercube coordinate of this node using hostname + //////////////////////////////////////////////////////////////// + // n runs 0...7 9...16 18...25 27...34 (8*4) 5 bits + // i runs 0..7 3 bits + // r runs 0..3 2 bits + // 2^10 = 1024 nodes + const int maxhdim = 10; + std::vector HyperCubeCoords(maxhdim,0); + std::vector RootHyperCubeCoords(maxhdim,0); + int R; + int I; + int N; + const int namelen = _POSIX_HOST_NAME_MAX; + char name[namelen]; + + // Parse ICE-XA hostname to get hypercube location + gethostname(name,namelen); + int nscan = sscanf(name,"r%di%dn%d",&R,&I,&N) ; + assert(nscan==3); + + int nlo = N%9; + int nhi = N/9; + uint32_t hypercoor = (R<<8)|(I<<5)|(nhi<<3)|nlo ; + uint32_t rootcoor = hypercoor; + + ////////////////////////////////////////////////////////////////// + // Print debug info + ////////////////////////////////////////////////////////////////// + for(int d=0;d>d)&0x1; + } + + std::cerr << " Hcoor (" ; + for(int d=0;d=0); + std::cerr << " WorldRank "<>d)&0x1; + } + + std::cerr << " rel Hcoor ("; + for(int d=0;d processor_coor(ndimension); + std::vector WorldDims = processors; std::vector ShmDims (ndimension,1); std::vector NodeDims (ndimension); + std::vector ShmCoor (ndimension); std::vector NodeCoor (ndimension); std::vector WorldCoor(ndimension); + std::vector HyperCoor(ndimension); + int dim = 0; + for(int l2=0;l2> bits; + } + //////////////////////////////////////////////////////////////// + // Check processor counts match + //////////////////////////////////////////////////////////////// + int Nprocessors=1; + for(int i=0;i &processors, ///////////////////////////////////////////////////////////////// int ierr= MPI_Comm_split(WorldComm,0,rank,&optimal_comm); assert(ierr==0); +#endif } +//////////////////////////////////////////////////////////////////////////////////////////// +// SHMGET +//////////////////////////////////////////////////////////////////////////////////////////// +#ifdef GRID_MPI3_SHMGET +void GlobalSharedMemory::SharedMemoryAllocate(uint64_t bytes, int flags) +{ + std::cout << "SharedMemoryAllocate "<< bytes<< " shmget implementation "< shmids(WorldShmSize); + + if ( WorldShmRank == 0 ) { + for(int r=0;r >& table,const La parallel_for(int i=0;i #define PARALLEL_FOR_LOOP _Pragma("omp parallel for schedule(static)") #define PARALLEL_FOR_LOOP_INTERN _Pragma("omp for schedule(static)") -#define PARALLEL_NESTED_LOOP2 _Pragma("omp parallel for schedule(static) collapse(2)") +#define PARALLEL_NESTED_LOOP2 _Pragma("omp parallel for collapse(2)") #define PARALLEL_REGION _Pragma("omp parallel") #define PARALLEL_CRITICAL _Pragma("omp critical") #else From a64497265d05583d1910b3c293e109c487aab2d6 Mon Sep 17 00:00:00 2001 From: Peter Boyle Date: Wed, 2 May 2018 14:07:28 +0100 Subject: [PATCH 03/15] TIming --- lib/stencil/Stencil.h | 9 ++++----- 1 file changed, 4 insertions(+), 5 deletions(-) diff --git a/lib/stencil/Stencil.h b/lib/stencil/Stencil.h index dc64f616..a79064eb 100644 --- a/lib/stencil/Stencil.h +++ b/lib/stencil/Stencil.h @@ -507,25 +507,24 @@ class CartesianStencil { // Stencil runs along coordinate axes only; NO diagonal template void CommsMerge(decompressor decompress,std::vector &mm,std::vector &dd) { + mergetime-=usecond(); for(int i=0;i Date: Wed, 2 May 2018 14:07:41 +0100 Subject: [PATCH 04/15] shmget reintroduce --- configure.ac | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/configure.ac b/configure.ac index 722b1678..819253a0 100644 --- a/configure.ac +++ b/configure.ac @@ -370,7 +370,7 @@ esac AC_ARG_ENABLE([shmpath],[AC_HELP_STRING([--enable-shmpath=path], [Select SHM mmap base path for hugetlbfs])], [ac_SHMPATH=${enable_shmpath}], - [ac_SHMPATH=/var/lib/hugetlbfs/pagesize-2MB/]) + [ac_SHMPATH=/var/lib/hugetlbfs/global/pagesize-2MB/]) AC_DEFINE_UNQUOTED([GRID_SHM_PATH],["$ac_SHMPATH"],[Path to a hugetlbfs filesystem for MMAPing]) ############### communication type selection From 12982a4455189cc568a18f1d6cffe819fc706c4e Mon Sep 17 00:00:00 2001 From: Peter Boyle Date: Wed, 2 May 2018 14:10:21 +0100 Subject: [PATCH 05/15] Hypercube optimisation --- lib/communicator/SharedMemoryMPI.cc | 28 +++++----------------------- 1 file changed, 5 insertions(+), 23 deletions(-) diff --git a/lib/communicator/SharedMemoryMPI.cc b/lib/communicator/SharedMemoryMPI.cc index 205d61a9..d55ff5df 100644 --- a/lib/communicator/SharedMemoryMPI.cc +++ b/lib/communicator/SharedMemoryMPI.cc @@ -173,15 +173,9 @@ void GlobalSharedMemory::OptimalCommunicator(const std::vector &processors, HyperCubeCoords[d] = (hypercoor>>d)&0x1; } - std::cerr << " Hcoor (" ; - for(int d=0;d &processors, hypercoor=hypercoor-rootcoor; assert(hypercoor=0); - std::cerr << " WorldRank "< &processors, HyperCubeCoords[d] = (hypercoor>>d)&0x1; } - std::cerr << " rel Hcoor ("; - for(int d=0;d &processors, int rank; Lexicographic::CoorFromIndexReversed(NodeCoor,WorldNode ,NodeDims); - std::cerr << "NodeCoor "; - for(int d=0;d Date: Wed, 2 May 2018 14:10:55 +0100 Subject: [PATCH 06/15] Revert to fast versoin --- lib/qcd/action/fermion/WilsonKernelsHand.cc | 616 +++++--------------- 1 file changed, 148 insertions(+), 468 deletions(-) diff --git a/lib/qcd/action/fermion/WilsonKernelsHand.cc b/lib/qcd/action/fermion/WilsonKernelsHand.cc index aa6b5f6b..50816495 100644 --- a/lib/qcd/action/fermion/WilsonKernelsHand.cc +++ b/lib/qcd/action/fermion/WilsonKernelsHand.cc @@ -30,181 +30,60 @@ Author: paboyle #define REGISTER -#define LOAD_CHIMU_BODY(F) \ - Chimu_00=ref(F)(0)(0); \ - Chimu_01=ref(F)(0)(1); \ - Chimu_02=ref(F)(0)(2); \ - Chimu_10=ref(F)(1)(0); \ - Chimu_11=ref(F)(1)(1); \ - Chimu_12=ref(F)(1)(2); \ - Chimu_20=ref(F)(2)(0); \ - Chimu_21=ref(F)(2)(1); \ - Chimu_22=ref(F)(2)(2); \ - Chimu_30=ref(F)(3)(0); \ - Chimu_31=ref(F)(3)(1); \ - Chimu_32=ref(F)(3)(2) +#define LOAD_CHIMU \ + {const SiteSpinor & ref (in._odata[offset]); \ + Chimu_00=ref()(0)(0);\ + Chimu_01=ref()(0)(1);\ + Chimu_02=ref()(0)(2);\ + Chimu_10=ref()(1)(0);\ + Chimu_11=ref()(1)(1);\ + Chimu_12=ref()(1)(2);\ + Chimu_20=ref()(2)(0);\ + Chimu_21=ref()(2)(1);\ + Chimu_22=ref()(2)(2);\ + Chimu_30=ref()(3)(0);\ + Chimu_31=ref()(3)(1);\ + Chimu_32=ref()(3)(2);} -#define LOAD_CHIMU(DIR,F,PERM) \ - { const SiteSpinor & ref (in._odata[offset]); LOAD_CHIMU_BODY(F); } - -#define LOAD_CHI_BODY(F) \ - Chi_00 = ref(F)(0)(0);\ - Chi_01 = ref(F)(0)(1);\ - Chi_02 = ref(F)(0)(2);\ - Chi_10 = ref(F)(1)(0);\ - Chi_11 = ref(F)(1)(1);\ - Chi_12 = ref(F)(1)(2) - -#define LOAD_CHI(DIR,F,PERM) \ - {const SiteHalfSpinor &ref(buf[offset]); LOAD_CHI_BODY(F); } - - -//G-parity implementations using in-place intrinsic ops - -//1l 1h -> 1h 1l -//0l 0h , 1h 1l -> 0l 1h 0h,1l -//0h,1l -> 1l,0h -//if( (distance == 1 && !perm_will_occur) || (distance == -1 && perm_will_occur) ) -//Pulled fermion through forwards face, GPBC on upper component -//Need 0= 0l 1h 1= 1l 0h -//else if( (distance == -1 && !perm) || (distance == 1 && perm) ) -//Pulled fermion through backwards face, GPBC on lower component -//Need 0= 1l 0h 1= 0l 1h - -//1l 1h -> 1h 1l -//0l 0h , 1h 1l -> 0l 1h 0h,1l -#define DO_TWIST_0L_1H(INTO,S,C,F, PERM, tmp1, tmp2, tmp3) \ - permute##PERM(tmp1, ref(1)(S)(C)); \ - exchange##PERM(tmp2,tmp3, ref(0)(S)(C), tmp1); \ - INTO = tmp2; - -//0l 0h -> 0h 0l -//1l 1h, 0h 0l -> 1l 0h, 1h 0l -#define DO_TWIST_1L_0H(INTO,S,C,F, PERM, tmp1, tmp2, tmp3) \ - permute##PERM(tmp1, ref(0)(S)(C)); \ - exchange##PERM(tmp2,tmp3, ref(1)(S)(C), tmp1); \ - INTO = tmp2; - - - - -#define LOAD_CHI_SETUP(DIR,F) \ - g = F; \ - direction = st._directions[DIR]; \ - distance = st._distances[DIR]; \ - sl = st._grid->_simd_layout[direction]; \ - inplace_twist = 0; \ - if(SE->_around_the_world && this->Params.twists[DIR % 4]){ \ - if(sl == 1){ \ - g = (F+1) % 2; \ - }else{ \ - inplace_twist = 1; \ - } \ - } - -#define LOAD_CHIMU_GPARITY_INPLACE_TWIST(DIR,F,PERM) \ - { const SiteSpinor &ref(in._odata[offset]); \ - LOAD_CHI_SETUP(DIR,F); \ - if(!inplace_twist){ \ - LOAD_CHIMU_BODY(g); \ - }else{ \ - if( ( F==0 && ((distance == 1 && !perm) || (distance == -1 && perm)) ) || \ - ( F==1 && ((distance == -1 && !perm) || (distance == 1 && perm)) ) ){ \ - DO_TWIST_0L_1H(Chimu_00,0,0,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_0L_1H(Chimu_01,0,1,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_0L_1H(Chimu_02,0,2,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_0L_1H(Chimu_10,1,0,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_0L_1H(Chimu_11,1,1,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_0L_1H(Chimu_12,1,2,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_0L_1H(Chimu_20,2,0,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_0L_1H(Chimu_21,2,1,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_0L_1H(Chimu_22,2,2,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_0L_1H(Chimu_30,3,0,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_0L_1H(Chimu_31,3,1,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_0L_1H(Chimu_32,3,2,F,PERM, U_11,U_20,U_21); \ - }else{ \ - DO_TWIST_1L_0H(Chimu_00,0,0,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_1L_0H(Chimu_01,0,1,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_1L_0H(Chimu_02,0,2,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_1L_0H(Chimu_10,1,0,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_1L_0H(Chimu_11,1,1,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_1L_0H(Chimu_12,1,2,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_1L_0H(Chimu_20,2,0,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_1L_0H(Chimu_21,2,1,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_1L_0H(Chimu_22,2,2,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_1L_0H(Chimu_30,3,0,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_1L_0H(Chimu_31,3,1,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_1L_0H(Chimu_32,3,2,F,PERM, U_11,U_20,U_21); \ - } \ - } \ - } - - -#define LOAD_CHI_GPARITY_INPLACE_TWIST(DIR,F,PERM) \ - { const SiteHalfSpinor &ref(buf[offset]); \ - LOAD_CHI_SETUP(DIR,F); \ - if(!inplace_twist){ \ - LOAD_CHI_BODY(g); \ - }else{ \ - if( ( F==0 && ((distance == 1 && !perm) || (distance == -1 && perm)) ) || \ - ( F==1 && ((distance == -1 && !perm) || (distance == 1 && perm)) ) ){ \ - DO_TWIST_0L_1H(Chi_00,0,0,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_0L_1H(Chi_01,0,1,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_0L_1H(Chi_02,0,2,F,PERM, UChi_00,UChi_01,UChi_02); \ - DO_TWIST_0L_1H(Chi_10,1,0,F,PERM, UChi_10,UChi_11,UChi_12); \ - DO_TWIST_0L_1H(Chi_11,1,1,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_0L_1H(Chi_12,1,2,F,PERM, U_11,U_20,U_21); \ - }else{ \ - DO_TWIST_1L_0H(Chi_00,0,0,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_1L_0H(Chi_01,0,1,F,PERM, U_11,U_20,U_21); \ - DO_TWIST_1L_0H(Chi_02,0,2,F,PERM, UChi_00,UChi_01,UChi_02); \ - DO_TWIST_1L_0H(Chi_10,1,0,F,PERM, UChi_10,UChi_11,UChi_12); \ - DO_TWIST_1L_0H(Chi_11,1,1,F,PERM, U_00,U_01,U_10); \ - DO_TWIST_1L_0H(Chi_12,1,2,F,PERM, U_11,U_20,U_21); \ - } \ - } \ - } - - -#define LOAD_CHI_GPARITY(DIR,F,PERM) LOAD_CHI_GPARITY_INPLACE_TWIST(DIR,F,PERM) -#define LOAD_CHIMU_GPARITY(DIR,F,PERM) LOAD_CHIMU_GPARITY_INPLACE_TWIST(DIR,F,PERM) +#define LOAD_CHI\ + {const SiteHalfSpinor &ref(buf[offset]); \ + Chi_00 = ref()(0)(0);\ + Chi_01 = ref()(0)(1);\ + Chi_02 = ref()(0)(2);\ + Chi_10 = ref()(1)(0);\ + Chi_11 = ref()(1)(1);\ + Chi_12 = ref()(1)(2);} // To splat or not to splat depends on the implementation -#define MULT_2SPIN_BODY \ - Impl::loadLinkElement(U_00,ref()(0,0)); \ - Impl::loadLinkElement(U_10,ref()(1,0)); \ - Impl::loadLinkElement(U_20,ref()(2,0)); \ - Impl::loadLinkElement(U_01,ref()(0,1)); \ - Impl::loadLinkElement(U_11,ref()(1,1)); \ - Impl::loadLinkElement(U_21,ref()(2,1)); \ - UChi_00 = U_00*Chi_00; \ - UChi_10 = U_00*Chi_10; \ - UChi_01 = U_10*Chi_00; \ - UChi_11 = U_10*Chi_10; \ - UChi_02 = U_20*Chi_00; \ - UChi_12 = U_20*Chi_10; \ - UChi_00+= U_01*Chi_01; \ - UChi_10+= U_01*Chi_11; \ - UChi_01+= U_11*Chi_01; \ - UChi_11+= U_11*Chi_11; \ - UChi_02+= U_21*Chi_01; \ - UChi_12+= U_21*Chi_11; \ - Impl::loadLinkElement(U_00,ref()(0,2)); \ - Impl::loadLinkElement(U_10,ref()(1,2)); \ - Impl::loadLinkElement(U_20,ref()(2,2)); \ - UChi_00+= U_00*Chi_02; \ - UChi_10+= U_00*Chi_12; \ - UChi_01+= U_10*Chi_02; \ - UChi_11+= U_10*Chi_12; \ - UChi_02+= U_20*Chi_02; \ - UChi_12+= U_20*Chi_12 - - -#define MULT_2SPIN(A,F) \ - {auto & ref(U._odata[sU](A)); MULT_2SPIN_BODY; } - -#define MULT_2SPIN_GPARITY(A,F) \ - {auto & ref(U._odata[sU](F)(A)); MULT_2SPIN_BODY; } +#define MULT_2SPIN(A)\ + {auto & ref(U._odata[sU](A)); \ + Impl::loadLinkElement(U_00,ref()(0,0)); \ + Impl::loadLinkElement(U_10,ref()(1,0)); \ + Impl::loadLinkElement(U_20,ref()(2,0)); \ + Impl::loadLinkElement(U_01,ref()(0,1)); \ + Impl::loadLinkElement(U_11,ref()(1,1)); \ + Impl::loadLinkElement(U_21,ref()(2,1)); \ + UChi_00 = U_00*Chi_00;\ + UChi_10 = U_00*Chi_10;\ + UChi_01 = U_10*Chi_00;\ + UChi_11 = U_10*Chi_10;\ + UChi_02 = U_20*Chi_00;\ + UChi_12 = U_20*Chi_10;\ + UChi_00+= U_01*Chi_01;\ + UChi_10+= U_01*Chi_11;\ + UChi_01+= U_11*Chi_01;\ + UChi_11+= U_11*Chi_11;\ + UChi_02+= U_21*Chi_01;\ + UChi_12+= U_21*Chi_11;\ + Impl::loadLinkElement(U_00,ref()(0,2)); \ + Impl::loadLinkElement(U_10,ref()(1,2)); \ + Impl::loadLinkElement(U_20,ref()(2,2)); \ + UChi_00+= U_00*Chi_02;\ + UChi_10+= U_00*Chi_12;\ + UChi_01+= U_10*Chi_02;\ + UChi_11+= U_10*Chi_12;\ + UChi_02+= U_20*Chi_02;\ + UChi_12+= U_20*Chi_12;} #define PERMUTE_DIR(dir) \ @@ -428,87 +307,84 @@ Author: paboyle result_31-= UChi_11; \ result_32-= UChi_12; -#define HAND_STENCIL_LEG(PROJ,PERM,DIR,RECON,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ +#define HAND_STENCIL_LEG(PROJ,PERM,DIR,RECON) \ SE=st.GetEntry(ptype,DIR,ss); \ offset = SE->_offset; \ local = SE->_is_local; \ perm = SE->_permute; \ if ( local ) { \ - LOAD_CHIMU_IMPL(DIR,F,PERM); \ + LOAD_CHIMU; \ PROJ; \ if ( perm) { \ PERMUTE_DIR(PERM); \ } \ } else { \ - LOAD_CHI_IMPL(DIR,F,PERM); \ + LOAD_CHI; \ } \ - MULT_2SPIN_IMPL(DIR,F); \ + MULT_2SPIN(DIR); \ RECON; - -#define HAND_STENCIL_LEG_INT(PROJ,PERM,DIR,RECON,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ +#define HAND_STENCIL_LEG_INT(PROJ,PERM,DIR,RECON) \ SE=st.GetEntry(ptype,DIR,ss); \ offset = SE->_offset; \ local = SE->_is_local; \ perm = SE->_permute; \ if ( local ) { \ - LOAD_CHIMU_IMPL(DIR,F,PERM); \ + LOAD_CHIMU; \ PROJ; \ if ( perm) { \ PERMUTE_DIR(PERM); \ } \ } else if ( st.same_node[DIR] ) { \ - LOAD_CHI_IMPL(DIR,F,PERM); \ + LOAD_CHI; \ } \ if (local || st.same_node[DIR] ) { \ - MULT_2SPIN_IMPL(DIR,F); \ + MULT_2SPIN(DIR); \ RECON; \ } -#define HAND_STENCIL_LEG_EXT(PROJ,PERM,DIR,RECON,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ +#define HAND_STENCIL_LEG_EXT(PROJ,PERM,DIR,RECON) \ SE=st.GetEntry(ptype,DIR,ss); \ offset = SE->_offset; \ - local = SE->_is_local; \ - perm = SE->_permute; \ if((!SE->_is_local)&&(!st.same_node[DIR]) ) { \ - LOAD_CHI_IMPL(DIR,F,PERM); \ - MULT_2SPIN_IMPL(DIR,F); \ + LOAD_CHI; \ + MULT_2SPIN(DIR); \ RECON; \ nmu++; \ } -#define HAND_RESULT(ss,F) \ +#define HAND_RESULT(ss) \ { \ SiteSpinor & ref (out._odata[ss]); \ - vstream(ref(F)(0)(0),result_00); \ - vstream(ref(F)(0)(1),result_01); \ - vstream(ref(F)(0)(2),result_02); \ - vstream(ref(F)(1)(0),result_10); \ - vstream(ref(F)(1)(1),result_11); \ - vstream(ref(F)(1)(2),result_12); \ - vstream(ref(F)(2)(0),result_20); \ - vstream(ref(F)(2)(1),result_21); \ - vstream(ref(F)(2)(2),result_22); \ - vstream(ref(F)(3)(0),result_30); \ - vstream(ref(F)(3)(1),result_31); \ - vstream(ref(F)(3)(2),result_32); \ + vstream(ref()(0)(0),result_00); \ + vstream(ref()(0)(1),result_01); \ + vstream(ref()(0)(2),result_02); \ + vstream(ref()(1)(0),result_10); \ + vstream(ref()(1)(1),result_11); \ + vstream(ref()(1)(2),result_12); \ + vstream(ref()(2)(0),result_20); \ + vstream(ref()(2)(1),result_21); \ + vstream(ref()(2)(2),result_22); \ + vstream(ref()(3)(0),result_30); \ + vstream(ref()(3)(1),result_31); \ + vstream(ref()(3)(2),result_32); \ } -#define HAND_RESULT_EXT(ss,F) \ +#define HAND_RESULT_EXT(ss) \ if (nmu){ \ SiteSpinor & ref (out._odata[ss]); \ - ref(F)(0)(0)+=result_00; \ - ref(F)(0)(1)+=result_01; \ - ref(F)(0)(2)+=result_02; \ - ref(F)(1)(0)+=result_10; \ - ref(F)(1)(1)+=result_11; \ - ref(F)(1)(2)+=result_12; \ - ref(F)(2)(0)+=result_20; \ - ref(F)(2)(1)+=result_21; \ - ref(F)(2)(2)+=result_22; \ - ref(F)(3)(0)+=result_30; \ - ref(F)(3)(1)+=result_31; \ - ref(F)(3)(2)+=result_32; \ + ref()(0)(0)+=result_00; \ + ref()(0)(1)+=result_01; \ + ref()(0)(2)+=result_02; \ + ref()(1)(0)+=result_10; \ + ref()(1)(1)+=result_11; \ + ref()(1)(2)+=result_12; \ + ref()(2)(0)+=result_20; \ + ref()(2)(1)+=result_21; \ + ref()(2)(2)+=result_22; \ + ref()(3)(0)+=result_30; \ + ref()(3)(1)+=result_31; \ + ref()(3)(2)+=result_32; \ } @@ -587,18 +463,15 @@ WilsonKernels::HandDhopSite(StencilImpl &st,LebesgueOrder &lo,DoubledGauge int offset,local,perm, ptype; StencilEntry *SE; -#define HAND_DOP_SITE(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ - HAND_STENCIL_LEG(XM_PROJ,3,Xp,XM_RECON,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(YM_PROJ,2,Yp,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(ZM_PROJ,1,Zp,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(TM_PROJ,0,Tp,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(XP_PROJ,3,Xm,XP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(YP_PROJ,2,Ym,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(ZP_PROJ,1,Zm,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(TP_PROJ,0,Tm,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_RESULT(ss,F) - - HAND_DOP_SITE(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); + HAND_STENCIL_LEG(XM_PROJ,3,Xp,XM_RECON); + HAND_STENCIL_LEG(YM_PROJ,2,Yp,YM_RECON_ACCUM); + HAND_STENCIL_LEG(ZM_PROJ,1,Zp,ZM_RECON_ACCUM); + HAND_STENCIL_LEG(TM_PROJ,0,Tp,TM_RECON_ACCUM); + HAND_STENCIL_LEG(XP_PROJ,3,Xm,XP_RECON_ACCUM); + HAND_STENCIL_LEG(YP_PROJ,2,Ym,YP_RECON_ACCUM); + HAND_STENCIL_LEG(ZP_PROJ,1,Zm,ZP_RECON_ACCUM); + HAND_STENCIL_LEG(TP_PROJ,0,Tm,TP_RECON_ACCUM); + HAND_RESULT(ss); } template @@ -612,19 +485,16 @@ void WilsonKernels::HandDhopSiteDag(StencilImpl &st,LebesgueOrder &lo,Doub StencilEntry *SE; int offset,local,perm, ptype; - -#define HAND_DOP_SITE_DAG(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ - HAND_STENCIL_LEG(XP_PROJ,3,Xp,XP_RECON,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(YP_PROJ,2,Yp,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(ZP_PROJ,1,Zp,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(TP_PROJ,0,Tp,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(XM_PROJ,3,Xm,XM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(YM_PROJ,2,Ym,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(ZM_PROJ,1,Zm,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG(TM_PROJ,0,Tm,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_RESULT(ss,F) - - HAND_DOP_SITE_DAG(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); + + HAND_STENCIL_LEG(XP_PROJ,3,Xp,XP_RECON); + HAND_STENCIL_LEG(YP_PROJ,2,Yp,YP_RECON_ACCUM); + HAND_STENCIL_LEG(ZP_PROJ,1,Zp,ZP_RECON_ACCUM); + HAND_STENCIL_LEG(TP_PROJ,0,Tp,TP_RECON_ACCUM); + HAND_STENCIL_LEG(XM_PROJ,3,Xm,XM_RECON_ACCUM); + HAND_STENCIL_LEG(YM_PROJ,2,Ym,YM_RECON_ACCUM); + HAND_STENCIL_LEG(ZM_PROJ,1,Zm,ZM_RECON_ACCUM); + HAND_STENCIL_LEG(TM_PROJ,0,Tm,TM_RECON_ACCUM); + HAND_RESULT(ss); } template void @@ -639,20 +509,16 @@ WilsonKernels::HandDhopSiteInt(StencilImpl &st,LebesgueOrder &lo,DoubledGa int offset,local,perm, ptype; StencilEntry *SE; - -#define HAND_DOP_SITE_INT(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ - ZERO_RESULT; \ - HAND_STENCIL_LEG_INT(XM_PROJ,3,Xp,XM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(YM_PROJ,2,Yp,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(ZM_PROJ,1,Zp,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(TM_PROJ,0,Tp,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(XP_PROJ,3,Xm,XP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(YP_PROJ,2,Ym,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(ZP_PROJ,1,Zm,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(TP_PROJ,0,Tm,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_RESULT(ss,F) - - HAND_DOP_SITE_INT(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); + ZERO_RESULT; + HAND_STENCIL_LEG_INT(XM_PROJ,3,Xp,XM_RECON_ACCUM); + HAND_STENCIL_LEG_INT(YM_PROJ,2,Yp,YM_RECON_ACCUM); + HAND_STENCIL_LEG_INT(ZM_PROJ,1,Zp,ZM_RECON_ACCUM); + HAND_STENCIL_LEG_INT(TM_PROJ,0,Tp,TM_RECON_ACCUM); + HAND_STENCIL_LEG_INT(XP_PROJ,3,Xm,XP_RECON_ACCUM); + HAND_STENCIL_LEG_INT(YP_PROJ,2,Ym,YP_RECON_ACCUM); + HAND_STENCIL_LEG_INT(ZP_PROJ,1,Zm,ZP_RECON_ACCUM); + HAND_STENCIL_LEG_INT(TP_PROJ,0,Tm,TP_RECON_ACCUM); + HAND_RESULT(ss); } template @@ -666,20 +532,16 @@ void WilsonKernels::HandDhopSiteDagInt(StencilImpl &st,LebesgueOrder &lo,D StencilEntry *SE; int offset,local,perm, ptype; - -#define HAND_DOP_SITE_DAG_INT(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ - ZERO_RESULT; \ - HAND_STENCIL_LEG_INT(XP_PROJ,3,Xp,XP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(YP_PROJ,2,Yp,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(ZP_PROJ,1,Zp,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(TP_PROJ,0,Tp,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(XM_PROJ,3,Xm,XM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(YM_PROJ,2,Ym,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(ZM_PROJ,1,Zm,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_INT(TM_PROJ,0,Tm,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_RESULT(ss,F) - - HAND_DOP_SITE_DAG_INT(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); + ZERO_RESULT; + HAND_STENCIL_LEG_INT(XP_PROJ,3,Xp,XP_RECON_ACCUM); + HAND_STENCIL_LEG_INT(YP_PROJ,2,Yp,YP_RECON_ACCUM); + HAND_STENCIL_LEG_INT(ZP_PROJ,1,Zp,ZP_RECON_ACCUM); + HAND_STENCIL_LEG_INT(TP_PROJ,0,Tp,TP_RECON_ACCUM); + HAND_STENCIL_LEG_INT(XM_PROJ,3,Xm,XM_RECON_ACCUM); + HAND_STENCIL_LEG_INT(YM_PROJ,2,Ym,YM_RECON_ACCUM); + HAND_STENCIL_LEG_INT(ZM_PROJ,1,Zm,ZM_RECON_ACCUM); + HAND_STENCIL_LEG_INT(TM_PROJ,0,Tm,TM_RECON_ACCUM); + HAND_RESULT(ss); } template void @@ -695,20 +557,16 @@ WilsonKernels::HandDhopSiteExt(StencilImpl &st,LebesgueOrder &lo,DoubledGa int offset,local,perm, ptype; StencilEntry *SE; int nmu=0; - -#define HAND_DOP_SITE_EXT(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ - ZERO_RESULT; \ - HAND_STENCIL_LEG_EXT(XM_PROJ,3,Xp,XM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(YM_PROJ,2,Yp,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(ZM_PROJ,1,Zp,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(TM_PROJ,0,Tp,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(XP_PROJ,3,Xm,XP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(YP_PROJ,2,Ym,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(ZP_PROJ,1,Zm,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(TP_PROJ,0,Tm,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_RESULT_EXT(ss,F) - - HAND_DOP_SITE_EXT(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); + ZERO_RESULT; + HAND_STENCIL_LEG_EXT(XM_PROJ,3,Xp,XM_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(YM_PROJ,2,Yp,YM_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(ZM_PROJ,1,Zp,ZM_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(TM_PROJ,0,Tp,TM_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(XP_PROJ,3,Xm,XP_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(YP_PROJ,2,Ym,YP_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(ZP_PROJ,1,Zm,ZP_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(TP_PROJ,0,Tm,TP_RECON_ACCUM); + HAND_RESULT_EXT(ss); } template @@ -723,193 +581,18 @@ void WilsonKernels::HandDhopSiteDagExt(StencilImpl &st,LebesgueOrder &lo,D StencilEntry *SE; int offset,local,perm, ptype; int nmu=0; - -#define HAND_DOP_SITE_DAG_EXT(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ - ZERO_RESULT; \ - HAND_STENCIL_LEG_EXT(XP_PROJ,3,Xp,XP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(YP_PROJ,2,Yp,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(ZP_PROJ,1,Zp,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(TP_PROJ,0,Tp,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(XM_PROJ,3,Xm,XM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(YM_PROJ,2,Ym,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(ZM_PROJ,1,Zm,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_STENCIL_LEG_EXT(TM_PROJ,0,Tm,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ - HAND_RESULT_EXT(ss,F) - - HAND_DOP_SITE_DAG_EXT(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); + ZERO_RESULT; + HAND_STENCIL_LEG_EXT(XP_PROJ,3,Xp,XP_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(YP_PROJ,2,Yp,YP_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(ZP_PROJ,1,Zp,ZP_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(TP_PROJ,0,Tp,TP_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(XM_PROJ,3,Xm,XM_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(YM_PROJ,2,Ym,YM_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(ZM_PROJ,1,Zm,ZM_RECON_ACCUM); + HAND_STENCIL_LEG_EXT(TM_PROJ,0,Tm,TM_RECON_ACCUM); + HAND_RESULT_EXT(ss); } - //////////////////////////////////////////////// - // Specialise Gparity to simple implementation - //////////////////////////////////////////////// -#define HAND_SPECIALISE_EMPTY(IMPL) \ - template<> void \ - WilsonKernels::HandDhopSite(StencilImpl &st, \ - LebesgueOrder &lo, \ - DoubledGaugeField &U, \ - SiteHalfSpinor *buf, \ - int sF,int sU, \ - const FermionField &in, \ - FermionField &out){ assert(0); } \ - template<> void \ - WilsonKernels::HandDhopSiteDag(StencilImpl &st, \ - LebesgueOrder &lo, \ - DoubledGaugeField &U, \ - SiteHalfSpinor *buf, \ - int sF,int sU, \ - const FermionField &in, \ - FermionField &out){ assert(0); } \ - template<> void \ - WilsonKernels::HandDhopSiteInt(StencilImpl &st, \ - LebesgueOrder &lo, \ - DoubledGaugeField &U, \ - SiteHalfSpinor *buf, \ - int sF,int sU, \ - const FermionField &in, \ - FermionField &out){ assert(0); } \ - template<> void \ - WilsonKernels::HandDhopSiteExt(StencilImpl &st, \ - LebesgueOrder &lo, \ - DoubledGaugeField &U, \ - SiteHalfSpinor *buf, \ - int sF,int sU, \ - const FermionField &in, \ - FermionField &out){ assert(0); } \ - template<> void \ - WilsonKernels::HandDhopSiteDagInt(StencilImpl &st, \ - LebesgueOrder &lo, \ - DoubledGaugeField &U, \ - SiteHalfSpinor *buf, \ - int sF,int sU, \ - const FermionField &in, \ - FermionField &out){ assert(0); } \ - template<> void \ - WilsonKernels::HandDhopSiteDagExt(StencilImpl &st, \ - LebesgueOrder &lo, \ - DoubledGaugeField &U, \ - SiteHalfSpinor *buf, \ - int sF,int sU, \ - const FermionField &in, \ - FermionField &out){ assert(0); } \ - - - -#define HAND_SPECIALISE_GPARITY(IMPL) \ - template<> void \ - WilsonKernels::HandDhopSite(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ - int ss,int sU,const FermionField &in, FermionField &out) \ - { \ - typedef IMPL Impl; \ - typedef typename Simd::scalar_type S; \ - typedef typename Simd::vector_type V; \ - \ - HAND_DECLARATIONS(ignore); \ - \ - int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ - StencilEntry *SE; \ - HAND_DOP_SITE(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - HAND_DOP_SITE(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - } \ - \ - template<> \ - void WilsonKernels::HandDhopSiteDag(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ - int ss,int sU,const FermionField &in, FermionField &out) \ - { \ - typedef IMPL Impl; \ - typedef typename Simd::scalar_type S; \ - typedef typename Simd::vector_type V; \ - \ - HAND_DECLARATIONS(ignore); \ - \ - StencilEntry *SE; \ - int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ - HAND_DOP_SITE_DAG(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - HAND_DOP_SITE_DAG(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - } \ - \ - template<> void \ - WilsonKernels::HandDhopSiteInt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ - int ss,int sU,const FermionField &in, FermionField &out) \ - { \ - typedef IMPL Impl; \ - typedef typename Simd::scalar_type S; \ - typedef typename Simd::vector_type V; \ - \ - HAND_DECLARATIONS(ignore); \ - \ - int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ - StencilEntry *SE; \ - HAND_DOP_SITE_INT(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - HAND_DOP_SITE_INT(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - } \ - \ - template<> \ - void WilsonKernels::HandDhopSiteDagInt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ - int ss,int sU,const FermionField &in, FermionField &out) \ - { \ - typedef IMPL Impl; \ - typedef typename Simd::scalar_type S; \ - typedef typename Simd::vector_type V; \ - \ - HAND_DECLARATIONS(ignore); \ - \ - StencilEntry *SE; \ - int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ - HAND_DOP_SITE_DAG_INT(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - HAND_DOP_SITE_DAG_INT(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - } \ - \ - template<> void \ - WilsonKernels::HandDhopSiteExt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ - int ss,int sU,const FermionField &in, FermionField &out) \ - { \ - typedef IMPL Impl; \ - typedef typename Simd::scalar_type S; \ - typedef typename Simd::vector_type V; \ - \ - HAND_DECLARATIONS(ignore); \ - \ - int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ - StencilEntry *SE; \ - int nmu=0; \ - HAND_DOP_SITE_EXT(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - nmu = 0; \ - HAND_DOP_SITE_EXT(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - } \ - template<> \ - void WilsonKernels::HandDhopSiteDagExt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ - int ss,int sU,const FermionField &in, FermionField &out) \ - { \ - typedef IMPL Impl; \ - typedef typename Simd::scalar_type S; \ - typedef typename Simd::vector_type V; \ - \ - HAND_DECLARATIONS(ignore); \ - \ - StencilEntry *SE; \ - int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ - int nmu=0; \ - HAND_DOP_SITE_DAG_EXT(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - nmu = 0; \ - HAND_DOP_SITE_DAG_EXT(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ - } - - -HAND_SPECIALISE_GPARITY(GparityWilsonImplF); -HAND_SPECIALISE_GPARITY(GparityWilsonImplD); -HAND_SPECIALISE_GPARITY(GparityWilsonImplFH); -HAND_SPECIALISE_GPARITY(GparityWilsonImplDF); - - - - - - - - - - - ////////////// Wilson ; uses this implementation ///////////////////// #define INSTANTIATE_THEM(A) \ @@ -930,8 +613,6 @@ INSTANTIATE_THEM(WilsonImplF); INSTANTIATE_THEM(WilsonImplD); INSTANTIATE_THEM(ZWilsonImplF); INSTANTIATE_THEM(ZWilsonImplD); -INSTANTIATE_THEM(GparityWilsonImplF); -INSTANTIATE_THEM(GparityWilsonImplD); INSTANTIATE_THEM(DomainWallVec5dImplF); INSTANTIATE_THEM(DomainWallVec5dImplD); INSTANTIATE_THEM(ZDomainWallVec5dImplF); @@ -940,12 +621,11 @@ INSTANTIATE_THEM(WilsonImplFH); INSTANTIATE_THEM(WilsonImplDF); INSTANTIATE_THEM(ZWilsonImplFH); INSTANTIATE_THEM(ZWilsonImplDF); -INSTANTIATE_THEM(GparityWilsonImplFH); -INSTANTIATE_THEM(GparityWilsonImplDF); INSTANTIATE_THEM(DomainWallVec5dImplFH); INSTANTIATE_THEM(DomainWallVec5dImplDF); INSTANTIATE_THEM(ZDomainWallVec5dImplFH); INSTANTIATE_THEM(ZDomainWallVec5dImplDF); INSTANTIATE_THEM(WilsonTwoIndexAntiSymmetricImplF); INSTANTIATE_THEM(WilsonTwoIndexAntiSymmetricImplD); + }} From 6f6c5c549a407441007e76a3f41c86bd0f3bf139 Mon Sep 17 00:00:00 2001 From: Peter Boyle Date: Wed, 2 May 2018 14:11:23 +0100 Subject: [PATCH 07/15] Split off gparity --- .../fermion/WilsonKernelsHandGparity.cc | 878 ++++++++++++++++++ 1 file changed, 878 insertions(+) create mode 100644 lib/qcd/action/fermion/WilsonKernelsHandGparity.cc diff --git a/lib/qcd/action/fermion/WilsonKernelsHandGparity.cc b/lib/qcd/action/fermion/WilsonKernelsHandGparity.cc new file mode 100644 index 00000000..3bf2f10a --- /dev/null +++ b/lib/qcd/action/fermion/WilsonKernelsHandGparity.cc @@ -0,0 +1,878 @@ + /************************************************************************************* + + Grid physics library, www.github.com/paboyle/Grid + + Source file: ./lib/qcd/action/fermion/WilsonKernelsHand.cc + + Copyright (C) 2015 + +Author: Peter Boyle +Author: paboyle + + This program is free software; you can redistribute it and/or modify + it under the terms of the GNU General Public License as published by + the Free Software Foundation; either version 2 of the License, or + (at your option) any later version. + + This program is distributed in the hope that it will be useful, + but WITHOUT ANY WARRANTY; without even the implied warranty of + MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the + GNU General Public License for more details. + + You should have received a copy of the GNU General Public License along + with this program; if not, write to the Free Software Foundation, Inc., + 51 Franklin Street, Fifth Floor, Boston, MA 02110-1301 USA. + + See the full license in the file "LICENSE" in the top level distribution directory + *************************************************************************************/ + /* END LEGAL */ +#include + +#define REGISTER + +#define LOAD_CHIMU_BODY(F) \ + Chimu_00=ref(F)(0)(0); \ + Chimu_01=ref(F)(0)(1); \ + Chimu_02=ref(F)(0)(2); \ + Chimu_10=ref(F)(1)(0); \ + Chimu_11=ref(F)(1)(1); \ + Chimu_12=ref(F)(1)(2); \ + Chimu_20=ref(F)(2)(0); \ + Chimu_21=ref(F)(2)(1); \ + Chimu_22=ref(F)(2)(2); \ + Chimu_30=ref(F)(3)(0); \ + Chimu_31=ref(F)(3)(1); \ + Chimu_32=ref(F)(3)(2) + +#define LOAD_CHIMU(DIR,F,PERM) \ + { const SiteSpinor & ref (in._odata[offset]); LOAD_CHIMU_BODY(F); } + +#define LOAD_CHI_BODY(F) \ + Chi_00 = ref(F)(0)(0);\ + Chi_01 = ref(F)(0)(1);\ + Chi_02 = ref(F)(0)(2);\ + Chi_10 = ref(F)(1)(0);\ + Chi_11 = ref(F)(1)(1);\ + Chi_12 = ref(F)(1)(2) + +#define LOAD_CHI(DIR,F,PERM) \ + {const SiteHalfSpinor &ref(buf[offset]); LOAD_CHI_BODY(F); } + + +//G-parity implementations using in-place intrinsic ops + +//1l 1h -> 1h 1l +//0l 0h , 1h 1l -> 0l 1h 0h,1l +//0h,1l -> 1l,0h +//if( (distance == 1 && !perm_will_occur) || (distance == -1 && perm_will_occur) ) +//Pulled fermion through forwards face, GPBC on upper component +//Need 0= 0l 1h 1= 1l 0h +//else if( (distance == -1 && !perm) || (distance == 1 && perm) ) +//Pulled fermion through backwards face, GPBC on lower component +//Need 0= 1l 0h 1= 0l 1h + +//1l 1h -> 1h 1l +//0l 0h , 1h 1l -> 0l 1h 0h,1l +#define DO_TWIST_0L_1H(INTO,S,C,F, PERM, tmp1, tmp2, tmp3) \ + permute##PERM(tmp1, ref(1)(S)(C)); \ + exchange##PERM(tmp2,tmp3, ref(0)(S)(C), tmp1); \ + INTO = tmp2; + +//0l 0h -> 0h 0l +//1l 1h, 0h 0l -> 1l 0h, 1h 0l +#define DO_TWIST_1L_0H(INTO,S,C,F, PERM, tmp1, tmp2, tmp3) \ + permute##PERM(tmp1, ref(0)(S)(C)); \ + exchange##PERM(tmp2,tmp3, ref(1)(S)(C), tmp1); \ + INTO = tmp2; + + + + +#define LOAD_CHI_SETUP(DIR,F) \ + g = F; \ + direction = st._directions[DIR]; \ + distance = st._distances[DIR]; \ + sl = st._grid->_simd_layout[direction]; \ + inplace_twist = 0; \ + if(SE->_around_the_world && this->Params.twists[DIR % 4]){ \ + if(sl == 1){ \ + g = (F+1) % 2; \ + }else{ \ + inplace_twist = 1; \ + } \ + } + +#define LOAD_CHIMU_GPARITY_INPLACE_TWIST(DIR,F,PERM) \ + { const SiteSpinor &ref(in._odata[offset]); \ + LOAD_CHI_SETUP(DIR,F); \ + if(!inplace_twist){ \ + LOAD_CHIMU_BODY(g); \ + }else{ \ + if( ( F==0 && ((distance == 1 && !perm) || (distance == -1 && perm)) ) || \ + ( F==1 && ((distance == -1 && !perm) || (distance == 1 && perm)) ) ){ \ + DO_TWIST_0L_1H(Chimu_00,0,0,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_0L_1H(Chimu_01,0,1,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_0L_1H(Chimu_02,0,2,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_0L_1H(Chimu_10,1,0,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_0L_1H(Chimu_11,1,1,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_0L_1H(Chimu_12,1,2,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_0L_1H(Chimu_20,2,0,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_0L_1H(Chimu_21,2,1,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_0L_1H(Chimu_22,2,2,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_0L_1H(Chimu_30,3,0,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_0L_1H(Chimu_31,3,1,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_0L_1H(Chimu_32,3,2,F,PERM, U_11,U_20,U_21); \ + }else{ \ + DO_TWIST_1L_0H(Chimu_00,0,0,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_1L_0H(Chimu_01,0,1,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_1L_0H(Chimu_02,0,2,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_1L_0H(Chimu_10,1,0,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_1L_0H(Chimu_11,1,1,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_1L_0H(Chimu_12,1,2,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_1L_0H(Chimu_20,2,0,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_1L_0H(Chimu_21,2,1,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_1L_0H(Chimu_22,2,2,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_1L_0H(Chimu_30,3,0,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_1L_0H(Chimu_31,3,1,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_1L_0H(Chimu_32,3,2,F,PERM, U_11,U_20,U_21); \ + } \ + } \ + } + + +#define LOAD_CHI_GPARITY_INPLACE_TWIST(DIR,F,PERM) \ + { const SiteHalfSpinor &ref(buf[offset]); \ + LOAD_CHI_SETUP(DIR,F); \ + if(!inplace_twist){ \ + LOAD_CHI_BODY(g); \ + }else{ \ + if( ( F==0 && ((distance == 1 && !perm) || (distance == -1 && perm)) ) || \ + ( F==1 && ((distance == -1 && !perm) || (distance == 1 && perm)) ) ){ \ + DO_TWIST_0L_1H(Chi_00,0,0,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_0L_1H(Chi_01,0,1,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_0L_1H(Chi_02,0,2,F,PERM, UChi_00,UChi_01,UChi_02); \ + DO_TWIST_0L_1H(Chi_10,1,0,F,PERM, UChi_10,UChi_11,UChi_12); \ + DO_TWIST_0L_1H(Chi_11,1,1,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_0L_1H(Chi_12,1,2,F,PERM, U_11,U_20,U_21); \ + }else{ \ + DO_TWIST_1L_0H(Chi_00,0,0,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_1L_0H(Chi_01,0,1,F,PERM, U_11,U_20,U_21); \ + DO_TWIST_1L_0H(Chi_02,0,2,F,PERM, UChi_00,UChi_01,UChi_02); \ + DO_TWIST_1L_0H(Chi_10,1,0,F,PERM, UChi_10,UChi_11,UChi_12); \ + DO_TWIST_1L_0H(Chi_11,1,1,F,PERM, U_00,U_01,U_10); \ + DO_TWIST_1L_0H(Chi_12,1,2,F,PERM, U_11,U_20,U_21); \ + } \ + } \ + } + + +#define LOAD_CHI_GPARITY(DIR,F,PERM) LOAD_CHI_GPARITY_INPLACE_TWIST(DIR,F,PERM) +#define LOAD_CHIMU_GPARITY(DIR,F,PERM) LOAD_CHIMU_GPARITY_INPLACE_TWIST(DIR,F,PERM) + +// To splat or not to splat depends on the implementation +#define MULT_2SPIN_BODY \ + Impl::loadLinkElement(U_00,ref()(0,0)); \ + Impl::loadLinkElement(U_10,ref()(1,0)); \ + Impl::loadLinkElement(U_20,ref()(2,0)); \ + Impl::loadLinkElement(U_01,ref()(0,1)); \ + Impl::loadLinkElement(U_11,ref()(1,1)); \ + Impl::loadLinkElement(U_21,ref()(2,1)); \ + UChi_00 = U_00*Chi_00; \ + UChi_10 = U_00*Chi_10; \ + UChi_01 = U_10*Chi_00; \ + UChi_11 = U_10*Chi_10; \ + UChi_02 = U_20*Chi_00; \ + UChi_12 = U_20*Chi_10; \ + UChi_00+= U_01*Chi_01; \ + UChi_10+= U_01*Chi_11; \ + UChi_01+= U_11*Chi_01; \ + UChi_11+= U_11*Chi_11; \ + UChi_02+= U_21*Chi_01; \ + UChi_12+= U_21*Chi_11; \ + Impl::loadLinkElement(U_00,ref()(0,2)); \ + Impl::loadLinkElement(U_10,ref()(1,2)); \ + Impl::loadLinkElement(U_20,ref()(2,2)); \ + UChi_00+= U_00*Chi_02; \ + UChi_10+= U_00*Chi_12; \ + UChi_01+= U_10*Chi_02; \ + UChi_11+= U_10*Chi_12; \ + UChi_02+= U_20*Chi_02; \ + UChi_12+= U_20*Chi_12 + + +#define MULT_2SPIN(A,F) \ + {auto & ref(U._odata[sU](A)); MULT_2SPIN_BODY; } + +#define MULT_2SPIN_GPARITY(A,F) \ + {auto & ref(U._odata[sU](F)(A)); MULT_2SPIN_BODY; } + + +#define PERMUTE_DIR(dir) \ + permute##dir(Chi_00,Chi_00);\ + permute##dir(Chi_01,Chi_01);\ + permute##dir(Chi_02,Chi_02);\ + permute##dir(Chi_10,Chi_10);\ + permute##dir(Chi_11,Chi_11);\ + permute##dir(Chi_12,Chi_12); + +// hspin(0)=fspin(0)+timesI(fspin(3)); +// hspin(1)=fspin(1)+timesI(fspin(2)); +#define XP_PROJ \ + Chi_00 = Chimu_00+timesI(Chimu_30);\ + Chi_01 = Chimu_01+timesI(Chimu_31);\ + Chi_02 = Chimu_02+timesI(Chimu_32);\ + Chi_10 = Chimu_10+timesI(Chimu_20);\ + Chi_11 = Chimu_11+timesI(Chimu_21);\ + Chi_12 = Chimu_12+timesI(Chimu_22); + +#define YP_PROJ \ + Chi_00 = Chimu_00-Chimu_30;\ + Chi_01 = Chimu_01-Chimu_31;\ + Chi_02 = Chimu_02-Chimu_32;\ + Chi_10 = Chimu_10+Chimu_20;\ + Chi_11 = Chimu_11+Chimu_21;\ + Chi_12 = Chimu_12+Chimu_22; + +#define ZP_PROJ \ + Chi_00 = Chimu_00+timesI(Chimu_20); \ + Chi_01 = Chimu_01+timesI(Chimu_21); \ + Chi_02 = Chimu_02+timesI(Chimu_22); \ + Chi_10 = Chimu_10-timesI(Chimu_30); \ + Chi_11 = Chimu_11-timesI(Chimu_31); \ + Chi_12 = Chimu_12-timesI(Chimu_32); + +#define TP_PROJ \ + Chi_00 = Chimu_00+Chimu_20; \ + Chi_01 = Chimu_01+Chimu_21; \ + Chi_02 = Chimu_02+Chimu_22; \ + Chi_10 = Chimu_10+Chimu_30; \ + Chi_11 = Chimu_11+Chimu_31; \ + Chi_12 = Chimu_12+Chimu_32; + + +// hspin(0)=fspin(0)-timesI(fspin(3)); +// hspin(1)=fspin(1)-timesI(fspin(2)); +#define XM_PROJ \ + Chi_00 = Chimu_00-timesI(Chimu_30);\ + Chi_01 = Chimu_01-timesI(Chimu_31);\ + Chi_02 = Chimu_02-timesI(Chimu_32);\ + Chi_10 = Chimu_10-timesI(Chimu_20);\ + Chi_11 = Chimu_11-timesI(Chimu_21);\ + Chi_12 = Chimu_12-timesI(Chimu_22); + +#define YM_PROJ \ + Chi_00 = Chimu_00+Chimu_30;\ + Chi_01 = Chimu_01+Chimu_31;\ + Chi_02 = Chimu_02+Chimu_32;\ + Chi_10 = Chimu_10-Chimu_20;\ + Chi_11 = Chimu_11-Chimu_21;\ + Chi_12 = Chimu_12-Chimu_22; + +#define ZM_PROJ \ + Chi_00 = Chimu_00-timesI(Chimu_20); \ + Chi_01 = Chimu_01-timesI(Chimu_21); \ + Chi_02 = Chimu_02-timesI(Chimu_22); \ + Chi_10 = Chimu_10+timesI(Chimu_30); \ + Chi_11 = Chimu_11+timesI(Chimu_31); \ + Chi_12 = Chimu_12+timesI(Chimu_32); + +#define TM_PROJ \ + Chi_00 = Chimu_00-Chimu_20; \ + Chi_01 = Chimu_01-Chimu_21; \ + Chi_02 = Chimu_02-Chimu_22; \ + Chi_10 = Chimu_10-Chimu_30; \ + Chi_11 = Chimu_11-Chimu_31; \ + Chi_12 = Chimu_12-Chimu_32; + +// fspin(0)=hspin(0); +// fspin(1)=hspin(1); +// fspin(2)=timesMinusI(hspin(1)); +// fspin(3)=timesMinusI(hspin(0)); +#define XP_RECON\ + result_00 = UChi_00;\ + result_01 = UChi_01;\ + result_02 = UChi_02;\ + result_10 = UChi_10;\ + result_11 = UChi_11;\ + result_12 = UChi_12;\ + result_20 = timesMinusI(UChi_10);\ + result_21 = timesMinusI(UChi_11);\ + result_22 = timesMinusI(UChi_12);\ + result_30 = timesMinusI(UChi_00);\ + result_31 = timesMinusI(UChi_01);\ + result_32 = timesMinusI(UChi_02); + +#define XP_RECON_ACCUM\ + result_00+=UChi_00;\ + result_01+=UChi_01;\ + result_02+=UChi_02;\ + result_10+=UChi_10;\ + result_11+=UChi_11;\ + result_12+=UChi_12;\ + result_20-=timesI(UChi_10);\ + result_21-=timesI(UChi_11);\ + result_22-=timesI(UChi_12);\ + result_30-=timesI(UChi_00);\ + result_31-=timesI(UChi_01);\ + result_32-=timesI(UChi_02); + +#define XM_RECON\ + result_00 = UChi_00;\ + result_01 = UChi_01;\ + result_02 = UChi_02;\ + result_10 = UChi_10;\ + result_11 = UChi_11;\ + result_12 = UChi_12;\ + result_20 = timesI(UChi_10);\ + result_21 = timesI(UChi_11);\ + result_22 = timesI(UChi_12);\ + result_30 = timesI(UChi_00);\ + result_31 = timesI(UChi_01);\ + result_32 = timesI(UChi_02); + +#define XM_RECON_ACCUM\ + result_00+= UChi_00;\ + result_01+= UChi_01;\ + result_02+= UChi_02;\ + result_10+= UChi_10;\ + result_11+= UChi_11;\ + result_12+= UChi_12;\ + result_20+= timesI(UChi_10);\ + result_21+= timesI(UChi_11);\ + result_22+= timesI(UChi_12);\ + result_30+= timesI(UChi_00);\ + result_31+= timesI(UChi_01);\ + result_32+= timesI(UChi_02); + +#define YP_RECON_ACCUM\ + result_00+= UChi_00;\ + result_01+= UChi_01;\ + result_02+= UChi_02;\ + result_10+= UChi_10;\ + result_11+= UChi_11;\ + result_12+= UChi_12;\ + result_20+= UChi_10;\ + result_21+= UChi_11;\ + result_22+= UChi_12;\ + result_30-= UChi_00;\ + result_31-= UChi_01;\ + result_32-= UChi_02; + +#define YM_RECON_ACCUM\ + result_00+= UChi_00;\ + result_01+= UChi_01;\ + result_02+= UChi_02;\ + result_10+= UChi_10;\ + result_11+= UChi_11;\ + result_12+= UChi_12;\ + result_20-= UChi_10;\ + result_21-= UChi_11;\ + result_22-= UChi_12;\ + result_30+= UChi_00;\ + result_31+= UChi_01;\ + result_32+= UChi_02; + +#define ZP_RECON_ACCUM\ + result_00+= UChi_00;\ + result_01+= UChi_01;\ + result_02+= UChi_02;\ + result_10+= UChi_10;\ + result_11+= UChi_11;\ + result_12+= UChi_12;\ + result_20-= timesI(UChi_00); \ + result_21-= timesI(UChi_01); \ + result_22-= timesI(UChi_02); \ + result_30+= timesI(UChi_10); \ + result_31+= timesI(UChi_11); \ + result_32+= timesI(UChi_12); + +#define ZM_RECON_ACCUM\ + result_00+= UChi_00;\ + result_01+= UChi_01;\ + result_02+= UChi_02;\ + result_10+= UChi_10;\ + result_11+= UChi_11;\ + result_12+= UChi_12;\ + result_20+= timesI(UChi_00); \ + result_21+= timesI(UChi_01); \ + result_22+= timesI(UChi_02); \ + result_30-= timesI(UChi_10); \ + result_31-= timesI(UChi_11); \ + result_32-= timesI(UChi_12); + +#define TP_RECON_ACCUM\ + result_00+= UChi_00;\ + result_01+= UChi_01;\ + result_02+= UChi_02;\ + result_10+= UChi_10;\ + result_11+= UChi_11;\ + result_12+= UChi_12;\ + result_20+= UChi_00; \ + result_21+= UChi_01; \ + result_22+= UChi_02; \ + result_30+= UChi_10; \ + result_31+= UChi_11; \ + result_32+= UChi_12; + +#define TM_RECON_ACCUM\ + result_00+= UChi_00;\ + result_01+= UChi_01;\ + result_02+= UChi_02;\ + result_10+= UChi_10;\ + result_11+= UChi_11;\ + result_12+= UChi_12;\ + result_20-= UChi_00; \ + result_21-= UChi_01; \ + result_22-= UChi_02; \ + result_30-= UChi_10; \ + result_31-= UChi_11; \ + result_32-= UChi_12; + +#define HAND_STENCIL_LEG(PROJ,PERM,DIR,RECON,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ + SE=st.GetEntry(ptype,DIR,ss); \ + offset = SE->_offset; \ + local = SE->_is_local; \ + perm = SE->_permute; \ + if ( local ) { \ + LOAD_CHIMU_IMPL(DIR,F,PERM); \ + PROJ; \ + if ( perm) { \ + PERMUTE_DIR(PERM); \ + } \ + } else { \ + LOAD_CHI_IMPL(DIR,F,PERM); \ + } \ + MULT_2SPIN_IMPL(DIR,F); \ + RECON; + + +#define HAND_STENCIL_LEG_INT(PROJ,PERM,DIR,RECON,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ + SE=st.GetEntry(ptype,DIR,ss); \ + offset = SE->_offset; \ + local = SE->_is_local; \ + perm = SE->_permute; \ + if ( local ) { \ + LOAD_CHIMU_IMPL(DIR,F,PERM); \ + PROJ; \ + if ( perm) { \ + PERMUTE_DIR(PERM); \ + } \ + } else if ( st.same_node[DIR] ) { \ + LOAD_CHI_IMPL(DIR,F,PERM); \ + } \ + if (local || st.same_node[DIR] ) { \ + MULT_2SPIN_IMPL(DIR,F); \ + RECON; \ + } + +#define HAND_STENCIL_LEG_EXT(PROJ,PERM,DIR,RECON,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ + SE=st.GetEntry(ptype,DIR,ss); \ + offset = SE->_offset; \ + local = SE->_is_local; \ + perm = SE->_permute; \ + if((!SE->_is_local)&&(!st.same_node[DIR]) ) { \ + LOAD_CHI_IMPL(DIR,F,PERM); \ + MULT_2SPIN_IMPL(DIR,F); \ + RECON; \ + nmu++; \ + } + +#define HAND_RESULT(ss,F) \ + { \ + SiteSpinor & ref (out._odata[ss]); \ + vstream(ref(F)(0)(0),result_00); \ + vstream(ref(F)(0)(1),result_01); \ + vstream(ref(F)(0)(2),result_02); \ + vstream(ref(F)(1)(0),result_10); \ + vstream(ref(F)(1)(1),result_11); \ + vstream(ref(F)(1)(2),result_12); \ + vstream(ref(F)(2)(0),result_20); \ + vstream(ref(F)(2)(1),result_21); \ + vstream(ref(F)(2)(2),result_22); \ + vstream(ref(F)(3)(0),result_30); \ + vstream(ref(F)(3)(1),result_31); \ + vstream(ref(F)(3)(2),result_32); \ + } + +#define HAND_RESULT_EXT(ss,F) \ + if (nmu){ \ + SiteSpinor & ref (out._odata[ss]); \ + ref(F)(0)(0)+=result_00; \ + ref(F)(0)(1)+=result_01; \ + ref(F)(0)(2)+=result_02; \ + ref(F)(1)(0)+=result_10; \ + ref(F)(1)(1)+=result_11; \ + ref(F)(1)(2)+=result_12; \ + ref(F)(2)(0)+=result_20; \ + ref(F)(2)(1)+=result_21; \ + ref(F)(2)(2)+=result_22; \ + ref(F)(3)(0)+=result_30; \ + ref(F)(3)(1)+=result_31; \ + ref(F)(3)(2)+=result_32; \ + } + + +#define HAND_DECLARATIONS(a) \ + Simd result_00; \ + Simd result_01; \ + Simd result_02; \ + Simd result_10; \ + Simd result_11; \ + Simd result_12; \ + Simd result_20; \ + Simd result_21; \ + Simd result_22; \ + Simd result_30; \ + Simd result_31; \ + Simd result_32; \ + Simd Chi_00; \ + Simd Chi_01; \ + Simd Chi_02; \ + Simd Chi_10; \ + Simd Chi_11; \ + Simd Chi_12; \ + Simd UChi_00; \ + Simd UChi_01; \ + Simd UChi_02; \ + Simd UChi_10; \ + Simd UChi_11; \ + Simd UChi_12; \ + Simd U_00; \ + Simd U_10; \ + Simd U_20; \ + Simd U_01; \ + Simd U_11; \ + Simd U_21; + +#define ZERO_RESULT \ + result_00=zero; \ + result_01=zero; \ + result_02=zero; \ + result_10=zero; \ + result_11=zero; \ + result_12=zero; \ + result_20=zero; \ + result_21=zero; \ + result_22=zero; \ + result_30=zero; \ + result_31=zero; \ + result_32=zero; + +#define Chimu_00 Chi_00 +#define Chimu_01 Chi_01 +#define Chimu_02 Chi_02 +#define Chimu_10 Chi_10 +#define Chimu_11 Chi_11 +#define Chimu_12 Chi_12 +#define Chimu_20 UChi_00 +#define Chimu_21 UChi_01 +#define Chimu_22 UChi_02 +#define Chimu_30 UChi_10 +#define Chimu_31 UChi_11 +#define Chimu_32 UChi_12 + +namespace Grid { +namespace QCD { + +template void +WilsonKernels::HandDhopSite(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, + int ss,int sU,const FermionField &in, FermionField &out) +{ +// T==0, Z==1, Y==2, Z==3 expect 1,2,2,2 simd layout etc... + typedef typename Simd::scalar_type S; + typedef typename Simd::vector_type V; + + HAND_DECLARATIONS(ignore); + + int offset,local,perm, ptype; + StencilEntry *SE; + +#define HAND_DOP_SITE(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ + HAND_STENCIL_LEG(XM_PROJ,3,Xp,XM_RECON,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(YM_PROJ,2,Yp,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(ZM_PROJ,1,Zp,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(TM_PROJ,0,Tp,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(XP_PROJ,3,Xm,XP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(YP_PROJ,2,Ym,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(ZP_PROJ,1,Zm,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(TP_PROJ,0,Tm,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_RESULT(ss,F) + + HAND_DOP_SITE(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); +} + +template +void WilsonKernels::HandDhopSiteDag(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, + int ss,int sU,const FermionField &in, FermionField &out) +{ + typedef typename Simd::scalar_type S; + typedef typename Simd::vector_type V; + + HAND_DECLARATIONS(ignore); + + StencilEntry *SE; + int offset,local,perm, ptype; + +#define HAND_DOP_SITE_DAG(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ + HAND_STENCIL_LEG(XP_PROJ,3,Xp,XP_RECON,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(YP_PROJ,2,Yp,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(ZP_PROJ,1,Zp,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(TP_PROJ,0,Tp,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(XM_PROJ,3,Xm,XM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(YM_PROJ,2,Ym,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(ZM_PROJ,1,Zm,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG(TM_PROJ,0,Tm,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_RESULT(ss,F) + + HAND_DOP_SITE_DAG(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); +} + +template void +WilsonKernels::HandDhopSiteInt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, + int ss,int sU,const FermionField &in, FermionField &out) +{ +// T==0, Z==1, Y==2, Z==3 expect 1,2,2,2 simd layout etc... + typedef typename Simd::scalar_type S; + typedef typename Simd::vector_type V; + + HAND_DECLARATIONS(ignore); + + int offset,local,perm, ptype; + StencilEntry *SE; + +#define HAND_DOP_SITE_INT(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ + ZERO_RESULT; \ + HAND_STENCIL_LEG_INT(XM_PROJ,3,Xp,XM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(YM_PROJ,2,Yp,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(ZM_PROJ,1,Zp,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(TM_PROJ,0,Tp,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(XP_PROJ,3,Xm,XP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(YP_PROJ,2,Ym,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(ZP_PROJ,1,Zm,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(TP_PROJ,0,Tm,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_RESULT(ss,F) + + HAND_DOP_SITE_INT(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); +} + +template +void WilsonKernels::HandDhopSiteDagInt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, + int ss,int sU,const FermionField &in, FermionField &out) +{ + typedef typename Simd::scalar_type S; + typedef typename Simd::vector_type V; + + HAND_DECLARATIONS(ignore); + + StencilEntry *SE; + int offset,local,perm, ptype; + +#define HAND_DOP_SITE_DAG_INT(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ + ZERO_RESULT; \ + HAND_STENCIL_LEG_INT(XP_PROJ,3,Xp,XP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(YP_PROJ,2,Yp,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(ZP_PROJ,1,Zp,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(TP_PROJ,0,Tp,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(XM_PROJ,3,Xm,XM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(YM_PROJ,2,Ym,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(ZM_PROJ,1,Zm,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_INT(TM_PROJ,0,Tm,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_RESULT(ss,F) + + HAND_DOP_SITE_DAG_INT(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); +} + +template void +WilsonKernels::HandDhopSiteExt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, + int ss,int sU,const FermionField &in, FermionField &out) +{ +// T==0, Z==1, Y==2, Z==3 expect 1,2,2,2 simd layout etc... + typedef typename Simd::scalar_type S; + typedef typename Simd::vector_type V; + + HAND_DECLARATIONS(ignore); + + int offset,local,perm, ptype; + StencilEntry *SE; + int nmu=0; + +#define HAND_DOP_SITE_EXT(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ + ZERO_RESULT; \ + HAND_STENCIL_LEG_EXT(XM_PROJ,3,Xp,XM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(YM_PROJ,2,Yp,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(ZM_PROJ,1,Zp,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(TM_PROJ,0,Tp,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(XP_PROJ,3,Xm,XP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(YP_PROJ,2,Ym,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(ZP_PROJ,1,Zm,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(TP_PROJ,0,Tm,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_RESULT_EXT(ss,F) + + HAND_DOP_SITE_EXT(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); +} + +template +void WilsonKernels::HandDhopSiteDagExt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, + int ss,int sU,const FermionField &in, FermionField &out) +{ + typedef typename Simd::scalar_type S; + typedef typename Simd::vector_type V; + + HAND_DECLARATIONS(ignore); + + StencilEntry *SE; + int offset,local,perm, ptype; + int nmu=0; + +#define HAND_DOP_SITE_DAG_EXT(F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL) \ + ZERO_RESULT; \ + HAND_STENCIL_LEG_EXT(XP_PROJ,3,Xp,XP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(YP_PROJ,2,Yp,YP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(ZP_PROJ,1,Zp,ZP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(TP_PROJ,0,Tp,TP_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(XM_PROJ,3,Xm,XM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(YM_PROJ,2,Ym,YM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(ZM_PROJ,1,Zm,ZM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_STENCIL_LEG_EXT(TM_PROJ,0,Tm,TM_RECON_ACCUM,F,LOAD_CHI_IMPL,LOAD_CHIMU_IMPL,MULT_2SPIN_IMPL); \ + HAND_RESULT_EXT(ss,F) + + HAND_DOP_SITE_DAG_EXT(, LOAD_CHI,LOAD_CHIMU,MULT_2SPIN); +} + +#define HAND_SPECIALISE_GPARITY(IMPL) \ + template<> void \ + WilsonKernels::HandDhopSite(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ + int ss,int sU,const FermionField &in, FermionField &out) \ + { \ + typedef IMPL Impl; \ + typedef typename Simd::scalar_type S; \ + typedef typename Simd::vector_type V; \ + \ + HAND_DECLARATIONS(ignore); \ + \ + int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ + StencilEntry *SE; \ + HAND_DOP_SITE(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + HAND_DOP_SITE(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + } \ + \ + template<> \ + void WilsonKernels::HandDhopSiteDag(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ + int ss,int sU,const FermionField &in, FermionField &out) \ + { \ + typedef IMPL Impl; \ + typedef typename Simd::scalar_type S; \ + typedef typename Simd::vector_type V; \ + \ + HAND_DECLARATIONS(ignore); \ + \ + StencilEntry *SE; \ + int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ + HAND_DOP_SITE_DAG(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + HAND_DOP_SITE_DAG(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + } \ + \ + template<> void \ + WilsonKernels::HandDhopSiteInt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ + int ss,int sU,const FermionField &in, FermionField &out) \ + { \ + typedef IMPL Impl; \ + typedef typename Simd::scalar_type S; \ + typedef typename Simd::vector_type V; \ + \ + HAND_DECLARATIONS(ignore); \ + \ + int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ + StencilEntry *SE; \ + HAND_DOP_SITE_INT(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + HAND_DOP_SITE_INT(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + } \ + \ + template<> \ + void WilsonKernels::HandDhopSiteDagInt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ + int ss,int sU,const FermionField &in, FermionField &out) \ + { \ + typedef IMPL Impl; \ + typedef typename Simd::scalar_type S; \ + typedef typename Simd::vector_type V; \ + \ + HAND_DECLARATIONS(ignore); \ + \ + StencilEntry *SE; \ + int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ + HAND_DOP_SITE_DAG_INT(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + HAND_DOP_SITE_DAG_INT(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + } \ + \ + template<> void \ + WilsonKernels::HandDhopSiteExt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ + int ss,int sU,const FermionField &in, FermionField &out) \ + { \ + typedef IMPL Impl; \ + typedef typename Simd::scalar_type S; \ + typedef typename Simd::vector_type V; \ + \ + HAND_DECLARATIONS(ignore); \ + \ + int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ + StencilEntry *SE; \ + int nmu=0; \ + HAND_DOP_SITE_EXT(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + nmu = 0; \ + HAND_DOP_SITE_EXT(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + } \ + template<> \ + void WilsonKernels::HandDhopSiteDagExt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ + int ss,int sU,const FermionField &in, FermionField &out) \ + { \ + typedef IMPL Impl; \ + typedef typename Simd::scalar_type S; \ + typedef typename Simd::vector_type V; \ + \ + HAND_DECLARATIONS(ignore); \ + \ + StencilEntry *SE; \ + int offset,local,perm, ptype, g, direction, distance, sl, inplace_twist; \ + int nmu=0; \ + HAND_DOP_SITE_DAG_EXT(0, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + nmu = 0; \ + HAND_DOP_SITE_DAG_EXT(1, LOAD_CHI_GPARITY,LOAD_CHIMU_GPARITY,MULT_2SPIN_GPARITY); \ + } + + +HAND_SPECIALISE_GPARITY(GparityWilsonImplF); +HAND_SPECIALISE_GPARITY(GparityWilsonImplD); +HAND_SPECIALISE_GPARITY(GparityWilsonImplFH); +HAND_SPECIALISE_GPARITY(GparityWilsonImplDF); + + + + + + + + + + + +////////////// Wilson ; uses this implementation ///////////////////// + +#define INSTANTIATE_THEM(A) \ +template void WilsonKernels::HandDhopSite(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf,\ + int ss,int sU,const FermionField &in, FermionField &out); \ +template void WilsonKernels::HandDhopSiteDag(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ + int ss,int sU,const FermionField &in, FermionField &out);\ +template void WilsonKernels::HandDhopSiteInt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf,\ + int ss,int sU,const FermionField &in, FermionField &out); \ +template void WilsonKernels::HandDhopSiteDagInt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ + int ss,int sU,const FermionField &in, FermionField &out); \ +template void WilsonKernels::HandDhopSiteExt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf,\ + int ss,int sU,const FermionField &in, FermionField &out); \ +template void WilsonKernels::HandDhopSiteDagExt(StencilImpl &st,LebesgueOrder &lo,DoubledGaugeField &U,SiteHalfSpinor *buf, \ + int ss,int sU,const FermionField &in, FermionField &out); + +INSTANTIATE_THEM(GparityWilsonImplF); +INSTANTIATE_THEM(GparityWilsonImplD); +INSTANTIATE_THEM(GparityWilsonImplFH); +INSTANTIATE_THEM(GparityWilsonImplDF); +}} From 6e7d5e22435eadba36fd8149821bd48e6aed0fc3 Mon Sep 17 00:00:00 2001 From: Guido Cossu Date: Wed, 2 May 2018 14:28:59 +0100 Subject: [PATCH 08/15] HMC: added Scidac checkpointer and support for metadata --- lib/qcd/hmc/HMCResourceManager.h | 19 ++- lib/qcd/hmc/checkpointers/BaseCheckpointer.h | 8 ++ .../hmc/checkpointers/BinaryCheckpointer.h | 3 + .../hmc/checkpointers/CheckPointerModules.h | 16 +++ lib/qcd/hmc/checkpointers/CheckPointers.h | 1 + lib/qcd/hmc/checkpointers/ILDGCheckpointer.h | 4 + lib/qcd/hmc/checkpointers/NerscCheckpointer.h | 3 + .../hmc/checkpointers/ScidacCheckpointer.h | 125 ++++++++++++++++++ tests/hmc/Test_hmc_WG_Production.cc | 11 +- 9 files changed, 185 insertions(+), 5 deletions(-) create mode 100644 lib/qcd/hmc/checkpointers/ScidacCheckpointer.h diff --git a/lib/qcd/hmc/HMCResourceManager.h b/lib/qcd/hmc/HMCResourceManager.h index 3e20a8c1..fcfaeaed 100644 --- a/lib/qcd/hmc/HMCResourceManager.h +++ b/lib/qcd/hmc/HMCResourceManager.h @@ -48,6 +48,22 @@ with this program; if not, write to the Free Software Foundation, Inc., } \ } +#define RegisterLoadCheckPointerMetadataFunction(NAME) \ + template < class Metadata > \ + void Load##NAME##Checkpointer(const CheckpointerParameters& Params_, const Metadata& M_) { \ + if (!have_CheckPointer) { \ + std::cout << GridLogDebug << "Loading Metadata Checkpointer " << #NAME \ + << std::endl; \ + CP = std::unique_ptr( \ + new NAME##CPModule(Params_, M_)); \ + have_CheckPointer = true; \ + } else { \ + std::cout << GridLogError << "Checkpointer already loaded " \ + << std::endl; \ + exit(1); \ + } \ + } + namespace Grid { namespace QCD { @@ -77,7 +93,7 @@ class HMCResourceManager { bool have_CheckPointer; // NOTE: operator << is not overloaded for std::vector - // so thsi function is necessary + // so this function is necessary void output_vector_string(const std::vector &vs){ for (auto &i: vs) std::cout << i << " "; @@ -254,6 +270,7 @@ class HMCResourceManager { RegisterLoadCheckPointerFunction(Nersc); #ifdef HAVE_LIME RegisterLoadCheckPointerFunction(ILDG); + RegisterLoadCheckPointerMetadataFunction(Scidac); #endif //////////////////////////////////////////////////////// diff --git a/lib/qcd/hmc/checkpointers/BaseCheckpointer.h b/lib/qcd/hmc/checkpointers/BaseCheckpointer.h index 9be9efca..f4ef252b 100644 --- a/lib/qcd/hmc/checkpointers/BaseCheckpointer.h +++ b/lib/qcd/hmc/checkpointers/BaseCheckpointer.h @@ -76,6 +76,14 @@ class BaseHmcCheckpointer : public HmcObservable { } } + void check_filename(const std::string &filename){ + std::ifstream f(filename.c_str()); + if(!f.good()){ + std::cout << GridLogError << "Filename " << filename << " not found. Aborting. " << std::endl; + abort(); + }; + } + virtual void initialize(const CheckpointerParameters &Params) = 0; virtual void CheckpointRestore(int traj, typename Impl::Field &U, diff --git a/lib/qcd/hmc/checkpointers/BinaryCheckpointer.h b/lib/qcd/hmc/checkpointers/BinaryCheckpointer.h index 59d655ad..025398eb 100644 --- a/lib/qcd/hmc/checkpointers/BinaryCheckpointer.h +++ b/lib/qcd/hmc/checkpointers/BinaryCheckpointer.h @@ -93,6 +93,9 @@ class BinaryHmcCheckpointer : public BaseHmcCheckpointer { void CheckpointRestore(int traj, Field &U, GridSerialRNG &sRNG, GridParallelRNG &pRNG) { std::string config, rng; this->build_filenames(traj, Params, config, rng); + this->check_filename(rng); + this->check_filename(config); + BinarySimpleMunger munge; diff --git a/lib/qcd/hmc/checkpointers/CheckPointerModules.h b/lib/qcd/hmc/checkpointers/CheckPointerModules.h index 5debedef..d49d6e72 100644 --- a/lib/qcd/hmc/checkpointers/CheckPointerModules.h +++ b/lib/qcd/hmc/checkpointers/CheckPointerModules.h @@ -136,6 +136,22 @@ class ILDGCPModule: public CheckPointerModule< ImplementationPolicy> { }; +template +class ScidacCPModule: public CheckPointerModule< ImplementationPolicy> { + typedef CheckPointerModule< ImplementationPolicy> CPBase; + Metadata M; + + //using CPBase::CPBase; // for constructors + + // acquire resource + virtual void initialize(){ + this->CheckPointPtr.reset(new ScidacHmcCheckpointer(this->Par_, M)); + } +public: + ScidacCPModule(typename CPBase::APar Par, Metadata M_):M(M_), CPBase(Par) {} + template + ScidacCPModule(Reader& Reader) : Parametrized(Reader){}; +}; #endif diff --git a/lib/qcd/hmc/checkpointers/CheckPointers.h b/lib/qcd/hmc/checkpointers/CheckPointers.h index 423ce45c..e7a5fa82 100644 --- a/lib/qcd/hmc/checkpointers/CheckPointers.h +++ b/lib/qcd/hmc/checkpointers/CheckPointers.h @@ -34,6 +34,7 @@ directory #include #include #include +#include //#include diff --git a/lib/qcd/hmc/checkpointers/ILDGCheckpointer.h b/lib/qcd/hmc/checkpointers/ILDGCheckpointer.h index 9bcc33df..f7e6b17e 100644 --- a/lib/qcd/hmc/checkpointers/ILDGCheckpointer.h +++ b/lib/qcd/hmc/checkpointers/ILDGCheckpointer.h @@ -95,6 +95,10 @@ class ILDGHmcCheckpointer : public BaseHmcCheckpointer { GridParallelRNG &pRNG) { std::string config, rng; this->build_filenames(traj, Params, config, rng); + this->check_filename(rng); + this->check_filename(config); + + uint32_t nersc_csum,scidac_csuma,scidac_csumb; BinaryIO::readRNG(sRNG, pRNG, rng, 0,nersc_csum,scidac_csuma,scidac_csumb); diff --git a/lib/qcd/hmc/checkpointers/NerscCheckpointer.h b/lib/qcd/hmc/checkpointers/NerscCheckpointer.h index a4b1b480..d452b994 100644 --- a/lib/qcd/hmc/checkpointers/NerscCheckpointer.h +++ b/lib/qcd/hmc/checkpointers/NerscCheckpointer.h @@ -69,6 +69,9 @@ class NerscHmcCheckpointer : public BaseHmcCheckpointer { GridParallelRNG &pRNG) { std::string config, rng; this->build_filenames(traj, Params, config, rng); + this->check_filename(rng); + this->check_filename(config); + FieldMetaData header; NerscIO::readRNGState(sRNG, pRNG, header, rng); diff --git a/lib/qcd/hmc/checkpointers/ScidacCheckpointer.h b/lib/qcd/hmc/checkpointers/ScidacCheckpointer.h new file mode 100644 index 00000000..0867b882 --- /dev/null +++ b/lib/qcd/hmc/checkpointers/ScidacCheckpointer.h @@ -0,0 +1,125 @@ +/************************************************************************************* + +Grid physics library, www.github.com/paboyle/Grid + +Source file: ./lib/qcd/hmc/ScidacCheckpointer.h + +Copyright (C) 2018 + +Author: Guido Cossu + +This program is free software; you can redistribute it and/or modify +it under the terms of the GNU General Public License as published by +the Free Software Foundation; either version 2 of the License, or +(at your option) any later version. + +This program is distributed in the hope that it will be useful, +but WITHOUT ANY WARRANTY; without even the implied warranty of +MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the +GNU General Public License for more details. + +You should have received a copy of the GNU General Public License along +with this program; if not, write to the Free Software Foundation, Inc., +51 Franklin Street, Fifth Floor, Boston, MA 02110-1301 USA. + +See the full license in the file "LICENSE" in the top level distribution +directory +*************************************************************************************/ +/* END LEGAL */ +#ifndef SCIDAC_CHECKPOINTER +#define SCIDAC_CHECKPOINTER + +#ifdef HAVE_LIME + +#include +#include +#include + +namespace Grid { +namespace QCD { + +// For generic fields +template +class ScidacHmcCheckpointer : public BaseHmcCheckpointer { + private: + CheckpointerParameters Params; + Metadata MData; + + typedef typename Implementation::Field Field; + + public: + //INHERIT_GIMPL_TYPES(Implementation); + + ScidacHmcCheckpointer(const CheckpointerParameters &Params_) { initialize(Params_); } + ScidacHmcCheckpointer(const CheckpointerParameters &Params_, const Metadata& M_):MData(M_) { initialize(Params_); } + + void initialize(const CheckpointerParameters &Params_) { + Params = Params_; + + // check here that the format is valid + int ieee32big = (Params.format == std::string("IEEE32BIG")); + int ieee32 = (Params.format == std::string("IEEE32")); + int ieee64big = (Params.format == std::string("IEEE64BIG")); + int ieee64 = (Params.format == std::string("IEEE64")); + + if (!(ieee64big || ieee32 || ieee32big || ieee64)) { + std::cout << GridLogError << "Unrecognized file format " << Params.format + << std::endl; + std::cout << GridLogError + << "Allowed: IEEE32BIG | IEEE32 | IEEE64BIG | IEEE64" + << std::endl; + + exit(1); + } + } + + void TrajectoryComplete(int traj, Field &U, GridSerialRNG &sRNG, + GridParallelRNG &pRNG) { + if ((traj % Params.saveInterval) == 0) { + std::string config, rng; + this->build_filenames(traj, Params, config, rng); + GridBase *grid = U._grid; + uint32_t nersc_csum,scidac_csuma,scidac_csumb; + BinaryIO::writeRNG(sRNG, pRNG, rng, 0,nersc_csum,scidac_csuma,scidac_csumb); + ScidacWriter _ScidacWriter(grid->IsBoss()); + _ScidacWriter.open(config); + _ScidacWriter.writeScidacFieldRecord(U, MData); + _ScidacWriter.close(); + + std::cout << GridLogMessage << "Written Scidac Configuration on " << config + << " checksum " << std::hex << nersc_csum<<"/" + << scidac_csuma<<"/" << scidac_csumb + << std::dec << std::endl; + } + }; + + void CheckpointRestore(int traj, Field &U, GridSerialRNG &sRNG, + GridParallelRNG &pRNG) { + std::string config, rng; + this->build_filenames(traj, Params, config, rng); + this->check_filename(rng); + this->check_filename(config); + + + uint32_t nersc_csum,scidac_csuma,scidac_csumb; + BinaryIO::readRNG(sRNG, pRNG, rng, 0,nersc_csum,scidac_csuma,scidac_csumb); + + Metadata md_content; + ScidacReader _ScidacReader; + _ScidacReader.open(config); + _ScidacReader.readScidacFieldRecord(U,md_content); // format from the header + _ScidacReader.close(); + + std::cout << GridLogMessage << "Read Scidac Configuration from " << config + << " checksum " << std::hex + << nersc_csum<<"/" + << scidac_csuma<<"/" + << scidac_csumb + << std::dec << std::endl; + }; +}; +} +} + +#endif // HAVE_LIME +#endif // ILDG_CHECKPOINTER diff --git a/tests/hmc/Test_hmc_WG_Production.cc b/tests/hmc/Test_hmc_WG_Production.cc index b99446d5..7f8d8124 100644 --- a/tests/hmc/Test_hmc_WG_Production.cc +++ b/tests/hmc/Test_hmc_WG_Production.cc @@ -33,6 +33,7 @@ namespace Grid{ GRID_SERIALIZABLE_CLASS_MEMBERS(ActionParameters, double, beta) + ActionParameters() = default; template ActionParameters(Reader& Reader){ @@ -68,11 +69,15 @@ int main(int argc, char **argv) { } Serialiser Reader(TheHMC.ParameterFile); - + // Read parameters from input file + ActionParameters WilsonPar(Reader); // Checkpointer definition CheckpointerParameters CPparams(Reader); - TheHMC.Resources.LoadNerscCheckpointer(CPparams); + //TheHMC.Resources.LoadNerscCheckpointer(CPparams); + + // Store metadata in the Scidac checkpointer + TheHMC.Resources.LoadScidacCheckpointer(CPparams, WilsonPar); RNGModuleParameters RNGpar(Reader); TheHMC.Resources.SetRNGSeeds(RNGpar); @@ -91,8 +96,6 @@ int main(int argc, char **argv) { // need wrappers of the fermionic classes // that have a complex construction // standard - ActionParameters WilsonPar(Reader); - //RealD beta = 6.4 ; WilsonGaugeActionR Waction(WilsonPar.beta); ActionLevel Level1(1); From b4583267447e3f9a68c2f689b757994bfbb8f25d Mon Sep 17 00:00:00 2001 From: Guido Cossu Date: Wed, 2 May 2018 14:29:22 +0100 Subject: [PATCH 09/15] Checkpointer module update --- lib/qcd/hmc/checkpointers/CheckPointerModules.h | 4 +--- 1 file changed, 1 insertion(+), 3 deletions(-) diff --git a/lib/qcd/hmc/checkpointers/CheckPointerModules.h b/lib/qcd/hmc/checkpointers/CheckPointerModules.h index d49d6e72..f17cd2c8 100644 --- a/lib/qcd/hmc/checkpointers/CheckPointerModules.h +++ b/lib/qcd/hmc/checkpointers/CheckPointerModules.h @@ -141,8 +141,6 @@ class ScidacCPModule: public CheckPointerModule< ImplementationPolicy> { typedef CheckPointerModule< ImplementationPolicy> CPBase; Metadata M; - //using CPBase::CPBase; // for constructors - // acquire resource virtual void initialize(){ this->CheckPointPtr.reset(new ScidacHmcCheckpointer(this->Par_, M)); @@ -150,7 +148,7 @@ class ScidacCPModule: public CheckPointerModule< ImplementationPolicy> { public: ScidacCPModule(typename CPBase::APar Par, Metadata M_):M(M_), CPBase(Par) {} template - ScidacCPModule(Reader& Reader) : Parametrized(Reader){}; + ScidacCPModule(Reader& Reader) : Parametrized(Reader), M(Reader){}; }; #endif From 4d4ac2517b3348c1cf517421ad3f3038fd623bea Mon Sep 17 00:00:00 2001 From: Guido Cossu Date: Wed, 2 May 2018 14:36:32 +0100 Subject: [PATCH 10/15] Adding Scalar field theory example for Scidac format --- tests/hmc/Test_hmc_ScalarActionNxN.cc | 8 ++++++-- 1 file changed, 6 insertions(+), 2 deletions(-) diff --git a/tests/hmc/Test_hmc_ScalarActionNxN.cc b/tests/hmc/Test_hmc_ScalarActionNxN.cc index 9e40beac..33e52d82 100644 --- a/tests/hmc/Test_hmc_ScalarActionNxN.cc +++ b/tests/hmc/Test_hmc_ScalarActionNxN.cc @@ -34,6 +34,8 @@ class ScalarActionParameters : Serializable { double, lambda, double, g); + ScalarActionParameters() = default; + template ScalarActionParameters(Reader& Reader){ read(Reader, "ScalarAction", *this); @@ -124,10 +126,13 @@ int main(int argc, char **argv) { ScalarGrid.set_rb(new GridRedBlackCartesian(ScalarGrid.get_full())); TheHMC.Resources.AddGrid("scalar", ScalarGrid); std::cout << "Lattice size : " << GridDefaultLatt() << std::endl; + + ScalarActionParameters SPar(Reader); // Checkpointer definition CheckpointerParameters CPparams(Reader); - TheHMC.Resources.LoadBinaryCheckpointer(CPparams); + //TheHMC.Resources.LoadBinaryCheckpointer(CPparams); + TheHMC.Resources.LoadScidacCheckpointer(CPparams, SPar); RNGModuleParameters RNGpar(Reader); TheHMC.Resources.SetRNGSeeds(RNGpar); @@ -140,7 +145,6 @@ int main(int argc, char **argv) { // Collect actions, here use more encapsulation // Scalar action in adjoint representation - ScalarActionParameters SPar(Reader); ScalarAction Saction(SPar.mass_squared, SPar.lambda, SPar.g); // Collect actions From ba37d51ee9a6f8ec3021b55db0543002e36441f3 Mon Sep 17 00:00:00 2001 From: Guido Cossu Date: Wed, 2 May 2018 15:32:06 +0100 Subject: [PATCH 11/15] Debugging the RNG IO --- lib/parallelIO/BinaryIO.h | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/lib/parallelIO/BinaryIO.h b/lib/parallelIO/BinaryIO.h index ce84fc81..08b7b9b4 100644 --- a/lib/parallelIO/BinaryIO.h +++ b/lib/parallelIO/BinaryIO.h @@ -263,7 +263,7 @@ PARALLEL_CRITICAL GridBase *grid, std::vector &iodata, std::string file, - uint64_t offset, + uint64_t& offset, const std::string &format, int control, uint32_t &nersc_csum, uint32_t &scidac_csuma, @@ -495,6 +495,7 @@ PARALLEL_CRITICAL exit(1); #endif } + offset = fout.tellp(); fout.close(); } timer.Stop(); @@ -699,7 +700,6 @@ PARALLEL_CRITICAL IOobject(w,grid,iodata,file,offset,format,BINARYIO_WRITE|BINARYIO_LEXICOGRAPHIC, nersc_csum,scidac_csuma,scidac_csumb); - iodata.resize(1); { std::vector tmp(RngStateCount); From 8c658de17997db1f3c7990687fc61932b1d82254 Mon Sep 17 00:00:00 2001 From: Peter Boyle Date: Wed, 2 May 2018 17:52:16 +0100 Subject: [PATCH 12/15] Compressor speed up (a little); streaming stores --- lib/qcd/action/fermion/WilsonCompressor.h | 32 ++++++++++++++--------- 1 file changed, 20 insertions(+), 12 deletions(-) diff --git a/lib/qcd/action/fermion/WilsonCompressor.h b/lib/qcd/action/fermion/WilsonCompressor.h index b47700ac..6ec2ede9 100644 --- a/lib/qcd/action/fermion/WilsonCompressor.h +++ b/lib/qcd/action/fermion/WilsonCompressor.h @@ -69,39 +69,47 @@ class WilsonCompressorTemplate< _HCspinor, _Hspinor, _Spinor, projector, /*****************************************************/ /* Compress includes precision change if mpi data is not same */ /*****************************************************/ - inline void Compress(SiteHalfSpinor *buf,Integer o,const SiteSpinor &in) { - projector::Proj(buf[o],in,mu,dag); + inline void Compress(SiteHalfSpinor * __restrict__ buf,Integer o,const SiteSpinor &in) { + SiteHalfSpinor tmp; + projector::Proj(tmp,in,mu,dag); + vstream(buf[o],tmp); } /*****************************************************/ /* Exchange includes precision change if mpi data is not same */ /*****************************************************/ - inline void Exchange(SiteHalfSpinor *mp, - SiteHalfSpinor *vp0, - SiteHalfSpinor *vp1, + inline void Exchange(SiteHalfSpinor * __restrict__ mp, + const SiteHalfSpinor * __restrict__ vp0, + const SiteHalfSpinor * __restrict__ vp1, Integer type,Integer o){ - exchange(mp[2*o],mp[2*o+1],vp0[o],vp1[o],type); + SiteHalfSpinor tmp1; + SiteHalfSpinor tmp2; + exchange(tmp1,tmp2,vp0[o],vp1[o],type); + vstream(mp[2*o ],tmp1); + vstream(mp[2*o+1],tmp2); } /*****************************************************/ /* Have a decompression step if mpi data is not same */ /*****************************************************/ - inline void Decompress(SiteHalfSpinor *out, - SiteHalfSpinor *in, Integer o) { + inline void Decompress(SiteHalfSpinor * __restrict__ out, + SiteHalfSpinor * __restrict__ in, Integer o) { assert(0); } /*****************************************************/ /* Compress Exchange */ /*****************************************************/ - inline void CompressExchange(SiteHalfSpinor *out0, - SiteHalfSpinor *out1, - const SiteSpinor *in, + inline void CompressExchange(SiteHalfSpinor * __restrict__ out0, + SiteHalfSpinor * __restrict__ out1, + const SiteSpinor * __restrict__ in, Integer j,Integer k, Integer m,Integer type){ SiteHalfSpinor temp1, temp2,temp3,temp4; projector::Proj(temp1,in[k],mu,dag); projector::Proj(temp2,in[m],mu,dag); - exchange(out0[j],out1[j],temp1,temp2,type); + exchange(temp3,temp4,temp1,temp2,type); + vstream(out0[j],temp3); + vstream(out1[j],temp4); } /*****************************************************/ From bfbf2f1fa046f95c1813c65e627dc16d5263cd6b Mon Sep 17 00:00:00 2001 From: Antonin Portelli Date: Thu, 3 May 2018 16:20:01 +0100 Subject: [PATCH 13/15] no threaded stencil benchmark if OpenMP is not supported --- benchmarks/Benchmark_comms.cc | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/benchmarks/Benchmark_comms.cc b/benchmarks/Benchmark_comms.cc index 304a09fc..6d95bbe2 100644 --- a/benchmarks/Benchmark_comms.cc +++ b/benchmarks/Benchmark_comms.cc @@ -450,7 +450,7 @@ int main (int argc, char ** argv) } - +#ifdef GRID_OMP std::cout< Date: Fri, 4 May 2018 14:13:35 +0100 Subject: [PATCH 14/15] Update --- tests/Test_compressed_lanczos_hot_start.cc | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/tests/Test_compressed_lanczos_hot_start.cc b/tests/Test_compressed_lanczos_hot_start.cc index 3276d0f8..293506c2 100644 --- a/tests/Test_compressed_lanczos_hot_start.cc +++ b/tests/Test_compressed_lanczos_hot_start.cc @@ -167,7 +167,7 @@ int main (int argc, char ** argv) { RealD mass = Params.mass; RealD M5 = Params.M5; std::vector blockSize = Params.blockSize; - std::vector latt({16,16,16,16}); + std::vector latt({32,32,16,16}); uint64_t vol = Ls*latt[0]*latt[1]*latt[2]*latt[3]; double mat_flop= 2.0*1320.0*vol; // Grids From 4ad0df6fde7b888cfb25e68af0a8b68aada0831f Mon Sep 17 00:00:00 2001 From: paboyle Date: Fri, 4 May 2018 17:33:23 +0100 Subject: [PATCH 15/15] Bump volume for Gerardo --- tests/Test_compressed_lanczos_hot_start.cc | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/tests/Test_compressed_lanczos_hot_start.cc b/tests/Test_compressed_lanczos_hot_start.cc index 293506c2..65f69c54 100644 --- a/tests/Test_compressed_lanczos_hot_start.cc +++ b/tests/Test_compressed_lanczos_hot_start.cc @@ -167,7 +167,7 @@ int main (int argc, char ** argv) { RealD mass = Params.mass; RealD M5 = Params.M5; std::vector blockSize = Params.blockSize; - std::vector latt({32,32,16,16}); + std::vector latt({32,32,32,32}); uint64_t vol = Ls*latt[0]*latt[1]*latt[2]*latt[3]; double mat_flop= 2.0*1320.0*vol; // Grids