Commit 88282e57 authored by Florent Pruvost's avatar Florent Pruvost
parents 0f9d37d0 2be6b19c
This diff is collapsed.
......@@ -18,6 +18,8 @@
#include "Utils/FPoint.hpp"
#include "Utils/FParameters.hpp"
#include "Utils/FGenerateDistribution.hpp"
#include "Files/FExportWriter.hpp"
#include "Utils/FParameterNames.hpp"
//
......@@ -35,7 +37,6 @@
//! \param -fout name: generic name for files (without extension) and save data
//! with following format in name.fma or name.bfma if -bin is set"
//! \param -bin save output in binary mode (name file name.bfma
//! \param -visufmt format for the visu file (vtk, vtp, cvs or cosmo). vtp is the default
//!
//!
//! \b examples
......@@ -44,6 +45,7 @@
//!
//! changeFormat -fin unitCubeXYZQ100.fma -fout unitCubeXYZQ100 -bin
// \param -visufmt format for the visu file (vtk, vtp, cvs or cosmo). vtp is the default
int main(int argc, char ** argv){
......@@ -51,8 +53,7 @@ int main(int argc, char ** argv){
"Driver to change the format of the input file. "
"fdlpoly is not supported for now.",
FParameterDefinitions::InputFile, FParameterDefinitions::OutputFile,
FParameterDefinitions::OutputVisuFile,FParameterDefinitions::FormatVisuFile,
FParameterDefinitions::OutputBinFormat);
FParameterDefinitions::OutputVisuFile);
FSize NbPoints;
FReal * particles = nullptr ;
......@@ -86,19 +87,8 @@ int main(int argc, char ** argv){
// Generate file for ScalFMM FMAGenericLoader
//
if(FParameters::existParameter(argc, argv, FParameterDefinitions::OutputFile.options)){
std::string name(FParameters::getStr(argc,argv,FParameterDefinitions::OutputFile.options, "output"));
std::string ext(".");
if(name.find(ext) !=std::string::npos) {
std::cout << "No file with extension permitted for output name : " << name << std::endl;
exit(-1);
}
if( FParameters::existParameter(argc, argv, FParameterDefinitions::OutputBinFormat.options)){
name += ".bfma";
}
else {
name += ".fma";
}
FFmaGenericWriter writer(name) ;
std::string name(FParameters::getStr(argc,argv,FParameterDefinitions::OutputFile.options, "output.fma"));
FFmaGenericWriter writer(name) ;
writer.writeHeader( loader->getCenterOfBox(), loader->getBoxWidth() , NbPoints, sizeof(FReal), nbData) ;
writer.writeArrayOfReal(particles, nbData, NbPoints);
}
......@@ -106,53 +96,8 @@ int main(int argc, char ** argv){
// Generate file for visualization purpose
//
if(FParameters::existParameter(argc, argv, FParameterDefinitions::OutputVisuFile.options)){
std::string outfilename(FParameters::getStr(argc,argv,FParameterDefinitions::OutputFile.options, "output"));
std::string visufile(""), fmt(FParameters::getStr(argc,argv,FParameterDefinitions::OutputVisuFile.options, "vtp"));
if( fmt == "vtp" ){
visufile = outfilename + ".vtp" ;
}
else if( fmt == "vtk" ){
visufile = outfilename + ".vtk" ;
}
else if( fmt == "cosmo" ){
if(nbData !=4) {
std::cerr << "Cosmos export accept only 4 data per particles. here: "<<nbData<<std::endl;
std::exit(EXIT_FAILURE);
}
visufile = outfilename + ".cosmo" ;
}
else {
visufile = outfilename + ".csv" ;
}
std::ofstream file( visufile, std::ofstream::out);
if(!file) {
std::cout << "Cannot open file."<< std::endl;
exit(-1) ;
} //
//
// Export data in cvs format
//
if( fmt == "vtp" ){
std::cout << "Writes in XML VTP format (visualization) in file "<< visufile <<std::endl ;
if(nbData==4){
exportVTKxml( file, particles, NbPoints) ;
}
else {
exportVTKxml( file, particles, NbPoints,nbData) ;
}
}
else if( fmt == "vtk" ){
std::cout << "Writes in VTK format (visualization) in file "<< visufile <<std::endl ;
exportVTK( file, particles, NbPoints,nbData) ;
}
else if( fmt == "cosmo" ){
std::cout << "Writes in COSMO format (visualization) in file "<< visufile <<std::endl ;
exportCOSMOS( file, particles, NbPoints) ;
}
else {
std::cout << "Writes in CVS format (visualization) in file "<<visufile<<std::endl ;
exportCVS( file, particles, NbPoints,nbData) ;
}
std::string outfilename(FParameters::getStr(argc,argv,FParameterDefinitions::OutputFile.options, "output.vtp"));
driverExportData(outfilename, particles , NbPoints,loader->getNbRecordPerline() );
}
//
delete particles ;
......
......@@ -16,6 +16,7 @@
#include "Utils/FPoint.hpp"
#include "Utils/FGenerateDistribution.hpp"
#include "Files/FFmaGenericLoader.hpp"
#include "Files/FExportWriter.hpp"
#include "Utils/FParameterNames.hpp"
......@@ -74,6 +75,10 @@
int main(int argc, char ** argv){
const FParameterNames LocalOptionEllipsoid = {
{"-ellipsoid"} ,
" non uniform distribution on an ellipsoid of aspect ratio given by \n -ar a:b:c with a, b and c > 0\n"
};
FHelpDescribeAndExit(argc, argv,
">> Driver to generate N points (non)uniformly distributed on a given geometry.\n"
"Options \n"
......@@ -97,12 +102,13 @@ int main(int argc, char ** argv){
" -radius R - default value 10.0\n"
" Physical values\n"
" -charge generate physical values between -1 and 1 otherwise generate between 0 and 1 \n"
" -zeromean the average of the physical values is zero \n"
" Output \n"
" -filename name: generic name for files (without extension) and save data\n"
" with following format in name.fma or name.bfma in -bin is set\n"
" -visufmt vtk, vtp, cosmo or cvs format.",
FParameterDefinitions::InputFile, FParameterDefinitions::OutputBinFormat, FParameterDefinitions::NbParticles);
" -zeromean the average of the physical values is zero \n",
// " Output \n"
// " -filename name: generic name for files (without extension) and save data\n"
// " with following format in name.fma or name.bfma in -bin is set\n"
// " -visufmt vtk, vtp, cosmo or cvs format.",
FParameterDefinitions::OutputFile,
FParameterDefinitions::NbParticles,FParameterDefinitions::OutputVisuFile,LocalOptionEllipsoid);
......@@ -188,14 +194,16 @@ int main(int argc, char ** argv){
std::cout << "A: "<<A<<" B "<< B << " C: " << C<<std::endl;
unifRandonPointsOnProlate(NbPoints,A,C,particles);
BoxWith = 2.0*C;
}
} //const int NbPoints = FParameters::getValue(argc,argv,FParameterDefinitions::NbParticles.options, 20000);
else if(FParameters::existParameter(argc, argv, "-ellipsoid")){
// else if(FParameters::existParameter(argc, argv, "-ellipsoid")){
std::string dd(":"),aspectRatio = FParameters::getStr(argc,argv,"-ar", "1:1:2");
// std::string dd(":"),aspectRatio = FParameters::getStr(argc,argv,"-ar", "1:1:2");
FReal A,B,C ;
size_t pos = aspectRatio.find(":"); aspectRatio.replace(pos,1," ");
pos = aspectRatio.find(":"); aspectRatio.replace(pos,1," ");
std::stringstream ss(aspectRatio); ss >>A >> B >> C ;
// std::cout << "A: "<<A<<" B "<< B << " C: " << C<<std::endl;
std::cout << "A: "<<A<<" B "<< B << " C: " << C<<std::endl;
nonunifRandonPointsOnElipsoid(NbPoints,A,B,C,particles);
BoxWith = 2.0*FMath::Max( A,FMath::Max( B,C)) ;
}
......@@ -220,14 +228,7 @@ int main(int argc, char ** argv){
BoxWith += 2*extraRadius ;
}
std::string name(genericFileName);
if( FParameters::existParameter(argc, argv, FParameterDefinitions::OutputBinFormat.options)){
name += ".bfma";
}
else {
name += ".fma";
}
std::cout << "Write "<< NbPoints <<" Particles in file " << name <<std::endl;
std::cout << "Write "<< NbPoints <<" Particles in file " << name << std::endl;
FFmaGenericWriter writer(name) ;
writer.writeHeader(Centre,BoxWith, NbPoints, *ppart) ;
writer.writeArrayOfParticles(ppart, NbPoints);
......@@ -237,43 +238,8 @@ int main(int argc, char ** argv){
// Generate file for visualization
//
if(FParameters::existParameter(argc, argv, FParameterDefinitions::OutputVisuFile.options)){
std::string visufile(""), fmt(FParameters::getStr(argc,argv,FParameterDefinitions::OutputVisuFile.options, "vtp"));
if( fmt == "vtp" ){
visufile = genericFileName + ".vtp" ;
}
else if( fmt == "vtk" ){
visufile = genericFileName + ".vtk" ;
}
else if( fmt == "cosmo" ){
visufile = genericFileName + ".cosmo" ;
}
else {
visufile = genericFileName + ".csv" ;
}
std::ofstream file( visufile.c_str(), std::ofstream::out);
if(!file) {
std::cout << "Cannot open file."<< std::endl;
exit(-1) ;
} //
//
// Export data in VTK format
//
if( fmt == "vtp" ){
std::cout << "Writes in XML VTP format (visualization) in file "<< visufile <<std::endl ;
exportVTKxml( file, particles, NbPoints) ;
}
else if( fmt == "vtk" ){
std::cout << "Writes in VTK format (visualization) in file "<< visufile <<std::endl ;
exportVTK( file, particles, NbPoints) ;
}
else if( fmt == "cosmo" ){
std::cout << "Writes in COSMO format (visualization) in file "<< visufile <<std::endl ;
exportCOSMOS( file, particles, NbPoints) ;
}
else {
std::cout << "Writes in CVS format (visualization) in file "<<visufile<<std::endl ;
exportCVS( file, particles, NbPoints) ;
}
std::string visufile(FParameters::getStr(argc,argv,FParameterDefinitions::OutputVisuFile.options, "output.vtp"));
driverExportData(visufile, particles , NbPoints);
}
//
delete [] particles ;
......
......@@ -6,10 +6,18 @@ Copyright (c) 2011-2014 Inria, All rights reserved.
This file contains the main features as well as overviews of specific
bug fixes (and other actions) for each version of ScalFMM since
version 1.1
1.3-
-----
- Add interpolation FMM based on uniform grid points.
- Add blocked version of the algorithm to increase the granularity (task-based approach)
-
1.2.1
-----
- Bug fix : Support for huge MPI message in tree construction and Parallel QuickSort (count can be greater than Int32.MaxValue)
- Bug fix : Data sharing attribute clauses for omp in Core/FAlgorithmThreadProc.hpp
1.2-
1.2
-----
- New FMA format to read/write particles
- Add examples repository
......@@ -23,7 +31,3 @@ version 1.1
- Add SSE and AVX code for 1/r kernel
- CMake improvements
1.2.1
-----
- Bug fix : Support for huge MPI message in tree construction and Parallel QuickSort (count can be greater than Int32.MaxValue)
- Bug fix : Data sharing attribute clauses for omp in Core/FAlgorithmThreadProc.hpp
......@@ -117,9 +117,9 @@ public:
// To get access to descriptor
friend struct FTestCellDescriptor;
friend struct FTestCell_Alignement;
};
#endif //FTESTCELL_HPP
This diff is collapsed.
......@@ -561,7 +561,7 @@ public:
FFmaGenericWriter(const std::string & filename): binaryFile(false) {
std::string ext(".bfma");
// open particle file
if(filename.find(ext) !=std::string::npos) {
if(filename.find(".bfma") !=std::string::npos) {
binaryFile = true;
this->file = new std::fstream (filename.c_str(),std::ifstream::out| std::ios::binary);
}
......
......@@ -18,7 +18,7 @@
template <class CellClass>
class FGroupOfCells {
/** One header is allocated at the beginning of each block */
struct BlockHeader{
struct alignas(1) BlockHeader{
MortonIndex startingIndex;
MortonIndex endingIndex;
int numberOfCellsInBlock;
......
......@@ -19,7 +19,7 @@
template <unsigned NbAttributesPerParticle, class AttributeClass = FReal>
class FGroupOfParticles {
/** One header is allocated at the beginning of each block */
struct BlockHeader{
struct alignas(1) BlockHeader{
MortonIndex startingIndex;
MortonIndex endingIndex;
int numberOfLeavesInBlock;
......@@ -36,7 +36,7 @@ class FGroupOfParticles {
};
/** Information about a leaf */
struct LeafHeader {
struct alignas(1) LeafHeader {
int nbParticles;
size_t offSet;
};
......
......@@ -162,7 +162,9 @@ public:
initCodelet();
FLOG(FLog::Controller << "FGroupTaskStarPUAlgorithm (Max Worker " << starpu_worker_get_count() << ")\n");
#ifdef STARPU_USE_CPU
FLOG(FLog::Controller << "FGroupTaskStarPUAlgorithm (Max CPU " << starpu_cpu_worker_get_count() << ")\n");
#endif
#ifdef ScalFMM_ENABLE_OPENCL_KERNEL
FLOG(FLog::Controller << "FGroupTaskStarPUAlgorithm (Max OpenCL " << starpu_opencl_worker_get_count() << ")\n");
#endif
......@@ -172,11 +174,40 @@ public:
}
~FGroupTaskStarPUAlgorithm(){
starpu_resume();
cleanHandle();
delete[] handles_up;
delete[] handles_down;
starpu_resume();
starpu_pthread_mutex_t releaseMutex;
starpu_pthread_mutex_init(&releaseMutex, NULL);
#ifdef STARPU_USE_CPU
FStarPUUtils::ExecOnWorkers(STARPU_CPU, [&](){
starpu_pthread_mutex_lock(&releaseMutex);
cpuWrapper.releaseKernel(starpu_worker_get_id());
starpu_pthread_mutex_unlock(&releaseMutex);
});
wrappers.set(FSTARPU_CPU_IDX, &cpuWrapper);
#endif
#ifdef ScalFMM_ENABLE_CUDA_KERNEL
FStarPUUtils::ExecOnWorkers(STARPU_CUDA, [&](){
starpu_pthread_mutex_lock(&releaseMutex);
cudaWrapper.releaseKernel(starpu_worker_get_id());
starpu_pthread_mutex_unlock(&releaseMutex);
});
wrappers.set(FSTARPU_CUDA_IDX, &cudaWrapper);
#endif
#ifdef ScalFMM_ENABLE_OPENCL_KERNEL
FStarPUUtils::ExecOnWorkers(STARPU_OPENCL, [&](){
starpu_pthread_mutex_lock(&releaseMutex);
openclWrapper.releaseKernel(starpu_worker_get_id());
starpu_pthread_mutex_unlock(&releaseMutex);
});
wrappers.set(FSTARPU_OPENCL_IDX, &openclWrapper);
#endif
starpu_pthread_mutex_destroy(&releaseMutex);
starpu_shutdown();
}
......
......@@ -180,7 +180,9 @@ public:
initCodeletMpi();
FLOG(FLog::Controller << "FGroupTaskStarPUAlgorithm (Max Worker " << starpu_worker_get_count() << ")\n");
#ifdef STARPU_USE_CPU
FLOG(FLog::Controller << "FGroupTaskStarPUAlgorithm (Max CPU " << starpu_cpu_worker_get_count() << ")\n");
#endif
#ifdef ScalFMM_ENABLE_OPENCL_KERNEL
FLOG(FLog::Controller << "FGroupTaskStarPUAlgorithm (Max OpenCL " << starpu_opencl_worker_get_count() << ")\n");
#endif
......@@ -190,12 +192,41 @@ public:
}
~FGroupTaskStarPUMpiAlgorithm(){
starpu_resume();
cleanHandle();
cleanHandleMpi();
delete[] handles_up;
delete[] handles_down;
starpu_resume();
starpu_pthread_mutex_t releaseMutex;
starpu_pthread_mutex_init(&releaseMutex, NULL);
#ifdef STARPU_USE_CPU
FStarPUUtils::ExecOnWorkers(STARPU_CPU, [&](){
starpu_pthread_mutex_lock(&releaseMutex);
cpuWrapper.releaseKernel(starpu_worker_get_id());
starpu_pthread_mutex_unlock(&releaseMutex);
});
wrappers.set(FSTARPU_CPU_IDX, &cpuWrapper);
#endif
#ifdef ScalFMM_ENABLE_CUDA_KERNEL
FStarPUUtils::ExecOnWorkers(STARPU_CUDA, [&](){
starpu_pthread_mutex_lock(&releaseMutex);
cudaWrapper.releaseKernel(starpu_worker_get_id());
starpu_pthread_mutex_unlock(&releaseMutex);
});
wrappers.set(FSTARPU_CUDA_IDX, &cudaWrapper);
#endif
#ifdef ScalFMM_ENABLE_OPENCL_KERNEL
FStarPUUtils::ExecOnWorkers(STARPU_OPENCL, [&](){
starpu_pthread_mutex_lock(&releaseMutex);
openclWrapper.releaseKernel(starpu_worker_get_id());
starpu_pthread_mutex_unlock(&releaseMutex);
});
wrappers.set(FSTARPU_OPENCL_IDX, &openclWrapper);
#endif
starpu_pthread_mutex_destroy(&releaseMutex);
starpu_mpi_shutdown();
starpu_shutdown();
}
......
......@@ -3,7 +3,7 @@
#include "../Utils/FGlobal.hpp"
struct OutOfBlockInteraction{
struct alignas(1) OutOfBlockInteraction{
MortonIndex outIndex;
MortonIndex insideIndex;
int outPosition;
......
......@@ -63,9 +63,14 @@ public:
kernels[workerId] = new KernelClass(*originalKernel);
}
void releaseKernel(const int workerId){
delete kernels[workerId];
kernels[workerId] = nullptr;
}
~FStarPUCpuWrapper(){
for(int idxKernel = 0 ; idxKernel < STARPU_MAXCPUS ; ++idxKernel ){
delete kernels[idxKernel];
FAssertLF(kernels[idxKernel] == nullptr);
}
}
......
......@@ -60,11 +60,14 @@ public:
kernels[workerId] = FCuda__BuildCudaKernel<CudaKernelClass>(originalKernel);
}
void releaseKernel(const int workerId){
FCuda__ReleaseCudaKernel(kernels[workerId]);
kernels[workerId] = nullptr;
}
~FStarPUCudaWrapper(){
for(int idxKernel = 0 ; idxKernel < STARPU_MAXCUDADEVS ; ++idxKernel ){
if(kernels[idxKernel]){
FCuda__ReleaseCudaKernel(kernels[idxKernel]);
}
FAssertLF(kernels[idxKernel] == nullptr);
}
}
......
......@@ -58,21 +58,28 @@ public:
kernels[workerId]->initDeviceFromKernel(*originalKernel);
}
void releaseKernel(const int workerId){
kernels[workerId]->releaseKernel();
delete kernels[workerId];
kernels[workerId] = nullptr;
}
~FStarPUOpenClWrapper(){
for(int idxKernel = 0 ; idxKernel < STARPU_MAXOPENCLDEVS ; ++idxKernel ){
delete kernels[idxKernel];
FAssertLF(kernels[idxKernel] == nullptr);
}
}
static void bottomPassCallback(void *buffers[], void *cl_arg){
cl_mem leafCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[0]));
cl_mem leafCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[0]));
size_t leafCellsSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[0]);
cl_mem containersPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[1]));
cl_mem containersPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[1]));
size_t containersSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[1]);
FStarPUPtrInterface* worker = nullptr;
starpu_codelet_unpack_args(cl_arg, &worker);
OpenCLKernelClass* kernel = worker->get<ThisClass>(FSTARPU_OPENCL_IDX)->kernels[starpu_worker_get_id()];
kernel->bottomPassPerform(leafCellsPtr, leafCellsSize, containersPtr, containersSize);
}
......@@ -81,7 +88,7 @@ public:
/////////////////////////////////////////////////////////////////////////////////////
static void upwardPassCallback(void *buffers[], void *cl_arg){
cl_mem currentCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[0]));
cl_mem currentCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[0]));
size_t currentCellsSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[0]);
FStarPUPtrInterface* worker = nullptr;
......@@ -94,7 +101,7 @@ public:
size_t subCellGroupsSize[9];
memset(subCellGroupsSize, 0, 9*sizeof(size_t));
for(int idxSubGroup = 0; idxSubGroup < nbSubCellGroups ; ++idxSubGroup){
subCellGroupsPtr[idxSubGroup] = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[idxSubGroup+1]));
subCellGroupsPtr[idxSubGroup] = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[idxSubGroup+1]));
subCellGroupsSize[idxSubGroup] = (STARPU_VARIABLE_GET_ELEMSIZE(buffers[idxSubGroup+1]));
}
......@@ -108,9 +115,9 @@ public:
/////////////////////////////////////////////////////////////////////////////////////
#ifdef STARPU_USE_MPI
static void transferInoutPassCallbackMpi(void *buffers[], void *cl_arg){
cl_mem currentCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[0]));
cl_mem currentCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[0]));
size_t currentCellsSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[0]);
cl_mem externalCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[1]));
cl_mem externalCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[1]));
size_t externalCellsSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[1]);
FStarPUPtrInterface* worker = nullptr;
......@@ -139,8 +146,7 @@ public:
/////////////////////////////////////////////////////////////////////////////////////
static void transferInPassCallback(void *buffers[], void *cl_arg){
FAssertLF(STARPU_VARIABLE_GET_PTR(buffers[0]) == STARPU_VARIABLE_GET_PTR(buffers[1]));
cl_mem currentCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[0]));
cl_mem currentCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[0]));
size_t currentCellsSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[0]);
FStarPUPtrInterface* worker = nullptr;
......@@ -153,12 +159,9 @@ public:
}
static void transferInoutPassCallback(void *buffers[], void *cl_arg){
FAssertLF(STARPU_VARIABLE_GET_PTR(buffers[0]) == STARPU_VARIABLE_GET_PTR(buffers[2]));
FAssertLF(STARPU_VARIABLE_GET_PTR(buffers[1]) == STARPU_VARIABLE_GET_PTR(buffers[3]));
cl_mem currentCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[0]));
cl_mem currentCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[0]));
size_t currentCellsSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[0]);
cl_mem externalCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[1]));
cl_mem externalCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[1]));
size_t externalCellsSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[1]);
FStarPUPtrInterface* worker = nullptr;
......@@ -186,7 +189,7 @@ public:
/// Downard Pass
/////////////////////////////////////////////////////////////////////////////////////
static void downardPassCallback(void *buffers[], void *cl_arg){
cl_mem currentCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[0]));
cl_mem currentCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[0]));
size_t currentCellsSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[0]);
FStarPUPtrInterface* worker = nullptr;
......@@ -199,7 +202,7 @@ public:
size_t subCellGroupsSize[9];
memset(subCellGroupsSize, 0, 9*sizeof(size_t));
for(int idxSubGroup = 0; idxSubGroup < nbSubCellGroups ; ++idxSubGroup){
subCellGroupsPtr[idxSubGroup] = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[idxSubGroup+1]));
subCellGroupsPtr[idxSubGroup] = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[idxSubGroup+1]));
subCellGroupsSize[idxSubGroup] = (STARPU_VARIABLE_GET_ELEMSIZE(buffers[idxSubGroup+1]));
}
......@@ -214,9 +217,9 @@ public:
#ifdef STARPU_USE_MPI
static void directInoutPassCallbackMpi(void *buffers[], void *cl_arg){
cl_mem containersPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[0]));
cl_mem containersPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[0]));
size_t containersSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[0]);
cl_mem externalContainersPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[1]));
cl_mem externalContainersPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[1]));
size_t externalContainersSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[1]);
FStarPUPtrInterface* worker = nullptr;
......@@ -243,7 +246,7 @@ public:
/////////////////////////////////////////////////////////////////////////////////////
static void directInPassCallback(void *buffers[], void *cl_arg){
cl_mem containersPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[0]));
cl_mem containersPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[0]));
size_t containerSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[0]);
FStarPUPtrInterface* worker = nullptr;
......@@ -253,9 +256,9 @@ public:
}
static void directInoutPassCallback(void *buffers[], void *cl_arg){
cl_mem containersPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[0]));
cl_mem containersPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[0]));
size_t containerSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[0]);
cl_mem externalContainersPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[1]));
cl_mem externalContainersPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[1]));
size_t externalContainersSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[1]);
FStarPUPtrInterface* worker = nullptr;
......@@ -282,9 +285,9 @@ public:
/////////////////////////////////////////////////////////////////////////////////////
static void mergePassCallback(void *buffers[], void *cl_arg){
cl_mem leafCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[0]));
cl_mem leafCellsPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[0]));
size_t leafCellsSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[0]);
cl_mem containersPtr = ((cl_mem)STARPU_VARIABLE_GET_PTR(buffers[1]));
cl_mem containersPtr = ((cl_mem)STARPU_VARIABLE_GET_DEV_HANDLE(buffers[1]));
size_t containersSize = STARPU_VARIABLE_GET_ELEMSIZE(buffers[1]);
FStarPUPtrInterface* worker = nullptr;
......
This diff is collapsed.
// @SCALFMM_PRIVATE
#ifndef FEMPTYOPENCLCODE_HPP
#define FEMPTYOPENCLCODE_HPP
// Return the same thing as FEmptyKernel.cl
class FEmptyOpenCLCode{
public:
FEmptyOpenCLCode(){
}
const char* getKernelCode(const int /*inDevId*/){
const char* kernelcode =
"typedef long long int MortonIndex; \
\
typedef struct OutOfBlockInteraction{ \
MortonIndex outIndex; \
MortonIndex insideIndex; \
int outPosition; \
} __attribute__ ((aligned (1))); \
struct Uptr9{ \
__global unsigned char* ptrs[9]; \
} __attribute__ ((aligned (1))); \
struct size_t9{ \
size_t v[9]; \
}__attribute__ ((aligned (1))); \
__kernel void FOpenCL__bottomPassPerform(__global unsigned char* leafCellsPtr, size_t leafCellsSize, \
__global unsigned char* containersPtr, size_t containersSize, \
__global void* userkernel ){ \
} \
__kernel void FOpenCL__upwardPassPerform(__global unsigned char* currentCellsPtr, size_t currentCellsSize, \
struct Uptr9 subCellGroupsPtr, struct size_t9 subCellGroupsSize, \
int nbSubCellGroups, int idxLevel, __global void* userkernel){ \
} \
__kernel void FOpenCL__transferInoutPassPerformMpi(__global unsigned char* currentCellsPtr, size_t currentCellsSize, \
__global unsigned char* externalCellsPtr, size_t externalCellsSize, \
int idxLevel, const __global struct OutOfBlockInteraction* outsideInteractions, \
size_t nbOutsideInteractions, __global void* userkernel){ \
} \
__kernel void FOpenCL__transferInPassPerform(__global unsigned char* currentCellsPtr, size_t currentCellsSize, \
int idxLevel, __global void* userkernel){ \
} \
__kernel void FOpenCL__transferInoutPassPerform(__global unsigned char* currentCellsPtr, size_t currentCellsSize, \
__global unsigned char* externalCellsPtr, size_t externalCellsSize, \
int idxLevel, const __global struct OutOfBlockInteraction* outsideInteractions, \
size_t nbOutsideInteractions, __global void* userkernel){ \
} \
__kernel void FOpenCL__downardPassPerform(__global unsigned char* currentCellsPtr, size_t currentCellsSize, \
struct Uptr9 subCellGroupsPtr, struct size_t9 subCellGroupsSize, \
int nbSubCellGroups, int idxLevel, __global void* userkernel){ \
} \
__kernel void FOpenCL__directInoutPassPerformMpi(__global unsigned char* containersPtr, size_t containersSize, \
__global unsigned char* externalContainersPtr, size_t externalContainersSize, \