changed made after merging with lastest dev version

3 files changed, 21 insertions, 20 deletions

diff --git a/Source/FieldSolver/SpectralSolver/SpectralAlgorithms/PsatdAlgorithm.H b/Source/FieldSolver/SpectralSolver/SpectralAlgorithms/PsatdAlgorithm.H
index 34743525e..36d5782e8 100644
--- a/Source/FieldSolver/SpectralSolver/SpectralAlgorithms/PsatdAlgorithm.H
+++ b/Source/FieldSolver/SpectralSolver/SpectralAlgorithms/PsatdAlgorithm.H
@@ -1,8 +1,8 @@
 #ifndef WARPX_PSATD_ALGORITHM_H_
 #define WARPX_PSATD_ALGORITHM_H_
 
-#include <SpectralKSpace.H>
-#include <SpectralFieldData.H>
+//#include <SpectralKSpace.H>
+//#include <SpectralFieldData.H>
 #include <SpectralBaseAlgorithm.H>
 
 /* \brief Class that updates the field in spectral space
@@ -17,19 +17,19 @@ class PsatdAlgorithm : public SpectralBaseAlgorithm
                          const amrex::DistributionMapping& dm,
                          const int norder_x, const int norder_y,
                          const int norder_z, const bool nodal, const amrex::Real dt);
-        PsatdAlgorithm() = default; // Default constructor
-        PsatdAlgorithm& operator=(PsatdAlgorithm&& algorithm) = default;
-        void pushSpectralFields(SpectralFieldData& f) const;
+        //PsatdAlgorithm() = default; // Default constructor
+        //PsatdAlgorithm& operator=(PsatdAlgorithm&& algorithm) = default;
         void InitializeCoefficience(const SpectralKSpace& spectral_kspace,
                                     const amrex::DistributionMapping& dm, 
                                     const amrex::Real dt);
+        void pushSpectralFields(SpectralFieldData& f) const override final;
 
     private:
         // Modified finite-order vectors
-        KVectorComponent modified_kx_vec, modified_kz_vec;
-#if (AMREX_SPACEDIM==3)
-        KVectorComponent modified_ky_vec;
-#endif
+//        KVectorComponent modified_kx_vec, modified_kz_vec;
+//#if (AMREX_SPACEDIM==3)
+//        KVectorComponent modified_ky_vec;
+//#endif
         SpectralCoefficients C_coef, S_ck_coef, X1_coef, X2_coef, X3_coef;
 };
 
diff --git a/Source/FieldSolver/SpectralSolver/SpectralAlgorithms/PsatdAlgorithm.cpp b/Source/FieldSolver/SpectralSolver/SpectralAlgorithms/PsatdAlgorithm.cpp
index 8dd2a830f..3da0ef453 100644
--- a/Source/FieldSolver/SpectralSolver/SpectralAlgorithms/PsatdAlgorithm.cpp
+++ b/Source/FieldSolver/SpectralSolver/SpectralAlgorithms/PsatdAlgorithm.cpp
@@ -23,6 +23,7 @@ PsatdAlgorithm::PsatdAlgorithm(const SpectralKSpace& spectral_kspace,
     X3_coef = SpectralCoefficients(ba, dm, 1, 0);
 
     InitializeCoefficience(spectral_kspace, dm, dt);
+}
 //    // Fill them with the right values:
 //    // Loop over boxes and allocate the corresponding coefficients
 //    // for each box owned by the local MPI proc
@@ -76,7 +77,6 @@ PsatdAlgorithm::PsatdAlgorithm(const SpectralKSpace& spectral_kspace,
 //            }
 //        });
 //    }
-};
 
 /* Advance the E and B field in spectral space (stored in `f`)
  * over one time step */
@@ -173,8 +173,7 @@ void PsatdAlgorithm::InitializeCoefficience(const SpectralKSpace& spectral_kspac
     // for each box owned by the local MPI proc
     for (MFIter mfi(ba, dm); mfi.isValid(); ++mfi){
 
-        //const Box& bx = ba[mfi];
-        const Box bx = ba[mfi];
+        const Box& bx = ba[mfi];
 
         // Extract pointers for the k vectors
         const Real* modified_kx = modified_kx_vec[mfi].dataPtr();
diff --git a/Source/FieldSolver/SpectralSolver/SpectralFieldData.cpp b/Source/FieldSolver/SpectralSolver/SpectralFieldData.cpp
index 5998bdd2b..ca9e87f0f 100644
--- a/Source/FieldSolver/SpectralSolver/SpectralFieldData.cpp
+++ b/Source/FieldSolver/SpectralSolver/SpectralFieldData.cpp
@@ -152,9 +152,10 @@ SpectralFieldData::ForwardTransform( const MultiFab& mf,
             [=] AMREX_GPU_DEVICE(int i, int j, int k) noexcept {
                 tmp_arr(i,j,k) = mf_arr(i,j,k,i_comp);                
             });
-//#ifdef AMREX_USE_GPU
-//            cudaDeviceSynchronize();
-//#endif
+#ifdef AMREX_USE_GPU
+            cudaDeviceSynchronize();
+#endif
+            amrex::Print() << " before forward transform " << mf_arr(15,15,15,0) << " " << mf_arr(15,15,15,1);
         }
 
         // Perform Fourier transform from `tmpRealField` to `tmpSpectralField`
@@ -165,7 +166,7 @@ SpectralFieldData::ForwardTransform( const MultiFab& mf,
         cufftResult result;
         cudaStream_t stream = amrex::Cuda::Device::cudaStream(); 
         amrex::Print() << " stream is " << stream << "\n";
-        cufftSetStream ( forward_plan[mfi], stream);
+//        cufftSetStream ( forward_plan[mfi], stream);
         result = cufftExecD2Z( forward_plan[mfi], 
                                tmpRealField[mfi].dataPtr(), 
                                reinterpret_cast<cuDoubleComplex*>(
@@ -268,7 +269,7 @@ SpectralFieldData::BackwardTransform( MultiFab& mf,
         cufftResult result;
         cudaStream_t stream = amrex::Cuda::Device::cudaStream(); 
         amrex::Print() << " stream is " << stream << "\n";
-        cufftSetStream ( backward_plan[mfi], stream);
+//        cufftSetStream ( backward_plan[mfi], stream);
         result = cufftExecZ2D( backward_plan[mfi], 
                                reinterpret_cast<cuDoubleComplex*>(
                                tmpSpectralField[mfi].dataPtr()),
@@ -298,9 +299,10 @@ SpectralFieldData::BackwardTransform( MultiFab& mf,
                 // Copy and normalize field
                 mf_arr(i,j,k,i_comp) = inv_N*tmp_arr(i,j,k);
             });
-//#ifdef AMREX_USE_GPU
-//            cudaDeviceSynchronize();
-//#endif
+#ifdef AMREX_USE_GPU
+            cudaDeviceSynchronize();
+#endif
+            amrex::Print() << " mf_arr after BT " << mf_arr(15,15,15,0) << " " << mf_arr(15,15,15,1) << "\n";
 
 #ifdef AMREX_USE_GPU
             cudaDeviceSynchronize();