diff --git a/core/example/cuda_uvm_example.cu b/core/example/cuda_uvm_example.cu
index 5ddb363b8a3195857ffbafd07dceaf64668433c5..67c356130724faae311b7bdaaff9daf28829600c 100644
--- a/core/example/cuda_uvm_example.cu
+++ b/core/example/cuda_uvm_example.cu
@@ -1,5 +1,6 @@
-#include "Cabana_AoSoA.hpp"
-#include "Cabana_Cuda.hpp"
+#include <Cabana_AoSoA.hpp>
+#include <Cabana_MemberDataTypes.hpp>
+#include <Cabana_Cuda.hpp>
#include <cstdlib>
#include <iostream>
@@ -60,7 +61,7 @@ __global__ void initializeParticles( ParticleList particles )
{
// Create a particle index. The Cuda block id is the struct id and the
// thread id in the block is the array id in the struct.
- ParticleList::Index idx( blockIdx.x, threadIdx.x );
+ Cabana::Index idx( array_size, blockIdx.x, threadIdx.x );
// Only do the operation if we have a valid particle.
if ( idx < particles.end() )
diff --git a/core/example/serial_example.cpp b/core/example/serial_example.cpp
index 3fd01e56dbeca6b48ea90ea6e093f0e1d500fdd5..ac978ac9b26009d58f6475032d6bb623659fc68c 100644
--- a/core/example/serial_example.cpp
+++ b/core/example/serial_example.cpp
@@ -1,5 +1,6 @@
-#include "Cabana_AoSoA.hpp"
-#include "Cabana_Serial.hpp"
+#include <Cabana_AoSoA.hpp>
+#include <Cabana_MemberDataTypes.hpp>
+#include <Cabana_Serial.hpp>
#include <cstdlib>
#include <iostream>
diff --git a/core/example/strided_example.cpp b/core/example/strided_example.cpp
deleted file mode 100644
index edc9444e73218e83c769a0d2fd8d91776b678fe3..0000000000000000000000000000000000000000
--- a/core/example/strided_example.cpp
+++ /dev/null
@@ -1,145 +0,0 @@
-#include "Cabana_AoSoA.hpp"
-#include "Cabana_Serial.hpp"
-
-#include <cstdlib>
-#include <iostream>
-
-//---------------------------------------------------------------------------//
-// Define particle data.
-//---------------------------------------------------------------------------//
-// Inner array size (the size of the arrays in the structs-of-arrays).
-//
-// NOTE: It would be nice to have a dynamic option for this if possible. This
-// would allow for a true SoA-only configuration where only enough memory for
-// all the particles is allocated. Although AoSoA with very large inner array
-// sizes may not perform any worse than SoA anyways so maybe this is not
-// needed.
-const std::size_t array_size = 10;
-
-// Spatial dimension.
-const std::size_t space_dim = 3;
-
-// User field enumeration. These will be used to index into the data set. Must
-// start at 0 and increment contiguously.
-//
-// NOTE: Users don't have to make this enum (or some other set of integral
-// constants) but it is a nice way to provide meaning to the different data
-// types and values assigned to the particles.
-//
-// NOTE: These enums are also ordered in the same way as the data in the
-// template parameters below.
-enum UserParticleFields
-{
- PositionX = 0,
- PositionY,
- PositionZ,
- Velocity,
- Stress,
- Status
-};
-
-// Designate the types that the particles will hold.
-using ParticleDataTypes =
- Cabana::MemberDataTypes<float, // (0) x-position type
- float, // (1) y-position type
- float, // (2) z-position type
- double[space_dim], // (3) velocity type
- double[space_dim][space_dim], // (4) stress type
- int // (5) status type
- >;
-
-// Set the type for the particle AoSoA.
-using ParticleList =
- Cabana::AoSoA<ParticleDataTypes,Cabana::Serial,array_size>;
-
-//---------------------------------------------------------------------------//
-// Helper functions.
-//---------------------------------------------------------------------------//
-// Function to intitialize the particles.
-void initializeParticles( ParticleList particles )
-{
- std::size_t num_s = particles.numSoA();
- for ( std::size_t s = 0; s < num_s; ++s )
- {
- auto pos_x = particles.array<PositionX>( s );
- auto pos_y = particles.array<PositionY>( s );
- auto pos_z = particles.array<PositionZ>( s );
- auto velocity = particles.array<Velocity>( s );
- auto stress = particles.array<Stress>( s );
- auto status = particles.array<Status>( s );
-
- std::size_t num_p = particles.arraySize( s );
- for ( int p = 0; p < num_p; ++p )
- {
- pos_x[p] = 1.1;
- pos_y[p] = 2.2;
- pos_z[p] = 3.3;
-
- for ( int d = 0; d < space_dim; ++d )
- velocity[p][d] = 1.1 * d;
-
- for ( int j = 0; j < space_dim; ++j )
- for ( int i = 0; i < space_dim; ++i )
- stress[p][i][j] = ( i == j ) ? 1.0 : 0.0;
-
- status[p] = 1;
- }
- }
-}
-
-// Function to print out the data for every particle.
-void printParticles( const ParticleList particles )
-{
- for ( auto idx = particles.begin();
- idx != particles.end();
- ++idx )
- {
- std::cout << std::endl;
-
- std::cout << "Struct id: " << idx.s() << std::endl;
- std::cout << "Struct offset: " << idx.i() << std::endl;
- std::cout << "Position: "
- << particles.get<PositionX>( idx ) << " "
- << particles.get<PositionY>( idx ) << " "
- << particles.get<PositionZ>( idx ) << std::endl;
-
- std::cout << "Velocity ";
- for ( int d = 0; d < space_dim; ++d )
- std::cout << particles.get<Velocity>( idx, d ) << " ";
- std::cout << std::endl;
-
- std::cout << "Stress ";
- for ( int j = 0; j < space_dim; ++j )
- {
- std::cout << "{ ";
- for ( int i = 0; i < space_dim; ++i )
- std::cout << particles.get<Stress>( idx, i, j ) << " " ;
- std::cout << "}";
- }
- std::cout << std::endl;
-
- std::cout << "Status " << particles.get<Status>(idx) << std::endl;
- }
-}
-
-//---------------------------------------------------------------------------//
-// Main.
-//---------------------------------------------------------------------------//
-int main()
-{
- // Declare a number of particles.
- int num_particle = 45;
-
- // Create the particle list.
- ParticleList particles( num_particle );
-
- // Initialize particles.
- initializeParticles( particles );
-
- // Print particles.
- printParticles( particles );
-
- return 0;
-}
-
-//---------------------------------------------------------------------------//
diff --git a/core/src/Cabana_AoSoA.hpp b/core/src/Cabana_AoSoA.hpp
index bf98fc4d92465bbfd3ec47a308c90aa96523c7eb..62266f5d85f2118bce3660153112e331e5f5a9ae 100644
--- a/core/src/Cabana_AoSoA.hpp
+++ b/core/src/Cabana_AoSoA.hpp
@@ -46,12 +46,6 @@ class AoSoA<MemberDataTypes<Types...>,Device,ArraySize>
// AoSoA type.
using aosoa_type = AoSoA<MemberDataTypes<Types...>,Device,ArraySize>;
- // Member data types.
- using member_types = MemberDataTypes<Types...>;
-
- // Number of member types.
- static constexpr std::size_t number_of_members = member_types::size;
-
// Device type.
using device_type = Device;
@@ -64,6 +58,15 @@ class AoSoA<MemberDataTypes<Types...>,Device,ArraySize>
// SoA type.
using soa_type = SoA<array_size,Types...>;
+ // Member data types.
+ using member_types = MemberDataTypes<Types...>;
+
+ // Number of member types.
+ static constexpr std::size_t number_of_members = member_types::size;
+
+ // The maximum rank supported for member types.
+ static constexpr std::size_t max_supported_rank = 4;
+
// Struct member array return type at a given index I.
template<std::size_t I>
using struct_member_array_type =
@@ -112,8 +115,10 @@ class AoSoA<MemberDataTypes<Types...>,Device,ArraySize>
, _capacity( 0 )
, _num_soa( 0 )
, _managed_data( nullptr )
- , _raw_data( nullptr )
- {}
+ {
+ storeRanksAndExtents(
+ std::integral_constant<std::size_t,number_of_members-1>() );
+ }
// Construct a container with n elements.
AoSoA( const std::size_t n )
@@ -121,9 +126,10 @@ class AoSoA<MemberDataTypes<Types...>,Device,ArraySize>
, _capacity( 0 )
, _num_soa( 0 )
, _managed_data( nullptr )
- , _raw_data( nullptr )
{
resize( _size );
+ storeRanksAndExtents(
+ std::integral_constant<std::size_t,number_of_members-1>() );
}
// Returns the number of elements in the container.
@@ -192,16 +198,18 @@ class AoSoA<MemberDataTypes<Types...>,Device,ArraySize>
if ( 0 < n % array_size ) ++num_soa_alloc;
_capacity = num_soa_alloc * array_size;
- soa_type* rp;
- memory_policy::allocate( rp, num_soa_alloc );
+ soa_type* data_block;
+ memory_policy::allocate( data_block, num_soa_alloc );
std::shared_ptr<soa_type> sp(
- rp, memory_policy::template deallocate<soa_type> );
+ data_block, memory_policy::template deallocate<soa_type> );
- if ( _raw_data != nullptr )
- memory_policy::copy( rp, _raw_data, _num_soa );
+ if ( _managed_data != nullptr )
+ memory_policy::copy( data_block, _managed_data.get(), _num_soa );
std::swap( _managed_data, sp );
- std::swap( _raw_data, rp );
+
+ storePointersAndStrides(
+ std::integral_constant<std::size_t,number_of_members-1>() );
}
// Get the number of structs-of-arrays in the array.
@@ -220,19 +228,17 @@ class AoSoA<MemberDataTypes<Types...>,Device,ArraySize>
// Member data type properties.
// Get the rank of the data for a given member at index I.
- template<std::size_t I>
CABANA_INLINE_FUNCTION
- constexpr std::size_t rank() const
+ std::size_t rank( const std::size_t I ) const
{
- return std::rank<struct_member_data_type<I> >::value;
+ return _ranks[I];
}
// Get the extent of a given member data dimension.
- template<std::size_t I, std::size_t DIM>
CABANA_INLINE_FUNCTION
- constexpr std::size_t extent() const
+ std::size_t extent( const std::size_t I, const std::size_t D ) const
{
- return std::extent<struct_member_data_type<I>,DIM>::value;
+ return _extents[I][D];
}
// -----------------------------
@@ -374,51 +380,148 @@ class AoSoA<MemberDataTypes<Types...>,Device,ArraySize>
// -------------------------------
// Raw data access.
- // Get the stride between SoA data for a given member at index I.
- template<std::size_t I>
+ // Get the stride between SoA data for a given member at index I. Note
+ // that this strides are computed in the context of the *value_type* for
+ // each member.
CABANA_INLINE_FUNCTION
- constexpr std::size_t stride() const
+ std::size_t stride( const std::size_t I ) const
{
- static_assert( 0 ==
- sizeof(soa_type) % sizeof(struct_member_value_type<I>),
- "Stride cannont be calculated for misaligned memory!" );
- return sizeof(soa_type) / sizeof(struct_member_value_type<I>);
+ return _strides[I];
}
- // Get a pointer to the data for a given member at index I.
- template<std::size_t I>
+ // Get an un-typed raw pointer to the data for a given member at index
+ // I. Users will need to cast this pointer to the appropriate type for the
+ // stride associated with this member to mean anything.
CABANA_INLINE_FUNCTION
- struct_member_pointer_type<I> pointer()
+ void* data( const std::size_t I )
{
- return static_cast<struct_member_pointer_type<I> >( array<I>(0) );
+ return _pointers[I];
}
- template<std::size_t I>
CABANA_INLINE_FUNCTION
- struct_member_const_pointer_type<I> pointer() const
+ const void* data( const std::size_t I ) const
{
- return static_cast<struct_member_pointer_type<I> >( array<I>(0) );
+ return _pointers[I];
}
private:
- // -------------------------------
- // Direct array data access within a struct
+ // Get a typed pointer to the data for a given member at index I.
+ template<std::size_t I>
+ CABANA_INLINE_FUNCTION
+ struct_member_pointer_type<I> typedPointer()
+ {
+ return static_cast<struct_member_pointer_type<I> >( _pointers[I] );
+ }
- // Access the data array at a given struct member index.
template<std::size_t I>
CABANA_INLINE_FUNCTION
- struct_member_array_type<I> array( const std::size_t s )
+ struct_member_const_pointer_type<I> typedPointer() const
{
+ return static_cast<struct_member_pointer_type<I> >( _pointers[I] );
+ }
- return getStructMember<I>( _raw_data[s] );
+ // Get the array at the given struct index.
+ template<std::size_t I>
+ CABANA_INLINE_FUNCTION
+ struct_member_array_type<I> array( const std::size_t s )
+ {
+ return reinterpret_cast<struct_member_array_type<I> >(
+ typedPointer<I>() + s * _strides[I] );
}
template<std::size_t I>
CABANA_INLINE_FUNCTION
struct_member_const_array_type<I> array( const std::size_t s ) const
{
- return getStructMember<I>( _raw_data[s] );
+ return reinterpret_cast<struct_member_array_type<I> >(
+ typedPointer<I>() + s * _strides[I] );
+ }
+
+ // Store the pointers and strides for each member element.
+ template<std::size_t N>
+ void assignPointersAndStrides()
+ {
+ static_assert( 0 <= N && N < number_of_members,
+ "Static loop out of bounds!" );
+ soa_type* data_block = _managed_data.get();
+ _pointers[N] =
+ static_cast<void*>( getStructMember<N>(data_block[0]) );
+ static_assert( 0 ==
+ sizeof(soa_type) % sizeof(struct_member_value_type<N>),
+ "Stride cannont be calculated for misaligned memory!" );
+ _strides[N] = sizeof(soa_type) / sizeof(struct_member_value_type<N>);
+ }
+
+ // Static loop through each member element to extract pointers and strides.
+ template<std::size_t N>
+ void storePointersAndStrides( std::integral_constant<std::size_t,N> )
+ {
+ assignPointersAndStrides<N>();
+ storePointersAndStrides( std::integral_constant<std::size_t,N-1>() );
+ }
+
+ void storePointersAndStrides( std::integral_constant<std::size_t,0> )
+ {
+ assignPointersAndStrides<0>();
+ }
+
+ // Store the extents of each of the member types.
+ template<std::size_t I, std::size_t N>
+ void assignExtents()
+ {
+ static_assert( 0 <= N && N < max_supported_rank,
+ "Static loop out of bounds!" );
+ _extents[I][N] = ( N < std::rank<struct_member_data_type<I> >::value )
+ ? std::extent<struct_member_data_type<I>,N>::value
+ : 0;
+ }
+
+ // Static loop over extents for each member element.
+ template<std::size_t I, std::size_t N>
+ void storeExtents( std::integral_constant<std::size_t,I>,
+ std::integral_constant<std::size_t,N> )
+ {
+ assignExtents<I,N>();
+ storeExtents( std::integral_constant<std::size_t,I>(),
+ std::integral_constant<std::size_t,N-1>() );
+ }
+
+ template<std::size_t I>
+ void storeExtents( std::integral_constant<std::size_t,I>,
+ std::integral_constant<std::size_t,0> )
+ {
+ assignExtents<I,0>();
+ }
+
+ // Store the rank for each member element type.
+ template<std::size_t N>
+ void assignRanks()
+ {
+ static_assert( std::rank<struct_member_data_type<N> >::value <=
+ max_supported_rank,
+ "Member type rank larger than max supported rank" );
+ static_assert( 0 <= N && N < number_of_members, "Static loop out of bounds!" );
+ _ranks[N] = std::rank<struct_member_data_type<N> >::value;
+ }
+
+ // Static loop over ranks and extents for each element.
+ template<std::size_t N>
+ void storeRanksAndExtents( std::integral_constant<std::size_t,N> )
+ {
+ assignRanks<N>();
+ storeExtents(
+ std::integral_constant<std::size_t,N>(),
+ std::integral_constant<std::size_t,max_supported_rank-1>() );
+ storeRanksAndExtents( std::integral_constant<std::size_t,N-1>() );
+ }
+
+ void storeRanksAndExtents( std::integral_constant<std::size_t,0> )
+ {
+ storeExtents(
+ std::integral_constant<std::size_t,0>(),
+ std::integral_constant<std::size_t,max_supported_rank-1>() );
+ assignRanks<0>();
}
private:
@@ -432,18 +535,24 @@ class AoSoA<MemberDataTypes<Types...>,Device,ArraySize>
// Number of structs-of-arrays in the array.
std::size_t _num_soa;
- // Structs-of-Arrays managed data. This shared pointer manages the memory
- // pointed to by _raw_data such that the copy constructor and assignment
- // operator for this class perform a shallow and reference counted copy of
- // the data.
+ // Structs-of-Arrays managed data. This shared pointer manages the block
+ // of memory owned by this class such that the copy constructor and
+ // assignment operator for this class perform a shallow and reference
+ // counted copy of the data.
std::shared_ptr<soa_type> _managed_data;
- // Structs-of-Arrays raw data. This data will be allocated per the
- // MemoryPolicy of the given device type on which the class is
- // templated. This pointer is managed by _managed_data and will be
- // deallocated per the MemoryPolicy when the last copy of this class
- // instance is destroyed.
- soa_type* _raw_data;
+ // Pointers to the first element of each member.
+ void* _pointers[number_of_members];
+
+ // Strides for each member. Note that these strides are computed in the
+ // context of the *value_type* of each member.
+ std::size_t _strides[number_of_members];
+
+ // The ranks of each of the data member types.
+ std::size_t _ranks[number_of_members];
+
+ // The extents of each of the data member type dimensions.
+ std::size_t _extents[number_of_members][max_supported_rank];
};
//---------------------------------------------------------------------------//
diff --git a/core/src/Cabana_MemberSlice.hpp b/core/src/Cabana_MemberSlice.hpp
index 566e765bd94f9b45ee2f11b55b3c50fc487ad800..22667cf0a50414832b6d5eed21c991da9af017ac 100644
--- a/core/src/Cabana_MemberSlice.hpp
+++ b/core/src/Cabana_MemberSlice.hpp
@@ -117,14 +117,13 @@ class MemberSlice
// Get the rank of the data for this member.
CABANA_INLINE_FUNCTION
- constexpr std::size_t rank() const
- { return _aosoa.template rank<I>(); }
+ std::size_t rank() const
+ { return _aosoa.rank(I); }
// Get the extent of a given member data dimension.
- template<std::size_t DIM>
CABANA_INLINE_FUNCTION
- constexpr std::size_t extent() const
- { return _aosoa.template extent<I,DIM>(); }
+ std::size_t extent( const std::size_t D ) const
+ { return _aosoa.extent(I,D); }
// -----------------------------
// Array range
@@ -293,23 +292,6 @@ class MemberSlice
return _aosoa.template get<J>(idx,d0,d1,d2,d3);
}
- // -------------------------------
- // Raw data access.
-
- // Get the stride between SoA data for a given member at index I.
- CABANA_INLINE_FUNCTION
- constexpr std::size_t stride() const
- { return _aosoa.template stride<I>(); }
-
- // Get a pointer to the data for a given member at index I.
- CABANA_INLINE_FUNCTION
- pointer_type pointer()
- { return _aosoa.template pointer<I>(); }
-
- CABANA_INLINE_FUNCTION
- const_pointer_type pointer() const
- { return _aosoa.template pointer<I>(); }
-
private:
// The array-of-structs-of-arrays this slice wraps. We hold this by value
diff --git a/core/unit_test/tstAoSoA.cpp b/core/unit_test/tstAoSoA.cpp
index 46b2c2be5f116c6970416a359100fa93072d1b2e..866809637308d01a9b274b7bda8cdd75364de993 100644
--- a/core/unit_test/tstAoSoA.cpp
+++ b/core/unit_test/tstAoSoA.cpp
@@ -73,48 +73,77 @@ BOOST_AUTO_TEST_CASE( aosoa_serial_api_test )
BOOST_TEST( Cabana::is_aosoa<AoSoA_t>::value );
// Create an AoSoA.
- std::size_t num_data = 35;
- AoSoA_t aosoa( num_data );
+ AoSoA_t aosoa;
// Check sizes.
- BOOST_TEST( aosoa.size() == std::size_t(35) );
- BOOST_TEST( aosoa.capacity() == std::size_t(40) );
- BOOST_TEST( aosoa.numSoA() == std::size_t(4) );
-
- BOOST_TEST( aosoa.arraySize(0) == std::size_t(10) );
- BOOST_TEST( aosoa.arraySize(1) == std::size_t(10) );
- BOOST_TEST( aosoa.arraySize(2) == std::size_t(10) );
- BOOST_TEST( aosoa.arraySize(3) == std::size_t(5) );
+ BOOST_TEST( aosoa.size() == std::size_t(0) );
+ BOOST_TEST( aosoa.capacity() == std::size_t(0) );
+ BOOST_TEST( aosoa.numSoA() == std::size_t(0) );
- BOOST_TEST( aosoa.rank<0>() == std::size_t(3) );
- std::size_t e00 = aosoa.extent<0,0>();
+ // Check member type properties.
+ BOOST_TEST( aosoa.rank(0) == std::size_t(3) );
+ std::size_t e00 = aosoa.extent(0,0);
BOOST_TEST( e00 == dim_1 );
- std::size_t e01 = aosoa.extent<0,1>();
+ std::size_t e01 = aosoa.extent(0,1);
BOOST_TEST( e01 == dim_2 );
- std::size_t e02 = aosoa.extent<0,2>();
+ std::size_t e02 = aosoa.extent(0,2);
BOOST_TEST( e02 == dim_3 );
-
- BOOST_TEST( aosoa.rank<1>() == std::size_t(0) );
-
- BOOST_TEST( aosoa.rank<2>() == std::size_t(4) );
- std::size_t e20 = aosoa.extent<2,0>();
+ std::size_t e03 = aosoa.extent(0,3);
+ BOOST_TEST( e03 == std::size_t(0) );
+
+ BOOST_TEST( aosoa.rank(1) == std::size_t(0) );
+ std::size_t e10 = aosoa.extent(1,0);
+ BOOST_TEST( e10 == std::size_t(0) );
+ std::size_t e11 = aosoa.extent(1,1);
+ BOOST_TEST( e11 == std::size_t(0) );
+ std::size_t e12 = aosoa.extent(1,2);
+ BOOST_TEST( e12 == std::size_t(0) );
+ std::size_t e13 = aosoa.extent(1,3);
+ BOOST_TEST( e13 == std::size_t(0) );
+
+ BOOST_TEST( aosoa.rank(2) == std::size_t(4) );
+ std::size_t e20 = aosoa.extent(2,0);
BOOST_TEST( e20 == dim_1 );
- std::size_t e21 = aosoa.extent<2,1>();
+ std::size_t e21 = aosoa.extent(2,1);
BOOST_TEST( e21 == dim_2 );
- std::size_t e22 = aosoa.extent<2,2>();
+ std::size_t e22 = aosoa.extent(2,2);
BOOST_TEST( e22 == dim_3 );
- std::size_t e23 = aosoa.extent<2,3>();
+ std::size_t e23 = aosoa.extent(2,3);
BOOST_TEST( e23 == dim_4 );
- BOOST_TEST( aosoa.rank<3>() == std::size_t(1) );
- std::size_t e30 = aosoa.extent<3,0>();
+ BOOST_TEST( aosoa.rank(3) == std::size_t(1) );
+ std::size_t e30 = aosoa.extent(3,0);
BOOST_TEST( e30 == dim_1 );
-
- BOOST_TEST( aosoa.rank<4>() == std::size_t(2) );
- std::size_t e40 = aosoa.extent<4,0>();
+ std::size_t e31 = aosoa.extent(3,1);
+ BOOST_TEST( e31 == std::size_t(0) );
+ std::size_t e32 = aosoa.extent(3,2);
+ BOOST_TEST( e32 == std::size_t(0) );
+ std::size_t e33 = aosoa.extent(3,3);
+ BOOST_TEST( e33 == std::size_t(0) );
+
+ BOOST_TEST( aosoa.rank(4) == std::size_t(2) );
+ std::size_t e40 = aosoa.extent(4,0);
BOOST_TEST( e40 == dim_1 );
- std::size_t e41 = aosoa.extent<4,1>();
+ std::size_t e41 = aosoa.extent(4,1);
BOOST_TEST( e41 == dim_2 );
+ std::size_t e42 = aosoa.extent(4,2);
+ BOOST_TEST( e42 == std::size_t(0) );
+ std::size_t e43 = aosoa.extent(4,3);
+ BOOST_TEST( e43 == std::size_t(0) );
+
+ // Resize
+ std::size_t num_data = 35;
+ aosoa.resize( num_data );
+
+ // Check sizes for the new allocation/size.
+ BOOST_TEST( aosoa.size() == std::size_t(35) );
+ BOOST_TEST( aosoa.capacity() == std::size_t(40) );
+ BOOST_TEST( aosoa.numSoA() == std::size_t(4) );
+
+ BOOST_TEST( aosoa.arraySize(0) == std::size_t(10) );
+ BOOST_TEST( aosoa.arraySize(1) == std::size_t(10) );
+ BOOST_TEST( aosoa.arraySize(2) == std::size_t(10) );
+ BOOST_TEST( aosoa.arraySize(3) == std::size_t(5) );
// Initialize data with the rank accessors.
float fval = 3.4;
@@ -193,16 +222,20 @@ BOOST_AUTO_TEST_CASE( aosoa_serial_api_test )
}
//---------------------------------------------------------------------------//
-BOOST_AUTO_TEST_CASE( aosoa_serial_pointer_stride_test )
+BOOST_AUTO_TEST_CASE( aosoa_raw_data_test )
{
// Inner array size.
const std::size_t array_size = 103;
+ // Multi dimensional member sizes.
+ const std::size_t dim_1 = 3;
+ const std::size_t dim_2 = 5;
+
// Declare data types. Note that this test only uses rank-0 data.
using DataTypes =
Cabana::MemberDataTypes<float,
int,
- double,
+ double[dim_1][dim_2],
int,
double
>;
@@ -211,40 +244,28 @@ BOOST_AUTO_TEST_CASE( aosoa_serial_pointer_stride_test )
using AoSoA_t = Cabana::AoSoA<DataTypes,Cabana::Serial,array_size>;
// Create an AoSoA using the default constructor.
- AoSoA_t aosoa;
-
- // Check sizes.
- BOOST_TEST( aosoa.size() == std::size_t(0) );
- BOOST_TEST( aosoa.capacity() == std::size_t(0) );
- BOOST_TEST( aosoa.numSoA() == std::size_t(0) );
-
- // Resize.
std::size_t num_data = 350;
- aosoa.resize( num_data );
-
- // Check sizes.
- BOOST_TEST( aosoa.size() == std::size_t(350) );
- BOOST_TEST( aosoa.capacity() == std::size_t(412) );
- BOOST_TEST( aosoa.numSoA() == std::size_t(4) );
-
- BOOST_TEST( aosoa.arraySize(0) == std::size_t(103) );
- BOOST_TEST( aosoa.arraySize(1) == std::size_t(103) );
- BOOST_TEST( aosoa.arraySize(2) == std::size_t(103) );
- BOOST_TEST( aosoa.arraySize(3) == std::size_t(41) );
+ AoSoA_t aosoa( num_data );
- // Get pointers to the data.
- float* p0 = aosoa.pointer<0>();
- int* p1 = aosoa.pointer<1>();
- double* p2 = aosoa.pointer<2>();
- int* p3 = aosoa.pointer<3>();
- double* p4 = aosoa.pointer<4>();
+ // Get raw pointers to the data as one would in a C interface (no templates).
+ float* p0 = (float*) aosoa.data(0);
+ int* p1 = (int*) aosoa.data(1);
+ double* p2 = (double*) aosoa.data(2);
+ int* p3 = (int*) aosoa.data(3);
+ double* p4 = (double*) aosoa.data(4);
// Get the strides between the member arrays.
- std::size_t st0 = aosoa.stride<0>();
- std::size_t st1 = aosoa.stride<1>();
- std::size_t st2 = aosoa.stride<2>();
- std::size_t st3 = aosoa.stride<3>();
- std::size_t st4 = aosoa.stride<4>();
+ std::size_t st0 = aosoa.stride(0);
+ std::size_t st1 = aosoa.stride(1);
+ std::size_t st2 = aosoa.stride(2);
+ std::size_t st3 = aosoa.stride(3);
+ std::size_t st4 = aosoa.stride(4);
+
+ // Member 2 is multidimensional so get its extents.
+ std::size_t m2e0 = aosoa.extent(2,0);
+ std::size_t m2e1 = aosoa.extent(2,1);
+ BOOST_TEST( m2e0 == dim_1 );
+ BOOST_TEST( m2e1 == dim_2 );
// Initialize the data with raw pointer/stride access. Start by looping
// over the structs. Each struct has a group of contiguous arrays of size
@@ -258,9 +279,15 @@ BOOST_AUTO_TEST_CASE( aosoa_serial_pointer_stride_test )
{
p0[ s * st0 + i ] = (s + i) * 1.0;
p1[ s * st1 + i ] = (s + i) * 2;
- p2[ s * st2 + i ] = (s + i) * 3.0;
p3[ s * st3 + i ] = (s + i) * 4;
p4[ s * st4 + i ] = (s + i) * 5.0;
+
+ // Member 2 has some extra dimensions so add those to the
+ // indexing.
+ for ( std::size_t j = 0; j < m2e0; ++j )
+ for ( std::size_t k = 0; k < m2e1; ++k )
+ p2[ s * st2 + i * m2e0 * m2e1 + j * m2e1 + k ] =
+ (s + i + j + k) * 3.0;
}
}
@@ -271,8 +298,12 @@ BOOST_AUTO_TEST_CASE( aosoa_serial_pointer_stride_test )
std::size_t i = idx.i();
BOOST_TEST( aosoa.get<0>(idx) == (s+i)*1.0 );
BOOST_TEST( aosoa.get<1>(idx) == int((s+i)*2) );
- BOOST_TEST( aosoa.get<2>(idx) == (s+i)*3.0 );
BOOST_TEST( aosoa.get<3>(idx) == int((s+i)*4) );
BOOST_TEST( aosoa.get<4>(idx) == (s+i)*5.0 );
+
+ // Member 2 has some extra dimensions so check those too.
+ for ( std::size_t j = 0; j < dim_1; ++j )
+ for ( std::size_t k = 0; k < dim_2; ++k )
+ BOOST_TEST( aosoa.get<2>(idx,j,k) == (s+i+j+k)*3.0 );
}
}
diff --git a/core/unit_test/tstMemberSlice.cpp b/core/unit_test/tstMemberSlice.cpp
index 89dfe04416486a773a69c3ec4e5a3ef3e0d4fa64..c87a4f8aa4f85aa42c0c64b572505b052dd4dc22 100644
--- a/core/unit_test/tstMemberSlice.cpp
+++ b/core/unit_test/tstMemberSlice.cpp
@@ -100,6 +100,56 @@ BOOST_AUTO_TEST_CASE( slice_serial_api_test )
BOOST_TEST( slice_0.arraySize(2) == std::size_t(10) );
BOOST_TEST( slice_0.arraySize(3) == std::size_t(5) );
+ BOOST_TEST( slice_0.rank() == std::size_t(3) );
+ std::size_t e00 = slice_0.extent(0);
+ BOOST_TEST( e00 == dim_1 );
+ std::size_t e01 = slice_0.extent(1);
+ BOOST_TEST( e01 == dim_2 );
+ std::size_t e02 = slice_0.extent(2);
+ BOOST_TEST( e02 == dim_3 );
+ std::size_t e03 = slice_0.extent(3);
+ BOOST_TEST( e03 == std::size_t(0) );
+
+ BOOST_TEST( slice_1.rank() == std::size_t(0) );
+ std::size_t e10 = slice_1.extent(0);
+ BOOST_TEST( e10 == std::size_t(0) );
+ std::size_t e11 = slice_1.extent(1);
+ BOOST_TEST( e11 == std::size_t(0) );
+ std::size_t e12 = slice_1.extent(2);
+ BOOST_TEST( e12 == std::size_t(0) );
+ std::size_t e13 = slice_1.extent(3);
+ BOOST_TEST( e13 == std::size_t(0) );
+
+ BOOST_TEST( slice_2.rank() == std::size_t(4) );
+ std::size_t e20 = slice_2.extent(0);
+ BOOST_TEST( e20 == dim_1 );
+ std::size_t e21 = slice_2.extent(1);
+ BOOST_TEST( e21 == dim_2 );
+ std::size_t e22 = slice_2.extent(2);
+ BOOST_TEST( e22 == dim_3 );
+ std::size_t e23 = slice_2.extent(3);
+ BOOST_TEST( e23 == dim_4 );
+
+ BOOST_TEST( slice_3.rank() == std::size_t(1) );
+ std::size_t e30 = slice_3.extent(0);
+ BOOST_TEST( e30 == dim_1 );
+ std::size_t e31 = slice_3.extent(1);
+ BOOST_TEST( e31 == std::size_t(0) );
+ std::size_t e32 = slice_3.extent(2);
+ BOOST_TEST( e32 == std::size_t(0) );
+ std::size_t e33 = slice_3.extent(3);
+ BOOST_TEST( e33 == std::size_t(0) );
+
+ BOOST_TEST( slice_4.rank() == std::size_t(2) );
+ std::size_t e40 = slice_4.extent(0);
+ BOOST_TEST( e40 == dim_1 );
+ std::size_t e41 = slice_4.extent(1);
+ BOOST_TEST( e41 == dim_2 );
+ std::size_t e42 = slice_4.extent(2);
+ BOOST_TEST( e42 == std::size_t(0) );
+ std::size_t e43 = slice_4.extent(3);
+ BOOST_TEST( e43 == std::size_t(0) );
+
// Initialize data with the rank accessors.
float fval = 3.4;
double dval = 1.23;
@@ -135,77 +185,3 @@ BOOST_AUTO_TEST_CASE( slice_serial_api_test )
// Check data members for proper initialization.
checkDataMembers( aosoa, fval, dval, ival, dim_1, dim_2, dim_3, dim_4 );
}
-
-//---------------------------------------------------------------------------//
-BOOST_AUTO_TEST_CASE( slice_serial_pointer_stride_test )
-{
- // Inner array size.
- const std::size_t array_size = 103;
-
- // Declare data types. Note that this test only uses rank-0 data.
- using DataTypes =
- Cabana::MemberDataTypes<float,
- int,
- double,
- int,
- double
- >;
-
- // Declare the AoSoA type.
- using AoSoA_t = Cabana::AoSoA<DataTypes,Cabana::Serial,array_size>;
-
- // Create an AoSoA.
- std::size_t num_data = 350;
- AoSoA_t aosoa( num_data );
-
- // Create some slices.
- auto slice_0 = Cabana::slice<0>( aosoa );
- auto slice_1 = Cabana::slice<1>( aosoa );
- auto slice_2 = Cabana::slice<2>( aosoa );
- auto slice_3 = Cabana::slice<3>( aosoa );
- auto slice_4 = Cabana::slice<4>( aosoa );
-
- // Get pointers to the data.
- float* p0 = slice_0.pointer();
- int* p1 = slice_1.pointer();
- double* p2 = slice_2.pointer();
- int* p3 = slice_3.pointer();
- double* p4 = slice_4.pointer();
-
- // Get the strides between the member arrays.
- std::size_t st0 = slice_0.stride();
- std::size_t st1 = slice_1.stride();
- std::size_t st2 = slice_2.stride();
- std::size_t st3 = slice_3.stride();
- std::size_t st4 = slice_4.stride();
-
- // Initialize the data with raw pointer/stride access. Start by looping
- // over the structs. Each struct has a group of contiguous arrays of size
- // array_size for each member.
- std::size_t num_soa = slice_0.numSoA();
- for ( std::size_t s = 0; s < num_soa; ++s )
- {
- // Loop over the array in each struct and set the values.
- std::size_t local_array_size = slice_0.arraySize( s );
- for ( std::size_t i = 0; i < local_array_size; ++i )
- {
- p0[ s * st0 + i ] = (s + i) * 1.0;
- p1[ s * st1 + i ] = (s + i) * 2;
- p2[ s * st2 + i ] = (s + i) * 3.0;
- p3[ s * st3 + i ] = (s + i) * 4;
- p4[ s * st4 + i ] = (s + i) * 5.0;
- }
- }
-
- // Check the results.
- for ( auto idx = aosoa.begin(); idx < aosoa.end(); ++idx )
- {
- std::size_t s = idx.s();
- std::size_t i = idx.i();
- BOOST_TEST( aosoa.get<0>(idx) == (s+i)*1.0 );
- BOOST_TEST( aosoa.get<1>(idx) == int((s+i)*2) );
- BOOST_TEST( aosoa.get<2>(idx) == (s+i)*3.0 );
- BOOST_TEST( aosoa.get<3>(idx) == int((s+i)*4) );
- BOOST_TEST( aosoa.get<4>(idx) == (s+i)*5.0 );
- }
-}