From 1b70db7dd1446eb4932627be0f6a3b00d0a23a78 Mon Sep 17 00:00:00 2001
From: Berenger Bramas <Berenger.Bramas@inria.fr>
Date: Wed, 25 Mar 2015 11:33:17 +0100
Subject: [PATCH] Update CUDA blocked part
---
Src/GroupTree/Core/FGroupTaskStarpuAlgorithm.hpp | 2 +-
Src/GroupTree/Cuda/FCudaEmptyKernel.hpp | 3 ++-
Src/GroupTree/TestKernel/FCudaTestKernels.hpp | 2 +-
Tests/noDist/testBlockedWithCudaAlgorithm.cpp | 6 +++---
4 files changed, 7 insertions(+), 6 deletions(-)
diff --git a/Src/GroupTree/Core/FGroupTaskStarpuAlgorithm.hpp b/Src/GroupTree/Core/FGroupTaskStarpuAlgorithm.hpp
index 03da05014..e3e6fae0d 100644
--- a/Src/GroupTree/Core/FGroupTaskStarpuAlgorithm.hpp
+++ b/Src/GroupTree/Core/FGroupTaskStarpuAlgorithm.hpp
@@ -41,7 +41,7 @@
template <class OctreeClass, class CellContainerClass, class KernelClass, class ParticleGroupClass, class StarPUCpuWrapperClass
#ifdef ScalFMM_ENABLE_CUDA_KERNEL
, class StarPUCudaWrapperClass = FStarPUCudaWrapper<KernelClass, FCudaEmptyCellSymb, int, int, FCudaGroupOfCells<FCudaEmptyCellSymb, int, int>,
- FCudaGroupOfParticles<0, 0, int>, FCudaGroupAttachedLeaf<0, 0, int>, FCudaEmptyKernel>
+ FCudaGroupOfParticles<int, 0, 0, int>, FCudaGroupAttachedLeaf<int, 0, 0, int>, FCudaEmptyKernel<int> >
#endif
#ifdef ScalFMM_ENABLE_OPENCL_KERNEL
, class StarPUOpenClWrapperClass = FStarPUOpenClWrapper<KernelClass, FOpenCLDeviceWrapper<KernelClass>>
diff --git a/Src/GroupTree/Cuda/FCudaEmptyKernel.hpp b/Src/GroupTree/Cuda/FCudaEmptyKernel.hpp
index 66bf7d42f..c5f6a8aee 100644
--- a/Src/GroupTree/Cuda/FCudaEmptyKernel.hpp
+++ b/Src/GroupTree/Cuda/FCudaEmptyKernel.hpp
@@ -10,10 +10,11 @@
/**
* This class defines what should be a Cuda kernel.
*/
+template <class FReal>
class FCudaEmptyKernel {
protected:
public:
- typedef FCudaGroupAttachedLeaf<0,0,int> ContainerClass;
+ typedef FCudaGroupAttachedLeaf<FReal,0,0,int> ContainerClass;
typedef FCudaCompositeCell<FCudaEmptyCellSymb,int,int> CellClass;
__device__ void P2M(CellClass /*pole*/, const ContainerClass* const /*particles*/) {
diff --git a/Src/GroupTree/TestKernel/FCudaTestKernels.hpp b/Src/GroupTree/TestKernel/FCudaTestKernels.hpp
index c18e3925d..1c205ef09 100644
--- a/Src/GroupTree/TestKernel/FCudaTestKernels.hpp
+++ b/Src/GroupTree/TestKernel/FCudaTestKernels.hpp
@@ -11,7 +11,7 @@
class FTestCudaKernels {
public:
typedef FCudaCompositeCell<FTestCellPODCore, FTestCellPODData, FTestCellPODData> CellClass;
- typedef FCudaGroupAttachedLeaf<0, 1, long long int> ContainerClass;
+ typedef FCudaGroupAttachedLeaf<FReal, 0, 1, long long int> ContainerClass;
/** Before upward */
__device__ void P2M(CellClass pole, const ContainerClass* const particles) {
diff --git a/Tests/noDist/testBlockedWithCudaAlgorithm.cpp b/Tests/noDist/testBlockedWithCudaAlgorithm.cpp
index e6446b0b8..c0a93d6cf 100644
--- a/Tests/noDist/testBlockedWithCudaAlgorithm.cpp
+++ b/Tests/noDist/testBlockedWithCudaAlgorithm.cpp
@@ -53,10 +53,10 @@
class FTestCudaKernels;
-template <unsigned NbSymbAttributes, unsigned NbAttributesPerParticle, class AttributeClass>
+template <class FReal, unsigned NbSymbAttributes, unsigned NbAttributesPerParticle, class AttributeClass>
class FCudaGroupAttachedLeaf;
-template <unsigned NbSymbAttributes, unsigned NbAttributesPerParticle, class AttributeClass>
+template <class FReal, unsigned NbSymbAttributes, unsigned NbAttributesPerParticle, class AttributeClass>
class FCudaGroupOfParticles;
template <class SymboleCellClass, class PoleCellClass, class LocalCellClass>
@@ -86,7 +86,7 @@ int main(int argc, char* argv[]){
typedef FStarPUCpuWrapper<typename GroupOctreeClass::CellGroupClass, GroupCellClass, GroupKernelClass, typename GroupOctreeClass::ParticleGroupClass, GroupContainerClass> GroupCpuWrapper;
typedef FStarPUCudaWrapper<GroupKernelClass, GroupCellSymbClass, GroupCellUpClass, GroupCellDownClass,
FCudaGroupOfCells<GroupCellSymbClass, GroupCellUpClass, GroupCellDownClass>,
- FCudaGroupOfParticles<0, 1, long long int>, FCudaGroupAttachedLeaf<0, 1, long long int>, FTestCudaKernels> GroupCudaWrapper;
+ FCudaGroupOfParticles<FReal, 0, 1, long long int>, FCudaGroupAttachedLeaf<FReal, 0, 1, long long int>, FTestCudaKernels > GroupCudaWrapper;
typedef FGroupTaskStarPUAlgorithm<GroupOctreeClass, typename GroupOctreeClass::CellGroupClass, GroupKernelClass, typename GroupOctreeClass::ParticleGroupClass,
GroupCpuWrapper, GroupCudaWrapper> GroupAlgorithm;
--
GitLab