ERFPhysBCFunct_w Class Reference

#include <ERF_PhysBCFunct.H>

Collaboration diagram for ERFPhysBCFunct_w:

Public Member Functions
	ERFPhysBCFunct_w (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, const TerrainType &terrain_type, amrex::Vector< std::unique_ptr< amrex::MultiFab >> &mapfac_lev, std::unique_ptr< amrex::MultiFab > &z_phys_nd, const bool use_real_bcs, amrex::Real *w_bc_data)
	~ERFPhysBCFunct_w ()
void	operator() (amrex::MultiFab &mf, amrex::MultiFab &xvel, amrex::MultiFab &yvel, amrex::IntVect const &nghost, const double time, int bccomp, bool do_fb)
void	impose_lateral_zvel_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, const amrex::Array4< amrex::Real const > &mf_u, const amrex::Array4< amrex::Real const > &mf_v, const amrex::Array4< amrex::Real const > &z_nd, const amrex::GpuArray< amrex::Real, AMREX_SPACEDIM > dxInv, TerrainType terrain_type, int bccomp, const double time)
void	impose_vertical_zvel_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, const amrex::Array4< amrex::Real const > &mf_u, const amrex::Array4< amrex::Real const > &mf_v, const amrex::Array4< amrex::Real const > &z_nd, const amrex::GpuArray< amrex::Real, AMREX_SPACEDIM > dxInv, int bccomp_u, int bccomp_v, int bccomp_w, TerrainType terrain_type, const double time)

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
TerrainType	m_terrain_type
amrex::MultiFab *	m_mapfac_u
amrex::MultiFab *	m_mapfac_v
amrex::MultiFab *	m_z_phys_nd
bool	m_use_real_bcs
amrex::Real *	m_w_bc_data

Constructor & Destructor Documentation

ERFPhysBCFunct_w()

ERFPhysBCFunct_w::ERFPhysBCFunct_w ( 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, const TerrainType &  terrain_type, amrex::Vector< std::unique_ptr< amrex::MultiFab >> &  mapfac_lev, std::unique_ptr< amrex::MultiFab > &  z_phys_nd, const bool  use_real_bcs, amrex::Real *  w_bc_data  )

inline

Construct the z-velocity physical boundary-condition functor.

Parameters

[in]	lev	level index
[in]	geom	geometry for the level
[in]	domain_bcs_type	host boundary-condition records
[in]	domain_bcs_type_d	device boundary-condition records
[in]	bc_extdir_vals	external Dirichlet values for boundary faces
[in]	bc_neumann_vals	Neumann values for boundary faces
[in]	terrain_type	terrain type for the level
[in]	mapfac_lev	map-factor MultiFabs for the level
[in]	z_phys_nd	nodal physical height MultiFab
[in]	use_real_bcs	whether real-data boundary conditions are active
[in]	w_bc_data	optional z-velocity boundary data

        : m_lev(lev),
          m_geom(geom), m_domain_bcs_type(domain_bcs_type),
          m_domain_bcs_type_d(domain_bcs_type_d),
          m_bc_extdir_vals(bc_extdir_vals),
          m_bc_neumann_vals(bc_neumann_vals),
          m_terrain_type(terrain_type),
          m_z_phys_nd(z_phys_nd.get()),
          m_use_real_bcs(use_real_bcs),
          m_w_bc_data(w_bc_data)
    {
          m_mapfac_u = mapfac_lev[MapFacType::u_x].get();
          m_mapfac_v = mapfac_lev[MapFacType::v_y].get();
    }

~ERFPhysBCFunct_w()

ERFPhysBCFunct_w::~ERFPhysBCFunct_w ( )

inline

Destroy the z-velocity physical boundary-condition functor.

{}

Member Function Documentation

impose_lateral_zvel_bcs()

void ERFPhysBCFunct_w::impose_lateral_zvel_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,
		const amrex::Array4< amrex::Real const > &	mf_u,
		const amrex::Array4< amrex::Real const > &	mf_v,
		const amrex::Array4< amrex::Real const > &	z_nd,
		const amrex::GpuArray< amrex::Real, AMREX_SPACEDIM >	dxInv,
		TerrainType	terrain_type,
		int	bccomp,
		const double	time
	)

Impose lateral boundary conditions on z velocity.

Parameters

[in,out]	dest_arr	z-velocity data to be filled
[in]	xvel_arr	x-velocity used to determine upwind inflow
[in]	yvel_arr	y-velocity used to determine upwind inflow
[in]	bx	box holding data to be filled
[in]	domain	simulation domain
[in]	mf_u	x map-factor field
[in]	mf_v	y map-factor field
[in]	z_nd	nodal physical height
[in]	dxInv	inverse cell size
[in]	terrain_type	terrain representation
[in]	bccomp	boundary-condition component
[in]	time	time at which data should be filled

Impose lateral boundary conditions on z-component of velocity

Parameters

[in,out]	dest_arr	Array4 of the quantity to be filled
[in]	xvel_arr	x-velocity used to determine upwind Dirichlet inflow
[in]	yvel_arr	y-velocity used to determine upwind Dirichlet inflow
[in]	bx	box associated with this data
[in]	domain	computational domain
[in]	mf_u	metric-aware x-velocity used to convert omega to w
[in]	mf_v	metric-aware y-velocity used to convert omega to w
[in]	z_phys_nd	height coordinate at nodes
[in]	dxInv	inverse cell size array
[in]	terrain_type	terrain type used to decide whether omega conversion is needed
[in]	bccomp	index into m_domain_bcs_type
[in]	time	time at which the data should be filled

{
    BL_PROFILE_VAR("impose_lateral_zvel_bcs()",impose_lateral_zvel_bcs);
    const auto& dom_lo = lbound(domain);
    const auto& dom_hi = ubound(domain);
 
    //
    // An ext_dir_upwind face decides inflow against the face velocity, but the box we
    // loop over does not always have a face velocity at every one of its cells: the
    // state carries one more lateral ghost than the velocities do (ngrow_state =
    // ComputeGhostCells()+1 against ngrow_vels = ComputeGhostCells()), and a velocity
    // box is nodal in its own direction where the array being read is cell centered.
    // Either way the outermost ghost cells have no velocity of their own, so sample
    // the nearest one that exists. Cells already inside the velocity array are
    // unaffected; only that outermost layer changes, and it previously read out of
    // bounds.
    //
    const auto& xvel_lo = lbound(xvel_arr);
    const auto& xvel_hi = ubound(xvel_arr);
    const auto& yvel_lo = lbound(yvel_arr);
    const auto& yvel_hi = ubound(yvel_arr);
 
    // Based on BCRec for the domain, we need to make BCRec for this Box
    // bccomp is used as starting index for m_domain_bcs_type
    //      0 is used as starting index for bcrs
    Vector<BCRec> bcrs_w(1);
    setBC(enclosedCells(bx), domain, bccomp, 0, 1, m_domain_bcs_type, bcrs_w);
 
    bool l_use_terrain_fitted_coords = ( (terrain_type == TerrainType::StaticFittedMesh) ||
                                         (terrain_type == TerrainType::MovingFittedMesh) );
 
    // xlo: ori = 0
    // ylo: ori = 1
    // zlo: ori = 2
    // xhi: ori = 3
    // yhi: ori = 4
    // zhi: ori = 5
 
    Gpu::DeviceVector<BCRec> bcrs_w_d(1);
    Gpu::copyAsync(Gpu::hostToDevice, bcrs_w.begin(), bcrs_w.end(), bcrs_w_d.begin());
    const BCRec* bc_ptr_w = bcrs_w_d.data();
 
    GpuArray<GpuArray<Real, AMREX_SPACEDIM*2>,1> l_bc_extdir_vals_d;
 
    for (int ori = 0; ori < 2*AMREX_SPACEDIM; ori++) {
        l_bc_extdir_vals_d[0][ori] = m_bc_extdir_vals[bccomp][ori];
    }
 
    GeometryData const& geomdata = m_geom.data();
    bool is_periodic_in_x = geomdata.isPeriodic(0);
    bool is_periodic_in_y = geomdata.isPeriodic(1);
 
    FArrayBox dhdtfab;
 
    // First do all ext_dir bcs
    if (!is_periodic_in_x)
    {
        Real* zvel_bc_ptr = m_w_bc_data;
        Box bx_xlo(bx);  bx_xlo.setBig  (0,dom_lo.x-1);
        Box bx_xhi(bx);  bx_xhi.setSmall(0,dom_hi.x+1);
        ParallelFor(bx_xlo, bx_xhi,
            [=] AMREX_GPU_DEVICE (int i, int j, int k) {
                int iflip = dom_lo.x - 1 - i;
                if ( (bc_ptr_w[0].lo(0) == ERFBCType::ext_dir) ||
                     (bc_ptr_w[0].lo(0) == ERFBCType::ext_dir_upwind && xvel_arr(dom_lo.x,Clamp(j,xvel_lo.y,xvel_hi.y),Clamp(k,xvel_lo.z,xvel_hi.z)) >= zero) )
                {
                    dest_arr(i,j,k) = (zvel_bc_ptr) ? zvel_bc_ptr[k] : l_bc_extdir_vals_d[0][0];
                    if (l_use_terrain_fitted_coords) {
                        dest_arr(i,j,k) = WFromOmega(i,j,k,dest_arr(i,j,k),
                                                     xvel_arr,yvel_arr,
                                                     mf_u,mf_v,z_phys_nd,dxInv);
                    }
                } else if (bc_ptr_w[0].lo(0) == ERFBCType::foextrap) {
                    dest_arr(i,j,k) =  dest_arr(dom_lo.x,j,k);
                } else if (bc_ptr_w[0].lo(0) == ERFBCType::open) {
                    dest_arr(i,j,k) =  dest_arr(dom_lo.x,j,k);
                } else if (bc_ptr_w[0].lo(0) == ERFBCType::reflect_even) {
                    dest_arr(i,j,k) =  dest_arr(iflip,j,k);
                } else if (bc_ptr_w[0].lo(0) == ERFBCType::reflect_odd) {
                    dest_arr(i,j,k) = -dest_arr(iflip,j,k);
                }
            },
            [=] AMREX_GPU_DEVICE (int i, int j, int k) {
                int iflip = 2*dom_hi.x + 1 - i;
                if ( (bc_ptr_w[0].hi(0) == ERFBCType::ext_dir) ||
                     (bc_ptr_w[0].hi(0) == ERFBCType::ext_dir_upwind && xvel_arr(dom_hi.x+1,Clamp(j,xvel_lo.y,xvel_hi.y),Clamp(k,xvel_lo.z,xvel_hi.z)) <= zero) )
                {
                    dest_arr(i,j,k) = (zvel_bc_ptr) ? zvel_bc_ptr[k] : l_bc_extdir_vals_d[0][3];
                    if (l_use_terrain_fitted_coords) {
                        dest_arr(i,j,k) = WFromOmega(i,j,k,dest_arr(i,j,k),
                                                     xvel_arr,yvel_arr,
                                                     mf_u,mf_v,z_phys_nd,dxInv);
                    }
                } else if (bc_ptr_w[0].hi(0) == ERFBCType::foextrap) {
                    dest_arr(i,j,k) =  dest_arr(dom_hi.x,j,k);
                } else if (bc_ptr_w[0].hi(0) == ERFBCType::open) {
                    dest_arr(i,j,k) =  dest_arr(dom_hi.x,j,k);
                } else if (bc_ptr_w[0].hi(0) == ERFBCType::reflect_even) {
                    dest_arr(i,j,k) =  dest_arr(iflip,j,k);
                } else if (bc_ptr_w[0].hi(0) == ERFBCType::reflect_odd) {
                    dest_arr(i,j,k) = -dest_arr(iflip,j,k);
                }
            }
        );
    }
 
    // First do all ext_dir bcs
    if (!is_periodic_in_y)
    {
        Real* zvel_bc_ptr = m_w_bc_data;
        Box bx_ylo(bx);  bx_ylo.setBig  (1,dom_lo.y-1);
        Box bx_yhi(bx);  bx_yhi.setSmall(1,dom_hi.y+1);
        ParallelFor(bx_ylo, bx_yhi,
            [=] AMREX_GPU_DEVICE (int i, int j, int k) {
                int jflip = dom_lo.y - 1 - j;
                if ( (bc_ptr_w[0].lo(1) == ERFBCType::ext_dir) ||
                     (bc_ptr_w[0].lo(1) == ERFBCType::ext_dir_upwind && yvel_arr(Clamp(i,yvel_lo.x,yvel_hi.x),dom_lo.y,Clamp(k,yvel_lo.z,yvel_hi.z)) >= zero) )
                {
                    dest_arr(i,j,k) = (zvel_bc_ptr) ? zvel_bc_ptr[k] : l_bc_extdir_vals_d[0][1];
                    if (l_use_terrain_fitted_coords) {
                        dest_arr(i,j,k) = WFromOmega(i,j,k,dest_arr(i,j,k),
                                                     xvel_arr,yvel_arr,
                                                     mf_u,mf_v,z_phys_nd,dxInv);
                    }
                } else if (bc_ptr_w[0].lo(1) == ERFBCType::foextrap) {
                    dest_arr(i,j,k) =  dest_arr(i,dom_lo.y,k);
                } else if (bc_ptr_w[0].lo(1) == ERFBCType::open) {
                    dest_arr(i,j,k) =  dest_arr(i,dom_lo.y,k);
                } else if (bc_ptr_w[0].lo(1) == ERFBCType::reflect_even) {
                    dest_arr(i,j,k) =  dest_arr(i,jflip,k);
                } else if (bc_ptr_w[0].lo(1) == ERFBCType::reflect_odd) {
                    dest_arr(i,j,k) = -dest_arr(i,jflip,k);
                }
            },
            [=] AMREX_GPU_DEVICE (int i, int j, int k) {
                int jflip =  2*dom_hi.y + 1 - j;
                if ( (bc_ptr_w[0].hi(1) == ERFBCType::ext_dir) ||
                     (bc_ptr_w[0].hi(1) == ERFBCType::ext_dir_upwind && yvel_arr(Clamp(i,yvel_lo.x,yvel_hi.x),dom_hi.y+1,Clamp(k,yvel_lo.z,yvel_hi.z)) <= zero) )
                {
                    dest_arr(i,j,k) = (zvel_bc_ptr) ? zvel_bc_ptr[k] : l_bc_extdir_vals_d[0][4];
                    if (l_use_terrain_fitted_coords) {
                        dest_arr(i,j,k) = WFromOmega(i,j,k,dest_arr(i,j,k),
                                                     xvel_arr,yvel_arr,
                                                     mf_u,mf_v,z_phys_nd,dxInv);
                    }
                } else if (bc_ptr_w[0].hi(1) == ERFBCType::foextrap) {
                    dest_arr(i,j,k) =  dest_arr(i,dom_hi.y,k);
                } else if (bc_ptr_w[0].hi(1) == ERFBCType::open) {
                    dest_arr(i,j,k) =  dest_arr(i,dom_hi.y,k);
                } else if (bc_ptr_w[0].hi(1) == ERFBCType::reflect_even) {
                    dest_arr(i,j,k) =  dest_arr(i,jflip,k);
                } else if (bc_ptr_w[0].hi(1) == ERFBCType::reflect_odd) {
                    dest_arr(i,j,k) = -dest_arr(i,jflip,k);
                }
            });
    }
    Gpu::streamSynchronize();
}

Here is the call graph for this function:

impose_vertical_zvel_bcs()

void ERFPhysBCFunct_w::impose_vertical_zvel_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,
		const amrex::Array4< amrex::Real const > &	mf_u,
		const amrex::Array4< amrex::Real const > &	mf_v,
		const amrex::Array4< amrex::Real const > &	z_nd,
		const amrex::GpuArray< amrex::Real, AMREX_SPACEDIM >	dxInv,
		int	bccomp_u,
		int	bccomp_v,
		int	bccomp_w,
		TerrainType	terrain_type,
		const double	time
	)

Impose vertical boundary conditions on z velocity.

Parameters

[in,out]	dest_arr	z-velocity data to be filled
[in]	xvel_arr	x-velocity used by terrain-aware vertical BCs
[in]	yvel_arr	y-velocity used by terrain-aware vertical BCs
[in]	bx	box holding data to be filled
[in]	domain	simulation domain
[in]	mf_u	x map-factor field
[in]	mf_v	y map-factor field
[in]	z_nd	nodal physical height
[in]	dxInv	inverse cell size
[in]	bccomp_u	x-velocity boundary-condition component
[in]	bccomp_v	y-velocity boundary-condition component
[in]	bccomp_w	z-velocity boundary-condition component
[in]	terrain_type	terrain representation
[in]	time	time at which data should be filled

Impose vertical boundary conditions on z-component of velocity

Parameters

[in,out]	dest_arr	Array4 of the quantity to be filled
[in]	xvel_arr	x-velocity used to convert omega to w
[in]	yvel_arr	y-velocity used to convert omega to w
[in]	bx	box associated with this data
[in]	domain	computational domain
[in]	mf_u	metric-aware x-velocity used to convert omega to w
[in]	mf_v	metric-aware y-velocity used to convert omega to w
[in]	z_phys_nd	height coordinate at nodes
[in]	dxInv	inverse cell size array
[in]	bccomp_u	index into m_domain_bcs_type corresponding to u
[in]	bccomp_v	index into m_domain_bcs_type corresponding to v
[in]	bccomp_w	index into m_domain_bcs_type corresponding to w
[in]	terrain_type	if MovingFittedMesh then the terrain is moving; otherwise fixed
[in]	time	time at which the data should be filled

{
    BL_PROFILE_VAR("impose_vertical_zvel_bcs()",impose_vertical_zvel_bcs);
    const auto& dom_lo = lbound(domain);
    const auto& dom_hi = ubound(domain);
 
    // xlo: ori = 0
    // ylo: ori = 1
    // zlo: ori = 2
    // xhi: ori = 3
    // yhi: ori = 4
    // zhi: ori = 5
 
    // Based on BCRec for the domain, we need to make BCRec for this Box
    // bccomp is used as starting index for m_domain_bcs_type
    //      0 is used as starting index for bcrs
    Vector<BCRec> bcrs_u(1), bcrs_v(1), bcrs_w(1);
    setBC(enclosedCells(bx), domain, bccomp_u, 0, 1, m_domain_bcs_type, bcrs_u);
    setBC(enclosedCells(bx), domain, bccomp_v, 0, 1, m_domain_bcs_type, bcrs_v);
    setBC(enclosedCells(bx), domain, bccomp_w, 0, 1, m_domain_bcs_type, bcrs_w);
 
    // We use these for the asserts below
    const BCRec* bc_ptr_u_h = bcrs_u.data();
    const BCRec* bc_ptr_v_h = bcrs_v.data();
    const BCRec* bc_ptr_w_h = bcrs_w.data();
 
    bool l_use_terrain_fitted_coords = ( (terrain_type == TerrainType::StaticFittedMesh) ||
                                         (terrain_type == TerrainType::MovingFittedMesh) );
    bool l_moving_terrain = (terrain_type == TerrainType::MovingFittedMesh);
 
    GpuArray<GpuArray<Real, AMREX_SPACEDIM*2>,1> l_bc_extdir_vals_d;
 
    for (int ori = 0; ori < 2*AMREX_SPACEDIM; ori++) {
        l_bc_extdir_vals_d[0][ori] = m_bc_extdir_vals[bccomp_w][ori];
    }
 
    // *******************************************************
    // Bottom boundary
    // *******************************************************
 
    // *******************************************************
    // Moving terrain
    // *******************************************************
    if (l_moving_terrain)
    {
        //************************************************************
        // NOTE: z_t depends on the time interval in which it is
        //       evaluated so we can't arbitrarily define it at a
        //       given time, we must specify an interval
        //************************************************************
 
    // Static terrain
    } else if (l_use_terrain_fitted_coords) {
 
        if (m_lev == 0) {
            AMREX_ALWAYS_ASSERT( (bc_ptr_u_h[0].lo(2) == ERFBCType::ext_dir && bc_ptr_v_h[0].lo(2) == ERFBCType::ext_dir) ||
                                 (bc_ptr_u_h[0].lo(2) != ERFBCType::ext_dir && bc_ptr_v_h[0].lo(2) != ERFBCType::ext_dir) );
        } else {
            // If we do not reach to the top or bottom boundary then the z-vel should be
            //    filled by interpolation from the coarser grid using ERF_FillPatcher.
        }
        if (bx.smallEnd(2) == dom_lo.z) {
            ParallelFor(makeSlab(bx,2,dom_lo.z), [=] AMREX_GPU_DEVICE (int i, int j, int k) {
                dest_arr(i,j,k) = WFromOmega(i,j,k,l_bc_extdir_vals_d[0][2],
                                             xvel_arr,yvel_arr,
                                             mf_u,mf_v,z_phys_nd,dxInv);
            });
        }
 
    // No terrain
    } else {
        if (bx.smallEnd(2) == dom_lo.z) {
            ParallelFor(makeSlab(bx,2,dom_lo.z), [=] AMREX_GPU_DEVICE (int i, int j, int k) {
               dest_arr(i,j,k) = l_bc_extdir_vals_d[0][2];
            });
        }
    }
 
    // *******************************************************
    // Top boundary
    // *******************************************************
 
    // NOTE: if we set SlipWall at top, that generates ERFBCType::ext_dir which sets w=0 here
    // NOTE: if we set  Outflow at top, that generates ERFBCType::foextrap which doesn't touch w here
    if (bx.bigEnd(2) == dom_hi.z+1) {
        if (bc_ptr_w_h[0].hi(2) == ERFBCType::ext_dir) {
            ParallelFor(makeSlab(bx,2,dom_hi.z+1), [=] AMREX_GPU_DEVICE (int i, int j, int k)
            {
                if (l_use_terrain_fitted_coords) {
                    dest_arr(i,j,k) = WFromOmega(i,j,k,l_bc_extdir_vals_d[0][5],
                                                 xvel_arr,yvel_arr,
                                                 mf_u,mf_v,z_phys_nd,dxInv);
                } else {
                    dest_arr(i,j,k) = l_bc_extdir_vals_d[0][5];
                }
            });
        } else if (bc_ptr_w_h[0].hi(2) == ERFBCType::neumann_int) {
            ParallelFor(makeSlab(bx,2,dom_hi.z+1), [=] AMREX_GPU_DEVICE (int i, int j, int k)
            {
                dest_arr(i,j,k) = (Real(4.0)*dest_arr(i,j,dom_hi.z) - dest_arr(i,j,dom_hi.z-1))/three;
            });
        }
    }
    Gpu::streamSynchronize();
}

Here is the call graph for this function:

operator()()

void ERFPhysBCFunct_w::operator()	(	amrex::MultiFab &	mf,
		amrex::MultiFab &	xvel,
		amrex::MultiFab &	yvel,
		amrex::IntVect const &	nghost,
		const double	time,
		int	bccomp,
		bool	do_fb
	)

Impose physical boundary conditions at domain boundaries

Parameters

[in,out]	mf	z-velocity MultiFab to be filled
[in]	xvel	x-velocity MultiFab used by upwind boundary conditions
[in]	yvel	y-velocity MultiFab used by upwind boundary conditions
[in]	nghost	number of ghost cells to be filled for velocity components
[in]	time	time at which the data should be filled
[in]	bccomp	index into m_domain_bcs_type for z velocity
[in]	do_fb	whether to fill periodic/interior ghost cells first

Impose physical boundary conditions on z velocity.

Parameters

[in,out]	mf	z-velocity MultiFab to be filled
[in]	xvel	x-velocity MultiFab used by upwind boundary conditions
[in]	yvel	y-velocity MultiFab used by upwind boundary conditions
[in]	nghost	number of ghost cells to be filled
[in]	time	time at which data should be filled
[in]	bccomp_w	index into m_domain_bcs_type for z velocity
[in]	do_fb	whether to fill periodic/interior ghost cells first

{
    BL_PROFILE("ERFPhysBCFunct_w::()");
 
    //
    // We fill all of the interior and periodic ghost cells first, so we can fill
    //    those directly inside the lateral and vertical calls.
    // If triply periodic this is all we do
    //
    if (do_fb) {
        mf.FillBoundary(m_geom.periodicity());
    }
 
    if (m_geom.isAllPeriodic()) return;
 
    int bccomp_u = BCVars::xvel_bc;
    int bccomp_v = BCVars::yvel_bc;
 
    const auto& domain = m_geom.Domain();
    const auto dxInv   = m_geom.InvCellSizeArray();
 
    Box gdomainz = surroundingNodes(domain,2);
    for (int i = 0; i < AMREX_SPACEDIM; ++i) {
        if (m_geom.isPeriodic(i)) {
            gdomainz.grow(i, nghost[i]);
        }
    }
    //
    // We want to make sure we impose the z-vels at k=0  if the box includes k=0
    //
    if (gdomainz.smallEnd(2) == 0) gdomainz.setSmall(2,1);
 
#ifdef AMREX_USE_OMP
#pragma omp parallel if (Gpu::notInLaunchRegion())
#endif
    {
        for (MFIter mfi(mf,false); mfi.isValid(); ++mfi)
        {
            //
            // This is the box we pass to the different routines
            // NOTE -- this is the full grid NOT the tile box
            //
            Box bx  = mfi.validbox();
 
            //
            // These are the boxes we use to test on relative to the domain
            //
            Box zbx = surroundingNodes(bx,2); zbx.grow(nghost);
            if (zbx.smallEnd(2) < domain.smallEnd(2)) zbx.setSmall(2,domain.smallEnd(2));
            if (zbx.bigEnd(2)   > domain.bigEnd(2))   {
                zbx.setBig(2,domain.bigEnd(2)+1);
            } else if (zbx.bigEnd(2) > xvel[mfi].box().bigEnd(2)) {
                // We do this so if a box at level > 0 doesn't reach the top boundary, we don't go out of bounds
                // (Note xvel is chosen arbitrarily; we could have used yvel instead)
                zbx.setBig(2,xvel[mfi].box().bigEnd(2));
            }
 
            Array4<const Real> z_nd_arr;
            const Array4<const Real>& mf_u = m_mapfac_u->const_array(mfi);
            const Array4<const Real>& mf_v = m_mapfac_v->const_array(mfi);
 
            if (m_z_phys_nd) {
                z_nd_arr = m_z_phys_nd->const_array(mfi);
            }
 
            //
            // Recall that gdomainz.smallEnd(2) = 1 not 0!
            //
            if (!gdomainz.contains(zbx)) {
                Array4<const Real> const& velx_arr = xvel.const_array(mfi);
                Array4<const Real> const& vely_arr = yvel.const_array(mfi);
                Array4<      Real> const& velz_arr = mf.array(mfi);
 
                if (!gdomainz.contains(zbx) && !m_use_real_bcs) {
                    impose_lateral_zvel_bcs(velz_arr,velx_arr,vely_arr,zbx,domain,
                                            mf_u,mf_v,z_nd_arr,dxInv,m_terrain_type,bccomp_w,time);
                }
 
                impose_vertical_zvel_bcs(velz_arr,velx_arr,vely_arr,zbx,domain,mf_u,mf_v,
                                         z_nd_arr,dxInv,bccomp_u, bccomp_v, bccomp_w, m_terrain_type, time);
            }
        } // MFIter
    } // OpenMP
} // operator()

Member Data Documentation

m_bc_extdir_vals

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

private

Referenced by impose_vertical_zvel_bcs().

m_bc_neumann_vals

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

private

m_domain_bcs_type

amrex::Vector<amrex::BCRec> ERFPhysBCFunct_w::m_domain_bcs_type

private

Referenced by impose_vertical_zvel_bcs().

m_domain_bcs_type_d

amrex::Gpu::DeviceVector<amrex::BCRec> ERFPhysBCFunct_w::m_domain_bcs_type_d

private

m_geom

amrex::Geometry ERFPhysBCFunct_w::m_geom

private

m_lev

int ERFPhysBCFunct_w::m_lev

private

Referenced by impose_vertical_zvel_bcs().

m_mapfac_u

amrex::MultiFab* ERFPhysBCFunct_w::m_mapfac_u

private

Referenced by ERFPhysBCFunct_w().

m_mapfac_v

amrex::MultiFab* ERFPhysBCFunct_w::m_mapfac_v

private

Referenced by ERFPhysBCFunct_w().

m_terrain_type

TerrainType ERFPhysBCFunct_w::m_terrain_type

private

m_use_real_bcs

bool ERFPhysBCFunct_w::m_use_real_bcs

private

m_w_bc_data

amrex::Real* ERFPhysBCFunct_w::m_w_bc_data

private

m_z_phys_nd

amrex::MultiFab* ERFPhysBCFunct_w::m_z_phys_nd

private

---

The documentation for this class was generated from the following files:

• Source/BoundaryConditions/ERF_PhysBCFunct.H
• Source/BoundaryConditions/ERF_BoundaryConditionsZvel.cpp
• Source/BoundaryConditions/ERF_PhysBCFunct.cpp