diff --git a/src/gpu/VirtualFluids_GPU/Visitor/ActuatorLine.h b/src/gpu/VirtualFluids_GPU/Visitor/ActuatorLine.h
index 6863840e3082b980a081a98d24a7a1b49fb79ba7..b458a9b50c189cd064bea53125bc79dc5c8a81ea 100644
--- a/src/gpu/VirtualFluids_GPU/Visitor/ActuatorLine.h
+++ b/src/gpu/VirtualFluids_GPU/Visitor/ActuatorLine.h
@@ -51,6 +51,15 @@ public:
uint getNBlades(){return this->nBlades;};
uint getNumberOfIndices(){return this->numberOfIndices;};
uint getNumberOfNodes(){return this->numberOfNodes;};
+ real* getBladeCoordsX(){return this->bladeCoordsXH;};
+ real* getBladeCoordsY(){return this->bladeCoordsYH;};
+ real* getBladeCoordsZ(){return this->bladeCoordsZH;};
+ real* getBladeVelocitiesX(){return this->bladeVelocitiesXH;};
+ real* getBladeVelocitiesY(){return this->bladeVelocitiesYH;};
+ real* getBladeVelocitiesZ(){return this->bladeVelocitiesZH;};
+ real* getBladeForcesX(){return this->bladeForcesXH;};
+ real* getBladeForcesY(){return this->bladeForcesYH;};
+ real* getBladeForcesZ(){return this->bladeForcesZH;};
private:
void initBoundingSphere(Parameter* para, CudaMemoryManager* cudaManager);
diff --git a/src/gpu/VirtualFluids_GPU/Visitor/Probe.cu b/src/gpu/VirtualFluids_GPU/Visitor/Probe.cu
index f6976c5680e65ec45f6fd5f723c1235e935c8dca..a517da72ff30bc3b8ec052fc88c9badfb59d788e 100644
--- a/src/gpu/VirtualFluids_GPU/Visitor/Probe.cu
+++ b/src/gpu/VirtualFluids_GPU/Visitor/Probe.cu
@@ -89,7 +89,8 @@ __global__ void interpQuantities( uint* pointIndices,
real* vx, real* vy, real* vz, real* rho,
uint* neighborX, uint* neighborY, uint* neighborZ,
bool* quantities,
- uint* quantityArrayOffsets, real* quantityArray
+ uint* quantityArrayOffsets, real* quantityArray,
+ bool interpolate
)
{
const uint x = threadIdx.x;
@@ -106,19 +107,29 @@ __global__ void interpQuantities( uint* pointIndices,
// Get indices of neighbor nodes.
// node referring to BSW cell as seen from probe point
uint k = pointIndices[node];
- uint ke, kn, kt, kne, kte, ktn, ktne;
- getNeighborIndicesBSW( k, ke, kn, kt, kne, kte, ktn, ktne, neighborX, neighborY, neighborZ);
+ real u_interpX, u_interpY, u_interpZ, rho_interp;
- // Trilinear interpolation of macroscopic quantities to probe point
- real dW, dE, dN, dS, dT, dB;
- getInterpolationWeights(dW, dE, dN, dS, dT, dB, distX[node], distY[node], distZ[node]);
+ if(interpolate)
+ {
+ uint ke, kn, kt, kne, kte, ktn, ktne;
+ getNeighborIndicesBSW( k, ke, kn, kt, kne, kte, ktn, ktne, neighborX, neighborY, neighborZ);
- real u_interpX, u_interpY, u_interpZ, rho_interp;
+ // Trilinear interpolation of macroscopic quantities to probe point
+ real dW, dE, dN, dS, dT, dB;
+ getInterpolationWeights(dW, dE, dN, dS, dT, dB, distX[node], distY[node], distZ[node]);
- u_interpX = trilinearInterpolation( dW, dE, dN, dS, dT, dB, k, ke, kn, kt, kne, kte, ktn, ktne, vx );
- u_interpY = trilinearInterpolation( dW, dE, dN, dS, dT, dB, k, ke, kn, kt, kne, kte, ktn, ktne, vy );
- u_interpZ = trilinearInterpolation( dW, dE, dN, dS, dT, dB, k, ke, kn, kt, kne, kte, ktn, ktne, vz );
- rho_interp = trilinearInterpolation( dW, dE, dN, dS, dT, dB, k, ke, kn, kt, kne, kte, ktn, ktne, rho );
+
+ u_interpX = trilinearInterpolation( dW, dE, dN, dS, dT, dB, k, ke, kn, kt, kne, kte, ktn, ktne, vx );
+ u_interpY = trilinearInterpolation( dW, dE, dN, dS, dT, dB, k, ke, kn, kt, kne, kte, ktn, ktne, vy );
+ u_interpZ = trilinearInterpolation( dW, dE, dN, dS, dT, dB, k, ke, kn, kt, kne, kte, ktn, ktne, vz );
+ rho_interp = trilinearInterpolation( dW, dE, dN, dS, dT, dB, k, ke, kn, kt, kne, kte, ktn, ktne, rho );
+ } else
+ {
+ u_interpX = vx[k];
+ u_interpY = vy[k];
+ u_interpZ = vz[k];
+ rho_interp = rho[k];
+ }
calculateQuantities(n, quantityArray, quantities, quantityArrayOffsets, nPoints, node, u_interpX, u_interpY, u_interpZ, rho_interp);
@@ -160,66 +171,77 @@ void Probe::init(Parameter* para, GridProvider* gridProvider, CudaMemoryManager*
pointCoordsX_level.push_back( pointCoordX );
pointCoordsY_level.push_back( pointCoordY );
pointCoordsZ_level.push_back( pointCoordZ );
+ break;
// printf("x %f y %f z %f", pointCoordX, pointCoordY, pointCoordZ);
}
}
}
- probeParams[level] = new ProbeStruct;
- probeParams[level]->vals = 1;
- probeParams[level]->nPoints = uint(probeIndices_level.size());
+ this->addProbeStruct(cudaManager, probeIndices_level,
+ distX_level, distY_level, distZ_level,
+ pointCoordsX_level, pointCoordsX_level, pointCoordsX_level,
+ level);
+ }
+}
- probeParams[level]->pointCoordsX = (real*)malloc(probeParams[level]->nPoints*sizeof(real));
- probeParams[level]->pointCoordsY = (real*)malloc(probeParams[level]->nPoints*sizeof(real));
- probeParams[level]->pointCoordsZ = (real*)malloc(probeParams[level]->nPoints*sizeof(real));
+void Probe::addProbeStruct(CudaMemoryManager* cudaManager, std::vector<int> probeIndices,
+ std::vector<real> distX, std::vector<real> distY, std::vector<real> distZ,
+ std::vector<real> pointCoordsX, std::vector<real> pointCoordsY, std::vector<real> pointCoordsZ,
+ int level)
+{
+ probeParams[level] = new ProbeStruct;
+ probeParams[level]->vals = 1;
+ probeParams[level]->nPoints = uint(probeIndices.size());
- std::copy(pointCoordsX_level.begin(), pointCoordsX_level.end(), probeParams[level]->pointCoordsX);
- std::copy(pointCoordsY_level.begin(), pointCoordsY_level.end(), probeParams[level]->pointCoordsY);
- std::copy(pointCoordsZ_level.begin(), pointCoordsZ_level.end(), probeParams[level]->pointCoordsZ);
+ probeParams[level]->pointCoordsX = (real*)malloc(probeParams[level]->nPoints*sizeof(real));
+ probeParams[level]->pointCoordsY = (real*)malloc(probeParams[level]->nPoints*sizeof(real));
+ probeParams[level]->pointCoordsZ = (real*)malloc(probeParams[level]->nPoints*sizeof(real));
- // Might have to catch nPoints=0 ?!?!
- cudaManager->cudaAllocProbeDistances(this, level);
- cudaManager->cudaAllocProbeIndices(this, level);
+ std::copy(pointCoordsX.begin(), pointCoordsX.end(), probeParams[level]->pointCoordsX);
+ std::copy(pointCoordsY.begin(), pointCoordsY.end(), probeParams[level]->pointCoordsY);
+ std::copy(pointCoordsZ.begin(), pointCoordsZ.end(), probeParams[level]->pointCoordsZ);
- std::copy(distX_level.begin(), distX_level.end(), probeParams[level]->distXH);
- std::copy(distY_level.begin(), distY_level.end(), probeParams[level]->distYH);
- std::copy(distZ_level.begin(), distZ_level.end(), probeParams[level]->distZH);
- std::copy(probeIndices_level.begin(), probeIndices_level.end(), probeParams[level]->pointIndicesH);
+ // Might have to catch nPoints=0 ?!?!
+ cudaManager->cudaAllocProbeDistances(this, level);
+ cudaManager->cudaAllocProbeIndices(this, level);
- cudaManager->cudaCopyProbeDistancesHtoD(this, level);
- cudaManager->cudaCopyProbeIndicesHtoD(this, level);
+ std::copy(distX.begin(), distX.end(), probeParams[level]->distXH);
+ std::copy(distY.begin(), distY.end(), probeParams[level]->distYH);
+ std::copy(distZ.begin(), distZ.end(), probeParams[level]->distZH);
+ std::copy(probeIndices.begin(), probeIndices.end(), probeParams[level]->pointIndicesH);
- uint arrOffset = 0;
+ cudaManager->cudaCopyProbeDistancesHtoD(this, level);
+ cudaManager->cudaCopyProbeIndicesHtoD(this, level);
- cudaManager->cudaAllocProbeQuantitiesAndOffsets(this, level);
+ uint arrOffset = 0;
- for( int var=0; var<int(PostProcessingVariable::LAST); var++){
- if(this->quantities[var])
- {
+ cudaManager->cudaAllocProbeQuantitiesAndOffsets(this, level);
- probeParams[level]->quantitiesH[var] = true;
- probeParams[level]->arrayOffsetsH[var] = arrOffset;
+ for( int var=0; var<int(PostProcessingVariable::LAST); var++){
+ if(this->quantities[var])
+ {
- arrOffset += getPostProcessingVariableNames(static_cast<PostProcessingVariable>(var)).size();
+ probeParams[level]->quantitiesH[var] = true;
+ probeParams[level]->arrayOffsetsH[var] = arrOffset;
- }}
-
- cudaManager->cudaCopyProbeQuantitiesAndOffsetsHtoD(this, level);
+ arrOffset += getPostProcessingVariableNames(static_cast<PostProcessingVariable>(var)).size();
- probeParams[level]->nArrays = arrOffset;
+ }}
+
+ cudaManager->cudaCopyProbeQuantitiesAndOffsetsHtoD(this, level);
- cudaManager->cudaAllocProbeQuantityArray(this, level);
+ probeParams[level]->nArrays = arrOffset;
- for(uint arr=0; arr<probeParams[level]->nArrays; arr++)
+ cudaManager->cudaAllocProbeQuantityArray(this, level);
+
+ for(uint arr=0; arr<probeParams[level]->nArrays; arr++)
+ {
+ for( uint point=0; point<probeParams[level]->nPoints; point++)
{
- for( uint point=0; point<probeParams[level]->nPoints; point++)
- {
- probeParams[level]->quantitiesArrayH[arr*probeParams[level]->nPoints+point] = 0.0f;
- }
+ probeParams[level]->quantitiesArrayH[arr*probeParams[level]->nPoints+point] = 0.0f;
}
- cudaManager->cudaCopyProbeQuantityArrayHtoD(this, level);
-
}
+ cudaManager->cudaCopyProbeQuantityArrayHtoD(this, level);
}
@@ -237,7 +259,7 @@ void Probe::visit(Parameter* para, CudaMemoryManager* cudaManager, int level, ui
probeStruct->distXD, probeStruct->distYD, probeStruct->distZD,
para->getParD(level)->vx_SP, para->getParD(level)->vy_SP, para->getParD(level)->vz_SP, para->getParD(level)->rho_SP,
para->getParD(level)->neighborX_SP, para->getParD(level)->neighborY_SP, para->getParD(level)->neighborZ_SP,
- probeStruct->quantitiesD, probeStruct->arrayOffsetsD, probeStruct->quantitiesArrayD );
+ probeStruct->quantitiesD, probeStruct->arrayOffsetsD, probeStruct->quantitiesArrayD, true);
probeStruct->vals++;
if(max(int(t) - int(this->tStartOut), -1) % this->tOut == 0)
diff --git a/src/gpu/VirtualFluids_GPU/Visitor/Probe.h b/src/gpu/VirtualFluids_GPU/Visitor/Probe.h
index 54a78d538bbfd418f6d6fc80368e275cf580c5cc..61ba927a1ee3111c5cb1b06c123f3d15d1f3734d 100644
--- a/src/gpu/VirtualFluids_GPU/Visitor/Probe.h
+++ b/src/gpu/VirtualFluids_GPU/Visitor/Probe.h
@@ -60,6 +60,12 @@ public:
void writeGridFiles(Parameter* para, int level, std::vector<std::string >& fnames, int t);
std::vector<std::string> getVarNames();
+private:
+ void addProbeStruct(CudaMemoryManager* cudaManager, std::vector<int> probeIndices,
+ std::vector<real> distX, std::vector<real> distY, std::vector<real> distZ,
+ std::vector<real> pointCoordsX, std::vector<real> pointCoordsY, std::vector<real> pointCoordsZ,
+ int level);
+
private:
const std::string probeName;