ERF
Energy Research and Forecasting: An Atmospheric Modeling Code
•All Classes Namespaces Files Functions Variables Typedefs Enumerations Enumerator Macros Pages
Public Member Functions | Private Attributes | List of all members
ERFPhysBCFunct_v Class Reference

#include <ERF_PhysBCFunct.H>

Collaboration diagram for ERFPhysBCFunct_v:
Collaboration graph
[legend]

Public Member Functions

 ERFPhysBCFunct_v (const int lev, const amrex::Geometry &geom, const amrex::Vector< amrex::BCRec > &domain_bcs_type, const amrex::Gpu::DeviceVector< amrex::BCRec > &domain_bcs_type_d, amrex::Array< amrex::Array< amrex::Real, AMREX_SPACEDIM *2 >, AMREX_SPACEDIM+NBCVAR_max > bc_extdir_vals, amrex::Array< amrex::Array< amrex::Real, AMREX_SPACEDIM *2 >, AMREX_SPACEDIM+NBCVAR_max > bc_neumann_vals, std::unique_ptr< amrex::MultiFab > &z_phys_nd, const bool use_real_bcs, amrex::Real *v_bc_data)
 
 ~ERFPhysBCFunct_v ()
 
void operator() (amrex::MultiFab &mf, amrex::MultiFab &xvel, amrex::MultiFab &yvel, amrex::IntVect const &nghost, const amrex::Real time, int bccomp, bool do_fb)
 
void impose_lateral_yvel_bcs (const amrex::Array4< amrex::Real > &dest_arr, const amrex::Array4< amrex::Real const > &xvel_arr, const amrex::Array4< amrex::Real const > &yvel_arr, const amrex::Box &bx, const amrex::Box &domain, int bccomp)
 
void impose_vertical_yvel_bcs (const amrex::Array4< amrex::Real > &dest_arr, const amrex::Box &bx, const amrex::Box &domain, const amrex::Array4< amrex::Real const > &z_nd, const amrex::GpuArray< amrex::Real, AMREX_SPACEDIM > dxInv, int bccomp)
 

Private Attributes

int m_lev
 
amrex::Geometry m_geom
 
amrex::Vector< amrex::BCRec > m_domain_bcs_type
 
amrex::Gpu::DeviceVector< amrex::BCRec > m_domain_bcs_type_d
 
amrex::Array< amrex::Array< amrex::Real, AMREX_SPACEDIM *2 >, AMREX_SPACEDIM+NBCVAR_max > m_bc_extdir_vals
 
amrex::Array< amrex::Array< amrex::Real, AMREX_SPACEDIM *2 >, AMREX_SPACEDIM+NBCVAR_max > m_bc_neumann_vals
 
amrex::MultiFab * m_z_phys_nd
 
bool m_use_real_bcs
 
amrex::Real * m_v_bc_data
 

Constructor & Destructor Documentation

◆ ERFPhysBCFunct_v()

ERFPhysBCFunct_v::ERFPhysBCFunct_v ( const int  lev,
const amrex::Geometry &  geom,
const amrex::Vector< amrex::BCRec > &  domain_bcs_type,
const amrex::Gpu::DeviceVector< amrex::BCRec > &  domain_bcs_type_d,
amrex::Array< amrex::Array< amrex::Real, AMREX_SPACEDIM *2 >, AMREX_SPACEDIM+NBCVAR_max >  bc_extdir_vals,
amrex::Array< amrex::Array< amrex::Real, AMREX_SPACEDIM *2 >, AMREX_SPACEDIM+NBCVAR_max >  bc_neumann_vals,
std::unique_ptr< amrex::MultiFab > &  z_phys_nd,
const bool  use_real_bcs,
amrex::Real *  v_bc_data 
)
inline
152  : m_lev(lev),
153  m_geom(geom), m_domain_bcs_type(domain_bcs_type),
154  m_domain_bcs_type_d(domain_bcs_type_d),
155  m_bc_extdir_vals(bc_extdir_vals),
156  m_bc_neumann_vals(bc_neumann_vals),
157  m_z_phys_nd(z_phys_nd.get()),
158  m_use_real_bcs(use_real_bcs),
159  m_v_bc_data(v_bc_data)
160  { }
ERFPhysBCFunct_v::m_geom
amrex::Geometry m_geom
Definition: ERF_PhysBCFunct.H:190
ERFPhysBCFunct_v::m_domain_bcs_type_d
amrex::Gpu::DeviceVector< amrex::BCRec > m_domain_bcs_type_d
Definition: ERF_PhysBCFunct.H:192
ERFPhysBCFunct_v::m_v_bc_data
amrex::Real * m_v_bc_data
Definition: ERF_PhysBCFunct.H:197
ERFPhysBCFunct_v::m_lev
int m_lev
Definition: ERF_PhysBCFunct.H:189
ERFPhysBCFunct_v::m_bc_neumann_vals
amrex::Array< amrex::Array< amrex::Real, AMREX_SPACEDIM *2 >, AMREX_SPACEDIM+NBCVAR_max > m_bc_neumann_vals
Definition: ERF_PhysBCFunct.H:194
ERFPhysBCFunct_v::m_use_real_bcs
bool m_use_real_bcs
Definition: ERF_PhysBCFunct.H:196
ERFPhysBCFunct_v::m_domain_bcs_type
amrex::Vector< amrex::BCRec > m_domain_bcs_type
Definition: ERF_PhysBCFunct.H:191
ERFPhysBCFunct_v::m_bc_extdir_vals
amrex::Array< amrex::Array< amrex::Real, AMREX_SPACEDIM *2 >, AMREX_SPACEDIM+NBCVAR_max > m_bc_extdir_vals
Definition: ERF_PhysBCFunct.H:193
ERFPhysBCFunct_v::m_z_phys_nd
amrex::MultiFab * m_z_phys_nd
Definition: ERF_PhysBCFunct.H:195

◆ ~ERFPhysBCFunct_v()

ERFPhysBCFunct_v::~ERFPhysBCFunct_v ( )
inline
162 {}

Member Function Documentation

◆ impose_lateral_yvel_bcs()

void ERFPhysBCFunct_v::impose_lateral_yvel_bcs ( const amrex::Array4< amrex::Real > &  dest_arr,
const amrex::Array4< amrex::Real const > &  xvel_arr,
const amrex::Array4< amrex::Real const > &  yvel_arr,
const amrex::Box &  bx,
const amrex::Box &  domain,
int  bccomp 
)
19 {
20  BL_PROFILE_VAR("impose_lateral_yvel_bcs()",impose_lateral_yvel_bcs);
21  const auto& dom_lo = lbound(domain);
22  const auto& dom_hi = ubound(domain);
23 
24  // Based on BCRec for the domain, we need to make BCRec for this Box
25  // bccomp is used as starting index for m_domain_bcs_type
26  // 0 is used as starting index for bcrs
27  Vector<BCRec> bcrs(1);
28  setBC(enclosedCells(bx), domain, bccomp, 0, 1, m_domain_bcs_type, bcrs);
29 
30  // xlo: ori = 0
31  // ylo: ori = 1
32  // zlo: ori = 2
33  // xhi: ori = 3
34  // yhi: ori = 4
35  // zhi: ori = 5
36 
37  Gpu::DeviceVector<BCRec> bcrs_d(1);
38  Gpu::copyAsync(Gpu::hostToDevice, bcrs.begin(), bcrs.end(), bcrs_d.begin());
39  const BCRec* bc_ptr = bcrs_d.data();
40 
41  GpuArray<GpuArray<Real, AMREX_SPACEDIM*2>, 1> l_bc_extdir_vals_d;
42 
43  for (int ori = 0; ori < 2*AMREX_SPACEDIM; ori++) {
44  l_bc_extdir_vals_d[0][ori] = m_bc_extdir_vals[bccomp][ori];
45  }
46 
47  GeometryData const& geomdata = m_geom.data();
48  bool is_periodic_in_x = geomdata.isPeriodic(0);
49  bool is_periodic_in_y = geomdata.isPeriodic(1);
50 
51  // First do all ext_dir bcs
52  if (!is_periodic_in_x)
53  {
54  // Populate ghost cells on lo-x and hi-x domain boundaries
55  Real* yvel_bc_ptr = m_v_bc_data;
56  Box bx_xlo(bx); bx_xlo.setBig (0,dom_lo.x-1);
57  Box bx_xhi(bx); bx_xhi.setSmall(0,dom_hi.x+1);
58  ParallelFor(bx_xlo, bx_xhi,
59  [=] AMREX_GPU_DEVICE (int i, int j, int k) {
60  int iflip = dom_lo.x - 1- i;
61  if (bc_ptr[0].lo(0) == ERFBCType::ext_dir) {
62  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][0];
63  } else if (bc_ptr[0].lo(0) == ERFBCType::ext_dir_upwind && xvel_arr(dom_lo.x,i,j) >= 0.) {
64  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][0];
65  } else if (bc_ptr[0].lo(0) == ERFBCType::foextrap) {
66  dest_arr(i,j,k) = dest_arr(dom_lo.x,j,k);
67  } else if (bc_ptr[0].lo(0) == ERFBCType::open) {
68  dest_arr(i,j,k) = dest_arr(dom_lo.x,j,k);
69  } else if (bc_ptr[0].lo(0) == ERFBCType::reflect_even) {
70  dest_arr(i,j,k) = dest_arr(iflip,j,k);
71  } else if (bc_ptr[0].lo(0) == ERFBCType::reflect_odd) {
72  dest_arr(i,j,k) = -dest_arr(iflip,j,k);
73  }
74  },
75  [=] AMREX_GPU_DEVICE (int i, int j, int k) {
76  int iflip = 2*dom_hi.x + 1 - i;
77  if (bc_ptr[0].hi(0) == ERFBCType::ext_dir) {
78  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][3];
79  } else if (bc_ptr[0].hi(0) == ERFBCType::ext_dir_upwind && xvel_arr(dom_hi.x+1,i,j) <= 0.) {
80  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][3];
81  } else if (bc_ptr[0].hi(0) == ERFBCType::foextrap) {
82  dest_arr(i,j,k) = dest_arr(dom_hi.x,j,k);
83  } else if (bc_ptr[0].hi(0) == ERFBCType::open) {
84  dest_arr(i,j,k) = dest_arr(dom_hi.x,j,k);
85  } else if (bc_ptr[0].hi(0) == ERFBCType::reflect_even) {
86  dest_arr(i,j,k) = dest_arr(iflip,j,k);
87  } else if (bc_ptr[0].hi(0) == ERFBCType::reflect_odd) {
88  dest_arr(i,j,k) = -dest_arr(iflip,j,k);
89  }
90  }
91  );
92  }
93 
94  if (!is_periodic_in_y)
95  {
96  // Populate ghost cells on lo-y and hi-y domain boundaries
97  Real* yvel_bc_ptr = m_v_bc_data;
98  Box bx_ylo(bx); bx_ylo.setBig (1,dom_lo.y-1);
99  Box bx_yhi(bx); bx_yhi.setSmall(1,dom_hi.y+2);
100  Box bx_ylo_face(bx); bx_ylo_face.setSmall(1,dom_lo.y ); bx_ylo_face.setBig(1,dom_lo.y );
101  Box bx_yhi_face(bx); bx_yhi_face.setSmall(1,dom_hi.y+1); bx_yhi_face.setBig(1,dom_hi.y+1);
102  ParallelFor(bx_ylo, bx_ylo_face,
103  [=] AMREX_GPU_DEVICE (int i, int j, int k)
104  {
105  int jflip = dom_lo.y-j;
106  if (bc_ptr[0].lo(1) == ERFBCType::ext_dir) {
107  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][1];
108  } else if (bc_ptr[0].lo(1) == ERFBCType::ext_dir_upwind && yvel_arr(i,dom_lo.y,k) >= 0.) {
109  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][1];
110  } else if (bc_ptr[0].lo(1) == ERFBCType::foextrap) {
111  dest_arr(i,j,k) = dest_arr(i,dom_lo.y,k);
112  } else if (bc_ptr[0].lo(1) == ERFBCType::open) {
113  dest_arr(i,j,k) = dest_arr(i,dom_lo.y,k);
114  } else if (bc_ptr[0].lo(1) == ERFBCType::reflect_even) {
115  dest_arr(i,j,k) = dest_arr(i,jflip,k);
116  } else if (bc_ptr[0].lo(1) == ERFBCType::reflect_odd) {
117  dest_arr(i,j,k) = -dest_arr(i,jflip,k);
118  } else if (bc_ptr[0].lo(1) == ERFBCType::neumann_int) {
119  dest_arr(i,j,k) = (4.0*dest_arr(i,dom_lo.y+1,k) - dest_arr(i,dom_lo.y+2,k))/3.0;
120  }
121  },
122  // We only set the values on the domain faces themselves if EXT_DIR or neumann_int
123  [=] AMREX_GPU_DEVICE (int i, int j, int k)
124  {
125  if (bc_ptr[0].lo(1) == ERFBCType::ext_dir) {
126  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][1];
127  } else if (bc_ptr[0].lo(1) == ERFBCType::ext_dir_upwind && yvel_arr(i,dom_lo.y,k) >= 0.) {
128  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][1];
129  } else if (bc_ptr[0].lo(1) == ERFBCType::neumann_int) {
130  dest_arr(i,j,k) = (4.0*dest_arr(i,dom_lo.y+1,k) - dest_arr(i,dom_lo.y+2,k))/3.0;
131  }
132  }
133  );
134  ParallelFor(bx_yhi, bx_yhi_face,
135  [=] AMREX_GPU_DEVICE (int i, int j, int k)
136  {
137  int jflip = 2*(dom_hi.y + 1) - j;
138  if (bc_ptr[0].hi(1) == ERFBCType::ext_dir) {
139  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][4];
140  } else if (bc_ptr[0].hi(1) == ERFBCType::ext_dir_upwind && yvel_arr(i,dom_hi.y+1,k) <= 0.) {
141  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][4];
142  } else if (bc_ptr[0].hi(1) == ERFBCType::foextrap) {
143  dest_arr(i,j,k) = dest_arr(i,dom_hi.y+1,k);
144  } else if (bc_ptr[0].hi(1) == ERFBCType::open) {
145  dest_arr(i,j,k) = dest_arr(i,dom_hi.y+1,k);
146  } else if (bc_ptr[0].hi(1) == ERFBCType::reflect_even) {
147  dest_arr(i,j,k) = dest_arr(i,jflip,k);
148  } else if (bc_ptr[0].hi(1) == ERFBCType::reflect_odd) {
149  dest_arr(i,j,k) = -dest_arr(i,jflip,k);
150  } else if (bc_ptr[0].hi(1) == ERFBCType::neumann_int) {
151  dest_arr(i,j,k) = (4.0*dest_arr(i,dom_hi.y,k) - dest_arr(i,dom_hi.y-1,k))/3.0;
152  }
153  },
154  // We only set the values on the domain faces themselves if EXT_DIR or neumann_int
155  [=] AMREX_GPU_DEVICE (int i, int j, int k)
156  {
157  if (bc_ptr[0].hi(1) == ERFBCType::ext_dir) {
158  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][4];
159  } else if (bc_ptr[0].hi(1) == ERFBCType::ext_dir_upwind && yvel_arr(i,dom_hi.y+1,k) <= 0.) {
160  dest_arr(i,j,k) = (yvel_bc_ptr) ? yvel_bc_ptr[k] : l_bc_extdir_vals_d[0][4];
161  } else if (bc_ptr[0].hi(1) == ERFBCType::neumann_int) {
162  dest_arr(i,j,k) = (4.0*dest_arr(i,dom_hi.y,k) - dest_arr(i,dom_hi.y-1,k))/3.0;
163  }
164  }
165  );
166  }
167  Gpu::streamSynchronize();
168 }
ERFPhysBCFunct_v::impose_lateral_yvel_bcs
void impose_lateral_yvel_bcs(const amrex::Array4< amrex::Real > &dest_arr, const amrex::Array4< amrex::Real const > &xvel_arr, const amrex::Array4< amrex::Real const > &yvel_arr, const amrex::Box &bx, const amrex::Box &domain, int bccomp)
Definition: ERF_BoundaryConditionsYvel.cpp:14
ERFBCType::open
@ open
Definition: ERF_IndexDefines.H:197
ERFBCType::reflect_odd
@ reflect_odd
Definition: ERF_IndexDefines.H:187
ERFBCType::foextrap
@ foextrap
Definition: ERF_IndexDefines.H:190
ERFBCType::ext_dir
@ ext_dir
Definition: ERF_IndexDefines.H:191
ERFBCType::ext_dir_upwind
@ ext_dir_upwind
Definition: ERF_IndexDefines.H:199
ERFBCType::neumann_int
@ neumann_int
Definition: ERF_IndexDefines.H:196
ERFBCType::reflect_even
@ reflect_even
Definition: ERF_IndexDefines.H:189

◆ impose_vertical_yvel_bcs()

void ERFPhysBCFunct_v::impose_vertical_yvel_bcs ( const amrex::Array4< amrex::Real > &  dest_arr,
const amrex::Box &  bx,
const amrex::Box &  domain,
const amrex::Array4< amrex::Real const > &  z_nd,
const amrex::GpuArray< amrex::Real, AMREX_SPACEDIM >  dxInv,
int  bccomp 
)
186 {
187  BL_PROFILE_VAR("impose_vertical_yvel_bcs()",impose_vertical_yvel_bcs);
188  const auto& dom_lo = lbound(domain);
189  const auto& dom_hi = ubound(domain);
190 
191  Box per_grown_domain(domain);
192  int growx = (m_geom.isPeriodic(0)) ? 1 : 0;
193  int growy = (m_geom.isPeriodic(1)) ? 1 : 0;
194  per_grown_domain.grow(IntVect(growx,growy,0));
195  const auto& perdom_lo = lbound(per_grown_domain);
196  const auto& perdom_hi = ubound(per_grown_domain);
197 
198  // Based on BCRec for the domain, we need to make BCRec for this Box
199  // bccomp is used as starting index for m_domain_bcs_type
200  // 0 is used as starting index for bcrs
201  Vector<BCRec> bcrs(1);
202  setBC(enclosedCells(bx), domain, bccomp, 0, 1, m_domain_bcs_type, bcrs);
203 
204  // xlo: ori = 0
205  // ylo: ori = 1
206  // zlo: ori = 2
207  // xhi: ori = 3
208  // yhi: ori = 4
209  // zhi: ori = 5
210 
211  Gpu::DeviceVector<BCRec> bcrs_d(1);
212  Gpu::copyAsync(Gpu::hostToDevice, bcrs.begin(), bcrs.end(), bcrs_d.begin());
213  const BCRec* bc_ptr = bcrs_d.data();
214 
215  GpuArray<GpuArray<Real, AMREX_SPACEDIM*2>, 1> l_bc_extdir_vals_d;
216 
217  for (int ori = 0; ori < 2*AMREX_SPACEDIM; ori++) {
218  l_bc_extdir_vals_d[0][ori] = m_bc_extdir_vals[bccomp][ori];
219  }
220 
221  GeometryData const& geomdata = m_geom.data();
222 
223  {
224  // Populate ghost cells on lo-z and hi-z domain boundaries
225  Box bx_zlo(bx); bx_zlo.setBig (2,dom_lo.z-1);
226  Box bx_zhi(bx); bx_zhi.setSmall(2,dom_hi.z+1);
227  ParallelFor(bx_zlo, bx_zhi,
228  [=] AMREX_GPU_DEVICE (int i, int j, int k) {
229  int kflip = dom_lo.z - 1 - k;
230  if (bc_ptr[0].lo(2) == ERFBCType::ext_dir) {
231  dest_arr(i,j,k) = l_bc_extdir_vals_d[0][2];
232  } else if (bc_ptr[0].lo(2) == ERFBCType::foextrap) {
233  dest_arr(i,j,k) = dest_arr(i,j,dom_lo.z);
234  } else if (bc_ptr[0].lo(2) == ERFBCType::open) {
235  dest_arr(i,j,k) = dest_arr(i,j,dom_lo.z);
236  } else if (bc_ptr[0].lo(2) == ERFBCType::reflect_even) {
237  dest_arr(i,j,k) = dest_arr(i,j,kflip);
238  } else if (bc_ptr[0].lo(2) == ERFBCType::reflect_odd) {
239  dest_arr(i,j,k) = -dest_arr(i,j,kflip);
240  }
241  },
242  [=] AMREX_GPU_DEVICE (int i, int j, int k) {
243  int kflip = 2*dom_hi.z + 1 - k;
244  if (bc_ptr[0].hi(2) == ERFBCType::ext_dir) {
245  dest_arr(i,j,k) = l_bc_extdir_vals_d[0][5];
246  } else if (bc_ptr[0].hi(2) == ERFBCType::foextrap) {
247  dest_arr(i,j,k) = dest_arr(i,j,dom_hi.z);
248  } else if (bc_ptr[0].hi(2) == ERFBCType::open) {
249  dest_arr(i,j,k) = dest_arr(i,j,dom_hi.z);
250  } else if (bc_ptr[0].hi(2) == ERFBCType::reflect_even) {
251  dest_arr(i,j,k) = dest_arr(i,j,kflip);
252  } else if (bc_ptr[0].hi(2) == ERFBCType::reflect_odd) {
253  dest_arr(i,j,k) = -dest_arr(i,j,kflip);
254  }
255  }
256  );
257  }
258 
259  if (m_z_phys_nd) {
260 
261  const auto& bx_lo = lbound(bx);
262  const auto& bx_hi = ubound(bx);
263 
264  const auto& zphys_lo = lbound(Box(z_phys_nd));
265  const auto& zphys_hi = ubound(Box(z_phys_nd));
266 
267  // Neumann conditions (d<var>/dn = 0) must be aware of the surface normal with terrain.
268  // An additional source term arises from d<var>/dx & d<var>/dy & met_h_xi/eta/zeta.
269  //=====================================================================================
270  // Only modify scalars, U, or V
271  // Loop over each component
272  // Hit for Neumann condition at kmin
273  if(bcrs[0].lo(2) == ERFBCType::foextrap) {
274  // Loop over ghost cells in bottom XY-plane (valid box)
275  Box xybx = bx;
276  xybx.setBig(2,-1);
277  xybx.setSmall(2,bx.smallEnd()[2]);
278  int k0 = 0;
279 
280  // Get the dz cell size
281  Real dz = geomdata.CellSize(2);
282 
283  // Fill all the Neumann srcs with terrain
284  ParallelFor(xybx, [=] AMREX_GPU_DEVICE (int i, int j, int k)
285  {
286  // Clip indices for ghost-cells
287  int ii = amrex::min(amrex::max(i,perdom_lo.x),perdom_hi.x);
288  ii = amrex::min(amrex::max(ii,zphys_lo.x),zphys_hi.x);
289  int jj = amrex::min(amrex::max(j,perdom_lo.y),perdom_hi.y);
290  jj = amrex::min(amrex::max(jj,zphys_lo.y),zphys_hi.y);
291 
292  // Get metrics
293  Real met_h_xi = Compute_h_xi_AtJface (ii, jj, k0, dxInv, z_phys_nd);
294  Real met_h_eta = Compute_h_eta_AtJface (ii, jj, k0, dxInv, z_phys_nd);
295  Real met_h_zeta = Compute_h_zeta_AtJface(ii, jj, k0, dxInv, z_phys_nd);
296 
297  // GradX at IJK location inside domain -- this relies on the assumption that we have
298  // used foextrap for cell-centered quantities outside the domain to define the gradient as zero
299  Real GradVarx, GradVary;
300  if ( i < dom_lo.x-1 || i > dom_hi.x+1 || (i+1 > bx_hi.x && i-1 < bx_lo.x) ) {
301  GradVarx = 0.0;
302  } else if (i+1 > bx_hi.x) {
303  GradVarx = dxInv[0] * (dest_arr(i ,j,k0) - dest_arr(i-1,j,k0));
304  } else if (i-1 < bx_lo.x) {
305  GradVarx = dxInv[0] * (dest_arr(i+1,j,k0) - dest_arr(i ,j,k0));
306  } else {
307  GradVarx = 0.5 * dxInv[0] * (dest_arr(i+1,j,k0) - dest_arr(i-1,j,k0));
308  }
309 
310  // GradY at IJK location inside domain -- this relies on the assumption that we have
311  // used foextrap for cell-centered quantities outside the domain to define the gradient as zero
312  if ( j < dom_lo.y-1 || j > dom_hi.y+1 || (j+1 > bx_hi.y && j-1 < bx_lo.y) ) {
313  GradVary = 0.0;
314  } else if (j+1 > bx_hi.y) {
315  GradVary = dxInv[1] * (dest_arr(i,j ,k0) - dest_arr(i,j-1,k0));
316  } else if (j-1 < bx_lo.y) {
317  GradVary = dxInv[1] * (dest_arr(i,j+1,k0) - dest_arr(i,j ,k0));
318  } else {
319  GradVary = 0.5 * dxInv[1] * (dest_arr(i,j+1,k0) - dest_arr(i,j-1,k0));
320  }
321 
322  // Prefactor
323  Real met_fac = met_h_zeta / ( met_h_xi*met_h_xi + met_h_eta*met_h_eta + 1. );
324 
325  // Accumulate in bottom ghost cell (EXTRAP already populated)
326  dest_arr(i,j,k) -= dz * met_fac * ( met_h_xi * GradVarx + met_h_eta * GradVary );
327  });
328  } // foextrap
329  } //m_z_phys_nd
330  Gpu::streamSynchronize();
331 }
Compute_h_xi_AtJface
AMREX_GPU_DEVICE AMREX_FORCE_INLINE amrex::Real Compute_h_xi_AtJface(const int &i, const int &j, const int &k, const amrex::GpuArray< amrex::Real, AMREX_SPACEDIM > &cellSizeInv, const amrex::Array4< const amrex::Real > &z_nd)
Definition: ERF_TerrainMetrics.H:151
Compute_h_zeta_AtJface
AMREX_GPU_DEVICE AMREX_FORCE_INLINE amrex::Real Compute_h_zeta_AtJface(const int &i, const int &j, const int &k, const amrex::GpuArray< amrex::Real, AMREX_SPACEDIM > &cellSizeInv, const amrex::Array4< const amrex::Real > &z_nd)
Definition: ERF_TerrainMetrics.H:137
Compute_h_eta_AtJface
AMREX_GPU_DEVICE AMREX_FORCE_INLINE amrex::Real Compute_h_eta_AtJface(const int &i, const int &j, const int &k, const amrex::GpuArray< amrex::Real, AMREX_SPACEDIM > &cellSizeInv, const amrex::Array4< const amrex::Real > &z_nd)
Definition: ERF_TerrainMetrics.H:165
ERFPhysBCFunct_v::impose_vertical_yvel_bcs
void impose_vertical_yvel_bcs(const amrex::Array4< amrex::Real > &dest_arr, const amrex::Box &bx, const amrex::Box &domain, const amrex::Array4< amrex::Real const > &z_nd, const amrex::GpuArray< amrex::Real, AMREX_SPACEDIM > dxInv, int bccomp)
Definition: ERF_BoundaryConditionsYvel.cpp:181
Here is the call graph for this function:

◆ operator()()

void ERFPhysBCFunct_v::operator() ( amrex::MultiFab &  mf,
amrex::MultiFab &  xvel,
amrex::MultiFab &  yvel,
amrex::IntVect const &  nghost,
const amrex::Real  time,
int  bccomp,
bool  do_fb 
)
166 {
167  BL_PROFILE("ERFPhysBCFunct_v::()");
168 
169  if (m_geom.isAllPeriodic()) return;
170 
171  const auto& domain = m_geom.Domain();
172  const auto dxInv = m_geom.InvCellSizeArray();
173 
174  Box gdomainy = surroundingNodes(domain,1);
175  for (int i = 0; i < AMREX_SPACEDIM; ++i) {
176  if (m_geom.isPeriodic(i)) {
177  gdomainy.grow(i, nghost[i]);
178  }
179  }
180 
181  //
182  // We fill all of the interior and periodic ghost cells first, so we can fill
183  // those directly inside the lateral and vertical calls.
184  //
185  if (do_fb) {
186  mf.FillBoundary(m_geom.periodicity());
187  }
188 
189 #ifdef AMREX_USE_OMP
190 #pragma omp parallel if (Gpu::notInLaunchRegion())
191 #endif
192  {
193  for (MFIter mfi(mf,false); mfi.isValid(); ++mfi)
194  {
195  //
196  // This is the box we pass to the different routines
197  // NOTE -- this is the full grid NOT the tile box
198  //
199  Box bx = mfi.validbox();
200 
201  //
202  // These are the boxes we use to test on relative to the domain
203  //
204  Box ybx1 = surroundingNodes(bx,1); ybx1.grow(nghost);
205  if (ybx1.smallEnd(2) < domain.smallEnd(2)) ybx1.setSmall(2,domain.smallEnd(2));
206  if (ybx1.bigEnd(2) > domain.bigEnd(2)) ybx1.setBig(2,domain.bigEnd(2));
207 
208  Box ybx2 = surroundingNodes(bx,1); ybx2.grow(nghost);
209 
210  Array4<const Real> z_nd_arr;
211 
212  if (m_z_phys_nd)
213  {
214  z_nd_arr = m_z_phys_nd->const_array(mfi);
215  }
216 
217  if (!gdomainy.contains(ybx2))
218  {
219  Array4< Real> const& dest_arr = mf.array(mfi);
220  Array4<const Real> const& velx_arr = xvel.const_array(mfi);
221  Array4<const Real> const& vely_arr = yvel.const_array(mfi);
222 
223  if (!m_use_real_bcs)
224  {
225  impose_lateral_yvel_bcs(dest_arr,velx_arr,vely_arr,ybx1,domain,bccomp);
226  }
227 
228  impose_vertical_yvel_bcs(dest_arr,ybx2,domain,z_nd_arr,dxInv,bccomp);
229  }
230 
231  } // MFIter
232  } // OpenMP
233 } // operator()
Vars::xvel
@ xvel
Definition: ERF_IndexDefines.H:141
Vars::yvel
@ yvel
Definition: ERF_IndexDefines.H:142

Member Data Documentation

◆ m_bc_extdir_vals

amrex::Array<amrex::Array<amrex::Real, AMREX_SPACEDIM*2>,AMREX_SPACEDIM+NBCVAR_max> ERFPhysBCFunct_v::m_bc_extdir_vals
private

◆ m_bc_neumann_vals

amrex::Array<amrex::Array<amrex::Real, AMREX_SPACEDIM*2>,AMREX_SPACEDIM+NBCVAR_max> ERFPhysBCFunct_v::m_bc_neumann_vals
private

◆ m_domain_bcs_type

amrex::Vector<amrex::BCRec> ERFPhysBCFunct_v::m_domain_bcs_type
private

◆ m_domain_bcs_type_d

amrex::Gpu::DeviceVector<amrex::BCRec> ERFPhysBCFunct_v::m_domain_bcs_type_d
private

◆ m_geom

amrex::Geometry ERFPhysBCFunct_v::m_geom
private

◆ m_lev

int ERFPhysBCFunct_v::m_lev
private

◆ m_use_real_bcs

bool ERFPhysBCFunct_v::m_use_real_bcs
private

◆ m_v_bc_data

amrex::Real* ERFPhysBCFunct_v::m_v_bc_data
private

◆ m_z_phys_nd

amrex::MultiFab* ERFPhysBCFunct_v::m_z_phys_nd
private
The documentation for this class was generated from the following files: