Physics objects
Description
Collaboration diagram for Physics objects:
Classes | |
| class | chrono::fsi::sph::SphBceManager |
| Manager for processing boundary condition enforcing (BCE) particle forces in an FSI system. More... | |
| class | chrono::fsi::sph::SphCollisionSystem |
| Base class for processing proximity computation in an FSI system. More... | |
| struct | chrono::fsi::sph::SphMarkerDataD |
| Struct to store the information of SPH particles on the device. More... | |
| struct | chrono::fsi::sph::SphMarkerDataH |
| Struct to store the information of SPH particles on the host. More... | |
| struct | chrono::fsi::sph::FsiBodyStateH |
| Rigid body states on host. More... | |
| struct | chrono::fsi::sph::FsiBodyStateD |
| Rigid body states on device. More... | |
| struct | chrono::fsi::sph::FsiMeshStateH |
| FEA mesh states on host. More... | |
| struct | chrono::fsi::sph::FsiMeshStateD |
| FEA mesh state on device. More... | |
| struct | chrono::fsi::sph::ProximityDataD |
| Struct to store neighbor search information on the device. More... | |
| struct | chrono::fsi::sph::GPUDeviceInfo |
| Struct to store GPU device information. More... | |
| struct | chrono::fsi::sph::Counters |
| Number of rigid and flexible solid bodies, fluid SPH particles, solid SPH particles, boundary SPH particles. More... | |
| struct | chrono::fsi::sph::FsiDataManager |
| Data manager for the SPH-based FSI system. More... | |
| class | chrono::fsi::sph::SphFluidDynamics |
| Class to represent the fluid/granular dynamics system. More... | |
| class | chrono::fsi::sph::SphForce |
| Base class to calculate force between SPH particles. More... | |
| class | chrono::fsi::sph::SphForceISPH |
| Inter-particle force calculation for the implicit SPH method. More... | |
| class | chrono::fsi::sph::SphForceWCSPH |
| Inter-particle force calculation for explicit SPH schemes. More... | |
Macros | |
| #define | INVPI Real(0.31830988618379) |
| #define | EPSILON Real(1e-8) |
| #define | CH_FSI_SPH_NOINLINE |
| Shift that brings marker b onto the periodic image of b closest to a, along one axis. More... | |
Typedefs | |
| typedef thrust::device_vector< Real3 >::iterator | chrono::fsi::sph::r3IterD |
| typedef device iterators for shorthand SPH operation of thrust vectors of Real3 | |
| typedef thrust::device_vector< Real4 >::iterator | chrono::fsi::sph::r4IterD |
| typedef device iterators for shorthand SPH operation of thrust vectors of Real4 | |
| typedef thrust::tuple< r4IterD, r3IterD, r4IterD, r3IterD, r3IterD, r3IterD > | chrono::fsi::sph::iterTupleSphD |
| typedef device tuple for holding SPH data pos,vel,[rho,pressure,mu,type],tauXxYyZz,tauXyXzYz,pc | |
| typedef thrust::zip_iterator< iterTupleSphD > | chrono::fsi::sph::zipIterSphD |
| typedef thrust::host_vector< Real3 >::iterator | chrono::fsi::sph::r3IterH |
| typedef host iterators for shorthand SPH operation of thrust vectors of Real3 | |
| typedef thrust::host_vector< Real4 >::iterator | chrono::fsi::sph::r4IterH |
| typedef host iterators for shorthand SPH operation of thrust vectors of Real4 | |
| typedef thrust::tuple< r4IterH, r3IterH, r4IterH, r3IterH, r3IterH, r3IterH > | chrono::fsi::sph::iterTupleH |
| typedef host tuple for holding SPH data pos,vel,[rho,pressure,mu,type],tauXxYyZz,tauXyXzYz,pc | |
| typedef thrust::zip_iterator< iterTupleH > | chrono::fsi::sph::zipIterSphH |
| typedef thrust::tuple< r3IterD, r3IterD, r3IterD, r4IterD, r3IterD, r3IterD > | chrono::fsi::sph::iterTupleRigidD |
| typedef device iterators for shorthand rigid body states: pos,lin_vel,lin_acc,rot,ang_Vel,ang_acc | |
| typedef thrust::zip_iterator< iterTupleRigidD > | chrono::fsi::sph::zipIterRigidD |
| typedef thrust::tuple< r3IterH, r3IterH, r3IterH, r4IterH, r3IterH, r3IterH > | chrono::fsi::sph::iterTupleRigidH |
| typedef host iterators for shorthand rigid body states: pos,lin_vel,lin_acc,rot,ang_Vel,ang_acc | |
| typedef thrust::zip_iterator< iterTupleRigidH > | chrono::fsi::sph::zipIterRigidH |
| typedef thrust::tuple< r3IterH, r3IterH, r3IterH, r4IterH, r3IterH, r3IterH > | chrono::fsi::sph::iterTupleChronoBodiesH |
| typedef device iterators for shorthand chrono bodies operations | |
| typedef thrust::zip_iterator< iterTupleChronoBodiesH > | chrono::fsi::sph::zipIterChronoBodiesH |
Enumerations | |
| enum | chrono::fsi::sph::MarkerGroup { FLUID, SOLID, BOUNDARY, NON_FLUID, NON_SOLID, NON_BOUNDARY, ALL } |
| Marker (SPH and BCE) groups. | |
| enum | chrono::fsi::sph::MarkerType { SPH_PARTICLE, SPH_HELPER, SPH_GHOST, BCE_WALL, BCE_RIGID, BCE_FLEX1D, BCE_FLEX2D } |
| Marker (SPH and BCE) type. | |
Functions | |
| void | chrono::fsi::sph::calculateDirectionsMesh1D (FsiDataManager &data_mgr) |
| Utility function to calculate (on the device) direction vectors at the flexible 1-D mesh nodes. More... | |
| void | chrono::fsi::sph::calculateDirectionsMesh2D (FsiDataManager &data_mgr) |
| Utility function to calculate (on the device) direction vectors at the flexible 2-D mesh nodes. More... | |
| void | chrono::fsi::sph::printDirectionsMesh1D (FsiDataManager &data_mgr) |
| void | chrono::fsi::sph::CopyParametersToDevice_SphBceManager (std::shared_ptr< ChFsiParamsSPH > paramsH, std::shared_ptr< Counters > countersH) |
| void | chrono::fsi::sph::CopyParametersToDevice_SphCollisionSystem (std::shared_ptr< ChFsiParamsSPH > paramsH, std::shared_ptr< Counters > countersH) |
| void | chrono::fsi::sph::CopyParametersToDevice_SphFluidDynamics (std::shared_ptr< ChFsiParamsSPH > paramsH, std::shared_ptr< Counters > countersH) |
| void | chrono::fsi::sph::CopyParametersToDevice_SphForceWCSPH (std::shared_ptr< ChFsiParamsSPH > paramsH, std::shared_ptr< Counters > countersH) |
| void | chrono::fsi::sph::CopyParametersToDevice_SphForceISPH (std::shared_ptr< ChFsiParamsSPH > paramsH, std::shared_ptr< Counters > countersH) |
| void | chrono::fsi::sph::CopyParametersToDevice (std::shared_ptr< ChFsiParamsSPH > paramsH, std::shared_ptr< Counters > countersH) |
| __host__ __device__ Real | chrono::fsi::sph::W3h_CubicSpline (Real d, Real invh) |
| __host__ __device__ Real3 | chrono::fsi::sph::GradW3h_CubicSpline (Real3 d, Real invh) |
| __host__ __device__ Real | chrono::fsi::sph::W3h_Quadratic (Real d, Real invh) |
| __host__ __device__ Real3 | chrono::fsi::sph::GradW3h_Quadratic (Real3 d, Real invh) |
| __host__ __device__ Real | chrono::fsi::sph::W3h_QuinticSpline (Real d, Real invh) |
| __host__ __device__ Real3 | chrono::fsi::sph::GradW3h_QuinticSpline (Real3 d, Real invh) |
| __host__ __device__ Real | chrono::fsi::sph::W3h_Wendland (Real d, Real invh) |
| __host__ __device__ Real3 | chrono::fsi::sph::GradW3h_Wendland (Real3 d, Real invh) |
| __host__ __device__ Real | chrono::fsi::sph::W3h (KernelType type, Real d, Real invh) |
| __host__ __device__ Real3 | chrono::fsi::sph::GradW3h (KernelType type, Real3 d, Real invh) |
| __device__ Real | chrono::fsi::sph::Eos (Real rho, EosType eos_type) |
| __device__ Real | chrono::fsi::sph::InvEos (Real pw, EosType eos_type) |
| __device__ Real | chrono::fsi::sph::FerrariCi (Real rho) |
| __device__ CH_FSI_SPH_NOINLINE Real | chrono::fsi::sph::MinimumImageShiftMultiPeriod (Real dist, Real period) |
| __device__ Real | chrono::fsi::sph::MinimumImageShift (Real dist, Real period) |
| __device__ Real3 | chrono::fsi::sph::Modify_Local_PosB (Real3 &b, Real3 a) |
| __device__ Real3 | chrono::fsi::sph::Distance (Real3 a, Real3 b) |
| __device__ void | chrono::fsi::sph::RotationMatrixFromQuaternion (Real3 &AD1, Real3 &AD2, Real3 &AD3, const Real4 &q) |
| __device__ Real3 | chrono::fsi::sph::InverseRotate_By_RotationMatrix_DeviceHost (const Real3 &A1, const Real3 &A2, const Real3 &A3, const Real3 &r3) |
| __device__ int3 | chrono::fsi::sph::calcGridPos (Real3 p) |
| __device__ int | chrono::fsi::sph::reduceGridIndex (int i, int n, bool periodic) |
| Reduce a bin index to the valid range [0, n). More... | |
| __device__ uint | chrono::fsi::sph::calcGridHash (int3 gridPos) |
| __device__ uint | chrono::fsi::sph::calcCellID (uint3 cellPos) |
| __device__ uint | chrono::fsi::sph::getCellPos (int trialCellPos, uint ub) |
| __device__ uint | chrono::fsi::sph::getCenterCellID (const uint *numPartsInCenterCells, const uint threadID) |
| __device__ Real | chrono::fsi::sph::Strain_Rate (Real3 grad_ux, Real3 grad_uy, Real3 grad_uz) |
| __device__ Real | chrono::fsi::sph::Tensor_Norm (Real *T) |
| __device__ Real | chrono::fsi::sph::Sym_Tensor_Norm (Real3 xx_yy_zz, Real3 xy_xz_yz) |
| __device__ Real | chrono::fsi::sph::Inertia_num (Real Strain_rate, Real rho, Real p, Real diam) |
| __device__ Real | chrono::fsi::sph::mu_I (Real Strain_rate, Real I) |
| __device__ Real | chrono::fsi::sph::mu_eff (Real Strain_rate, Real p, Real mu_I) |
| __device__ Real | chrono::fsi::sph::Herschel_Bulkley_stress (Real Strain_rate, Real k, Real n, Real tau0) |
| __device__ Real | chrono::fsi::sph::Herschel_Bulkley_mu_eff (Real Strain_rate, Real k, Real n, Real tau0) |
| __global__ void | chrono::fsi::sph::calc_A_tensor (Real *A_tensor, Real *G_tensor, Real4 *sortedPosRad, Real4 *sortedRhoPreMu, Real *sumWij_inv, uint *cellStart, uint *cellEnd, volatile bool *error_flag) |
| __global__ void | chrono::fsi::sph::calc_L_tensor (Real *A_tensor, Real *L_tensor, Real *G_tensor, Real4 *sortedPosRad, Real4 *sortedRhoPreMu, Real *sumWij_inv, uint *cellStart, uint *cellEnd, volatile bool *error_flag) |
| __global__ void | chrono::fsi::sph::calcRho_kernel (Real4 *sortedPosRad, Real4 *sortedRhoPreMu, Real *sumWij_inv, const uint *neighborList, const uint *mynumContacts, volatile bool *error_flag) |
| __global__ void | chrono::fsi::sph::calcNormalizedRho_Gi_fillInMatrixIndices (Real4 *sortedPosRad, Real3 *sortedVelMas, Real4 *sortedRhoPreMu, Real *sumWij_inv, Real *G_i, Real3 *normals, const uint *csrColInd, const uint *numContacts, volatile bool *error_flag) |
| __global__ void | chrono::fsi::sph::Function_Gradient_Laplacian_Operator (Real4 *sortedPosRad, Real3 *sortedVelMas, Real4 *sortedRhoPreMu, Real *sumWij_inv, Real *G_tensor, Real *L_tensor, Real *A_L, Real3 *A_G, Real *A_f, uint *csrColInd, uint *numContacts, volatile bool *error_flag) |
| __global__ void | chrono::fsi::sph::Jacobi_SOR_Iter (Real4 *sortedRhoPreMu, Real *A_Matrix, Real3 *V_old, Real3 *V_new, Real3 *b3vec, Real *q_old, Real *q_new, Real *b1vec, const uint *csrColInd, const uint *numContacts, bool _3dvector, volatile bool *error_flag) |
| __global__ void | chrono::fsi::sph::Update_AND_Calc_Res (Real4 *sortedRhoPreMu, Real3 *V_old, Real3 *V_new, Real *q_old, Real *q_new, Real *Residuals, bool _3dvector, volatile bool *error_flag) |
| __global__ void | chrono::fsi::sph::Initialize_Variables (Real4 *sortedRhoPreMu, Real *p_old, Real3 *sortedVelMas, Real3 *V_new, volatile bool *error_flag) |
| __global__ void | chrono::fsi::sph::UpdateDensity (Real3 *vis_vel, Real3 *XSPH_Vel, Real3 *new_vel, Real4 *sortedPosRad, Real4 *sortedRhoPreMu, Real *sumWij_inv, uint *cellStart, uint *cellEnd, volatile bool *error_flag) |
| __host__ __device__ bool | chrono::fsi::sph::IsInMarkerGroup (MarkerGroup group, Real code) |
| __host__ __device__ MarkerType | chrono::fsi::sph::GetMarkerType (Real code) |
| __host__ __device__ Real | chrono::fsi::sph::GetMarkerCode (MarkerType type) |
| __host__ __device__ bool | chrono::fsi::sph::IsSphParticle (Real code) |
| __host__ __device__ bool | chrono::fsi::sph::IsFluidParticle (Real code) |
| __host__ __device__ bool | chrono::fsi::sph::IsBceMarker (Real code) |
| __host__ __device__ bool | chrono::fsi::sph::IsBceWallMarker (Real code) |
| __host__ __device__ bool | chrono::fsi::sph::IsBceSolidMarker (Real code) |
Macro Definition Documentation
◆ CH_FSI_SPH_NOINLINE
| #define CH_FSI_SPH_NOINLINE |
Shift that brings marker b onto the periodic image of b closest to a, along one axis.
period is the box length on an axis carrying periodic boundary conditions, and ZERO on a non-periodic one. A zero period disables the shift, which is the correct treatment and is the point of the first branch: on a non-periodic axis the separation between two markers is physical, so folding it by a box length does not select a nearer image, it fabricates an interaction between markers a box length apart. That is reachable, because calcGridHash reduces a marker lying outside a non-periodic axis into the edge bin of that axis, which makes it a neighbor candidate of the markers genuinely there.
The reduction is TOTAL: a marker any number of periods outside the domain is imaged correctly. That matters because the bin reduction in calcGridHash is also total, and the two must agree. A single-period shift would leave a marker two or more periods out binned as its own image yet measured a whole box away, so it would silently fail to interact. Rigid-body markers reach that state, because ApplyPeriodicBoundary*_D deliberately does not wrap them, so a body driving through a periodic channel accumulates unbounded offset.
Within one period the result is identical to a single-period shift, and exactly so: rint returns 0 or +/-1 there, and multiplying a length by those is exact in IEEE 754.
The division is deliberately in an out-of-line callee, and the ordering of this function is deliberate too. Measured on a 106k-particle non-periodic settling case, interleaved repeats against libraries differing by one thing each: inlining the division cost 4% even though that case never executes it, while this form runs 2.5% faster than the single-period version it supersedes and is at worst neutral against unpatched code, because a non-periodic axis now leaves after one comparison. Register counts, occupancy and spill counts are identical for all of those variants across every kernel of both translation units that include this header, so register pressure is NOT the explanation for the inline form's cost, and none is claimed.
Function Documentation
◆ calculateDirectionsMesh1D()
| void chrono::fsi::sph::calculateDirectionsMesh1D | ( | FsiDataManager & | data_mgr | ) |
Utility function to calculate (on the device) direction vectors at the flexible 1-D mesh nodes.
For 1-D meshes, these are averages of the segment direction vectors of adjacent segments.
◆ calculateDirectionsMesh2D()
| void chrono::fsi::sph::calculateDirectionsMesh2D | ( | FsiDataManager & | data_mgr | ) |
Utility function to calculate (on the device) direction vectors at the flexible 2-D mesh nodes.
For 2-D meshes, these are averages of the face normals of adjacent faces.
◆ reduceGridIndex()
|
inline |
Reduce a bin index to the valid range [0, n).
A periodic axis wraps, a non-periodic axis clamps to its edge bin. The reduction is total: no bin index, however far out, can produce a hash outside the cell arrays. For anything within one period of the domain this is identical to adjusting by a single period.
