Shuenhoy/vtk-m
1
0
Fork 0
mirror of https://gitlab.kitware.com/vtk/vtk-m synced 2024-09-16 17:22:55 +00:00
Code Issues 2 Actions Packages Projects Releases Wiki Activity

Merge branch 'master' of https://gitlab.kitware.com/vtk/vtk-m into pics_unittests

Browse Source
This commit is contained in:
Dave Pugmire 2017-08-10 16:18:18 -04:00
commit b23564a526
182 changed files with 6577 additions and 1522 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

34
CMake/VTKmBuildType.cmake Normal file
View File

@ -0,0 +1,34 @@
##============================================================================
## Copyright (c) Kitware, Inc.
## All rights reserved.
## See LICENSE.txt for details.
## This software is distributed WITHOUT ANY WARRANTY; without even
## the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
## PURPOSE. See the above copyright notice for more information.
##
## Copyright 2014 Sandia Corporation.
## Copyright 2014 UT-Battelle, LLC.
## Copyright 2014 Los Alamos National Security.
##
## Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
## the U.S. Government retains certain rights in this software.
##
## Under the terms of Contract DE-AC52-06NA25396 with Los Alamos National
## Laboratory (LANL), the U.S. Government retains certain rights in
## this software.
##============================================================================
# Set a default build type if none was specified
set(default_build_type "Release")
if(EXISTS "${CMAKE_SOURCE_DIR}/.git")
set(default_build_type "Debug")
endif()
if(NOT CMAKE_BUILD_TYPE AND NOT CMAKE_CONFIGURATION_TYPES)
message(STATUS "Setting build type to '${default_build_type}' as none was specified.")
set(CMAKE_BUILD_TYPE "${default_build_type}" CACHE
STRING "Choose the type of build." FORCE)
# Set the possible values of build type for cmake-gui
set_property(CACHE CMAKE_BUILD_TYPE PROPERTY STRINGS "Debug" "Release"
"MinSizeRel" "RelWithDebInfo")
endif()

9
CMake/VTKmConfigureComponents.cmake
View File

@ -336,7 +336,9 @@ macro(vtkm_configure_component_CUDA)
# 5 - pascal
# - Uses: --generate-code=arch=compute_60,code=compute_60
# - Uses: --generate-code=arch=compute_61,code=compute_61
# 6 - all
# 6 - volta
# - Uses: --generate-code=arch=compute_70,code=compute_70
# 7 - all
# - Uses: --generate-code=arch=compute_20,code=compute_20
# - Uses: --generate-code=arch=compute_30,code=compute_30
# - Uses: --generate-code=arch=compute_35,code=compute_35
@ -344,11 +346,12 @@ macro(vtkm_configure_component_CUDA)
# - Uses: --generate-code=arch=compute_52,code=compute_52
# - Uses: --generate-code=arch=compute_60,code=compute_60
# - Uses: --generate-code=arch=compute_61,code=compute_61
# - Uses: --generate-code=arch=compute_70,code=compute_70
#
#specify the property
set(VTKm_CUDA_Architecture "native" CACHE STRING "Which GPU Architecture(s) to compile for")
set_property(CACHE VTKm_CUDA_Architecture PROPERTY STRINGS native fermi kepler maxwell pascal all)
set_property(CACHE VTKm_CUDA_Architecture PROPERTY STRINGS native fermi kepler maxwell pascal volta all)
#detect what the propery is set too
if(VTKm_CUDA_Architecture STREQUAL "native")
@ -403,6 +406,8 @@ macro(vtkm_configure_component_CUDA)
elseif(VTKm_CUDA_Architecture STREQUAL "pascal")
list(APPEND CUDA_NVCC_FLAGS "--generate-code=arch=compute_60,code=compute_60")
list(APPEND CUDA_NVCC_FLAGS "--generate-code=arch=compute_61,code=compute_61")
elseif(VTKm_CUDA_Architecture STREQUAL "volta")
list(APPEND CUDA_NVCC_FLAGS "--generate-code=arch=compute_70,code=compute_70")
elseif(VTKm_CUDA_Architecture STREQUAL "all")
list(APPEND CUDA_NVCC_FLAGS "--generate-code=arch=compute_20,code=compute_20")
list(APPEND CUDA_NVCC_FLAGS "--generate-code=arch=compute_30,code=compute_30")

16
CMakeLists.txt
View File

@ -38,9 +38,13 @@ set(VTKm_EXPORT_NAME "VTKmTargets")
set(VTKm_CMAKE_MODULE_PATH ${VTKm_SOURCE_DIR}/CMake)
set(CMAKE_MODULE_PATH ${CMAKE_MODULE_PATH} ${VTKm_CMAKE_MODULE_PATH})
# Setup default build types
include(VTKmBuildType)
# Determine VTK-m version
include(Utilities/Git/Git.cmake)
include(VTKmDetermineVersion)
# Load hardcoded version in case this is not a Git repository
file(STRINGS version.txt version_txt)
extract_version_components("${version_txt}" "VTKm")
@ -108,18 +112,6 @@ option(VTKm_USE_DOUBLE_PRECISION
)
option(VTKm_USE_64BIT_IDS "Use 64-bit indices." ON)
if (VTKm_ENABLE_CUDA)
option(VTKm_USE_UNIFIED_MEMORY
"Use CUDA unified memory"
OFF
)
endif (VTKm_ENABLE_CUDA)
if (VTKm_USE_UNIFIED_MEMORY)
set(CMAKE_CXX_FLAGS "-DVTKM_USE_UNIFIED_MEMORY ${CMAKE_CXX_FLAGS}")
endif()
option(BUILD_SHARED_LIBS "Build VTK-m with shared libraries" ON)
set(VTKm_BUILD_SHARED_LIBS ${BUILD_SHARED_LIBS})

2
README.md
View File

@ -165,7 +165,7 @@ relevant options
| VTKm_ENABLE_EXAMPLES | Disabled by default. Turn on building of simple examples of using VTK-m. |
| VTKm_ENABLE_BENCHMARKS | Disabled by default. Turn on additional timing tests. |
| VTKm_ENABLE_CUDA | Disabled by default. Enable CUDA backend. |
| VTKm_CUDA_Architecture | Defaults to native. Specify what GPU architecture(s) to build CUDA code for, options include native, fermi, kepler, maxwell, and pascal. |
| VTKm_CUDA_Architecture | Defaults to native. Specify what GPU architecture(s) to build CUDA code for, options include native, fermi, kepler, maxwell, pascal, and volta. |
| VTKm_ENABLE_TBB | Disabled by default. Enable Intel Threading Building Blocks backend. |
| VTKm_ENABLE_TESTING | Enabled by default. Turn on header, unit, worklet, and filter tests. |
| VTKm_ENABLE_RENDERING | Enabled by default. Turn on the rendering module. |

7
examples/hello_world/HelloWorld.cxx
View File

@ -86,8 +86,10 @@ struct HelloVTKMInterop
{
for (int j = 0; j < dim; ++j)
{
this->InputData.push_back(vtkm::Vec<T, 3>(
2.f * static_cast<T>(i / dim) - 1.f, 0.f, 2.f * static_cast<T>(j / dim) - 1.f));
vtkm::Vec<T, 3> t(2.f * (static_cast<T>(i) / static_cast<T>(dim)) - 1.f,
0.f,
2.f * (static_cast<T>(j) / static_cast<T>(dim)) - 1.f);
this->InputData.push_back(t);
}
}
@ -169,7 +171,6 @@ struct HelloVTKMInterop
vtkm::interop::TransferToOpenGL(this->OutCoords, this->VBOState, DeviceAdapter());
vtkm::interop::TransferToOpenGL(this->OutColors, this->ColorState, DeviceAdapter());
this->render();
if (t > 10)
{

53
examples/unified_memory/UnifiedMemory.cu
View File

@ -30,6 +30,7 @@
#include <vtkm/cont/CellSetExplicit.h>
#include <vtkm/cont/DataSet.h>
#include <vtkm/cont/Timer.h>
#include <vtkm/cont/cuda/internal/CudaAllocator.h>
namespace
{
@ -92,8 +93,8 @@ vtkm::cont::DataSet MakeIsosurfaceTestDataSet(vtkm::Id3 dims)
vtkm::Float32 maxs[3] = { 1.0f, 1.0f, 1.0f };
vtkm::cont::ArrayHandle<vtkm::Float32> fieldArray;
vtkm::cont::ArrayHandleCounting<vtkm::Id> vertexCountImplicitArray(0, 1, vdims[0] * vdims[1] *
vdims[2]);
vtkm::cont::ArrayHandleCounting<vtkm::Id> vertexCountImplicitArray(
0, 1, vdims[0] * vdims[1] * vdims[2]);
vtkm::worklet::DispatcherMapField<TangleField> tangleFieldDispatcher(
TangleField(vdims, mins, maxs));
tangleFieldDispatcher.Invoke(vertexCountImplicitArray, fieldArray);
@ -153,37 +154,39 @@ int main(int argc, char* argv[])
typedef vtkm::cont::DeviceAdapterAlgorithm<VTKM_DEFAULT_DEVICE_ADAPTER_TAG> DeviceAlgorithms;
vtkm::worklet::SineWorklet sineWorklet;
#ifdef VTKM_USE_UNIFIED_MEMORY
std::cout << "Testing with unified memory" << std::endl;
bool usingManagedMemory = vtkm::cont::cuda::internal::CudaAllocator::UsingManagedMemory();
vtkm::worklet::DispatcherMapField<vtkm::worklet::SineWorklet> dispatcher(sineWorklet);
if (usingManagedMemory)
{
std::cout << "Testing with unified memory" << std::endl;
vtkm::cont::Timer<> timer;
vtkm::worklet::DispatcherMapField<vtkm::worklet::SineWorklet> dispatcher(sineWorklet);
dispatcher.Invoke(input, output);
std::cout << output.GetPortalConstControl().Get(output.GetNumberOfValues() - 1) << std::endl;
vtkm::cont::Timer<> timer;
vtkm::Float64 elapsedTime = timer.GetElapsedTime();
std::cout << "Time: " << elapsedTime << std::endl;
dispatcher.Invoke(input, output);
std::cout << output.GetPortalConstControl().Get(output.GetNumberOfValues() - 1) << std::endl;
#else
vtkm::Float64 elapsedTime = timer.GetElapsedTime();
std::cout << "Time: " << elapsedTime << std::endl;
}
else
{
vtkm::worklet::DispatcherStreamingMapField<vtkm::worklet::SineWorklet> dispatcher(sineWorklet);
vtkm::Id NBlocks = N / (1024 * 1024 * 1024);
NBlocks *= 2;
dispatcher.SetNumberOfBlocks(NBlocks);
std::cout << "Testing with streaming (without unified memory) with " << NBlocks << " blocks"
<< std::endl;
vtkm::worklet::DispatcherStreamingMapField<vtkm::worklet::SineWorklet> dispatcher(sineWorklet);
vtkm::Id NBlocks = N / (1024 * 1024 * 1024);
NBlocks *= 2;
dispatcher.SetNumberOfBlocks(NBlocks);
std::cout << "Testing with streaming (without unified memory) with " << NBlocks << " blocks"
<< std::endl;
vtkm::cont::Timer<> timer;
vtkm::cont::Timer<> timer;
dispatcher.Invoke(input, output);
std::cout << output.GetPortalConstControl().Get(output.GetNumberOfValues() - 1) << std::endl;
dispatcher.Invoke(input, output);
std::cout << output.GetPortalConstControl().Get(output.GetNumberOfValues() - 1) << std::endl;
vtkm::Float64 elapsedTime = timer.GetElapsedTime();
std::cout << "Time: " << elapsedTime << std::endl;
#endif
vtkm::Float64 elapsedTime = timer.GetElapsedTime();
std::cout << "Time: " << elapsedTime << std::endl;
}
int dim = 128;
if (argc > 2)

8
vtkm/CellShape.h
View File

@ -87,7 +87,7 @@ struct CellShapeIdToTag
// probably means you are using an ID that does not have a defined cell
// shape.
typedef std::false_type valid;
using valid = std::false_type;
};
// Define a tag for each cell shape as well as the support structs to go
@ -112,8 +112,8 @@ struct CellShapeIdToTag
template <> \
struct CellShapeIdToTag<vtkm::idname> \
{ \
typedef std::true_type valid; \
typedef vtkm::CellShapeTag##name Tag; \
using valid = std::true_type; \
using Tag = vtkm::CellShapeTag##name; \
}
VTKM_DEFINE_CELL_TAG(Empty, CELL_SHAPE_EMPTY);
@ -153,7 +153,7 @@ struct CellShapeTagGeneric
#define vtkmGenericCellShapeMacroCase(cellShapeId, call) \
case vtkm::cellShapeId: \
{ \
typedef vtkm::CellShapeIdToTag<vtkm::cellShapeId>::Tag CellShapeTag; \
using CellShapeTag = vtkm::CellShapeIdToTag<vtkm::cellShapeId>::Tag; \
call; \
} \
break

16
vtkm/Math.h
View File

@ -100,13 +100,13 @@ struct FloatingPointReturnCondition
template <typename T, typename = void>
struct FloatingPointReturnType
{
typedef vtkm::Float64 Type;
using Type = vtkm::Float64;
};
template <typename T>
struct FloatingPointReturnType<T, typename FloatingPointReturnCondition<T>::type>
{
typedef vtkm::Float32 Type;
using Type = vtkm::Float32;
};
} // namespace detail
@ -1731,7 +1731,7 @@ static inline VTKM_EXEC_CONT T Max(T x, T y, vtkm::TypeTraitsScalarTag)
template <typename T>
static inline VTKM_EXEC_CONT T Max(const T& x, const T& y, vtkm::TypeTraitsVectorTag)
{
typedef vtkm::VecTraits<T> Traits;
using Traits = vtkm::VecTraits<T>;
T result;
for (vtkm::IdComponent index = 0; index < Traits::NUM_COMPONENTS; index++)
{
@ -1750,7 +1750,7 @@ static inline VTKM_EXEC_CONT T Min(T x, T y, vtkm::TypeTraitsScalarTag)
template <typename T>
static inline VTKM_EXEC_CONT T Min(const T& x, const T& y, vtkm::TypeTraitsVectorTag)
{
typedef vtkm::VecTraits<T> Traits;
using Traits = vtkm::VecTraits<T>;
T result;
for (vtkm::IdComponent index = 0; index < Traits::NUM_COMPONENTS; index++)
{
@ -1813,7 +1813,7 @@ struct FloatLimits;
template <>
struct FloatLimits<vtkm::Float32>
{
typedef vtkm::detail::IEEE754Bits32 BitsType;
using BitsType = vtkm::detail::IEEE754Bits32;
VTKM_EXEC_CONT
static vtkm::Float32 Nan()
@ -1843,7 +1843,7 @@ struct FloatLimits<vtkm::Float32>
template <int N>
struct FloatLimits<vtkm::Vec<vtkm::Float32, N>>
{
typedef vtkm::detail::IEEE754Bits32 BitsType;
using BitsType = vtkm::detail::IEEE754Bits32;
VTKM_EXEC_CONT
static vtkm::Vec<vtkm::Float32, N> Nan()
@ -1876,7 +1876,7 @@ struct FloatLimits<vtkm::Vec<vtkm::Float32, N>>
template <>
struct FloatLimits<vtkm::Float64>
{
typedef vtkm::detail::IEEE754Bits64 BitsType;
using BitsType = vtkm::detail::IEEE754Bits64;
VTKM_EXEC_CONT
static vtkm::Float64 Nan()
@ -1906,7 +1906,7 @@ struct FloatLimits<vtkm::Float64>
template <int N>
struct FloatLimits<vtkm::Vec<vtkm::Float64, N>>
{
typedef vtkm::detail::IEEE754Bits64 BitsType;
using BitsType = vtkm::detail::IEEE754Bits64;
VTKM_EXEC_CONT
static vtkm::Vec<vtkm::Float64, N> Nan()

16
vtkm/Math.h.in
View File

@ -247,13 +247,13 @@ struct FloatingPointReturnCondition
template <typename T, typename = void>
struct FloatingPointReturnType
{
typedef vtkm::Float64 Type;
using Type = vtkm::Float64;
};
template <typename T>
struct FloatingPointReturnType<T, typename FloatingPointReturnCondition<T>::type>
{
typedef vtkm::Float32 Type;
using Type = vtkm::Float32;
};
} // namespace detail
@ -513,7 +513,7 @@ static inline VTKM_EXEC_CONT T Max(T x, T y, vtkm::TypeTraitsScalarTag)
template <typename T>
static inline VTKM_EXEC_CONT T Max(const T& x, const T& y, vtkm::TypeTraitsVectorTag)
{
typedef vtkm::VecTraits<T> Traits;
using Traits = vtkm::VecTraits<T>;
T result;
for (vtkm::IdComponent index = 0; index < Traits::NUM_COMPONENTS; index++)
{
@ -532,7 +532,7 @@ static inline VTKM_EXEC_CONT T Min(T x, T y, vtkm::TypeTraitsScalarTag)
template <typename T>
static inline VTKM_EXEC_CONT T Min(const T& x, const T& y, vtkm::TypeTraitsVectorTag)
{
typedef vtkm::VecTraits<T> Traits;
using Traits = vtkm::VecTraits<T>;
T result;
for (vtkm::IdComponent index = 0; index < Traits::NUM_COMPONENTS; index++)
{
@ -595,7 +595,7 @@ struct FloatLimits;
template <>
struct FloatLimits<vtkm::Float32>
{
typedef vtkm::detail::IEEE754Bits32 BitsType;
using BitsType = vtkm::detail::IEEE754Bits32;
VTKM_EXEC_CONT
static vtkm::Float32 Nan()
@ -625,7 +625,7 @@ struct FloatLimits<vtkm::Float32>
template <int N>
struct FloatLimits<vtkm::Vec<vtkm::Float32, N>>
{
typedef vtkm::detail::IEEE754Bits32 BitsType;
using BitsType = vtkm::detail::IEEE754Bits32;
VTKM_EXEC_CONT
static vtkm::Vec<vtkm::Float32, N> Nan()
@ -658,7 +658,7 @@ struct FloatLimits<vtkm::Vec<vtkm::Float32, N>>
template <>
struct FloatLimits<vtkm::Float64>
{
typedef vtkm::detail::IEEE754Bits64 BitsType;
using BitsType = vtkm::detail::IEEE754Bits64;
VTKM_EXEC_CONT
static vtkm::Float64 Nan()
@ -688,7 +688,7 @@ struct FloatLimits<vtkm::Float64>
template <int N>
struct FloatLimits<vtkm::Vec<vtkm::Float64, N>>
{
typedef vtkm::detail::IEEE754Bits64 BitsType;
using BitsType = vtkm::detail::IEEE754Bits64;
VTKM_EXEC_CONT
static vtkm::Vec<vtkm::Float64, N> Nan()

14
vtkm/Matrix.h
View File

@ -45,7 +45,7 @@ template <typename T, vtkm::IdComponent NumRow, vtkm::IdComponent NumCol>
class Matrix
{
public:
typedef T ComponentType;
using ComponentType = T;
static const vtkm::IdComponent NUM_ROWS = NumRow;
static const vtkm::IdComponent NUM_COLUMNS = NumCol;
@ -533,8 +533,8 @@ struct TypeTraitsMatrixTag
template <typename T, vtkm::IdComponent NumRow, vtkm::IdComponent NumCol>
struct TypeTraits<vtkm::Matrix<T, NumRow, NumCol>>
{
typedef typename TypeTraits<T>::NumericTag NumericTag;
typedef TypeTraitsMatrixTag DimensionalityTag;
using NumericTag = typename TypeTraits<T>::NumericTag;
using DimensionalityTag = vtkm::TypeTraitsMatrixTag;
};
/// A matrix has vector traits to implement component-wise operations.
@ -543,13 +543,13 @@ template <typename T, vtkm::IdComponent NumRow, vtkm::IdComponent NumCol>
struct VecTraits<vtkm::Matrix<T, NumRow, NumCol>>
{
private:
typedef vtkm::Matrix<T, NumRow, NumCol> MatrixType;
using MatrixType = vtkm::Matrix<T, NumRow, NumCol>;
public:
typedef T ComponentType;
using ComponentType = T;
static const vtkm::IdComponent NUM_COMPONENTS = NumRow * NumCol;
typedef vtkm::VecTraitsTagMultipleComponents HasMultipleComponents;
typedef vtkm::VecTraitsTagSizeStatic IsSizeStatic;
using HasMultipleComponents = vtkm::VecTraitsTagMultipleComponents;
using IsSizeStatic = vtkm::VecTraitsTagSizeStatic;
VTKM_EXEC_CONT
static vtkm::IdComponent GetNumberOfComponents(const MatrixType&) { return NUM_COMPONENTS; }

8
vtkm/Pair.h
View File

@ -40,19 +40,19 @@ struct Pair
{
/// The type of the first object.
///
typedef T1 FirstType;
using FirstType = T1;
/// The type of the second object.
///
typedef T2 SecondType;
using SecondType = T2;
/// The same as FirstType, but follows the naming convention of std::pair.
///
typedef FirstType first_type;
using first_type = FirstType;
/// The same as SecondType, but follows the naming convention of std::pair.
///
typedef SecondType second_type;
using second_type = SecondType;
/// The pair's first object. Note that this field breaks VTK-m's naming
/// conventions to make vtkm::Pair more compatible with std::pair.

22
vtkm/TypeTraits.h
View File

@ -74,13 +74,13 @@ public:
/// \brief A tag to determing whether the type is integer or real.
///
/// This tag is either TypeTraitsRealTag or TypeTraitsIntegerTag.
typedef TypeTraitsUnknownTag NumericTag;
using NumericTag = vtkm::TypeTraitsUnknownTag;
/// \brief A tag to determine whether the type has multiple components.
///
/// This tag is either TypeTraitsScalarTag or TypeTraitsVectorTag. Scalars can
/// also be treated as vectors.
typedef TypeTraitsUnknownTag DimensionalityTag;
using DimensionalityTag = vtkm::TypeTraitsUnknownTag;
VTKM_EXEC_CONT static T ZeroInitialization() { return T(); }
};
@ -96,8 +96,8 @@ struct TypeTraits<const T> : TypeTraits<T>
template <> \
struct TypeTraits<T> \
{ \
typedef TypeTraitsRealTag NumericTag; \
typedef TypeTraitsScalarTag DimensionalityTag; \
using NumericTag = TypeTraitsRealTag; \
using DimensionalityTag = TypeTraitsScalarTag; \
VTKM_EXEC_CONT static T ZeroInitialization() { return T(); } \
};
@ -105,11 +105,11 @@ struct TypeTraits<const T> : TypeTraits<T>
template <> \
struct TypeTraits<T> \
{ \
typedef TypeTraitsIntegerTag NumericTag; \
typedef TypeTraitsScalarTag DimensionalityTag; \
using NumericTag = TypeTraitsIntegerTag; \
using DimensionalityTag = TypeTraitsScalarTag; \
VTKM_EXEC_CONT static T ZeroInitialization() \
{ \
typedef T ReturnType; \
using ReturnType = T; \
return ReturnType(); \
} \
};
@ -140,8 +140,8 @@ VTKM_BASIC_INTEGER_TYPE(unsigned long long)
template <typename T, vtkm::IdComponent Size>
struct TypeTraits<vtkm::Vec<T, Size>>
{
typedef typename vtkm::TypeTraits<T>::NumericTag NumericTag;
typedef TypeTraitsVectorTag DimensionalityTag;
using NumericTag = typename vtkm::TypeTraits<T>::NumericTag;
using DimensionalityTag = vtkm::TypeTraitsVectorTag;
VTKM_EXEC_CONT
static vtkm::Vec<T, Size> ZeroInitialization()
@ -179,8 +179,8 @@ struct TypeTraits<vtkm::VecC<T>>
template <typename T, typename U>
struct TypeTraits<vtkm::Pair<T, U>>
{
typedef TypeTraitsUnknownTag NumericTag;
typedef TypeTraitsScalarTag DimensionalityTag;
using NumericTag = vtkm::TypeTraitsUnknownTag;
using DimensionalityTag = vtkm::TypeTraitsScalarTag;
VTKM_EXEC_CONT
static vtkm::Pair<T, U> ZeroInitialization()

60
vtkm/Types.h
View File

@ -145,45 +145,45 @@ namespace vtkm
//*****************************************************************************
#if VTKM_SIZE_FLOAT == 4
typedef float Float32;
using Float32 = float;
#else
#error Could not find a 32-bit float.
#endif
#if VTKM_SIZE_DOUBLE == 8
typedef double Float64;
using Float64 = double;
#else
#error Could not find a 64-bit float.
#endif
#if VTKM_SIZE_CHAR == 1
typedef signed char Int8;
typedef unsigned char UInt8;
using Int8 = signed char;
using UInt8 = unsigned char;
#else
#error Could not find an 8-bit integer.
#endif
#if VTKM_SIZE_SHORT == 2
typedef signed short Int16;
typedef unsigned short UInt16;
using Int16 = short;
using UInt16 = unsigned short;
#else
#error Could not find a 16-bit integer.
#endif
#if VTKM_SIZE_INT == 4
typedef signed int Int32;
typedef unsigned int UInt32;
using Int32 = int;
using UInt32 = unsigned int;
#else
#error Could not find a 32-bit integer.
#endif
//In this order so that we exactly match the logic that exists in VTK
#if VTKM_SIZE_LONG_LONG == 8
typedef signed long long Int64;
typedef unsigned long long UInt64;
using Int64 = long long;
using UInt64 = unsigned long long;
#elif VTKM_SIZE_LONG == 8
typedef signed long Int64;
typedef unsigned long UInt64;
using Int64 = signed long;
using UInt64 = unsigned long;
#else
#error Could not find a 64-bit integer.
#endif
@ -193,12 +193,12 @@ typedef unsigned long UInt64;
#if VTKM_SIZE_ID == 4
/// Represents an ID (index into arrays).
typedef vtkm::Int32 Id;
using Id = vtkm::Int32;
#elif VTKM_SIZE_ID == 8
/// Represents an ID.
typedef vtkm::Int64 Id;
using Id = vtkm::Int64;
#else
#error Unknown Id Size
@ -209,17 +209,17 @@ typedef vtkm::Int64 Id;
/// to be quite small. However, we are currently using a 32-bit width
/// integer because modern processors tend to access them more efficiently
/// than smaller widths.
typedef vtkm::Int32 IdComponent;
using IdComponent = vtkm::Int32;
#ifdef VTKM_USE_DOUBLE_PRECISION
/// The floating point type to use when no other precision is specified.
typedef vtkm::Float64 FloatDefault;
using FloatDefault = vtkm::Float64;
#else //VTKM_USE_DOUBLE_PRECISION
/// The floating point type to use when no other precision is specified.
typedef vtkm::Float32 FloatDefault;
using FloatDefault = vtkm::Float32;
#endif //VTKM_USE_DOUBLE_PRECISION
@ -434,7 +434,7 @@ template <typename T, typename DerivedClass>
class VTKM_ALWAYS_EXPORT VecBaseCommon
{
public:
typedef T ComponentType;
using ComponentType = T;
protected:
VTKM_EXEC_CONT
@ -661,7 +661,7 @@ template <typename T, vtkm::IdComponent Size, typename DerivedClass>
class VTKM_ALWAYS_EXPORT VecBase : public vtkm::detail::VecBaseCommon<T, DerivedClass>
{
public:
typedef T ComponentType;
using ComponentType = T;
static const vtkm::IdComponent NUM_COMPONENTS = Size;
protected:
@ -816,11 +816,11 @@ protected:
template <typename T, vtkm::IdComponent Size>
class VTKM_ALWAYS_EXPORT Vec : public detail::VecBase<T, Size, Vec<T, Size>>
{
typedef detail::VecBase<T, Size, Vec<T, Size>> Superclass;
using Superclass = detail::VecBase<T, Size, Vec<T, Size>>;
public:
#ifdef VTKM_DOXYGEN_ONLY
typedef T ComponentType;
using ComponentType = T;
static const vtkm::IdComponent NUM_COMPONENTS = Size;
#endif
@ -848,7 +848,7 @@ template <typename T>
class VTKM_ALWAYS_EXPORT Vec<T, 0>
{
public:
typedef T ComponentType;
using ComponentType = T;
static const vtkm::IdComponent NUM_COMPONENTS = 0;
VTKM_EXEC_CONT Vec() {}
@ -879,7 +879,7 @@ public:
template <typename T>
class VTKM_ALWAYS_EXPORT Vec<T, 1> : public detail::VecBase<T, 1, Vec<T, 1>>
{
typedef detail::VecBase<T, 1, Vec<T, 1>> Superclass;
using Superclass = detail::VecBase<T, 1, Vec<T, 1>>;
public:
VTKM_EXEC_CONT Vec() {}
@ -909,7 +909,7 @@ public:
template <typename T>
class VTKM_ALWAYS_EXPORT Vec<T, 2> : public detail::VecBase<T, 2, Vec<T, 2>>
{
typedef detail::VecBase<T, 2, Vec<T, 2>> Superclass;
using Superclass = detail::VecBase<T, 2, Vec<T, 2>>;
public:
VTKM_EXEC_CONT Vec() {}
@ -933,12 +933,12 @@ public:
};
/// Id2 corresponds to a 2-dimensional index
typedef vtkm::Vec<vtkm::Id, 2> Id2;
using Id2 = vtkm::Vec<vtkm::Id, 2>;
template <typename T>
class VTKM_ALWAYS_EXPORT Vec<T, 3> : public detail::VecBase<T, 3, Vec<T, 3>>
{
typedef detail::VecBase<T, 3, Vec<T, 3>> Superclass;
using Superclass = detail::VecBase<T, 3, Vec<T, 3>>;
public:
VTKM_EXEC_CONT Vec() {}
@ -964,12 +964,12 @@ public:
/// Id3 corresponds to a 3-dimensional index for 3d arrays. Note that
/// the precision of each index may be less than vtkm::Id.
typedef vtkm::Vec<vtkm::Id, 3> Id3;
using Id3 = vtkm::Vec<vtkm::Id, 3>;
template <typename T>
class VTKM_ALWAYS_EXPORT Vec<T, 4> : public detail::VecBase<T, 4, Vec<T, 4>>
{
typedef detail::VecBase<T, 4, Vec<T, 4>> Superclass;
using Superclass = detail::VecBase<T, 4, Vec<T, 4>>;
public:
VTKM_EXEC_CONT Vec() {}
@ -1001,12 +1001,12 @@ public:
template <typename ComponentType, vtkm::IdComponent NumComponents>
struct VecOrScalar
{
typedef vtkm::Vec<ComponentType, NumComponents> Type;
using Type = vtkm::Vec<ComponentType, NumComponents>;
};
template <typename ComponentType>
struct VecOrScalar<ComponentType, 1>
{
typedef ComponentType Type;
using Type = ComponentType;
};
/// Initializes and returns a Vec of length 2.

28
vtkm/VecTraits.h
View File

@ -58,13 +58,13 @@ namespace internal
template <vtkm::IdComponent numComponents>
struct VecTraitsMultipleComponentChooser
{
typedef VecTraitsTagMultipleComponents Type;
using Type = vtkm::VecTraitsTagMultipleComponents;
};
template <>
struct VecTraitsMultipleComponentChooser<1>
{
typedef VecTraitsTagSingleComponent Type;
using Type = vtkm::VecTraitsTagSingleComponent;
};
} // namespace detail
@ -78,7 +78,7 @@ struct VecTraits
{
/// Type of the components in the vector.
///
typedef typename VecType::ComponentType ComponentType;
using ComponentType = typename VecType::ComponentType;
/// \brief Number of components in the vector.
///
@ -95,8 +95,8 @@ struct VecTraits
/// This tag can be useful for creating specialized functions when a vector
/// is really just a scalar.
///
typedef typename internal::VecTraitsMultipleComponentChooser<NUM_COMPONENTS>::Type
HasMultipleComponents;
using HasMultipleComponents =
typename internal::VecTraitsMultipleComponentChooser<NUM_COMPONENTS>::Type;
/// \brief A tag specifying whether the size of this vector is known at compile time.
///
@ -104,7 +104,7 @@ struct VecTraits
/// set to \c VecTraitsTagSizeVariable, then the number of components is not
/// known at compile time and must be queried with \c GetNumberOfComponents.
///
typedef vtkm::VecTraitsTagSizeStatic IsSizeStatic;
using IsSizeStatic = vtkm::VecTraitsTagSizeStatic;
/// Returns the value in a given component of the vector.
///
@ -141,11 +141,11 @@ struct VecTraits<const T> : VecTraits<T>
template <typename T, vtkm::IdComponent Size>
struct VecTraits<vtkm::Vec<T, Size>>
{
typedef vtkm::Vec<T, Size> VecType;
using VecType = vtkm::Vec<T, Size>;
/// Type of the components in the vector.
///
typedef typename VecType::ComponentType ComponentType;
using ComponentType = typename VecType::ComponentType;
/// Number of components in the vector.
///
@ -160,8 +160,8 @@ struct VecTraits<vtkm::Vec<T, Size>>
/// "real" vector). This tag can be useful for creating specialized functions
/// when a vector is really just a scalar.
///
typedef typename internal::VecTraitsMultipleComponentChooser<NUM_COMPONENTS>::Type
HasMultipleComponents;
using HasMultipleComponents =
typename internal::VecTraitsMultipleComponentChooser<NUM_COMPONENTS>::Type;
/// A tag specifying whether the size of this vector is known at compile
/// time. If set to \c VecTraitsTagSizeStatic, then \c NUM_COMPONENTS is set.
@ -169,7 +169,7 @@ struct VecTraits<vtkm::Vec<T, Size>>
/// not known at compile time and must be queried with \c
/// GetNumberOfComponents.
///
typedef vtkm::VecTraitsTagSizeStatic IsSizeStatic;
using IsSizeStatic = vtkm::VecTraitsTagSizeStatic;
/// Returns the value in a given component of the vector.
///
@ -334,10 +334,10 @@ namespace internal
template <typename ScalarType>
struct VecTraitsBasic
{
typedef ScalarType ComponentType;
using ComponentType = ScalarType;
static const vtkm::IdComponent NUM_COMPONENTS = 1;
typedef VecTraitsTagSingleComponent HasMultipleComponents;
typedef vtkm::VecTraitsTagSizeStatic IsSizeStatic;
using HasMultipleComponents = vtkm::VecTraitsTagSingleComponent;
using IsSizeStatic = vtkm::VecTraitsTagSizeStatic;
VTKM_EXEC_CONT
static vtkm::IdComponent GetNumberOfComponents(const ScalarType&) { return 1; }

11
vtkm/benchmarking/BenchmarkTopologyAlgorithms.cxx
View File

@ -88,12 +88,15 @@ public:
const vtkm::IdComponent& numCellIDs) const
{
//simple functor that returns the average cell Value.
avgVal = static_cast<OutType>(cellValues[0]);
for (vtkm::IdComponent cellIndex = 1; cellIndex < numCellIDs; ++cellIndex)
avgVal = vtkm::TypeTraits<OutType>::ZeroInitialization();
if (numCellIDs != 0)
{
avgVal += static_cast<OutType>(cellValues[cellIndex]);
for (vtkm::IdComponent cellIndex = 0; cellIndex < numCellIDs; ++cellIndex)
{
avgVal += static_cast<OutType>(cellValues[cellIndex]);
}
avgVal = avgVal / static_cast<OutType>(numCellIDs);
}
avgVal = avgVal / static_cast<OutType>(numCellIDs);
}
};

68
vtkm/cont/ArrayHandle.h
View File

@ -128,15 +128,15 @@ struct GetTypeInParentheses;
template <typename T>
struct GetTypeInParentheses<void(T)>
{
typedef T type;
using type = T;
};
} // namespace detail
// Implementation for VTKM_ARRAY_HANDLE_SUBCLASS macros
#define VTK_M_ARRAY_HANDLE_SUBCLASS_IMPL(classname, fullclasstype, superclass, typename__) \
typedef typename__ vtkm::cont::detail::GetTypeInParentheses<void fullclasstype>::type Thisclass; \
typedef typename__ vtkm::cont::detail::GetTypeInParentheses<void superclass>::type Superclass; \
using Thisclass = typename__ vtkm::cont::detail::GetTypeInParentheses<void fullclasstype>::type; \
using Superclass = typename__ vtkm::cont::detail::GetTypeInParentheses<void superclass>::type; \
\
VTKM_IS_ARRAY_HANDLE(Superclass); \
\
@ -166,8 +166,8 @@ struct GetTypeInParentheses<void(T)>
return *this; \
} \
\
typedef typename__ Superclass::ValueType ValueType; \
typedef typename__ Superclass::StorageTag StorageTag
using ValueType = typename__ Superclass::ValueType; \
using StorageTag = typename__ Superclass::StorageTag
/// \brief Macro to make default methods in ArrayHandle subclasses.
///
@ -250,17 +250,17 @@ private:
vtkm::cont::internal::ArrayHandleExecutionManagerBase<T, StorageTag_>;
public:
typedef vtkm::cont::internal::Storage<T, StorageTag_> StorageType;
typedef T ValueType;
typedef StorageTag_ StorageTag;
typedef typename StorageType::PortalType PortalControl;
typedef typename StorageType::PortalConstType PortalConstControl;
using StorageType = vtkm::cont::internal::Storage<T, StorageTag_>;
using ValueType = T;
using StorageTag = StorageTag_;
using PortalControl = typename StorageType::PortalType;
using PortalConstControl = typename StorageType::PortalConstType;
template <typename DeviceAdapterTag>
struct ExecutionTypes
{
typedef typename ExecutionManagerType::template ExecutionTypes<DeviceAdapterTag>::Portal Portal;
typedef typename ExecutionManagerType::template ExecutionTypes<DeviceAdapterTag>::PortalConst
PortalConst;
using Portal = typename ExecutionManagerType::template ExecutionTypes<DeviceAdapterTag>::Portal;
using PortalConst =
typename ExecutionManagerType::template ExecutionTypes<DeviceAdapterTag>::PortalConst;
};
/// Constructs an empty ArrayHandle. Typically used for output or
@ -488,8 +488,8 @@ template <typename T>
VTKM_CONT vtkm::cont::ArrayHandle<T, vtkm::cont::StorageTagBasic> make_ArrayHandle(const T* array,
vtkm::Id length)
{
typedef vtkm::cont::ArrayHandle<T, vtkm::cont::StorageTagBasic> ArrayHandleType;
typedef vtkm::cont::internal::Storage<T, vtkm::cont::StorageTagBasic> StorageType;
using ArrayHandleType = vtkm::cont::ArrayHandle<T, vtkm::cont::StorageTagBasic>;
using StorageType = vtkm::cont::internal::Storage<T, vtkm::cont::StorageTagBasic>;
return ArrayHandleType(StorageType(array, length));
}
@ -520,6 +520,24 @@ printSummary_ArrayHandle_Value(const T& value, std::ostream& out, vtkm::VecTrait
out << value;
}
VTKM_NEVER_EXPORT
VTKM_CONT
inline void printSummary_ArrayHandle_Value(vtkm::UInt8 value,
std::ostream& out,
vtkm::VecTraitsTagSingleComponent)
{
out << static_cast<int>(value);
}
VTKM_NEVER_EXPORT
VTKM_CONT
inline void printSummary_ArrayHandle_Value(vtkm::Int8 value,
std::ostream& out,
vtkm::VecTraitsTagSingleComponent)
{
out << static_cast<int>(value);
}
template <typename T>
VTKM_NEVER_EXPORT VTKM_CONT inline void printSummary_ArrayHandle_Value(
const T& value,
@ -540,24 +558,6 @@ VTKM_NEVER_EXPORT VTKM_CONT inline void printSummary_ArrayHandle_Value(
out << ")";
}
VTKM_NEVER_EXPORT
VTKM_CONT
inline void printSummary_ArrayHandle_Value(UInt8 value,
std::ostream& out,
vtkm::VecTraitsTagSingleComponent)
{
out << static_cast<int>(value);
}
VTKM_NEVER_EXPORT
VTKM_CONT
inline void printSummary_ArrayHandle_Value(Int8 value,
std::ostream& out,
vtkm::VecTraitsTagSingleComponent)
{
out << static_cast<int>(value);
}
template <typename T1, typename T2>
VTKM_NEVER_EXPORT VTKM_CONT inline void printSummary_ArrayHandle_Value(
const vtkm::Pair<T1, T2>& value,
@ -573,6 +573,8 @@ VTKM_NEVER_EXPORT VTKM_CONT inline void printSummary_ArrayHandle_Value(
out << "}";
}
} // namespace detail
template <typename T, typename StorageT>

4
vtkm/cont/ArrayHandle.hxx
View File

@ -175,8 +175,8 @@ void ArrayHandle<T, S>::CopyInto(IteratorType dest, DeviceAdapterTag) const
/// class and call CopyInto. The dynamic conversion will be sucessful
/// becuase the check to ensure the ExecutionArray is of the type
/// DeviceAdapterTag has already passed
typedef vtkm::cont::internal::ArrayHandleExecutionManager<T, StorageTag, DeviceAdapterTag>
ConcreteType;
using ConcreteType =
vtkm::cont::internal::ArrayHandleExecutionManager<T, StorageTag, DeviceAdapterTag>;
ConcreteType* ConcreteExecutionArray =
dynamic_cast<ConcreteType*>(this->Internals->ExecutionArray.get());

18
vtkm/cont/ArrayHandleCartesianProduct.h
View File

@ -41,11 +41,11 @@ template <typename ValueType_,
class VTKM_ALWAYS_EXPORT ArrayPortalCartesianProduct
{
public:
typedef ValueType_ ValueType;
typedef ValueType_ IteratorType;
typedef PortalTypeFirst_ PortalTypeFirst;
typedef PortalTypeSecond_ PortalTypeSecond;
typedef PortalTypeThird_ PortalTypeThird;
using ValueType = ValueType_;
using IteratorType = ValueType_;
using PortalTypeFirst = PortalTypeFirst_;
using PortalTypeSecond = PortalTypeSecond_;
using PortalTypeThird = PortalTypeThird_;
VTKM_SUPPRESS_EXEC_WARNINGS
VTKM_EXEC_CONT
@ -189,7 +189,7 @@ class Storage<T, StorageTagCartesianProduct<FirstHandleType, SecondHandleType, T
VTKM_IS_ARRAY_HANDLE(ThirdHandleType);
public:
typedef T ValueType;
using ValueType = T;
typedef vtkm::exec::internal::ArrayPortalCartesianProduct<
ValueType,
@ -292,10 +292,10 @@ class ArrayTransfer<T,
typedef vtkm::cont::internal::Storage<T, StorageTag> StorageType;
public:
typedef T ValueType;
using ValueType = T;
typedef typename StorageType::PortalType PortalControl;
typedef typename StorageType::PortalConstType PortalConstControl;
using PortalControl = typename StorageType::PortalType;
using PortalConstControl = typename StorageType::PortalConstType;
typedef vtkm::exec::internal::ArrayPortalCartesianProduct<
ValueType,

87
vtkm/cont/ArrayHandleCompositeVector.h
View File

@ -46,7 +46,7 @@ template <typename ValueType>
struct VTKM_ALWAYS_EXPORT CompositeVectorSwizzleFunctor
{
static const vtkm::IdComponent NUM_COMPONENTS = vtkm::VecTraits<ValueType>::NUM_COMPONENTS;
typedef vtkm::Vec<vtkm::IdComponent, NUM_COMPONENTS> ComponentMapType;
using ComponentMapType = vtkm::Vec<vtkm::IdComponent, NUM_COMPONENTS>;
// Caution! This is a reference.
const ComponentMapType& SourceComponents;
@ -118,7 +118,7 @@ struct CompositeVectorArrayToPortalCont
template <typename ArrayHandleType, vtkm::IdComponent Index>
struct ReturnType
{
typedef typename ArrayHandleType::PortalConstControl type;
using type = typename ArrayHandleType::PortalConstControl;
};
template <typename ArrayHandleType, vtkm::IdComponent Index>
@ -136,7 +136,7 @@ struct CompositeVectorArrayToPortalExec
template <typename ArrayHandleType, vtkm::IdComponent Index>
struct ReturnType
{
typedef typename ArrayHandleType::template ExecutionTypes<DeviceAdapterTag>::PortalConst type;
using type = typename ArrayHandleType::template ExecutionTypes<DeviceAdapterTag>::PortalConst;
};
template <typename ArrayHandleType, vtkm::IdComponent Index>
@ -179,14 +179,14 @@ struct CheckArraySizeFunctor
template <typename SignatureWithPortals>
class VTKM_ALWAYS_EXPORT ArrayPortalCompositeVector
{
typedef vtkm::internal::FunctionInterface<SignatureWithPortals> PortalTypes;
using PortalTypes = vtkm::internal::FunctionInterface<SignatureWithPortals>;
public:
typedef typename PortalTypes::ResultType ValueType;
using ValueType = typename PortalTypes::ResultType;
static const vtkm::IdComponent NUM_COMPONENTS = vtkm::VecTraits<ValueType>::NUM_COMPONENTS;
// Used internally.
typedef vtkm::Vec<vtkm::IdComponent, NUM_COMPONENTS> ComponentMapType;
using ComponentMapType = vtkm::Vec<vtkm::IdComponent, NUM_COMPONENTS>;
VTKM_STATIC_ASSERT(NUM_COMPONENTS == PortalTypes::ARITY);
@ -248,10 +248,10 @@ struct VTKM_ALWAYS_EXPORT StorageTagCompositeVector
template <typename SignatureWithArrays>
struct ArrayHandleCompositeVectorTraits
{
typedef vtkm::cont::internal::StorageTagCompositeVector<SignatureWithArrays> Tag;
typedef typename vtkm::internal::FunctionInterface<SignatureWithArrays>::ResultType ValueType;
typedef vtkm::cont::internal::Storage<ValueType, Tag> StorageType;
typedef vtkm::cont::ArrayHandle<ValueType, Tag> Superclass;
using Tag = vtkm::cont::internal::StorageTagCompositeVector<SignatureWithArrays>;
using ValueType = typename vtkm::internal::FunctionInterface<SignatureWithArrays>::ResultType;
using StorageType = vtkm::cont::internal::Storage<ValueType, Tag>;
using Superclass = vtkm::cont::ArrayHandle<ValueType, Tag>;
};
// It may seem weird that this specialization throws an exception for
@ -261,18 +261,19 @@ template <typename SignatureWithArrays>
class Storage<typename ArrayHandleCompositeVectorTraits<SignatureWithArrays>::ValueType,
vtkm::cont::internal::StorageTagCompositeVector<SignatureWithArrays>>
{
typedef vtkm::internal::FunctionInterface<SignatureWithArrays> FunctionInterfaceWithArrays;
using FunctionInterfaceWithArrays = vtkm::internal::FunctionInterface<SignatureWithArrays>;
static const vtkm::IdComponent NUM_COMPONENTS = FunctionInterfaceWithArrays::ARITY;
typedef vtkm::Vec<vtkm::IdComponent, NUM_COMPONENTS> ComponentMapType;
using ComponentMapType = vtkm::Vec<vtkm::IdComponent, NUM_COMPONENTS>;
typedef typename FunctionInterfaceWithArrays::template StaticTransformType<
detail::CompositeVectorArrayToPortalCont>::type FunctionInterfaceWithPortals;
typedef typename FunctionInterfaceWithPortals::Signature SignatureWithPortals;
using FunctionInterfaceWithPortals =
typename FunctionInterfaceWithArrays::template StaticTransformType<
detail::CompositeVectorArrayToPortalCont>::type;
using SignatureWithPortals = typename FunctionInterfaceWithPortals::Signature;
public:
typedef ArrayPortalCompositeVector<SignatureWithPortals> PortalType;
typedef PortalType PortalConstType;
typedef typename PortalType::ValueType ValueType;
using PortalType = ArrayPortalCompositeVector<SignatureWithPortals>;
using PortalConstType = PortalType;
using ValueType = typename PortalType::ValueType;
VTKM_CONT
Storage()
@ -373,22 +374,22 @@ class ArrayTransfer<typename ArrayHandleCompositeVectorTraits<SignatureWithArray
{
VTKM_IS_DEVICE_ADAPTER_TAG(DeviceAdapterTag);
typedef typename ArrayHandleCompositeVectorTraits<SignatureWithArrays>::StorageType StorageType;
using StorageType = typename ArrayHandleCompositeVectorTraits<SignatureWithArrays>::StorageType;
typedef vtkm::internal::FunctionInterface<SignatureWithArrays> FunctionWithArrays;
typedef typename FunctionWithArrays::template StaticTransformType<
detail::CompositeVectorArrayToPortalExec<DeviceAdapterTag>>::type FunctionWithPortals;
typedef typename FunctionWithPortals::Signature SignatureWithPortals;
using FunctionWithArrays = vtkm::internal::FunctionInterface<SignatureWithArrays>;
using FunctionWithPortals = typename FunctionWithArrays::template StaticTransformType<
detail::CompositeVectorArrayToPortalExec<DeviceAdapterTag>>::type;
using SignatureWithPortals = typename FunctionWithPortals::Signature;
public:
typedef typename ArrayHandleCompositeVectorTraits<SignatureWithArrays>::ValueType ValueType;
using ValueType = typename ArrayHandleCompositeVectorTraits<SignatureWithArrays>::ValueType;
// These are not currently fully implemented.
typedef typename StorageType::PortalType PortalControl;
typedef typename StorageType::PortalConstType PortalConstControl;
using PortalControl = typename StorageType::PortalType;
using PortalConstControl = typename StorageType::PortalConstType;
typedef ArrayPortalCompositeVector<SignatureWithPortals> PortalExecution;
typedef ArrayPortalCompositeVector<SignatureWithPortals> PortalConstExecution;
using PortalExecution = ArrayPortalCompositeVector<SignatureWithPortals>;
using PortalConstExecution = ArrayPortalCompositeVector<SignatureWithPortals>;
VTKM_CONT
ArrayTransfer(StorageType* storage)
@ -460,9 +461,9 @@ template <typename Signature>
class ArrayHandleCompositeVector
: public internal::ArrayHandleCompositeVectorTraits<Signature>::Superclass
{
typedef typename internal::ArrayHandleCompositeVectorTraits<Signature>::StorageType StorageType;
typedef
typename internal::ArrayPortalCompositeVector<Signature>::ComponentMapType ComponentMapType;
using StorageType = typename internal::ArrayHandleCompositeVectorTraits<Signature>::StorageType;
using ComponentMapType =
typename internal::ArrayPortalCompositeVector<Signature>::ComponentMapType;
public:
VTKM_ARRAY_HANDLE_SUBCLASS(
@ -553,15 +554,15 @@ struct ArrayHandleCompositeVectorType
VTKM_IS_ARRAY_HANDLE(ArrayHandleType4);
private:
typedef
typename vtkm::VecTraits<typename ArrayHandleType1::ValueType>::ComponentType ComponentType;
using ComponentType =
typename vtkm::VecTraits<typename ArrayHandleType1::ValueType>::ComponentType;
typedef vtkm::Vec<ComponentType, 4> Signature(ArrayHandleType1,
ArrayHandleType2,
ArrayHandleType3,
ArrayHandleType4);
public:
typedef vtkm::cont::ArrayHandleCompositeVector<Signature> type;
using type = vtkm::cont::ArrayHandleCompositeVector<Signature>;
};
template <typename ArrayHandleType1, typename ArrayHandleType2, typename ArrayHandleType3>
@ -572,14 +573,14 @@ struct ArrayHandleCompositeVectorType<ArrayHandleType1, ArrayHandleType2, ArrayH
VTKM_IS_ARRAY_HANDLE(ArrayHandleType3);
private:
typedef
typename vtkm::VecTraits<typename ArrayHandleType1::ValueType>::ComponentType ComponentType;
using ComponentType =
typename vtkm::VecTraits<typename ArrayHandleType1::ValueType>::ComponentType;
typedef vtkm::Vec<ComponentType, 3> Signature(ArrayHandleType1,
ArrayHandleType2,
ArrayHandleType3);
public:
typedef vtkm::cont::ArrayHandleCompositeVector<Signature> type;
using type = vtkm::cont::ArrayHandleCompositeVector<Signature>;
};
template <typename ArrayHandleType1, typename ArrayHandleType2>
@ -589,12 +590,12 @@ struct ArrayHandleCompositeVectorType<ArrayHandleType1, ArrayHandleType2>
VTKM_IS_ARRAY_HANDLE(ArrayHandleType2);
private:
typedef
typename vtkm::VecTraits<typename ArrayHandleType1::ValueType>::ComponentType ComponentType;
using ComponentType =
typename vtkm::VecTraits<typename ArrayHandleType1::ValueType>::ComponentType;
typedef vtkm::Vec<ComponentType, 2> Signature(ArrayHandleType1, ArrayHandleType2);
public:
typedef vtkm::cont::ArrayHandleCompositeVector<Signature> type;
using type = vtkm::cont::ArrayHandleCompositeVector<Signature>;
};
template <typename ArrayHandleType1>
@ -603,12 +604,12 @@ struct ArrayHandleCompositeVectorType<ArrayHandleType1>
VTKM_IS_ARRAY_HANDLE(ArrayHandleType1);
private:
typedef
typename vtkm::VecTraits<typename ArrayHandleType1::ValueType>::ComponentType ComponentType;
using ComponentType =
typename vtkm::VecTraits<typename ArrayHandleType1::ValueType>::ComponentType;
typedef ComponentType Signature(ArrayHandleType1);
public:
typedef vtkm::cont::ArrayHandleCompositeVector<Signature> type;
using type = vtkm::cont::ArrayHandleCompositeVector<Signature>;
};
// clang-format off

9
vtkm/cont/ArrayHandleCounting.h
View File

@ -37,10 +37,10 @@ namespace internal
template <class CountingValueType>
class VTKM_ALWAYS_EXPORT ArrayPortalCounting
{
typedef typename vtkm::VecTraits<CountingValueType>::ComponentType ComponentType;
using ComponentType = typename vtkm::VecTraits<CountingValueType>::ComponentType;
public:
typedef CountingValueType ValueType;
using ValueType = CountingValueType;
VTKM_EXEC_CONT
ArrayPortalCounting()
@ -102,9 +102,8 @@ private:
template <typename ConstantValueType>
struct ArrayHandleCountingTraits
{
typedef vtkm::cont::StorageTagImplicit<
vtkm::cont::internal::ArrayPortalCounting<ConstantValueType>>
Tag;
using Tag =
vtkm::cont::StorageTagImplicit<vtkm::cont::internal::ArrayPortalCounting<ConstantValueType>>;
};
} // namespace internal

303
vtkm/cont/ArrayHandleExtractComponent.h Normal file
View File

@ -0,0 +1,303 @@
//=============================================================================
//
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
//
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2017 Sandia Corporation.
// Copyright 2017 UT-Battelle, LLC.
// Copyright 2017 Los Alamos National Security.
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
// Under the terms of Contract DE-AC52-06NA25396 with Los Alamos National
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//
//=============================================================================
#ifndef vtk_m_cont_ArrayHandleExtractComponent_h
#define vtk_m_cont_ArrayHandleExtractComponent_h
#include <vtkm/VecTraits.h>
#include <vtkm/cont/ArrayHandle.h>
namespace vtkm
{
namespace cont
{
namespace internal
{
template <typename PortalType, vtkm::IdComponent Component>
class VTKM_ALWAYS_EXPORT ArrayPortalExtractComponent
{
public:
using VectorType = typename PortalType::ValueType;
using Traits = vtkm::VecTraits<VectorType>;
using ComponentType = typename Traits::ComponentType;
static const vtkm::IdComponent COMPONENT = Component;
VTKM_EXEC_CONT
ArrayPortalExtractComponent()
: Portal()
{
}
VTKM_EXEC_CONT
ArrayPortalExtractComponent(const PortalType& portal)
: Portal(portal)
{
}
// Copy constructor
VTKM_EXEC_CONT ArrayPortalExtractComponent(
const ArrayPortalExtractComponent<PortalType, Component>& src)
: Portal(src.GetPortal())
{
}
VTKM_EXEC_CONT
vtkm::Id GetNumberOfValues() const { return this->Portal.GetNumberOfValues(); }
VTKM_EXEC_CONT
ComponentType Get(vtkm::Id index) const
{
return Traits::GetComponent(this->Portal.Get(index), Component);
}
VTKM_EXEC_CONT
void Set(vtkm::Id index, const ComponentType& value) const
{
VectorType vec = this->Portal.Get(index);
Traits::SetComponent(vec, Component, value);
this->Portal.Set(index, vec);
}
VTKM_EXEC_CONT
const PortalType& GetPortal() const { return this->Portal; }
private:
PortalType Portal;
}; // class ArrayPortalExtractComponent
} // namespace internal
template <typename ArrayHandleType, vtkm::IdComponent Component>
class StorageTagExtractComponent
{
static const vtkm::IdComponent COMPONENT = Component;
};
namespace internal
{
template <typename ArrayHandleType, vtkm::IdComponent Component>
class Storage<typename vtkm::VecTraits<typename ArrayHandleType::ValueType>::ComponentType,
StorageTagExtractComponent<ArrayHandleType, Component>>
{
public:
using PortalType =
ArrayPortalExtractComponent<typename ArrayHandleType::PortalControl, Component>;
using PortalConstType =
ArrayPortalExtractComponent<typename ArrayHandleType::PortalConstControl, Component>;
using ValueType = typename PortalType::ComponentType;
VTKM_CONT
Storage()
: Valid(false)
{
}
VTKM_CONT
Storage(const ArrayHandleType& array)
: Array(array)
, Valid(true)
{
}
VTKM_CONT
PortalConstType GetPortalConst() const
{
VTKM_ASSERT(this->Valid);
return PortalConstType(this->Array.GetPortalConstControl());
}
VTKM_CONT
PortalType GetPortal()
{
VTKM_ASSERT(this->Valid);
return PortalType(this->Array.GetPortalControl());
}
VTKM_CONT
vtkm::Id GetNumberOfValues() const
{
VTKM_ASSERT(this->Valid);
return this->Array.GetNumberOfValues();
}
VTKM_CONT
void Allocate(vtkm::Id numberOfValues)
{
VTKM_ASSERT(this->Valid);
this->Array.Allocate(numberOfValues);
}
VTKM_CONT
void Shrink(vtkm::Id numberOfValues)
{
VTKM_ASSERT(this->Valid);
this->Array.Shrink(numberOfValues);
}
VTKM_CONT
void ReleaseResources()
{
VTKM_ASSERT(this->Valid);
this->Array.ReleaseResources();
}
VTKM_CONT
const ArrayHandleType& GetArray() const
{
VTKM_ASSERT(this->Valid);
return this->Array;
}
private:
ArrayHandleType Array;
bool Valid;
}; // class Storage
template <typename ArrayHandleType, vtkm::IdComponent Component, typename Device>
class ArrayTransfer<typename vtkm::VecTraits<typename ArrayHandleType::ValueType>::ComponentType,
StorageTagExtractComponent<ArrayHandleType, Component>,
Device>
{
public:
using ValueType = typename vtkm::VecTraits<typename ArrayHandleType::ValueType>::ComponentType;
private:
using StorageTag = StorageTagExtractComponent<ArrayHandleType, Component>;
using StorageType = vtkm::cont::internal::Storage<ValueType, StorageTag>;
using ArrayValueType = typename ArrayHandleType::ValueType;
using ArrayStorageTag = typename ArrayHandleType::StorageTag;
using ArrayStorageType =
vtkm::cont::internal::Storage<typename ArrayHandleType::ValueType, ArrayStorageTag>;
public:
using PortalControl = typename StorageType::PortalType;
using PortalConstControl = typename StorageType::PortalConstType;
using ExecutionTypes = typename ArrayHandleType::template ExecutionTypes<Device>;
using PortalExecution = ArrayPortalExtractComponent<typename ExecutionTypes::Portal, Component>;
using PortalConstExecution =
ArrayPortalExtractComponent<typename ExecutionTypes::PortalConst, Component>;
VTKM_CONT
ArrayTransfer(StorageType* storage)
: Array(storage->GetArray())
{
}
VTKM_CONT
vtkm::Id GetNumberOfValues() const { return this->Array.GetNumberOfValues(); }
VTKM_CONT
PortalConstExecution PrepareForInput(bool vtkmNotUsed(updateData))
{
return PortalConstExecution(this->Array.PrepareForInput(Device()));
}
VTKM_CONT
PortalExecution PrepareForInPlace(bool vtkmNotUsed(updateData))
{
return PortalExecution(this->Array.PrepareForInPlace(Device()));
}
VTKM_CONT
PortalExecution PrepareForOutput(vtkm::Id numberOfValues)
{
return PortalExecution(this->Array.PrepareForOutput(numberOfValues, Device()));
}
VTKM_CONT
void RetrieveOutputData(StorageType* vtkmNotUsed(storage)) const
{
// Implementation of this method should be unnecessary. The internal
// array handle should automatically retrieve the output data as
// necessary.
}
VTKM_CONT
void Shrink(vtkm::Id numberOfValues) { this->Array.Shrink(numberOfValues); }
VTKM_CONT
void ReleaseResources() { this->Array.ReleaseResourcesExecution(); }
private:
ArrayHandleType Array;
};
}
}
} // namespace vtkm::cont::internal
namespace vtkm
{
namespace cont
{
/// \brief A fancy ArrayHandle that turns a vector array into a scalar array by
/// slicing out a single component of each vector.
///
/// ArrayHandleExtractComponent is a specialization of ArrayHandle. It takes an
/// input ArrayHandle with a vtkm::Vec ValueType and a component index
/// and uses this information to expose a scalar array consisting of the
/// specified component across all vectors in the input ArrayHandle. So for a
/// given index i, ArrayHandleExtractComponent looks up the i-th vtkm::Vec in
/// the index array and reads or writes to the specified component, leave all
/// other components unmodified. This is done on the fly rather than creating a
/// copy of the array.
template <typename ArrayHandleType, vtkm::IdComponent Component>
class ArrayHandleExtractComponent
: public vtkm::cont::ArrayHandle<
typename vtkm::VecTraits<typename ArrayHandleType::ValueType>::ComponentType,
StorageTagExtractComponent<ArrayHandleType, Component>>
{
public:
static const vtkm::IdComponent COMPONENT = Component;
VTKM_ARRAY_HANDLE_SUBCLASS(
ArrayHandleExtractComponent,
(ArrayHandleExtractComponent<ArrayHandleType, Component>),
(vtkm::cont::ArrayHandle<
typename vtkm::VecTraits<typename ArrayHandleType::ValueType>::ComponentType,
StorageTagExtractComponent<ArrayHandleType, Component>>));
protected:
using StorageType = vtkm::cont::internal::Storage<ValueType, StorageTag>;
public:
VTKM_CONT
ArrayHandleExtractComponent(const ArrayHandleType& array)
: Superclass(StorageType(array))
{
}
};
/// make_ArrayHandleExtractComponent is convenience function to generate an
/// ArrayHandleExtractComponent.
template <vtkm::IdComponent Component, typename ArrayHandleType>
VTKM_CONT ArrayHandleExtractComponent<ArrayHandleType, Component> make_ArrayHandleExtractComponent(
const ArrayHandleType& array)
{
return ArrayHandleExtractComponent<ArrayHandleType, Component>(array);
}
}
} // namespace vtkm::cont
#endif // vtk_m_cont_ArrayHandleExtractComponent_h

6
vtkm/cont/ArrayHandleImplicit.h
View File

@ -91,8 +91,8 @@ public:
#endif
}
typedef vtkm::cont::internal::IteratorFromArrayPortal<ArrayPortalImplicit<FunctorType>>
IteratorType;
using IteratorType =
vtkm::cont::internal::IteratorFromArrayPortal<ArrayPortalImplicit<FunctorType>>;
VTKM_CONT
IteratorType GetIteratorBegin() const { return IteratorType(*this); }
@ -115,7 +115,7 @@ template <class FunctorType>
class ArrayHandleImplicit : public detail::ArrayHandleImplicitTraits<FunctorType>::Superclass
{
private:
typedef typename detail::ArrayHandleImplicitTraits<FunctorType> ArrayTraits;
using ArrayTraits = typename detail::ArrayHandleImplicitTraits<FunctorType>;
public:
VTKM_ARRAY_HANDLE_SUBCLASS(ArrayHandleImplicit,

10
vtkm/cont/ArrayHandlePermutation.h
View File

@ -37,7 +37,7 @@ template <typename IndexPortalType, typename ValuePortalType>
class VTKM_ALWAYS_EXPORT ArrayPortalPermutation
{
public:
typedef typename ValuePortalType::ValueType ValueType;
using ValueType = typename ValuePortalType::ValueType;
VTKM_SUPPRESS_EXEC_WARNINGS
VTKM_EXEC_CONT
@ -125,7 +125,7 @@ class Storage<typename ValueArrayType::ValueType,
VTKM_IS_ARRAY_HANDLE(ValueArrayType);
public:
typedef typename ValueArrayType::ValueType ValueType;
using ValueType = typename ValueArrayType::ValueType;
typedef vtkm::exec::internal::ArrayPortalPermutation<typename IndexArrayType::PortalConstControl,
typename ValueArrayType::PortalControl>
@ -209,15 +209,15 @@ class ArrayTransfer<typename ValueArrayType::ValueType,
Device>
{
public:
typedef typename ValueArrayType::ValueType ValueType;
using ValueType = typename ValueArrayType::ValueType;
private:
typedef StorageTagPermutation<IndexArrayType, ValueArrayType> StorageTag;
typedef vtkm::cont::internal::Storage<ValueType, StorageTag> StorageType;
public:
typedef typename StorageType::PortalType PortalControl;
typedef typename StorageType::PortalConstType PortalConstControl;
using PortalControl = typename StorageType::PortalType;
using PortalConstControl = typename StorageType::PortalConstType;
typedef vtkm::exec::internal::ArrayPortalPermutation<
typename IndexArrayType::template ExecutionTypes<Device>::PortalConst,

2
vtkm/cont/ArrayHandleReverse.h
View File

@ -39,7 +39,7 @@ template <typename PortalType>
class VTKM_ALWAYS_EXPORT ArrayPortalReverse
{
public:
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
VTKM_EXEC_CONT
ArrayPortalReverse()

15
vtkm/cont/ArrayHandleStreaming.h
View File

@ -34,8 +34,8 @@ template <typename P>
class VTKM_ALWAYS_EXPORT ArrayPortalStreaming
{
public:
typedef P PortalType;
typedef typename PortalType::ValueType ValueType;
using PortalType = P;
using ValueType = typename PortalType::ValueType;
VTKM_CONT
ArrayPortalStreaming()
@ -124,13 +124,12 @@ template <typename ArrayHandleInputType>
class Storage<typename ArrayHandleInputType::ValueType, StorageTagStreaming<ArrayHandleInputType>>
{
public:
typedef typename ArrayHandleInputType::ValueType ValueType;
using ValueType = typename ArrayHandleInputType::ValueType;
typedef vtkm::cont::internal::ArrayPortalStreaming<typename ArrayHandleInputType::PortalControl>
PortalType;
typedef vtkm::cont::internal::ArrayPortalStreaming<
typename ArrayHandleInputType::PortalConstControl>
PortalConstType;
using PortalType =
vtkm::cont::internal::ArrayPortalStreaming<typename ArrayHandleInputType::PortalControl>;
using PortalConstType =
vtkm::cont::internal::ArrayPortalStreaming<typename ArrayHandleInputType::PortalConstControl>;
VTKM_CONT
Storage()

525
vtkm/cont/ArrayHandleSwizzle.h Normal file
View File

@ -0,0 +1,525 @@
//=============================================================================
//
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
//
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2017 Sandia Corporation.
// Copyright 2017 UT-Battelle, LLC.
// Copyright 2017 Los Alamos National Security.
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
// Under the terms of Contract DE-AC52-06NA25396 with Los Alamos National
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//
//=============================================================================
#ifndef vtk_m_cont_ArrayHandleSwizzle_h
#define vtk_m_cont_ArrayHandleSwizzle_h
#include <vtkm/VecTraits.h>
#include <vtkm/cont/ArrayHandle.h>
#include <array>
namespace vtkm
{
namespace cont
{
namespace internal
{
// If TestValue appears more than once in ComponentMap, IsUnique will be false,
// but true if TestValue is unique.
template <vtkm::IdComponent TestValue, vtkm::IdComponent... ComponentMap>
struct ComponentIsUnique;
// Terminal case:
template <vtkm::IdComponent TestValue, vtkm::IdComponent Head>
struct ComponentIsUnique<TestValue, Head>
{
const static bool IsUnique = TestValue != Head;
};
// Recursive case:
template <vtkm::IdComponent TestValue, vtkm::IdComponent Head, vtkm::IdComponent... Tail>
struct ComponentIsUnique<TestValue, Head, Tail...>
{
using Next = ComponentIsUnique<TestValue, Tail...>;
const static bool IsUnique = TestValue != Head && Next::IsUnique;
};
// Validate the component map.
// All elements must be (1) unique, (2) >= 0, and (3) < InputSize
template <vtkm::IdComponent InputSize, vtkm::IdComponent... ComponentMap>
struct ValidateComponentMap;
// Terminal impl:
template <vtkm::IdComponent InputSize, vtkm::IdComponent Head>
struct ValidateComponentMap<InputSize, Head>
{
static const bool Valid = Head >= 0 && Head < InputSize;
};
// Recursive impl:
template <vtkm::IdComponent InputSize, vtkm::IdComponent Head, vtkm::IdComponent... Tail>
struct ValidateComponentMap<InputSize, Head, Tail...>
{
using Next = ValidateComponentMap<InputSize, Tail...>;
static const bool IsUnique = ComponentIsUnique<Head, Tail...>::IsUnique;
static const bool Valid = Head >= 0 && Head < InputSize && IsUnique && Next::Valid;
};
} // end namespace internal
/// This class collects metadata for an ArrayHandleSwizzle.
template <typename InputValueType, vtkm::IdComponent... ComponentMap>
struct ArrayHandleSwizzleTraits
{
/// The number of elements in the ComponentMap.
static const vtkm::IdComponent COUNT = static_cast<vtkm::IdComponent>(sizeof...(ComponentMap));
VTKM_STATIC_ASSERT_MSG(COUNT > 0, "Invalid ComponentMap: Cannot swizzle zero components.");
/// A std::array containing the ComponentMap for runtime querying.
using RuntimeComponentMapType = std::array<vtkm::IdComponent, COUNT>;
static VTKM_CONSTEXPR RuntimeComponentMapType GenerateRuntimeComponentMap()
{
return RuntimeComponentMapType{ { ComponentMap... } };
}
/// The ValueType of the ArrayHandleSwizzle's internal ArrayHandle.
using InputType = InputValueType;
/// The VecTraits for InputType.
using InputTraits = VecTraits<InputType>;
using Validator = internal::ValidateComponentMap<InputTraits::NUM_COMPONENTS, ComponentMap...>;
VTKM_STATIC_ASSERT_MSG(Validator::Valid,
"Invalid ComponentMap: Ids in ComponentMap must be unique, positive, and "
"less than the number of input components.");
/// The ComponentType of the ArrayHandleSwizzle.
using ComponentType = typename InputTraits::ComponentType;
/// The ValueType of the ArrayHandleSwizzle.
using OutputType = vtkm::Vec<ComponentType, COUNT>;
// The VecTraits for OutputType.
using OutputTraits = VecTraits<OutputType>;
/// If true, we use all components in the input vector. If false, we'll need
/// to make sure to preserve existing values on write.
static const bool ALL_COMPS_USED = InputTraits::NUM_COMPONENTS == COUNT;
private:
template <vtkm::IdComponent OutputIndex, vtkm::IdComponent... Map>
struct GetImpl;
// Terminal case:
template <vtkm::IdComponent OutputIndex, vtkm::IdComponent Head>
struct GetImpl<OutputIndex, Head>
{
VTKM_CONSTEXPR vtkm::IdComponent operator()() const { return OutputIndex == 0 ? Head : -1; }
};
// Recursive case:
template <vtkm::IdComponent OutputIndex, vtkm::IdComponent Head, vtkm::IdComponent... Tail>
struct GetImpl<OutputIndex, Head, Tail...>
{
using Next = GetImpl<OutputIndex - 1, Tail...>;
VTKM_CONSTEXPR vtkm::IdComponent operator()() const
{
return OutputIndex == 0 ? Head : Next()();
}
};
public:
/// Get the component from ComponentMap at the specified index as a
/// compile-time constant:
template <vtkm::IdComponent OutputIndex>
static VTKM_CONSTEXPR vtkm::IdComponent Get()
{
return GetImpl<OutputIndex, ComponentMap...>()();
}
private:
template <vtkm::IdComponent OutputIndex, vtkm::IdComponent... Map>
struct SwizzleImpl;
// Terminal case:
template <vtkm::IdComponent OutputIndex, vtkm::IdComponent Head>
struct SwizzleImpl<OutputIndex, Head>
{
static const vtkm::IdComponent InputIndex = Head;
void operator()(const InputType& in, OutputType& out) const
{
OutputTraits::SetComponent(out, OutputIndex, InputTraits::GetComponent(in, InputIndex));
}
};
// Recursive case:
template <vtkm::IdComponent OutputIndex, vtkm::IdComponent Head, vtkm::IdComponent... Tail>
struct SwizzleImpl<OutputIndex, Head, Tail...>
{
using Next = SwizzleImpl<OutputIndex + 1, Tail...>;
static const vtkm::IdComponent InputIndex = Head;
void operator()(const InputType& in, OutputType& out) const
{
OutputTraits::SetComponent(out, OutputIndex, InputTraits::GetComponent(in, InputIndex));
Next()(in, out);
}
};
public:
/// Swizzle the input type into the output type.
static void Swizzle(const InputType& in, OutputType& out)
{
SwizzleImpl<0, ComponentMap...>()(in, out);
}
// UnSwizzle output type --> input type
private:
template <vtkm::IdComponent OutputIndex, vtkm::IdComponent... Map>
struct UnSwizzleImpl;
// Terminal case:
template <vtkm::IdComponent OutputIndex, vtkm::IdComponent Head>
struct UnSwizzleImpl<OutputIndex, Head>
{
static const vtkm::IdComponent InputIndex = Head;
void operator()(const OutputType& out, InputType& in) const
{
InputTraits::SetComponent(in, InputIndex, OutputTraits::GetComponent(out, OutputIndex));
}
};
// Recursive case:
template <vtkm::IdComponent OutputIndex, vtkm::IdComponent Head, vtkm::IdComponent... Tail>
struct UnSwizzleImpl<OutputIndex, Head, Tail...>
{
using Next = UnSwizzleImpl<OutputIndex + 1, Tail...>;
static const vtkm::IdComponent InputIndex = Head;
void operator()(const OutputType& out, InputType& in) const
{
InputTraits::SetComponent(in, InputIndex, OutputTraits::GetComponent(out, OutputIndex));
Next()(out, in);
}
};
// Entry point:
public:
/// Unswizzle the output type back into the input type.
/// @warning If the entire vector is not used, there may be uninitialized
/// data in the resulting InputType vector. See ALL_COMPS_USED flag.
static void UnSwizzle(const OutputType& out, InputType& in)
{
UnSwizzleImpl<0, ComponentMap...>()(out, in);
}
};
namespace internal
{
template <typename PortalType, vtkm::IdComponent... ComponentMap>
class VTKM_ALWAYS_EXPORT ArrayPortalSwizzle
{
using Traits = ArrayHandleSwizzleTraits<typename PortalType::ValueType, ComponentMap...>;
public:
using ValueType = typename Traits::OutputType;
VTKM_EXEC_CONT
ArrayPortalSwizzle()
: Portal()
{
}
VTKM_EXEC_CONT
ArrayPortalSwizzle(const PortalType& portal)
: Portal(portal)
{
}
// Copy constructor
VTKM_EXEC_CONT ArrayPortalSwizzle(const ArrayPortalSwizzle<PortalType, ComponentMap...>& src)
: Portal(src.GetPortal())
{
}
VTKM_EXEC_CONT
vtkm::Id GetNumberOfValues() const { return this->Portal.GetNumberOfValues(); }
VTKM_EXEC_CONT
ValueType Get(vtkm::Id index) const
{
typename Traits::OutputType result;
Traits::Swizzle(this->Portal.Get(index), result);
return result;
}
VTKM_EXEC_CONT
void Set(vtkm::Id index, const ValueType& value) const
{
SetImpl<!Traits::ALL_COMPS_USED>(this->Portal)(index, value);
}
private:
// If NeedsRead is true, we need to initialize the InputType vector we write
// with the current values at @a index to avoid overwriting unused components.
template <bool NeedsRead>
struct SetImpl
{
const PortalType& Portal;
SetImpl(const PortalType& portal)
: Portal(portal)
{
}
void operator()(const vtkm::Id& index, const ValueType& value)
{
typename Traits::InputType in;
if (NeedsRead)
{
in = this->Portal.Get(index);
}
Traits::UnSwizzle(value, in);
this->Portal.Set(index, in);
}
};
public:
VTKM_EXEC_CONT
const PortalType& GetPortal() const { return this->Portal; }
private:
PortalType Portal;
}; // class ArrayPortalSwizzle
} // namespace internal
template <typename ArrayHandleType, vtkm::IdComponent... ComponentMap>
class StorageTagSwizzle
{
};
namespace internal
{
template <typename ArrayHandleType, vtkm::IdComponent... ComponentMap>
class Storage<typename ArrayHandleSwizzleTraits<typename ArrayHandleType::ValueType,
ComponentMap...>::OutputType,
StorageTagSwizzle<ArrayHandleType, ComponentMap...>>
{
public:
using PortalType = ArrayPortalSwizzle<typename ArrayHandleType::PortalControl, ComponentMap...>;
using PortalConstType =
ArrayPortalSwizzle<typename ArrayHandleType::PortalConstControl, ComponentMap...>;
using ValueType = typename PortalType::ValueType;
VTKM_CONT
Storage()
: Valid(false)
{
}
VTKM_CONT
Storage(const ArrayHandleType& array)
: Array(array)
, Valid(true)
{
}
VTKM_CONT
PortalConstType GetPortalConst() const
{
VTKM_ASSERT(this->Valid);
return PortalConstType(this->Array.GetPortalConstControl());
}
VTKM_CONT
PortalType GetPortal()
{
VTKM_ASSERT(this->Valid);
return PortalType(this->Array.GetPortalControl());
}
VTKM_CONT
vtkm::Id GetNumberOfValues() const
{
VTKM_ASSERT(this->Valid);
return this->Array.GetNumberOfValues();
}
VTKM_CONT
void Allocate(vtkm::Id numberOfValues)
{
VTKM_ASSERT(this->Valid);
this->Array.Allocate(numberOfValues);
}
VTKM_CONT
void Shrink(vtkm::Id numberOfValues)
{
VTKM_ASSERT(this->Valid);
this->Array.Shrink(numberOfValues);
}
VTKM_CONT
void ReleaseResources()
{
VTKM_ASSERT(this->Valid);
this->Array.ReleaseResources();
}
VTKM_CONT
const ArrayHandleType& GetArray() const
{
VTKM_ASSERT(this->Valid);
return this->Array;
}
private:
ArrayHandleType Array;
bool Valid;
}; // class Storage
template <typename ArrayHandleType, vtkm::IdComponent... ComponentMap, typename Device>
class ArrayTransfer<typename ArrayHandleSwizzleTraits<typename ArrayHandleType::ValueType,
ComponentMap...>::OutputType,
StorageTagSwizzle<ArrayHandleType, ComponentMap...>,
Device>
{
using ArrayExecutionTypes = typename ArrayHandleType::template ExecutionTypes<Device>;
using StorageTag = StorageTagSwizzle<ArrayHandleType, ComponentMap...>;
public:
using SwizzleTraits =
ArrayHandleSwizzleTraits<typename ArrayHandleType::ValueType, ComponentMap...>;
using ValueType = typename SwizzleTraits::OutputType;
private:
using StorageType = vtkm::cont::internal::Storage<ValueType, StorageTag>;
public:
using PortalControl = typename StorageType::PortalType;
using PortalConstControl = typename StorageType::PortalConstType;
using PortalExecution = ArrayPortalSwizzle<typename ArrayExecutionTypes::Portal, ComponentMap...>;
using PortalConstExecution =
ArrayPortalSwizzle<typename ArrayExecutionTypes::PortalConst, ComponentMap...>;
VTKM_CONT
ArrayTransfer(StorageType* storage)
: Array(storage->GetArray())
{
}
VTKM_CONT
vtkm::Id GetNumberOfValues() const { return this->Array.GetNumberOfValues(); }
VTKM_CONT
PortalConstExecution PrepareForInput(bool vtkmNotUsed(updateData))
{
return PortalConstExecution(this->Array.PrepareForInput(Device()));
}
VTKM_CONT
PortalExecution PrepareForInPlace(bool vtkmNotUsed(updateData))
{
return PortalExecution(this->Array.PrepareForInPlace(Device()));
}
VTKM_CONT
PortalExecution PrepareForOutput(vtkm::Id numberOfValues)
{
return PortalExecution(this->Array.PrepareForOutput(numberOfValues, Device()));
}
VTKM_CONT
void RetrieveOutputData(StorageType* vtkmNotUsed(storage)) const
{
// Implementation of this method should be unnecessary. The internal
// array handle should automatically retrieve the output data as
// necessary.
}
VTKM_CONT
void Shrink(vtkm::Id numberOfValues) { this->Array.Shrink(numberOfValues); }
VTKM_CONT
void ReleaseResources() { this->Array.ReleaseResourcesExecution(); }
private:
ArrayHandleType Array;
};
}
}
} // namespace vtkm::cont::internal
namespace vtkm
{
namespace cont
{
/// \brief A fancy ArrayHandle that rearranges and/or removes components of an
/// ArrayHandle with a vtkm::Vec ValueType.
///
/// ArrayHandleSwizzle is a specialization of ArrayHandle. It takes an
/// input ArrayHandle with a vtkm::Vec ValueType and a compile-time component
/// map and uses this information to create a new array consisting of the
/// specified components of the input ArrayHandle in the specified order. So for
/// a given index i, ArrayHandleSwizzle looks up the i-th vtkm::Vec in
/// the index array and reads or writes to the specified components, leaving all
/// other components unmodified. This is done on the fly rather than creating a
/// copy of the array.
template <typename ArrayHandleType, vtkm::IdComponent... ComponentMap>
class ArrayHandleSwizzle : public vtkm::cont::ArrayHandle<
typename ArrayHandleSwizzleTraits<typename ArrayHandleType::ValueType,
ComponentMap...>::OutputType,
StorageTagSwizzle<ArrayHandleType, ComponentMap...>>
{
public:
VTKM_ARRAY_HANDLE_SUBCLASS(
ArrayHandleSwizzle,
(ArrayHandleSwizzle<ArrayHandleType, ComponentMap...>),
(vtkm::cont::ArrayHandle<typename ArrayHandleSwizzleTraits<typename ArrayHandleType::ValueType,
ComponentMap...>::OutputType,
StorageTagSwizzle<ArrayHandleType, ComponentMap...>>));
using SwizzleTraits =
ArrayHandleSwizzleTraits<typename ArrayHandleType::ValueType, ComponentMap...>;
protected:
using StorageType = vtkm::cont::internal::Storage<ValueType, StorageTag>;
public:
VTKM_CONT
ArrayHandleSwizzle(const ArrayHandleType& array)
: Superclass(StorageType(array))
{
}
};
/// make_ArrayHandleSwizzle is convenience function to generate an
/// ArrayHandleSwizzle.
template <vtkm::IdComponent... ComponentMap, typename ArrayHandleType>
VTKM_CONT ArrayHandleSwizzle<ArrayHandleType, ComponentMap...> make_ArrayHandleSwizzle(
const ArrayHandleType& array)
{
return ArrayHandleSwizzle<ArrayHandleType, ComponentMap...>(array);
}
}
} // namespace vtkm::cont
#endif // vtk_m_cont_ArrayHandleSwizzle_h

16
vtkm/cont/ArrayHandleZip.h
View File

@ -36,10 +36,10 @@ template <typename ValueType_, typename PortalTypeFirst_, typename PortalTypeSec
class ArrayPortalZip
{
public:
typedef ValueType_ ValueType;
typedef ValueType_ IteratorType;
typedef PortalTypeFirst_ PortalTypeFirst;
typedef PortalTypeSecond_ PortalTypeSecond;
using ValueType = ValueType_;
using IteratorType = ValueType_;
using PortalTypeFirst = PortalTypeFirst_;
using PortalTypeSecond = PortalTypeSecond_;
VTKM_SUPPRESS_EXEC_WARNINGS
VTKM_EXEC_CONT
@ -137,7 +137,7 @@ class Storage<T, StorageTagZip<FirstHandleType, SecondHandleType>>
VTKM_IS_ARRAY_HANDLE(SecondHandleType);
public:
typedef T ValueType;
using ValueType = T;
typedef vtkm::exec::internal::ArrayPortalZip<ValueType,
typename FirstHandleType::PortalControl,
@ -221,10 +221,10 @@ class ArrayTransfer<T, StorageTagZip<FirstHandleType, SecondHandleType>, Device>
typedef vtkm::cont::internal::Storage<T, StorageTag> StorageType;
public:
typedef T ValueType;
using ValueType = T;
typedef typename StorageType::PortalType PortalControl;
typedef typename StorageType::PortalConstType PortalConstControl;
using PortalControl = typename StorageType::PortalType;
using PortalConstControl = typename StorageType::PortalConstType;
typedef vtkm::exec::internal::ArrayPortalZip<
ValueType,

2
vtkm/cont/ArrayPortalToIterators.h
View File

@ -57,7 +57,7 @@ public:
/// The type of the iterator.
///
typedef vtkm::cont::internal::IteratorFromArrayPortal<PortalType> IteratorType;
using IteratorType = vtkm::cont::internal::IteratorFromArrayPortal<PortalType>;
/// Returns an iterator pointing to the beginning of the ArrayPortal.
///

2
vtkm/cont/CMakeLists.txt
View File

@ -25,6 +25,7 @@ set(headers
ArrayHandleCompositeVector.h
ArrayHandleConstant.h
ArrayHandleCounting.h
ArrayHandleExtractComponent.h
ArrayHandleDiscard.h
ArrayHandleGroupVec.h
ArrayHandleGroupVecVariable.h
@ -33,6 +34,7 @@ set(headers
ArrayHandlePermutation.h
ArrayHandleReverse.h
ArrayHandleStreaming.h
ArrayHandleSwizzle.h
ArrayHandleTransform.h
ArrayHandleUniformPointCoordinates.h
ArrayHandleZip.h

95
vtkm/cont/CellSetExplicit.h
View File

@ -22,16 +22,11 @@
#include <vtkm/CellShape.h>
#include <vtkm/TopologyElementTag.h>
#include <vtkm/cont/ArrayHandleConstant.h>
#include <vtkm/cont/CellSet.h>
#include <vtkm/cont/internal/ConnectivityExplicitInternals.h>
#include <vtkm/exec/ConnectivityExplicit.h>
#include <vtkm/cont/ArrayHandleConstant.h>
#include <vtkm/cont/ArrayHandleCounting.h>
#include <vtkm/exec/ExecutionWholeArray.h>
#include <vtkm/worklet/DispatcherMapField.h>
#include <vtkm/worklet/WorkletMapField.h>
#include <map>
#include <utility>
@ -348,98 +343,14 @@ public:
// nothing to do
}
// Worklet to expand the PointToCell numIndices array by repeating cell index
class ExpandIndices : public vtkm::worklet::WorkletMapField
{
public:
typedef void ControlSignature(FieldIn<> cellIndex,
FieldIn<> offset,
FieldIn<> numIndices,
WholeArrayOut<> cellIndices);
typedef void ExecutionSignature(_1, _2, _3, _4);
typedef _1 InputDomain;
VTKM_CONT
ExpandIndices() {}
template <typename PortalType>
VTKM_EXEC void operator()(const vtkm::Id& cellIndex,
const vtkm::Id& offset,
const vtkm::Id& numIndices,
const PortalType& cellIndices) const
{
VTKM_ASSERT(cellIndices.GetNumberOfValues() >= offset + numIndices);
vtkm::Id startIndex = offset;
for (vtkm::Id i = 0; i < numIndices; i++)
{
cellIndices.Set(startIndex++, cellIndex);
}
}
};
template <typename Device>
VTKM_CONT void CreateConnectivity(Device,
vtkm::TopologyElementTagCell,
vtkm::TopologyElementTagPoint)
{
// PointToCell connectivity array (point indices) will be
// transformed into the CellToPoint numIndices array using reduction
//
// PointToCell numIndices array using expansion will be
// transformed into the CellToPoint connectivity array
if (this->CellToPoint.ElementsValid)
{
return;
}
typedef vtkm::cont::DeviceAdapterAlgorithm<Device> Algorithm;
// Sizes of the PointToCell information
vtkm::Id numberOfCells = this->GetNumberOfCells();
vtkm::Id connectivityLength = this->PointToCell.Connectivity.GetNumberOfValues();
// PointToCell connectivity will be basis of CellToPoint numIndices
vtkm::cont::ArrayHandle<vtkm::Id> pointIndices;
Algorithm::Copy(this->PointToCell.Connectivity, pointIndices);
// PointToCell numIndices will be basis of CellToPoint connectivity
this->CellToPoint.Connectivity.Allocate(connectivityLength);
vtkm::cont::ArrayHandleCounting<vtkm::Id> index(0, 1, numberOfCells);
this->PointToCell.BuildIndexOffsets(Device());
vtkm::worklet::DispatcherMapField<ExpandIndices, Device> expandDispatcher;
expandDispatcher.Invoke(index,
this->PointToCell.IndexOffsets,
this->PointToCell.NumIndices,
this->CellToPoint.Connectivity);
// SortByKey where key is PointToCell connectivity and value is the expanded cellIndex
Algorithm::SortByKey(pointIndices, this->CellToPoint.Connectivity);
if (this->GetNumberOfPoints() <= 0 && connectivityLength > 0)
{
this->NumberOfPoints = pointIndices.GetPortalControl().Get(connectivityLength - 1) + 1;
}
vtkm::Id numberOfPoints = this->GetNumberOfPoints();
// CellToPoint numIndices from the now sorted PointToCell connectivity
vtkm::cont::ArrayHandleConstant<vtkm::Id> numArray(1, connectivityLength);
vtkm::cont::ArrayHandle<vtkm::Id> uniquePoints;
vtkm::cont::ArrayHandle<vtkm::Id> numIndices;
uniquePoints.Allocate(numberOfPoints);
numIndices.Allocate(numberOfPoints);
Algorithm::ReduceByKey(pointIndices, numArray, uniquePoints, numIndices, vtkm::Add());
// Set the CellToPoint information
this->CellToPoint.Shapes = vtkm::cont::make_ArrayHandleConstant(
static_cast<vtkm::UInt8>(CELL_SHAPE_VERTEX), numberOfPoints);
Algorithm::Copy(numIndices, this->CellToPoint.NumIndices);
this->CellToPoint.ElementsValid = true;
this->CellToPoint.IndexOffsetsValid = false;
internal::ComputeCellToPointConnectivity(
this->CellToPoint, this->PointToCell, this->GetNumberOfPoints(), Device());
}
void PrintSummary(std::ostream& out) const VTKM_OVERRIDE

335
vtkm/cont/CellSetPermutation.h
View File

@ -22,30 +22,157 @@
#include <vtkm/CellShape.h>
#include <vtkm/CellTraits.h>
#include <vtkm/cont/ArrayHandleCast.h>
#include <vtkm/cont/ArrayHandlePermutation.h>
#include <vtkm/cont/CellSet.h>
#include <vtkm/cont/internal/ConnectivityExplicitInternals.h>
#include <vtkm/internal/ConnectivityStructuredInternals.h>
#include <vtkm/worklet/DispatcherMapTopology.h>
#include <vtkm/worklet/WorkletMapTopology.h>
#include <vtkm/exec/ConnectivityPermuted.h>
#ifndef VTKM_DEFAULT_CELLSET_PERMUTATION_STORAGE_TAG
#define VTKM_DEFAULT_CELLSET_PERMUTATION_STORAGE_TAG VTKM_DEFAULT_STORAGE_TAG
#endif
namespace vtkm
{
namespace cont
{
template <typename OriginalCellSet, typename ValidCellArrayHandleType>
class CellSetPermutation;
namespace internal
{
template <typename OriginalCellSet, typename PermutationArrayHandleType>
class CellSetGeneralPermutation : public CellSet
struct WriteConnectivity : public vtkm::worklet::WorkletMapPointToCell
{
typedef void ControlSignature(CellSetIn cellset,
FieldInCell<IdType> offset,
WholeArrayOut<> connectivity);
typedef void ExecutionSignature(PointCount, PointIndices, _2, _3);
typedef _1 InputDomain;
template <typename PointIndicesType, typename OutPortalType>
VTKM_EXEC void operator()(vtkm::IdComponent pointcount,
const PointIndicesType& pointIndices,
vtkm::Id offset,
OutPortalType& connectivity) const
{
for (vtkm::IdComponent i = 0; i < pointcount; ++i)
{
connectivity.Set(offset++, pointIndices[i]);
}
}
};
// default for CellSetExplicit and CellSetSingleType
template <typename CellSetPermutationType>
class PointToCell
{
private:
using NumIndicesArrayType = decltype(make_ArrayHandlePermutation(
std::declval<CellSetPermutationType>().GetValidCellIds(),
std::declval<CellSetPermutationType>().GetFullCellSet().GetNumIndicesArray(
vtkm::TopologyElementTagPoint(),
vtkm::TopologyElementTagCell())));
public:
using ConnectivityArrays = vtkm::cont::internal::ConnectivityExplicitInternals<
VTKM_DEFAULT_STORAGE_TAG, // dummy, shapes array is not used
typename NumIndicesArrayType::StorageTag>;
template <typename Device>
static ConnectivityArrays Get(const CellSetPermutationType& cellset, Device)
{
ConnectivityArrays conn;
conn.NumIndices =
NumIndicesArrayType(cellset.GetValidCellIds(),
cellset.GetFullCellSet().GetNumIndicesArray(
vtkm::TopologyElementTagPoint(), vtkm::TopologyElementTagCell()));
vtkm::Id connectivityLength = vtkm::cont::DeviceAdapterAlgorithm<Device>::ScanExclusive(
vtkm::cont::make_ArrayHandleCast(conn.NumIndices, vtkm::Id()), conn.IndexOffsets);
conn.Connectivity.Allocate(connectivityLength);
vtkm::worklet::DispatcherMapTopology<WriteConnectivity, Device>().Invoke(
cellset, conn.IndexOffsets, conn.Connectivity);
return conn;
}
};
// specialization for CellSetStructured
template <vtkm::IdComponent DIMENSION, typename PermutationArrayHandleType>
class PointToCell<CellSetPermutation<CellSetStructured<DIMENSION>, PermutationArrayHandleType>>
{
private:
using CellSetPermutationType =
CellSetPermutation<CellSetStructured<DIMENSION>, PermutationArrayHandleType>;
public:
using ConnectivityArrays = vtkm::cont::internal::ConnectivityExplicitInternals<
VTKM_DEFAULT_STORAGE_TAG, // dummy, shapes array is not used
typename vtkm::cont::ArrayHandleConstant<vtkm::IdComponent>::StorageTag,
VTKM_DEFAULT_STORAGE_TAG,
typename vtkm::cont::ArrayHandleCounting<vtkm::Id>::StorageTag>;
template <typename Device>
static ConnectivityArrays Get(const CellSetPermutationType& cellset, Device)
{
vtkm::Id numberOfCells = cellset.GetNumberOfCells();
vtkm::IdComponent numPointsInCell =
vtkm::internal::ConnectivityStructuredInternals<DIMENSION>::NUM_POINTS_IN_CELL;
vtkm::Id connectivityLength = numberOfCells * numPointsInCell;
ConnectivityArrays conn;
conn.NumIndices = make_ArrayHandleConstant(numPointsInCell, numberOfCells);
conn.IndexOffsets = ArrayHandleCounting<vtkm::Id>(0, numPointsInCell, numberOfCells);
conn.Connectivity.Allocate(connectivityLength);
vtkm::worklet::DispatcherMapTopology<WriteConnectivity, Device>().Invoke(
cellset, conn.IndexOffsets, conn.Connectivity);
return conn;
}
};
template <typename CellSetPermutationType>
struct CellSetPermutationTraits;
template <typename OriginalCellSet_, typename PermutationArrayHandleType_>
struct CellSetPermutationTraits<CellSetPermutation<OriginalCellSet_, PermutationArrayHandleType_>>
{
using OriginalCellSet = OriginalCellSet_;
using PermutationArrayHandleType = PermutationArrayHandleType_;
};
template <typename OriginalCellSet_,
typename OriginalPermutationArrayHandleType,
typename PermutationArrayHandleType_>
struct CellSetPermutationTraits<
CellSetPermutation<CellSetPermutation<OriginalCellSet_, OriginalPermutationArrayHandleType>,
PermutationArrayHandleType_>>
{
using PreviousCellSet = CellSetPermutation<OriginalCellSet_, OriginalPermutationArrayHandleType>;
using PermutationArrayHandleType = vtkm::cont::ArrayHandlePermutation<
PermutationArrayHandleType_,
typename CellSetPermutationTraits<PreviousCellSet>::PermutationArrayHandleType>;
using OriginalCellSet = typename CellSetPermutationTraits<PreviousCellSet>::OriginalCellSet;
using Superclass = CellSetPermutation<OriginalCellSet, PermutationArrayHandleType>;
};
} // internal
template <typename OriginalCellSetType_,
typename PermutationArrayHandleType_ =
vtkm::cont::ArrayHandle<vtkm::Id, VTKM_DEFAULT_CELLSET_PERMUTATION_STORAGE_TAG>>
class CellSetPermutation : public CellSet
{
public:
using OriginalCellSetType = OriginalCellSetType_;
using PermutationArrayHandleType = PermutationArrayHandleType_;
VTKM_CONT
CellSetGeneralPermutation(const PermutationArrayHandleType& validCellIds,
const OriginalCellSet& cellset,
const std::string& name)
CellSetPermutation(const PermutationArrayHandleType& validCellIds,
const OriginalCellSetType& cellset,
const std::string& name = std::string())
: CellSet(name)
, ValidCellIds(validCellIds)
, FullCellSet(cellset)
@ -53,7 +180,7 @@ public:
}
VTKM_CONT
CellSetGeneralPermutation(const std::string& name)
CellSetPermutation(const std::string& name = std::string())
: CellSet(name)
, ValidCellIds()
, FullCellSet()
@ -61,108 +188,174 @@ public:
}
VTKM_CONT
vtkm::Id GetNumberOfCells() const VTKM_OVERRIDE { return this->ValidCellIds.GetNumberOfValues(); }
const OriginalCellSetType& GetFullCellSet() const { return this->FullCellSet; }
VTKM_CONT
vtkm::Id GetNumberOfPoints() const VTKM_OVERRIDE { return this->FullCellSet.GetNumberOfPoints(); }
const PermutationArrayHandleType& GetValidCellIds() const { return this->ValidCellIds; }
vtkm::Id GetNumberOfFaces() const VTKM_OVERRIDE { return -1; }
VTKM_CONT
vtkm::Id GetNumberOfCells() const override { return this->ValidCellIds.GetNumberOfValues(); }
vtkm::Id GetNumberOfEdges() const VTKM_OVERRIDE { return -1; }
VTKM_CONT
vtkm::Id GetNumberOfPoints() const override { return this->FullCellSet.GetNumberOfPoints(); }
VTKM_CONT
vtkm::Id GetNumberOfFaces() const override { return -1; }
VTKM_CONT
vtkm::Id GetNumberOfEdges() const override { return -1; }
VTKM_CONT
vtkm::IdComponent GetNumberOfPointsInCell(vtkm::Id cellIndex) const
{
return this->FullCellSet.GetNumberOfPointsInCell(
this->ValidCellIds.GetPortalConstControl().Get(cellIndex));
}
//This is the way you can fill the memory from another system without copying
VTKM_CONT
void Fill(const PermutationArrayHandleType& validCellIds, const OriginalCellSet& cellset)
void Fill(const PermutationArrayHandleType& validCellIds, const OriginalCellSetType& cellset)
{
ValidCellIds = validCellIds;
FullCellSet = cellset;
this->ValidCellIds = validCellIds;
this->FullCellSet = cellset;
}
template <typename TopologyElement>
VTKM_CONT vtkm::Id GetSchedulingRange(TopologyElement) const
VTKM_CONT vtkm::Id GetSchedulingRange(vtkm::TopologyElementTagCell) const
{
VTKM_IS_TOPOLOGY_ELEMENT_TAG(TopologyElement);
return this->ValidCellIds.GetNumberOfValues();
}
template <typename Device, typename FromTopology, typename ToTopology>
struct ExecutionTypes
VTKM_CONT vtkm::Id GetSchedulingRange(vtkm::TopologyElementTagPoint) const
{
VTKM_IS_DEVICE_ADAPTER_TAG(Device);
VTKM_IS_TOPOLOGY_ELEMENT_TAG(FromTopology);
VTKM_IS_TOPOLOGY_ELEMENT_TAG(ToTopology);
typedef typename PermutationArrayHandleType::template ExecutionTypes<Device>::PortalConst
ExecPortalType;
typedef typename OriginalCellSet::template ExecutionTypes<Device, FromTopology, ToTopology>::
ExecObjectType OrigExecObjectType;
typedef vtkm::exec::ConnectivityPermuted<ExecPortalType, OrigExecObjectType> ExecObjectType;
};
template <typename Device, typename FromTopology, typename ToTopology>
typename ExecutionTypes<Device, FromTopology, ToTopology>::ExecObjectType
PrepareForInput(Device d, FromTopology f, ToTopology t) const
{
typedef
typename ExecutionTypes<Device, FromTopology, ToTopology>::ExecObjectType ConnectivityType;
return ConnectivityType(this->ValidCellIds.PrepareForInput(d),
this->FullCellSet.PrepareForInput(d, f, t));
return this->FullCellSet.GetNumberOfPoints();
}
void PrintSummary(std::ostream& out) const VTKM_OVERRIDE
template <typename Device, typename FromTopology, typename ToTopology>
struct ExecutionTypes;
template <typename Device>
struct ExecutionTypes<Device, vtkm::TopologyElementTagPoint, vtkm::TopologyElementTagCell>
{
out << " CellSetGeneralPermutation of: " << std::endl;
VTKM_IS_DEVICE_ADAPTER_TAG(Device);
using ExecPortalType =
typename PermutationArrayHandleType::template ExecutionTypes<Device>::PortalConst;
using OrigExecObjectType = typename OriginalCellSetType::template ExecutionTypes<
Device,
vtkm::TopologyElementTagPoint,
vtkm::TopologyElementTagCell>::ExecObjectType;
using ExecObjectType =
vtkm::exec::ConnectivityPermutedPointToCell<ExecPortalType, OrigExecObjectType>;
};
template <typename Device>
struct ExecutionTypes<Device, vtkm::TopologyElementTagCell, vtkm::TopologyElementTagPoint>
{
VTKM_IS_DEVICE_ADAPTER_TAG(Device);
using ConnectiviyPortalType =
typename vtkm::cont::ArrayHandle<vtkm::Id>::template ExecutionTypes<Device>::PortalConst;
using NumIndicesPortalType = typename vtkm::cont::ArrayHandle<
vtkm::IdComponent>::template ExecutionTypes<Device>::PortalConst;
using IndexOffsetPortalType =
typename vtkm::cont::ArrayHandle<vtkm::Id>::template ExecutionTypes<Device>::PortalConst;
using ExecObjectType = vtkm::exec::ConnectivityPermutedCellToPoint<ConnectiviyPortalType,
NumIndicesPortalType,
IndexOffsetPortalType>;
};
template <typename Device>
VTKM_CONT typename ExecutionTypes<Device,
vtkm::TopologyElementTagPoint,
vtkm::TopologyElementTagCell>::ExecObjectType
PrepareForInput(Device device,
vtkm::TopologyElementTagPoint from,
vtkm::TopologyElementTagCell to) const
{
using ConnectivityType = typename ExecutionTypes<Device,
vtkm::TopologyElementTagPoint,
vtkm::TopologyElementTagCell>::ExecObjectType;
return ConnectivityType(this->ValidCellIds.PrepareForInput(device),
this->FullCellSet.PrepareForInput(device, from, to));
}
template <typename Device>
VTKM_CONT typename ExecutionTypes<Device,
vtkm::TopologyElementTagCell,
vtkm::TopologyElementTagPoint>::ExecObjectType
PrepareForInput(Device device, vtkm::TopologyElementTagCell, vtkm::TopologyElementTagPoint) const
{
if (!this->CellToPoint.ElementsValid)
{
auto pointToCell = internal::PointToCell<CellSetPermutation>::Get(*this, device);
internal::ComputeCellToPointConnectivity(
this->CellToPoint, pointToCell, this->GetNumberOfPoints(), device);
this->CellToPoint.BuildIndexOffsets(device);
}
using ConnectivityType = typename ExecutionTypes<Device,
vtkm::TopologyElementTagCell,
vtkm::TopologyElementTagPoint>::ExecObjectType;
return ConnectivityType(this->CellToPoint.Connectivity.PrepareForInput(device),
this->CellToPoint.NumIndices.PrepareForInput(device),
this->CellToPoint.IndexOffsets.PrepareForInput(device));
}
VTKM_CONT
void PrintSummary(std::ostream& out) const override
{
out << "CellSetPermutation of: " << std::endl;
this->FullCellSet.PrintSummary(out);
out << "Permutation Array: " << std::endl;
vtkm::cont::printSummary_ArrayHandle(this->ValidCellIds, out);
}
private:
using CellToPointConnectivity = vtkm::cont::internal::ConnectivityExplicitInternals<
typename ArrayHandleConstant<vtkm::UInt8>::StorageTag>;
PermutationArrayHandleType ValidCellIds;
OriginalCellSet FullCellSet;
OriginalCellSetType FullCellSet;
mutable CellToPointConnectivity CellToPoint;
};
} //namespace internal
#ifndef VTKM_DEFAULT_CELLSET_PERMUTATION_STORAGE_TAG
#define VTKM_DEFAULT_CELLSET_PERMUTATION_STORAGE_TAG VTKM_DEFAULT_STORAGE_TAG
#endif
template <typename OriginalCellSet,
typename PermutationArrayHandleType =
vtkm::cont::ArrayHandle<vtkm::Id, VTKM_DEFAULT_CELLSET_PERMUTATION_STORAGE_TAG>>
class CellSetPermutation
: public vtkm::cont::internal::CellSetGeneralPermutation<OriginalCellSet,
PermutationArrayHandleType>
template <typename CellSetType,
typename PermutationArrayHandleType1,
typename PermutationArrayHandleType2>
class CellSetPermutation<CellSetPermutation<CellSetType, PermutationArrayHandleType1>,
PermutationArrayHandleType2>
: public internal::CellSetPermutationTraits<
CellSetPermutation<CellSetPermutation<CellSetType, PermutationArrayHandleType1>,
PermutationArrayHandleType2>>::Superclass
{
VTKM_IS_CELL_SET(OriginalCellSet);
VTKM_IS_ARRAY_HANDLE(PermutationArrayHandleType);
typedef typename vtkm::cont::internal::CellSetGeneralPermutation<OriginalCellSet,
PermutationArrayHandleType>
ParentType;
private:
using Superclass = typename internal::CellSetPermutationTraits<CellSetPermutation>::Superclass;
public:
VTKM_CONT
CellSetPermutation(const PermutationArrayHandleType& validCellIds,
const OriginalCellSet& cellset,
CellSetPermutation(const PermutationArrayHandleType2& validCellIds,
const CellSetPermutation<CellSetType, PermutationArrayHandleType1>& cellset,
const std::string& name = std::string())
: ParentType(validCellIds, cellset, name)
: Superclass(vtkm::cont::make_ArrayHandlePermutation(validCellIds, cellset.GetValidCellIds()),
cellset.GetFullCellSet(),
name)
{
}
VTKM_CONT
CellSetPermutation(const std::string& name = std::string())
: ParentType(name)
: Superclass(name)
{
}
VTKM_CONT
CellSetPermutation<OriginalCellSet, PermutationArrayHandleType>& operator=(
const CellSetPermutation<OriginalCellSet, PermutationArrayHandleType>& src)
void Fill(const PermutationArrayHandleType2& validCellIds,
const CellSetPermutation<CellSetType, PermutationArrayHandleType1>& cellset)
{
ParentType::operator=(src);
return *this;
this->ValidCellIds = make_ArrayHandlePermutation(validCellIds, cellset.GetValidCellIds());
this->FullCellSet = cellset.GetFullCellSet();
}
};

2
vtkm/cont/DeviceAdapterAlgorithm.h
View File

@ -583,7 +583,7 @@ public:
///
VTKM_CONT bool Exists() const
{
typedef vtkm::cont::DeviceAdapterTraits<DeviceAdapterTag> DeviceAdapterTraits;
using DeviceAdapterTraits = vtkm::cont::DeviceAdapterTraits<DeviceAdapterTag>;
return DeviceAdapterTraits::Valid;
}
};

10
vtkm/cont/DynamicArrayHandle.h
View File

@ -238,8 +238,8 @@ public:
VTKM_CONT bool IsType()
{
VTKM_IS_ARRAY_HANDLE(ArrayHandleType);
typedef typename ArrayHandleType::ValueType ValueType;
typedef typename ArrayHandleType::StorageTag StorageTag;
using ValueType = typename ArrayHandleType::ValueType;
using StorageTag = typename ArrayHandleType::StorageTag;
return this->IsTypeAndStorage<ValueType, StorageTag>();
}
@ -281,8 +281,8 @@ public:
VTKM_CONT ArrayHandleType Cast() const
{
VTKM_IS_ARRAY_HANDLE(ArrayHandleType);
typedef typename ArrayHandleType::ValueType ValueType;
typedef typename ArrayHandleType::StorageTag StorageTag;
using ValueType = typename ArrayHandleType::ValueType;
using StorageTag = typename ArrayHandleType::StorageTag;
// Technically, this method returns a copy of the \c ArrayHandle. But
// because \c ArrayHandle acts like a shared pointer, it is valid to
// do the copy.
@ -545,7 +545,7 @@ namespace internal
template <typename TypeList, typename StorageList>
struct DynamicTransformTraits<vtkm::cont::DynamicArrayHandleBase<TypeList, StorageList>>
{
typedef vtkm::cont::internal::DynamicTransformTagCastAndCall DynamicTag;
using DynamicTag = vtkm::cont::internal::DynamicTransformTagCastAndCall;
};
} // namespace internal

2
vtkm/cont/Field.h
View File

@ -412,7 +412,7 @@ namespace internal
template <>
struct DynamicTransformTraits<vtkm::cont::Field>
{
typedef vtkm::cont::internal::DynamicTransformTagCastAndCall DynamicTag;
using DynamicTag = vtkm::cont::internal::DynamicTransformTagCastAndCall;
};
} // namespace internal

4
vtkm/cont/Storage.h
View File

@ -98,8 +98,8 @@ class Storage
: public vtkm::cont::internal::UndefinedStorage
{
public:
typedef vtkm::cont::internal::detail::UndefinedArrayPortal<T> PortalType;
typedef vtkm::cont::internal::detail::UndefinedArrayPortal<T> PortalConstType;
using PortalType = vtkm::cont::internal::detail::UndefinedArrayPortal<T>;
using PortalConstType = vtkm::cont::internal::detail::UndefinedArrayPortal<T>;
};
#else //VTKM_DOXYGEN_ONLY
{

24
vtkm/cont/StorageBasic.h
View File

@ -59,15 +59,15 @@ void free_aligned(void* mem);
template <typename T, size_t Alignment>
struct AlignedAllocator
{
typedef T value_type;
typedef T& reference;
typedef const T& const_reference;
typedef T* pointer;
typedef const T* const_pointer;
typedef void* void_pointer;
typedef const void* const_void_pointer;
typedef std::ptrdiff_t difference_type;
typedef std::size_t size_type;
using value_type = T;
using reference = T&;
using const_reference = const T&;
using pointer = T*;
using const_pointer = const T*;
using void_pointer = void*;
using const_void_pointer = const void*;
using difference_type = std::ptrdiff_t;
using size_type = std::size_t;
template <typename U>
struct rebind
@ -172,9 +172,9 @@ template <typename ValueT>
class VTKM_ALWAYS_EXPORT Storage<ValueT, vtkm::cont::StorageTagBasic> : public StorageBasicBase
{
public:
typedef ValueT ValueType;
typedef vtkm::cont::internal::ArrayPortalFromIterators<ValueType*> PortalType;
typedef vtkm::cont::internal::ArrayPortalFromIterators<const ValueType*> PortalConstType;
using ValueType = ValueT;
using PortalType = vtkm::cont::internal::ArrayPortalFromIterators<ValueType*>;
using PortalConstType = vtkm::cont::internal::ArrayPortalFromIterators<const ValueType*>;
/// The original design of this class provided an allocator as a template
/// parameters. That messed things up, though, because other templated

24
vtkm/cont/StorageImplicit.h
View File

@ -46,7 +46,7 @@ namespace cont
template <class ArrayPortalType>
struct VTKM_ALWAYS_EXPORT StorageTagImplicit
{
typedef ArrayPortalType PortalType;
using PortalType = ArrayPortalType;
};
namespace internal
@ -56,15 +56,15 @@ template <class ArrayPortalType>
class Storage<typename ArrayPortalType::ValueType, StorageTagImplicit<ArrayPortalType>>
{
public:
typedef typename ArrayPortalType::ValueType ValueType;
typedef ArrayPortalType PortalConstType;
using ValueType = typename ArrayPortalType::ValueType;
using PortalConstType = ArrayPortalType;
// This is meant to be invalid. Because implicit arrays are read only, you
// should only be able to use the const version.
struct PortalType
{
typedef void* ValueType;
typedef void* IteratorType;
using ValueType = void*;
using IteratorType = void*;
};
VTKM_CONT
@ -101,16 +101,16 @@ template <typename T, class ArrayPortalType, class DeviceAdapterTag>
class ArrayTransfer<T, StorageTagImplicit<ArrayPortalType>, DeviceAdapterTag>
{
private:
typedef StorageTagImplicit<ArrayPortalType> StorageTag;
using StorageTag = StorageTagImplicit<ArrayPortalType>;
typedef vtkm::cont::internal::Storage<T, StorageTag> StorageType;
public:
typedef T ValueType;
using ValueType = T;
typedef typename StorageType::PortalType PortalControl;
typedef typename StorageType::PortalConstType PortalConstControl;
typedef PortalControl PortalExecution;
typedef PortalConstControl PortalConstExecution;
using PortalControl = typename StorageType::PortalType;
using PortalConstControl = typename StorageType::PortalConstType;
using PortalExecution = PortalControl;
using PortalConstExecution = PortalConstControl;
VTKM_CONT
ArrayTransfer(StorageType* storage)
@ -147,7 +147,7 @@ public:
template <class IteratorTypeControl>
VTKM_CONT void CopyInto(IteratorTypeControl dest) const
{
typedef typename StorageType::PortalConstType PortalType;
using PortalType = typename StorageType::PortalConstType;
PortalType portal = this->Storage->GetPortalConst();
std::copy(vtkm::cont::ArrayPortalToIteratorBegin(portal),

4
vtkm/cont/arg/Transport.h
View File

@ -55,8 +55,8 @@ struct Transport
/// For example, for an \c ArrayHandle, the \c ExecObjectType is the portal
/// used in the execution environment.
///
typedef
typename ContObjectType::template ExecutionTypes<DeviceAdapterTag>::PortalConst ExecObjectType;
using ExecObjectType =
typename ContObjectType::template ExecutionTypes<DeviceAdapterTag>::PortalConst;
/// \brief Send data to the execution environment.
///

2
vtkm/cont/arg/TransportTagArrayIn.h
View File

@ -47,7 +47,7 @@ struct Transport<vtkm::cont::arg::TransportTagArrayIn, ContObjectType, Device>
{
VTKM_IS_ARRAY_HANDLE(ContObjectType);
typedef typename ContObjectType::template ExecutionTypes<Device>::PortalConst ExecObjectType;
using ExecObjectType = typename ContObjectType::template ExecutionTypes<Device>::PortalConst;
template <typename InputDomainType>
VTKM_CONT ExecObjectType operator()(const ContObjectType& object,

2
vtkm/cont/arg/TransportTagArrayInOut.h
View File

@ -50,7 +50,7 @@ struct Transport<vtkm::cont::arg::TransportTagArrayInOut, ContObjectType, Device
// is not an array handle as an argument that is expected to be one.
VTKM_IS_ARRAY_HANDLE(ContObjectType);
typedef typename ContObjectType::template ExecutionTypes<Device>::Portal ExecObjectType;
using ExecObjectType = typename ContObjectType::template ExecutionTypes<Device>::Portal;
template <typename InputDomainType>
VTKM_CONT ExecObjectType operator()(ContObjectType object,

2
vtkm/cont/arg/TransportTagArrayOut.h
View File

@ -49,7 +49,7 @@ struct Transport<vtkm::cont::arg::TransportTagArrayOut, ContObjectType, Device>
// is not an array handle as an argument that is expected to be one.
VTKM_IS_ARRAY_HANDLE(ContObjectType);
typedef typename ContObjectType::template ExecutionTypes<Device>::Portal ExecObjectType;
using ExecObjectType = typename ContObjectType::template ExecutionTypes<Device>::Portal;
template <typename InputDomainType>
VTKM_CONT ExecObjectType operator()(ContObjectType object,

2
vtkm/cont/arg/TransportTagAtomicArray.h
View File

@ -53,7 +53,7 @@ struct Transport<vtkm::cont::arg::TransportTagAtomicArray,
vtkm::cont::ArrayHandle<T, vtkm::cont::StorageTagBasic>,
Device>
{
typedef vtkm::exec::AtomicArray<T, Device> ExecObjectType;
using ExecObjectType = vtkm::exec::AtomicArray<T, Device>;
template <typename InputDomainType>
VTKM_CONT ExecObjectType operator()(vtkm::cont::ArrayHandle<T, vtkm::cont::StorageTagBasic> array,

4
vtkm/cont/arg/TransportTagCellSetIn.h
View File

@ -50,10 +50,10 @@ struct Transport<vtkm::cont::arg::TransportTagCellSetIn<FromTopology, ToTopology
{
VTKM_IS_CELL_SET(ContObjectType);
typedef
using ExecObjectType =
typename ContObjectType::template ExecutionTypes<Device,
FromTopology,
ToTopology>::ExecObjectType ExecObjectType;
ToTopology>::ExecObjectType;
template <typename InputDomainType>
VTKM_CONT ExecObjectType

2
vtkm/cont/arg/TransportTagExecObject.h
View File

@ -53,7 +53,7 @@ struct Transport<vtkm::cont::arg::TransportTagExecObject, ContObjectType, Device
(std::is_base_of<vtkm::exec::ExecutionObjectBase, ContObjectType>::value),
"All execution objects are expected to inherit from vtkm::exec::ExecutionObjectBase");
typedef ContObjectType ExecObjectType;
using ExecObjectType = ContObjectType;
template <typename InputDomainType>
VTKM_CONT ExecObjectType

2
vtkm/cont/arg/TransportTagTopologyFieldIn.h
View File

@ -87,7 +87,7 @@ struct Transport<vtkm::cont::arg::TransportTagTopologyFieldIn<TopologyElementTag
{
VTKM_IS_ARRAY_HANDLE(ContObjectType);
typedef typename ContObjectType::template ExecutionTypes<Device>::PortalConst ExecObjectType;
using ExecObjectType = typename ContObjectType::template ExecutionTypes<Device>::PortalConst;
VTKM_CONT
ExecObjectType operator()(const ContObjectType& object,

6
vtkm/cont/arg/TransportTagWholeArrayIn.h
View File

@ -54,10 +54,10 @@ struct Transport<vtkm::cont::arg::TransportTagWholeArrayIn, ContObjectType, Devi
// is not an array handle as an argument that is expected to be one.
VTKM_IS_ARRAY_HANDLE(ContObjectType);
typedef typename ContObjectType::ValueType ValueType;
typedef typename ContObjectType::StorageTag StorageTag;
using ValueType = typename ContObjectType::ValueType;
using StorageTag = typename ContObjectType::StorageTag;
typedef vtkm::exec::ExecutionWholeArrayConst<ValueType, StorageTag, Device> ExecObjectType;
using ExecObjectType = vtkm::exec::ExecutionWholeArrayConst<ValueType, StorageTag, Device>;
template <typename InputDomainType>
VTKM_CONT ExecObjectType

6
vtkm/cont/arg/TransportTagWholeArrayInOut.h
View File

@ -56,10 +56,10 @@ struct Transport<vtkm::cont::arg::TransportTagWholeArrayInOut, ContObjectType, D
// is not an array handle as an argument that is expected to be one.
VTKM_IS_ARRAY_HANDLE(ContObjectType);
typedef typename ContObjectType::ValueType ValueType;
typedef typename ContObjectType::StorageTag StorageTag;
using ValueType = typename ContObjectType::ValueType;
using StorageTag = typename ContObjectType::StorageTag;
typedef vtkm::exec::ExecutionWholeArray<ValueType, StorageTag, Device> ExecObjectType;
using ExecObjectType = vtkm::exec::ExecutionWholeArray<ValueType, StorageTag, Device>;
template <typename InputDomainType>
VTKM_CONT ExecObjectType

6
vtkm/cont/arg/TransportTagWholeArrayOut.h
View File

@ -56,10 +56,10 @@ struct Transport<vtkm::cont::arg::TransportTagWholeArrayOut, ContObjectType, Dev
// is not an array handle as an argument that is expected to be one.
VTKM_IS_ARRAY_HANDLE(ContObjectType);
typedef typename ContObjectType::ValueType ValueType;
typedef typename ContObjectType::StorageTag StorageTag;
using ValueType = typename ContObjectType::ValueType;
using StorageTag = typename ContObjectType::StorageTag;
typedef vtkm::exec::ExecutionWholeArray<ValueType, StorageTag, Device> ExecObjectType;
using ExecObjectType = vtkm::exec::ExecutionWholeArray<ValueType, StorageTag, Device>;
template <typename InputDomainType>
VTKM_CONT ExecObjectType

6
vtkm/cont/arg/testing/UnitTestTransportArrayIn.cxx
View File

@ -40,7 +40,7 @@ struct TestKernelIn : public vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
if (!test_equal(this->Portal.Get(index), TestValue(index, ValueType())))
{
this->RaiseError("Got bad execution object.");
@ -60,10 +60,10 @@ struct TryArrayInType
array[index] = TestValue(index, T());
}
typedef vtkm::cont::ArrayHandle<T> ArrayHandleType;
using ArrayHandleType = vtkm::cont::ArrayHandle<T>;
ArrayHandleType handle = vtkm::cont::make_ArrayHandle(array, ARRAY_SIZE);
typedef typename ArrayHandleType::template ExecutionTypes<Device>::PortalConst PortalType;
using PortalType = typename ArrayHandleType::template ExecutionTypes<Device>::PortalConst;
vtkm::cont::arg::Transport<vtkm::cont::arg::TransportTagArrayIn, ArrayHandleType, Device>
transport;

6
vtkm/cont/arg/testing/UnitTestTransportArrayInOut.cxx
View File

@ -40,7 +40,7 @@ struct TestKernelInOut : public vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
ValueType inValue = this->Portal.Get(index);
this->Portal.Set(index, inValue + inValue);
}
@ -58,10 +58,10 @@ struct TryArrayInOutType
array[index] = TestValue(index, T());
}
typedef vtkm::cont::ArrayHandle<T> ArrayHandleType;
using ArrayHandleType = vtkm::cont::ArrayHandle<T>;
ArrayHandleType handle = vtkm::cont::make_ArrayHandle(array, ARRAY_SIZE);
typedef typename ArrayHandleType::template ExecutionTypes<Device>::Portal PortalType;
using PortalType = typename ArrayHandleType::template ExecutionTypes<Device>::Portal;
vtkm::cont::arg::Transport<vtkm::cont::arg::TransportTagArrayInOut, ArrayHandleType, Device>
transport;

6
vtkm/cont/arg/testing/UnitTestTransportArrayOut.cxx
View File

@ -41,7 +41,7 @@ struct TestKernelOut : public vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
this->Portal.Set(index, TestValue(index, ValueType()));
}
};
@ -52,10 +52,10 @@ struct TryArrayOutType
template <typename T>
void operator()(T) const
{
typedef vtkm::cont::ArrayHandle<T> ArrayHandleType;
using ArrayHandleType = vtkm::cont::ArrayHandle<T>;
ArrayHandleType handle;
typedef typename ArrayHandleType::template ExecutionTypes<Device>::Portal PortalType;
using PortalType = typename ArrayHandleType::template ExecutionTypes<Device>::Portal;
vtkm::cont::arg::Transport<vtkm::cont::arg::TransportTagArrayOut, ArrayHandleType, Device>
transport;

8
vtkm/cont/arg/testing/UnitTestTransportWholeArray.cxx
View File

@ -50,7 +50,7 @@ struct TestOutKernel : public vtkm::exec::FunctorBase
{
this->RaiseError("Out whole array has wrong size.");
}
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
this->Portal.Set(index, TestValue(index, ValueType()));
}
};
@ -67,7 +67,7 @@ struct TestInKernel : public vtkm::exec::FunctorBase
{
this->RaiseError("In whole array has wrong size.");
}
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
if (!test_equal(this->Portal.Get(index), TestValue(index, ValueType())))
{
this->RaiseError("Got bad execution object.");
@ -87,7 +87,7 @@ struct TestInOutKernel : public vtkm::exec::FunctorBase
{
this->RaiseError("In/Out whole array has wrong size.");
}
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
this->Portal.Set(index, this->Portal.Get(index) + ValueType(OFFSET));
}
};
@ -106,7 +106,7 @@ struct TestAtomicKernel : public vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename AtomicType::ValueType ValueType;
using ValueType = typename AtomicType::ValueType;
this->AtomicArray.Add(0, static_cast<ValueType>(index));
}
};

72
vtkm/cont/cuda/internal/ArrayManagerExecutionCuda.cu
View File

@ -21,6 +21,9 @@
#define vtk_m_cont_cuda_internal_ArrayManagerExecutionCuda_cu
#include <vtkm/cont/cuda/internal/ArrayManagerExecutionCuda.h>
#include <vtkm/cont/cuda/internal/CudaAllocator.h>
using vtkm::cont::cuda::internal::CudaAllocator;
namespace vtkm
{
@ -43,6 +46,25 @@ DeviceAdapterId ExecutionArrayInterfaceBasic<DeviceAdapterTagCuda>::GetDeviceId(
void ExecutionArrayInterfaceBasic<DeviceAdapterTagCuda>::Allocate(TypelessExecutionArray& execArray,
vtkm::Id numBytes) const
{
// Detect if we can reuse a device-accessible pointer from the control env:
if (CudaAllocator::IsDevicePointer(execArray.ArrayControl))
{
const vtkm::Id managedCapacity = static_cast<const char*>(execArray.ArrayControlCapacity) -
static_cast<const char*>(execArray.ArrayControl);
if (managedCapacity >= numBytes)
{
if (execArray.Array && execArray.Array != execArray.ArrayControl)
{
this->Free(execArray);
}
execArray.Array = const_cast<void*>(execArray.ArrayControl);
execArray.ArrayEnd = static_cast<char*>(execArray.Array) + numBytes;
execArray.ArrayCapacity = const_cast<void*>(execArray.ArrayControlCapacity);
return;
}
}
if (execArray.Array != nullptr)
{
const vtkm::Id cap =
@ -64,18 +86,8 @@ void ExecutionArrayInterfaceBasic<DeviceAdapterTagCuda>::Allocate(TypelessExecut
// Attempt to allocate:
try
{
char* tmp;
#ifdef VTKM_USE_UNIFIED_MEMORY
int dev;
VTKM_CUDA_CALL(cudaGetDevice(&dev));
VTKM_CUDA_CALL(cudaMallocManaged(&tmp, static_cast<std::size_t>(numBytes)));
VTKM_CUDA_CALL(cudaMemAdvise(tmp, numBytes, cudaMemAdviseSetPreferredLocation, dev));
VTKM_CUDA_CALL(cudaMemPrefetchAsync(tmp, numBytes, dev, 0));
VTKM_CUDA_CALL(cudaStreamSynchronize(0));
#else
VTKM_CUDA_CALL(cudaMalloc(&tmp, static_cast<std::size_t>(numBytes)));
#endif
// Cast to char* so that the pointer math below will work.
char* tmp = static_cast<char*>(CudaAllocator::Allocate(static_cast<size_t>(numBytes)));
execArray.Array = tmp;
execArray.ArrayEnd = tmp + numBytes;
execArray.ArrayCapacity = tmp + numBytes;
@ -91,9 +103,20 @@ void ExecutionArrayInterfaceBasic<DeviceAdapterTagCuda>::Allocate(TypelessExecut
void ExecutionArrayInterfaceBasic<DeviceAdapterTagCuda>::Free(
TypelessExecutionArray& execArray) const
{
// If we're sharing a device-accessible pointer between control/exec, don't
// actually free it -- just null the pointers here:
if (execArray.Array == execArray.ArrayControl &&
CudaAllocator::IsDevicePointer(execArray.ArrayControl))
{
execArray.Array = nullptr;
execArray.ArrayEnd = nullptr;
execArray.ArrayCapacity = nullptr;
return;
}
if (execArray.Array != nullptr)
{
VTKM_CUDA_CALL(cudaFree(execArray.Array));
CudaAllocator::Free(execArray.Array);
execArray.Array = nullptr;
execArray.ArrayEnd = nullptr;
execArray.ArrayCapacity = nullptr;
@ -104,6 +127,13 @@ void ExecutionArrayInterfaceBasic<DeviceAdapterTagCuda>::CopyFromControl(const v
void* executionPtr,
vtkm::Id numBytes) const
{
// Do nothing if we're sharing a device-accessible pointer between control and
// execution:
if (controlPtr == executionPtr && CudaAllocator::IsDevicePointer(controlPtr))
{
return;
}
VTKM_CUDA_CALL(cudaMemcpy(
executionPtr, controlPtr, static_cast<std::size_t>(numBytes), cudaMemcpyHostToDevice));
}
@ -112,6 +142,22 @@ void ExecutionArrayInterfaceBasic<DeviceAdapterTagCuda>::CopyToControl(const voi
void* controlPtr,
vtkm::Id numBytes) const
{
// Do nothing if we're sharing a cuda managed pointer between control and execution:
if (controlPtr == executionPtr && CudaAllocator::IsDevicePointer(controlPtr))
{
// If we're trying to copy a shared, non-managed device pointer back to
// control throw an exception -- the pointer cannot be read from control,
// so this operation is invalid.
if (!CudaAllocator::IsManagedPointer(controlPtr))
{
throw vtkm::cont::ErrorBadValue(
"Control pointer is a CUDA device pointer that does not supported managed access.");
}
// If it is managed, just return and let CUDA handle the migration for us.
return;
}
VTKM_CUDA_CALL(cudaMemcpy(
controlPtr, executionPtr, static_cast<std::size_t>(numBytes), cudaMemcpyDeviceToHost));
}

19
vtkm/cont/cuda/internal/ArrayManagerExecutionThrustDevice.h
View File

@ -35,6 +35,7 @@ VTKM_THIRDPARTY_PRE_INCLUDE
VTKM_THIRDPARTY_POST_INCLUDE
#include <vtkm/cont/cuda/ErrorCuda.h>
#include <vtkm/cont/cuda/internal/CudaAllocator.h>
#include <vtkm/cont/cuda/internal/ThrustExceptionHandler.h>
#include <vtkm/exec/cuda/internal/ArrayPortalFromThrust.h>
@ -172,17 +173,8 @@ public:
// Attempt to allocate:
try
{
ValueType* tmp;
#ifdef VTKM_USE_UNIFIED_MEMORY
int dev;
VTKM_CUDA_CALL(cudaGetDevice(&dev));
VTKM_CUDA_CALL(cudaMallocManaged(&tmp, bufferSize));
VTKM_CUDA_CALL(cudaMemAdvise(tmp, bufferSize, cudaMemAdviseSetPreferredLocation, dev));
VTKM_CUDA_CALL(cudaMemPrefetchAsync(tmp, bufferSize, dev, 0));
VTKM_CUDA_CALL(cudaStreamSynchronize(0));
#else
VTKM_CUDA_CALL(cudaMalloc(&tmp, bufferSize));
#endif
ValueType* tmp =
static_cast<ValueType*>(vtkm::cont::cuda::internal::CudaAllocator::Allocate(bufferSize));
this->Begin = PointerType(tmp);
}
catch (const std::exception& error)
@ -214,9 +206,6 @@ public:
storage->Allocate(this->GetNumberOfValues());
try
{
#ifdef VTKM_USE_UNIFIED_MEMORY
cudaDeviceSynchronize();
#endif
::thrust::copy(
this->Begin, this->End, vtkm::cont::ArrayPortalToIteratorBegin(storage->GetPortal()));
}
@ -254,7 +243,7 @@ public:
{
if (this->Begin.get() != nullptr)
{
VTKM_CUDA_CALL(cudaFree(this->Begin.get()));
vtkm::cont::cuda::internal::CudaAllocator::Free(this->Begin.get());
this->Begin = PointerType(static_cast<ValueType*>(nullptr));
this->End = PointerType(static_cast<ValueType*>(nullptr));
this->Capacity = PointerType(static_cast<ValueType*>(nullptr));

8
vtkm/cont/cuda/internal/CMakeLists.txt
View File

@ -21,6 +21,7 @@
set(headers
ArrayManagerExecutionCuda.h
ArrayManagerExecutionThrustDevice.h
CudaAllocator.h
DeviceAdapterAlgorithmCuda.h
DeviceAdapterAlgorithmThrust.h
DeviceAdapterTagCuda.h
@ -30,6 +31,11 @@ set(headers
VirtualObjectTransferCuda.h
)
set(sources
ArrayManagerExecutionCuda.cu
CudaAllocator.cu
)
vtkm_declare_headers(CUDA ${headers} TESTABLE ${VTKm_ENABLE_CUDA})
#-----------------------------------------------------------------------------
@ -49,7 +55,7 @@ if (VTKm_ENABLE_CUDA)
list(APPEND compile_options -Xcompiler=-fPIC)
endif()
cuda_compile(vtkm_cont_cuda_object_files ArrayManagerExecutionCuda.cu
cuda_compile(vtkm_cont_cuda_object_files ${sources}
OPTIONS "${compile_options}")
#Setup the dependency chain for the custom object build so that

204
vtkm/cont/cuda/internal/CudaAllocator.cu Normal file
View File

@ -0,0 +1,204 @@
//============================================================================
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2017 Sandia Corporation.
// Copyright 2017 UT-Battelle, LLC.
// Copyright 2017 Los Alamos National Security.
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
//
// Under the terms of Contract DE-AC52-06NA25396 with Los Alamos National
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//============================================================================
#include <vtkm/cont/cuda/ErrorCuda.h>
#include <vtkm/cont/cuda/internal/CudaAllocator.h>
#include <cuda_runtime.h>
// These static vars are in an anon namespace to work around MSVC linker issues.
namespace
{
// Has CudaAllocator::Initialize been called?
static bool IsInitialized = false;
// True if all devices support concurrent pagable managed memory.
static bool ManagedMemorySupported = false;
}
namespace vtkm
{
namespace cont
{
namespace cuda
{
namespace internal
{
bool CudaAllocator::UsingManagedMemory()
{
CudaAllocator::Initialize();
return ManagedMemorySupported;
}
bool CudaAllocator::IsDevicePointer(const void* ptr)
{
if (!ptr)
{
return false;
}
cudaPointerAttributes attr;
cudaError_t err = cudaPointerGetAttributes(&attr, ptr);
// This function will return invalid value if the pointer is unknown to the
// cuda runtime. Manually catch this value since it's not really an error.
if (err == cudaErrorInvalidValue)
{
cudaGetLastError(); // Clear the error so we don't raise it later...
return false;
}
VTKM_CUDA_CALL(err /*= cudaPointerGetAttributes(&attr, ptr)*/);
return attr.devicePointer == ptr;
}
bool CudaAllocator::IsManagedPointer(const void* ptr)
{
if (!ptr)
{
return false;
}
cudaPointerAttributes attr;
cudaError_t err = cudaPointerGetAttributes(&attr, ptr);
// This function will return invalid value if the pointer is unknown to the
// cuda runtime. Manually catch this value since it's not really an error.
if (err == cudaErrorInvalidValue)
{
cudaGetLastError(); // Clear the error so we don't raise it later...
return false;
}
VTKM_CUDA_CALL(err /*= cudaPointerGetAttributes(&attr, ptr)*/);
return attr.isManaged != 0;
}
void* CudaAllocator::Allocate(std::size_t numBytes)
{
CudaAllocator::Initialize();
void* ptr = nullptr;
if (ManagedMemorySupported)
{
VTKM_CUDA_CALL(cudaMallocManaged(&ptr, numBytes));
}
else
{
VTKM_CUDA_CALL(cudaMalloc(&ptr, numBytes));
}
return ptr;
}
void CudaAllocator::Free(void* ptr)
{
CudaAllocator::Initialize();
VTKM_CUDA_CALL(cudaFree(ptr));
}
void CudaAllocator::PrepareForControl(const void* ptr, std::size_t numBytes)
{
CudaAllocator::Initialize();
if (ManagedMemorySupported)
{
// TODO these hints need to be benchmarked and adjusted once we start
// sharing the pointers between cont/exec
VTKM_CUDA_CALL(
cudaMemAdvise(ptr, numBytes, cudaMemAdviseSetPreferredLocation, cudaCpuDeviceId));
VTKM_CUDA_CALL(cudaMemAdvise(ptr, numBytes, cudaMemAdviseUnsetReadMostly, cudaCpuDeviceId));
VTKM_CUDA_CALL(cudaMemPrefetchAsync(ptr, numBytes, cudaCpuDeviceId, 0));
}
}
void CudaAllocator::PrepareForInput(const void* ptr, std::size_t numBytes)
{
CudaAllocator::Initialize();
if (ManagedMemorySupported)
{
int dev;
VTKM_CUDA_CALL(cudaGetDevice(&dev));
VTKM_CUDA_CALL(cudaMemAdvise(ptr, numBytes, cudaMemAdviseSetPreferredLocation, dev));
VTKM_CUDA_CALL(cudaMemAdvise(ptr, numBytes, cudaMemAdviseSetReadMostly, dev));
VTKM_CUDA_CALL(cudaMemPrefetchAsync(ptr, numBytes, dev, 0));
}
}
void CudaAllocator::PrepareForOutput(const void* ptr, std::size_t numBytes)
{
CudaAllocator::Initialize();
if (ManagedMemorySupported)
{
int dev;
VTKM_CUDA_CALL(cudaGetDevice(&dev));
VTKM_CUDA_CALL(cudaMemAdvise(ptr, numBytes, cudaMemAdviseSetPreferredLocation, dev));
VTKM_CUDA_CALL(cudaMemAdvise(ptr, numBytes, cudaMemAdviseUnsetReadMostly, dev));
VTKM_CUDA_CALL(cudaMemPrefetchAsync(ptr, numBytes, dev, 0));
}
}
void CudaAllocator::PrepareForInPlace(const void* ptr, std::size_t numBytes)
{
CudaAllocator::Initialize();
if (ManagedMemorySupported)
{
int dev;
VTKM_CUDA_CALL(cudaGetDevice(&dev));
VTKM_CUDA_CALL(cudaMemAdvise(ptr, numBytes, cudaMemAdviseSetPreferredLocation, dev));
VTKM_CUDA_CALL(cudaMemAdvise(ptr, numBytes, cudaMemAdviseUnsetReadMostly, dev));
VTKM_CUDA_CALL(cudaMemPrefetchAsync(ptr, numBytes, dev, 0));
}
}
void CudaAllocator::Initialize()
{
if (!IsInitialized)
{
int numDevices;
VTKM_CUDA_CALL(cudaGetDeviceCount(&numDevices));
if (numDevices == 0)
{
ManagedMemorySupported = false;
IsInitialized = true;
return;
}
// Check all devices, use the feature set supported by all
bool managed = true;
cudaDeviceProp prop;
for (int i = 0; i < numDevices && managed; ++i)
{
VTKM_CUDA_CALL(cudaGetDeviceProperties(&prop, i));
// We check for concurrentManagedAccess, as devices with only the
// managedAccess property have extra synchronization requirements.
managed = managed && prop.concurrentManagedAccess;
}
ManagedMemorySupported = managed;
IsInitialized = true;
}
}
}
}
}
} // end namespace vtkm::cont::cuda::internal

69
vtkm/cont/cuda/internal/CudaAllocator.h Normal file
View File

@ -0,0 +1,69 @@
//============================================================================
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2017 Sandia Corporation.
// Copyright 2017 UT-Battelle, LLC.
// Copyright 2017 Los Alamos National Security.
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
//
// Under the terms of Contract DE-AC52-06NA25396 with Los Alamos National
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//============================================================================
#ifndef vtk_m_cont_cuda_internal_CudaAllocator_h
#define vtk_m_cont_cuda_internal_CudaAllocator_h
#include <vtkm/cont/vtkm_cont_export.h>
#include <vtkm/internal/ExportMacros.h>
#include <cstddef>
namespace vtkm
{
namespace cont
{
namespace cuda
{
namespace internal
{
/// Collection of cuda-specific memory management operations.
struct VTKM_CONT_EXPORT CudaAllocator
{
/// Returns true if all detected CUDA devices support pageable managed memory
/// that can be accessed concurrently by the CPU and GPUs.
static VTKM_CONT bool UsingManagedMemory();
/// Returns true if the pointer is accessable from a CUDA device.
static VTKM_CONT bool IsDevicePointer(const void* ptr);
/// Returns true if the pointer is a CUDA pointer allocated with
/// cudaMallocManaged.
static VTKM_CONT bool IsManagedPointer(const void* ptr);
static VTKM_CONT void* Allocate(std::size_t numBytes);
static VTKM_CONT void Free(void* ptr);
static VTKM_CONT void PrepareForControl(const void* ptr, std::size_t numBytes);
static VTKM_CONT void PrepareForInput(const void* ptr, std::size_t numBytes);
static VTKM_CONT void PrepareForOutput(const void* ptr, std::size_t numBytes);
static VTKM_CONT void PrepareForInPlace(const void* ptr, std::size_t numBytes);
private:
static VTKM_CONT void Initialize();
};
}
}
}
} // end namespace vtkm::cont::cuda::internal
#endif // vtk_m_cont_cuda_internal_CudaAllocator_h

1
vtkm/cont/cuda/testing/CMakeLists.txt
View File

@ -27,6 +27,7 @@ set(unit_tests
UnitTestCudaDeviceAdapter.cu
UnitTestCudaImplicitFunction.cu
UnitTestCudaMath.cu
UnitTestCudaShareUserProvidedManagedMemory.cu
UnitTestCudaVirtualObjectCache.cu
)
vtkm_unit_tests(CUDA SOURCES ${unit_tests})

286
vtkm/cont/cuda/testing/UnitTestCudaShareUserProvidedManagedMemory.cu Normal file
View File

@ -0,0 +1,286 @@
//============================================================================
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2017 Sandia Corporation.
// Copyright 2017 UT-Battelle, LLC.
// Copyright 2017 Los Alamos National Security.
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
//
// Under the terms of Contract DE-AC52-06NA25396 with Los Alamos National
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//============================================================================
#ifdef VTKM_DEVICE_ADAPTER
#undef VTKM_DEVICE_ADAPTER
#endif
#define VTKM_DEVICE_ADAPTER VTKM_DEVICE_ADAPTER_ERROR
#include <vtkm/cont/cuda/internal/testing/Testing.h>
#include <vtkm/cont/ArrayHandle.h>
#include <vtkm/cont/cuda/DeviceAdapterCuda.h>
#include <vtkm/cont/cuda/ErrorCuda.h>
#include <vtkm/cont/cuda/internal/CudaAllocator.h>
#include <vtkm/cont/cuda/internal/testing/Testing.h>
#include <cuda_runtime.h>
using vtkm::cont::cuda::internal::CudaAllocator;
namespace
{
template <typename ValueType>
ValueType* AllocateManagedPointer(vtkm::Id numValues)
{
void* result;
VTKM_CUDA_CALL(cudaMallocManaged(&result, static_cast<size_t>(numValues) * sizeof(ValueType)));
// Some sanity checks:
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(result),
"Allocated pointer is not a device pointer.");
VTKM_TEST_ASSERT(CudaAllocator::IsManagedPointer(result), "Allocated pointer is not managed.");
return static_cast<ValueType*>(result);
}
template <typename ValueType>
ValueType* AllocateDevicePointer(vtkm::Id numValues)
{
void* result;
VTKM_CUDA_CALL(cudaMalloc(&result, static_cast<size_t>(numValues) * sizeof(ValueType)));
// Some sanity checks:
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(result),
"Allocated pointer is not a device pointer.");
VTKM_TEST_ASSERT(!CudaAllocator::IsManagedPointer(result), "Allocated pointer is managed.");
return static_cast<ValueType*>(result);
}
template <typename ValueType>
vtkm::cont::ArrayHandle<ValueType> CreateArrayHandle(vtkm::Id numValues, bool managed)
{
ValueType* ptr = managed ? AllocateManagedPointer<ValueType>(numValues)
: AllocateDevicePointer<ValueType>(numValues);
return vtkm::cont::make_ArrayHandle(ptr, numValues);
}
template <typename ValueType>
void TestPrepareForInput(bool managed)
{
vtkm::cont::ArrayHandle<ValueType> handle = CreateArrayHandle<ValueType>(32, managed);
handle.PrepareForInput(vtkm::cont::DeviceAdapterTagCuda());
ValueType* contArray = handle.Internals->ControlArray.GetArray();
ValueType* execArray = handle.Internals->ExecutionArray;
VTKM_TEST_ASSERT(contArray != nullptr, "No control array after PrepareForInput.");
VTKM_TEST_ASSERT(execArray != nullptr, "No execution array after PrepareForInput.");
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(execArray),
"PrepareForInput execution array not device pointer.");
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(contArray),
"PrepareForInput control array not device pointer.");
if (managed)
{
VTKM_TEST_ASSERT(CudaAllocator::IsManagedPointer(execArray),
"PrepareForInput execution array unmanaged.");
VTKM_TEST_ASSERT(CudaAllocator::IsManagedPointer(contArray),
"PrepareForInput control array unmanaged.");
}
VTKM_TEST_ASSERT(execArray == contArray, "PrepareForInput managed arrays not shared.");
}
template <typename ValueType>
void TestPrepareForInPlace(bool managed)
{
vtkm::cont::ArrayHandle<ValueType> handle = CreateArrayHandle<ValueType>(32, managed);
handle.PrepareForInPlace(vtkm::cont::DeviceAdapterTagCuda());
ValueType* contArray = handle.Internals->ControlArray.GetArray();
ValueType* execArray = handle.Internals->ExecutionArray;
VTKM_TEST_ASSERT(contArray != nullptr, "No control array after PrepareForInPlace.");
VTKM_TEST_ASSERT(execArray != nullptr, "No execution array after PrepareForInPlace.");
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(execArray),
"PrepareForInPlace execution array not device pointer.");
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(contArray),
"PrepareForInPlace control array not device pointer.");
if (managed)
{
VTKM_TEST_ASSERT(CudaAllocator::IsManagedPointer(execArray),
"PrepareForInPlace execution array unmanaged.");
VTKM_TEST_ASSERT(CudaAllocator::IsManagedPointer(contArray),
"PrepareForInPlace control array unmanaged.");
}
VTKM_TEST_ASSERT(execArray == contArray, "PrepareForInPlace managed arrays not shared.");
}
template <typename ValueType>
void TestPrepareForOutput(bool managed)
{
// Should reuse a managed control pointer if buffer is large enough.
vtkm::cont::ArrayHandle<ValueType> handle = CreateArrayHandle<ValueType>(32, managed);
handle.PrepareForOutput(32, vtkm::cont::DeviceAdapterTagCuda());
ValueType* contArray = handle.Internals->ControlArray.GetArray();
ValueType* execArray = handle.Internals->ExecutionArray;
VTKM_TEST_ASSERT(contArray != nullptr, "No control array after PrepareForOutput.");
VTKM_TEST_ASSERT(execArray != nullptr, "No execution array after PrepareForOutput.");
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(execArray),
"PrepareForOutput execution array not device pointer.");
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(contArray),
"PrepareForOutput control array not device pointer.");
if (managed)
{
VTKM_TEST_ASSERT(CudaAllocator::IsManagedPointer(execArray),
"PrepareForOutput execution array unmanaged.");
VTKM_TEST_ASSERT(CudaAllocator::IsManagedPointer(contArray),
"PrepareForOutput control array unmanaged.");
}
VTKM_TEST_ASSERT(execArray == contArray, "PrepareForOutput managed arrays not shared.");
}
template <typename ValueType>
void TestReleaseResourcesExecution(bool managed)
{
vtkm::cont::ArrayHandle<ValueType> handle = CreateArrayHandle<ValueType>(32, managed);
handle.PrepareForInput(vtkm::cont::DeviceAdapterTagCuda());
ValueType* origArray = handle.Internals->ExecutionArray;
handle.ReleaseResourcesExecution();
ValueType* contArray = handle.Internals->ControlArray.GetArray();
ValueType* execArray = handle.Internals->ExecutionArray;
VTKM_TEST_ASSERT(contArray != nullptr, "No control array after ReleaseResourcesExecution.");
VTKM_TEST_ASSERT(execArray == nullptr,
"Execution array not cleared after ReleaseResourcesExecution.");
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(contArray),
"ReleaseResourcesExecution control array not device pointer.");
if (managed)
{
VTKM_TEST_ASSERT(CudaAllocator::IsManagedPointer(contArray),
"ReleaseResourcesExecution control array unmanaged.");
}
VTKM_TEST_ASSERT(origArray == contArray,
"Control array changed after ReleaseResourcesExecution.");
}
template <typename ValueType>
void TestRoundTrip(bool managed)
{
vtkm::cont::ArrayHandle<ValueType> handle = CreateArrayHandle<ValueType>(32, managed);
handle.PrepareForOutput(32, vtkm::cont::DeviceAdapterTagCuda());
ValueType* origContArray = handle.Internals->ControlArray.GetArray();
{
ValueType* contArray = handle.Internals->ControlArray.GetArray();
ValueType* execArray = handle.Internals->ExecutionArray;
VTKM_TEST_ASSERT(contArray != nullptr, "No control array after PrepareForOutput.");
VTKM_TEST_ASSERT(execArray != nullptr, "No execution array after PrepareForOutput.");
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(execArray),
"PrepareForOutput execution array not device pointer.");
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(contArray),
"PrepareForOutput control array not device pointer.");
if (managed)
{
VTKM_TEST_ASSERT(CudaAllocator::IsManagedPointer(execArray),
"PrepareForOutput execution array unmanaged.");
VTKM_TEST_ASSERT(CudaAllocator::IsManagedPointer(contArray),
"PrepareForOutput control array unmanaged.");
}
VTKM_TEST_ASSERT(execArray == contArray, "PrepareForOutput managed arrays not shared.");
}
try
{
handle.GetPortalControl();
}
catch (vtkm::cont::ErrorBadValue&)
{
if (managed)
{
throw; // Exception is unexpected
}
// If !managed, this exception is intentional to indicate that the control
// array is a non-managed device pointer, and thus unaccessible from the
// control environment. Return because we've already validated correct
// behavior by catching this exception.
return;
}
if (!managed)
{
VTKM_TEST_FAIL("Expected exception not thrown.");
}
{
ValueType* contArray = handle.Internals->ControlArray.GetArray();
ValueType* execArray = handle.Internals->ExecutionArray;
VTKM_TEST_ASSERT(contArray != nullptr, "No control array after GetPortalConst.");
VTKM_TEST_ASSERT(execArray == nullptr, "Execution array not cleared after GetPortalConst.");
VTKM_TEST_ASSERT(CudaAllocator::IsDevicePointer(contArray),
"GetPortalConst control array not device pointer.");
VTKM_TEST_ASSERT(origContArray == contArray, "GetPortalConst changed control array.");
}
}
template <typename ValueType>
void DoTests()
{
TestPrepareForInput<ValueType>(false);
TestPrepareForInPlace<ValueType>(false);
TestPrepareForOutput<ValueType>(false);
TestReleaseResourcesExecution<ValueType>(false);
TestRoundTrip<ValueType>(false);
// If this device does not support managed memory, skip the managed tests.
if (!CudaAllocator::UsingManagedMemory())
{
std::cerr << "Skipping some tests -- device does not support managed memory.\n";
}
else
{
TestPrepareForInput<ValueType>(true);
TestPrepareForInPlace<ValueType>(true);
TestPrepareForOutput<ValueType>(true);
TestReleaseResourcesExecution<ValueType>(true);
TestRoundTrip<ValueType>(true);
}
}
struct ArgToTemplateType
{
template <typename ValueType>
void operator()(ValueType) const
{
DoTests<ValueType>();
}
};
void Launch()
{
using Types = vtkm::ListTagBase<vtkm::UInt8,
vtkm::Vec<vtkm::UInt8, 3>,
vtkm::Float32,
vtkm::Vec<vtkm::Float32, 4>,
vtkm::Float64,
vtkm::Vec<vtkm::Float64, 4>>;
vtkm::testing::Testing::TryTypes(ArgToTemplateType(), Types());
}
} // end anon namespace
int UnitTestCudaShareUserProvidedManagedMemory(int, char* [])
{
int ret = vtkm::cont::testing::Testing::Run(Launch);
return vtkm::cont::cuda::internal::Testing::CheckCudaBeforeExit(ret);
}

15
vtkm/cont/internal/ArrayHandleBasicImpl.h
View File

@ -39,16 +39,25 @@ namespace internal
struct VTKM_ALWAYS_EXPORT TypelessExecutionArray
{
VTKM_CONT
TypelessExecutionArray(void*& array, void*& arrayEnd, void*& arrayCapacity)
TypelessExecutionArray(void*& array,
void*& arrayEnd,
void*& arrayCapacity,
const void* arrayControl,
const void* arrayControlCapacity)
: Array(array)
, ArrayEnd(arrayEnd)
, ArrayCapacity(arrayCapacity)
, ArrayControl(arrayControl)
, ArrayControlCapacity(arrayControlCapacity)
{
}
void*& Array;
void*& ArrayEnd;
void*& ArrayCapacity;
// Used by cuda to detect and share managed memory allocations.
const void* ArrayControl;
const void* ArrayControlCapacity;
};
/// Factory that generates execution portals for basic storage.
@ -219,7 +228,9 @@ public:
internal::TypelessExecutionArray execArray(
reinterpret_cast<void*&>(this->ExecutionArray),
reinterpret_cast<void*&>(this->ExecutionArrayEnd),
reinterpret_cast<void*&>(this->ExecutionArrayCapacity));
reinterpret_cast<void*&>(this->ExecutionArrayCapacity),
this->ControlArray.GetBasePointer(),
this->ControlArray.GetCapacityPointer());
this->ExecutionInterface->Free(execArray);
}

25
vtkm/cont/internal/ArrayHandleBasicImpl.hxx
View File

@ -302,7 +302,9 @@ ArrayHandle<T, StorageTagBasic>::PrepareForInput(DeviceAdapterTag device) const
internal::TypelessExecutionArray execArray(
reinterpret_cast<void*&>(priv->ExecutionArray),
reinterpret_cast<void*&>(priv->ExecutionArrayEnd),
reinterpret_cast<void*&>(priv->ExecutionArrayCapacity));
reinterpret_cast<void*&>(priv->ExecutionArrayCapacity),
this->Internals->ControlArray.GetBasePointer(),
this->Internals->ControlArray.GetCapacityPointer());
const vtkm::Id numBytes =
static_cast<vtkm::Id>(sizeof(ValueType)) * this->GetStorage().GetNumberOfValues();
@ -334,10 +336,11 @@ ArrayHandle<T, StorageTagBasic>::PrepareForOutput(vtkm::Id numVals, DeviceAdapte
// the execution environment.
this->Internals->ControlArrayValid = false;
internal::TypelessExecutionArray execArray(
reinterpret_cast<void*&>(priv->ExecutionArray),
reinterpret_cast<void*&>(priv->ExecutionArrayEnd),
reinterpret_cast<void*&>(priv->ExecutionArrayCapacity));
internal::TypelessExecutionArray execArray(reinterpret_cast<void*&>(priv->ExecutionArray),
reinterpret_cast<void*&>(priv->ExecutionArrayEnd),
reinterpret_cast<void*&>(priv->ExecutionArrayCapacity),
this->Internals->ControlArray.GetBasePointer(),
this->Internals->ControlArray.GetCapacityPointer());
this->Internals->ExecutionInterface->Allocate(execArray,
static_cast<vtkm::Id>(sizeof(ValueType)) * numVals);
@ -370,7 +373,9 @@ ArrayHandle<T, StorageTagBasic>::PrepareForInPlace(DeviceAdapterTag device)
internal::TypelessExecutionArray execArray(
reinterpret_cast<void*&>(this->Internals->ExecutionArray),
reinterpret_cast<void*&>(this->Internals->ExecutionArrayEnd),
reinterpret_cast<void*&>(this->Internals->ExecutionArrayCapacity));
reinterpret_cast<void*&>(this->Internals->ExecutionArrayCapacity),
this->Internals->ControlArray.GetBasePointer(),
this->Internals->ControlArray.GetCapacityPointer());
vtkm::Id numBytes =
static_cast<vtkm::Id>(sizeof(ValueType)) * this->GetStorage().GetNumberOfValues();
@ -413,7 +418,9 @@ void ArrayHandle<T, StorageTagBasic>::PrepareForDevice(DeviceAdapterTag) const
internal::TypelessExecutionArray execArray(
reinterpret_cast<void*&>(priv->ExecutionArray),
reinterpret_cast<void*&>(priv->ExecutionArrayEnd),
reinterpret_cast<void*&>(priv->ExecutionArrayCapacity));
reinterpret_cast<void*&>(priv->ExecutionArrayCapacity),
this->Internals->ControlArray.GetBasePointer(),
this->Internals->ControlArray.GetCapacityPointer());
priv->ExecutionInterface->Free(execArray);
delete priv->ExecutionInterface;
priv->ExecutionInterface = nullptr;
@ -465,7 +472,9 @@ void ArrayHandle<T, StorageTagBasic>::ReleaseResourcesExecutionInternal()
internal::TypelessExecutionArray execArray(
reinterpret_cast<void*&>(this->Internals->ExecutionArray),
reinterpret_cast<void*&>(this->Internals->ExecutionArrayEnd),
reinterpret_cast<void*&>(this->Internals->ExecutionArrayCapacity));
reinterpret_cast<void*&>(this->Internals->ExecutionArrayCapacity),
this->Internals->ControlArray.GetBasePointer(),
this->Internals->ControlArray.GetCapacityPointer());
this->Internals->ExecutionInterface->Free(execArray);
this->Internals->ExecutionArrayValid = false;
}

18
vtkm/cont/internal/ArrayHandleExecutionManager.h
View File

@ -49,18 +49,18 @@ public:
typedef vtkm::cont::internal::ArrayTransfer<T, Storage, DeviceAdapter> ArrayTransferType;
public:
typedef typename ArrayTransferType::PortalExecution Portal;
typedef typename ArrayTransferType::PortalConstExecution PortalConst;
using Portal = typename ArrayTransferType::PortalExecution;
using PortalConst = typename ArrayTransferType::PortalConstExecution;
};
/// The type of value held in the array (vtkm::FloatDefault, vtkm::Vec, etc.)
///
typedef T ValueType;
using ValueType = T;
/// An array portal that can be used in the control environment.
///
typedef typename StorageType::PortalType PortalControl;
typedef typename StorageType::PortalConstType PortalConstControl;
using PortalControl = typename StorageType::PortalType;
using PortalConstControl = typename StorageType::PortalConstType;
VTKM_CONT
virtual ~ArrayHandleExecutionManagerBase() {}
@ -200,11 +200,11 @@ class ArrayHandleExecutionManager : public ArrayHandleExecutionManagerBase<T, St
typedef vtkm::cont::internal::Storage<T, Storage> StorageType;
public:
typedef typename ArrayTransferType::PortalControl PortalControl;
typedef typename ArrayTransferType::PortalConstControl PortalConstControl;
using PortalControl = typename ArrayTransferType::PortalControl;
using PortalConstControl = typename ArrayTransferType::PortalConstControl;
typedef typename ArrayTransferType::PortalExecution PortalExecution;
typedef typename ArrayTransferType::PortalConstExecution PortalConstExecution;
using PortalExecution = typename ArrayTransferType::PortalExecution;
using PortalConstExecution = typename ArrayTransferType::PortalConstExecution;
VTKM_CONT
ArrayHandleExecutionManager(StorageType* storage)

8
vtkm/cont/internal/ArrayManagerExecutionShareWithControl.h
View File

@ -44,10 +44,10 @@ template <typename T, class StorageTag>
class ArrayManagerExecutionShareWithControl
{
public:
typedef T ValueType;
using ValueType = T;
typedef vtkm::cont::internal::Storage<ValueType, StorageTag> StorageType;
typedef typename StorageType::PortalType PortalType;
typedef typename StorageType::PortalConstType PortalConstType;
using PortalType = typename StorageType::PortalType;
using PortalConstType = typename StorageType::PortalConstType;
VTKM_CONT
ArrayManagerExecutionShareWithControl(StorageType* storage)
@ -99,7 +99,7 @@ public:
template <class IteratorTypeControl>
VTKM_CONT void CopyInto(IteratorTypeControl dest) const
{
typedef typename StorageType::PortalConstType::IteratorType IteratorType;
using IteratorType = typename StorageType::PortalConstType::IteratorType;
IteratorType beginIterator = this->Storage->GetPortalConst().GetIteratorBegin();
std::copy(beginIterator, beginIterator + this->Storage->GetNumberOfValues(), dest);

14
vtkm/cont/internal/ArrayPortalFromIterators.h
View File

@ -50,8 +50,8 @@ class ArrayPortalFromIterators<IteratorT,
typename std::remove_pointer<IteratorT>::type>::value>::type>
{
public:
typedef typename std::iterator_traits<IteratorT>::value_type ValueType;
typedef IteratorT IteratorType;
using ValueType = typename std::iterator_traits<IteratorT>::value_type;
using IteratorType = IteratorT;
VTKM_SUPPRESS_EXEC_WARNINGS
VTKM_CONT
@ -124,8 +124,8 @@ class ArrayPortalFromIterators<IteratorT,
typename std::remove_pointer<IteratorT>::type>::value>::type>
{
public:
typedef typename std::iterator_traits<IteratorT>::value_type ValueType;
typedef IteratorT IteratorType;
using ValueType = typename std::iterator_traits<IteratorT>::value_type;
using IteratorType = IteratorT;
VTKM_SUPPRESS_EXEC_WARNINGS
VTKM_CONT
@ -216,11 +216,11 @@ namespace cont
template <typename _IteratorType>
class ArrayPortalToIterators<vtkm::cont::internal::ArrayPortalFromIterators<_IteratorType>>
{
typedef vtkm::cont::internal::ArrayPortalFromIterators<_IteratorType> PortalType;
using PortalType = vtkm::cont::internal::ArrayPortalFromIterators<_IteratorType>;
public:
#if !defined(VTKM_MSVC) || (defined(_ITERATOR_DEBUG_LEVEL) && _ITERATOR_DEBUG_LEVEL == 0)
typedef _IteratorType IteratorType;
using IteratorType = _IteratorType;
VTKM_SUPPRESS_EXEC_WARNINGS
VTKM_EXEC_CONT
@ -255,7 +255,7 @@ public:
IteratorType GetEnd() const
{
IteratorType iterator = this->Iterator;
typedef typename std::iterator_traits<IteratorType>::difference_type difference_type;
using difference_type = typename std::iterator_traits<IteratorType>::difference_type;
#if !defined(VTKM_MSVC) || (defined(_ITERATOR_DEBUG_LEVEL) && _ITERATOR_DEBUG_LEVEL == 0)
std::advance(iterator, static_cast<difference_type>(this->NumberOfValues));

10
vtkm/cont/internal/ArrayTransfer.h
View File

@ -57,17 +57,17 @@ private:
public:
/// The type of value held in the array (vtkm::FloatDefault, vtkm::Vec, etc.)
///
typedef T ValueType;
using ValueType = T;
/// An array portal that can be used in the control environment.
///
typedef typename StorageType::PortalType PortalControl;
typedef typename StorageType::PortalConstType PortalConstControl;
using PortalControl = typename StorageType::PortalType;
using PortalConstControl = typename StorageType::PortalConstType;
/// An array portal that can be used in the execution environment.
///
typedef typename ArrayManagerType::PortalType PortalExecution;
typedef typename ArrayManagerType::PortalConstType PortalConstExecution;
using PortalExecution = typename ArrayManagerType::PortalType;
using PortalConstExecution = typename ArrayManagerType::PortalConstType;
VTKM_CONT
ArrayTransfer(StorageType* storage)

115
vtkm/cont/internal/ConnectivityExplicitInternals.h
View File

@ -20,10 +20,16 @@
#ifndef vtk_m_cont_internal_ConnectivityExplicitInternals_h
#define vtk_m_cont_internal_ConnectivityExplicitInternals_h
#include <vtkm/CellShape.h>
#include <vtkm/cont/ArrayHandle.h>
#include <vtkm/cont/ArrayHandleCast.h>
#include <vtkm/cont/ArrayHandleConstant.h>
#include <vtkm/cont/ArrayHandleCounting.h>
#include <vtkm/cont/DeviceAdapterAlgorithm.h>
#include <vtkm/cont/internal/DeviceAdapterError.h>
#include <vtkm/exec/ExecutionWholeArray.h>
#include <vtkm/worklet/DispatcherMapField.h>
#include <vtkm/worklet/WorkletMapField.h>
namespace vtkm
{
@ -164,6 +170,115 @@ struct ConnectivityExplicitInternals
}
}
};
// Worklet to expand the PointToCell numIndices array by repeating cell index
class ExpandIndices : public vtkm::worklet::WorkletMapField
{
public:
typedef void ControlSignature(FieldIn<> cellIndex,
FieldIn<> offset,
FieldIn<> numIndices,
WholeArrayOut<> cellIndices);
typedef void ExecutionSignature(_1, _2, _3, _4);
using InputDomain = _1;
VTKM_CONT
ExpandIndices() {}
template <typename PortalType>
VTKM_EXEC void operator()(const vtkm::Id& cellIndex,
const vtkm::Id& offset,
const vtkm::Id& numIndices,
const PortalType& cellIndices) const
{
VTKM_ASSERT(cellIndices.GetNumberOfValues() >= offset + numIndices);
vtkm::Id startIndex = offset;
for (vtkm::Id i = 0; i < numIndices; i++)
{
cellIndices.Set(startIndex++, cellIndex);
}
}
};
class ScatterValues : public vtkm::worklet::WorkletMapField
{
public:
typedef void ControlSignature(FieldIn<> index, FieldIn<> value, WholeArrayOut<> output);
typedef void ExecutionSignature(_1, _2, _3);
using InputDomain = _1;
template <typename T, typename PortalType>
VTKM_EXEC void operator()(const vtkm::Id& index, const T& value, const PortalType& output) const
{
output.Set(index, value);
}
};
template <typename PointToCell, typename C2PShapeStorageTag, typename Device>
void ComputeCellToPointConnectivity(ConnectivityExplicitInternals<C2PShapeStorageTag>& cell2Point,
const PointToCell& point2Cell,
vtkm::Id numberOfPoints,
Device)
{
// PointToCell connectivity array (point indices) will be
// transformed into the CellToPoint numIndices array using reduction
//
// PointToCell numIndices array using expansion will be
// transformed into the CellToPoint connectivity array
if (cell2Point.ElementsValid)
{
return;
}
using Algorithm = vtkm::cont::DeviceAdapterAlgorithm<Device>;
// Sizes of the PointToCell information
vtkm::Id numberOfCells = point2Cell.NumIndices.GetNumberOfValues();
vtkm::Id connectivityLength = point2Cell.Connectivity.GetNumberOfValues();
// PointToCell connectivity will be basis of CellToPoint numIndices
vtkm::cont::ArrayHandle<vtkm::Id> pointIndices;
Algorithm::Copy(point2Cell.Connectivity, pointIndices);
// PointToCell numIndices will be basis of CellToPoint connectivity
cell2Point.Connectivity.Allocate(connectivityLength);
vtkm::cont::ArrayHandleCounting<vtkm::Id> index(0, 1, numberOfCells);
vtkm::worklet::DispatcherMapField<ExpandIndices, Device> expandDispatcher;
expandDispatcher.Invoke(
index, point2Cell.IndexOffsets, point2Cell.NumIndices, cell2Point.Connectivity);
// SortByKey where key is PointToCell connectivity and value is the expanded cellIndex
Algorithm::SortByKey(pointIndices, cell2Point.Connectivity);
// CellToPoint numIndices from the now sorted PointToCell connectivity
vtkm::cont::ArrayHandleConstant<vtkm::IdComponent> numArray(1, connectivityLength);
vtkm::cont::ArrayHandle<vtkm::Id> uniquePoints;
vtkm::cont::ArrayHandle<vtkm::IdComponent> numIndices;
Algorithm::ReduceByKey(pointIndices, numArray, uniquePoints, numIndices, vtkm::Add());
// if not all the points have a cell
if (uniquePoints.GetNumberOfValues() < numberOfPoints)
{
vtkm::cont::ArrayHandle<vtkm::IdComponent> fullNumIndices;
Algorithm::Copy(vtkm::cont::ArrayHandleConstant<vtkm::IdComponent>(0, numberOfPoints),
fullNumIndices);
vtkm::worklet::DispatcherMapField<ScatterValues, Device>().Invoke(
uniquePoints, numIndices, fullNumIndices);
numIndices = fullNumIndices;
}
// Set the CellToPoint information
cell2Point.Shapes = vtkm::cont::make_ArrayHandleConstant(
static_cast<vtkm::UInt8>(CELL_SHAPE_VERTEX), numberOfPoints);
cell2Point.NumIndices = numIndices;
cell2Point.ElementsValid = true;
cell2Point.IndexOffsetsValid = false;
}
}
}
} // namespace vtkm::cont::internal

50
vtkm/cont/internal/DeviceAdapterAlgorithmGeneral.h
View File

@ -155,8 +155,8 @@ public:
DeviceAdapterTag>::PortalConst StencilPortalType;
StencilPortalType stencilPortal = stencil.PrepareForInput(DeviceAdapterTag());
typedef
typename IndexArrayType::template ExecutionTypes<DeviceAdapterTag>::Portal IndexPortalType;
using IndexPortalType =
typename IndexArrayType::template ExecutionTypes<DeviceAdapterTag>::Portal;
IndexPortalType indexPortal = indices.PrepareForOutput(arrayLength, DeviceAdapterTag());
StencilToIndexFlagKernel<StencilPortalType, IndexPortalType, UnaryPredicate> indexKernel(
@ -328,7 +328,7 @@ public:
typedef ReduceKernel<InputPortalType, U, BinaryFunctor> ReduceKernelType;
typedef vtkm::cont::ArrayHandleImplicit<ReduceKernelType> ReduceHandleType;
using ReduceHandleType = vtkm::cont::ArrayHandleImplicit<ReduceKernelType>;
typedef vtkm::cont::ArrayHandle<U, vtkm::cont::StorageTagBasic> TempArrayType;
ReduceKernelType kernel(
@ -418,8 +418,8 @@ public:
typedef typename vtkm::cont::ArrayHandle<T, KIn>::template ExecutionTypes<
DeviceAdapterTag>::PortalConst InputPortalType;
typedef typename vtkm::cont::ArrayHandle<ReduceKeySeriesStates>::template ExecutionTypes<
DeviceAdapterTag>::Portal KeyStatePortalType;
using KeyStatePortalType = typename vtkm::cont::ArrayHandle<
ReduceKeySeriesStates>::template ExecutionTypes<DeviceAdapterTag>::Portal;
InputPortalType inputPortal = keys.PrepareForInput(DeviceAdapterTag());
KeyStatePortalType keyStatePortal =
@ -438,7 +438,7 @@ public:
{
typedef vtkm::cont::ArrayHandle<U, VIn> ValueInHandleType;
typedef vtkm::cont::ArrayHandle<U, VOut> ValueOutHandleType;
typedef vtkm::cont::ArrayHandle<ReduceKeySeriesStates> StencilHandleType;
using StencilHandleType = vtkm::cont::ArrayHandle<ReduceKeySeriesStates>;
typedef vtkm::cont::ArrayHandleZip<ValueInHandleType, StencilHandleType> ZipInHandleType;
typedef vtkm::cont::ArrayHandleZip<ValueOutHandleType, StencilHandleType> ZipOutHandleType;
@ -477,10 +477,10 @@ public:
typedef vtkm::cont::ArrayHandle<T, vtkm::cont::StorageTagBasic> TempArrayType;
typedef vtkm::cont::ArrayHandle<T, COut> OutputArrayType;
typedef
typename TempArrayType::template ExecutionTypes<DeviceAdapterTag>::PortalConst SrcPortalType;
typedef
typename OutputArrayType::template ExecutionTypes<DeviceAdapterTag>::Portal DestPortalType;
using SrcPortalType =
typename TempArrayType::template ExecutionTypes<DeviceAdapterTag>::PortalConst;
using DestPortalType =
typename OutputArrayType::template ExecutionTypes<DeviceAdapterTag>::Portal;
vtkm::Id numValues = input.GetNumberOfValues();
if (numValues <= 0)
@ -545,8 +545,8 @@ public:
typedef typename vtkm::cont::ArrayHandle<T, KIn>::template ExecutionTypes<
DeviceAdapterTag>::PortalConst InputPortalType;
typedef typename vtkm::cont::ArrayHandle<ReduceKeySeriesStates>::template ExecutionTypes<
DeviceAdapterTag>::Portal KeyStatePortalType;
using KeyStatePortalType = typename vtkm::cont::ArrayHandle<
ReduceKeySeriesStates>::template ExecutionTypes<DeviceAdapterTag>::Portal;
InputPortalType inputPortal = keys.PrepareForInput(DeviceAdapterTag());
KeyStatePortalType keyStatePortal =
@ -566,8 +566,8 @@ public:
typedef typename vtkm::cont::ArrayHandle<T, KIn>::template ExecutionTypes<
DeviceAdapterTag>::PortalConst InputPortalType;
typedef typename vtkm::cont::ArrayHandle<ReduceKeySeriesStates>::template ExecutionTypes<
DeviceAdapterTag>::PortalConst KeyStatePortalType;
using KeyStatePortalType = typename vtkm::cont::ArrayHandle<
ReduceKeySeriesStates>::template ExecutionTypes<DeviceAdapterTag>::PortalConst;
InputPortalType inputPortal = values.PrepareForInput(DeviceAdapterTag());
KeyStatePortalType keyStatePortal = keystate.PrepareForInput(DeviceAdapterTag());
@ -724,8 +724,8 @@ public:
typedef typename vtkm::cont::ArrayHandle<T, KIn>::template ExecutionTypes<
DeviceAdapterTag>::PortalConst InputPortalType;
typedef typename vtkm::cont::ArrayHandle<ReduceKeySeriesStates>::template ExecutionTypes<
DeviceAdapterTag>::Portal KeyStatePortalType;
using KeyStatePortalType = typename vtkm::cont::ArrayHandle<
ReduceKeySeriesStates>::template ExecutionTypes<DeviceAdapterTag>::Portal;
InputPortalType inputPortal = keys.PrepareForInput(DeviceAdapterTag());
KeyStatePortalType keyStatePortal =
@ -744,7 +744,7 @@ public:
{
typedef vtkm::cont::ArrayHandle<U, VIn> ValueInHandleType;
typedef vtkm::cont::ArrayHandle<U, VOut> ValueOutHandleType;
typedef vtkm::cont::ArrayHandle<ReduceKeySeriesStates> StencilHandleType;
using StencilHandleType = vtkm::cont::ArrayHandle<ReduceKeySeriesStates>;
typedef vtkm::cont::ArrayHandleZip<ValueInHandleType, StencilHandleType> ZipInHandleType;
typedef vtkm::cont::ArrayHandleZip<ValueOutHandleType, StencilHandleType> ZipOutHandleType;
@ -770,8 +770,8 @@ public:
VTKM_CONT static void Sort(vtkm::cont::ArrayHandle<T, Storage>& values,
BinaryCompare binary_compare)
{
typedef typename vtkm::cont::ArrayHandle<T, Storage> ArrayType;
typedef typename ArrayType::template ExecutionTypes<DeviceAdapterTag>::Portal PortalType;
using ArrayType = typename vtkm::cont::ArrayHandle<T, Storage>;
using PortalType = typename ArrayType::template ExecutionTypes<DeviceAdapterTag>::Portal;
vtkm::Id numValues = values.GetNumberOfValues();
if (numValues < 2)
@ -818,8 +818,9 @@ public:
//combine the keys and values into a ZipArrayHandle
//we than need to specify a custom compare function wrapper
//that only checks for key side of the pair, using a custom compare functor.
typedef vtkm::cont::ArrayHandle<T, StorageT> KeyType;
typedef vtkm::cont::ArrayHandle<U, StorageU> ValueType;
using KeyType = vtkm::cont::ArrayHandle<T, StorageT>;
;
using ValueType = vtkm::cont::ArrayHandle<U, StorageU>;
typedef vtkm::cont::ArrayHandleZip<KeyType, ValueType> ZipHandleType;
ZipHandleType zipHandle = vtkm::cont::make_ArrayHandleZip(keys, values);
@ -835,8 +836,9 @@ public:
//we than need to specify a custom compare function wrapper
//that only checks for key side of the pair, using the custom compare
//functor that the user passed in
typedef vtkm::cont::ArrayHandle<T, StorageT> KeyType;
typedef vtkm::cont::ArrayHandle<U, StorageU> ValueType;
using KeyType = vtkm::cont::ArrayHandle<T, StorageT>;
;
using ValueType = vtkm::cont::ArrayHandle<U, StorageU>;
typedef vtkm::cont::ArrayHandleZip<KeyType, ValueType> ZipHandleType;
ZipHandleType zipHandle = vtkm::cont::make_ArrayHandleZip(keys, values);
@ -997,7 +999,7 @@ public:
private:
typedef typename vtkm::cont::ArrayHandle<T, vtkm::cont::StorageTagBasic>::template ExecutionTypes<
DeviceTag>::Portal PortalType;
typedef vtkm::cont::ArrayPortalToIterators<PortalType> IteratorsType;
using IteratorsType = vtkm::cont::ArrayPortalToIterators<PortalType>;
IteratorsType Iterators;
#if defined(VTKM_MSVC) //MSVC atomics

4
vtkm/cont/internal/DeviceAdapterTag.h
View File

@ -38,8 +38,8 @@ namespace vtkm
namespace cont
{
typedef vtkm::Int8 DeviceAdapterId;
typedef std::string DeviceAdapterNameType;
using DeviceAdapterId = vtkm::Int8;
using DeviceAdapterNameType = std::string;
template <typename DeviceAdapter>
struct DeviceAdapterTraits;

2
vtkm/cont/internal/DynamicTransform.h
View File

@ -129,7 +129,7 @@ struct DynamicTransformTraits
/// should specialize this class redefine it to \c
/// DynamicTransformTagCastAndCall.
///
typedef vtkm::cont::internal::DynamicTransformTagStatic DynamicTag;
using DynamicTag = vtkm::cont::internal::DynamicTransformTagStatic;
};
/// This functor can be used as the transform in the \c DynamicTransformCont

34
vtkm/cont/internal/FunctorsGeneral.h
View File

@ -211,8 +211,8 @@ struct ReduceStencilGeneration : vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id centerIndex) const
{
typedef typename InputPortalType::ValueType ValueType;
typedef typename KeyStatePortalType::ValueType KeyStateType;
using ValueType = typename InputPortalType::ValueType;
using KeyStateType = typename KeyStatePortalType::ValueType;
const vtkm::Id leftIndex = centerIndex - 1;
const vtkm::Id rightIndex = centerIndex + 1;
@ -356,7 +356,7 @@ struct CopyKernel
VTKM_EXEC_CONT
void operator()(vtkm::Id index) const
{
typedef typename OutputPortalType::ValueType ValueType;
using ValueType = typename OutputPortalType::ValueType;
this->OutputPortal.Set(
index + this->OutputOffset,
static_cast<ValueType>(this->InputPortal.Get(index + this->InputOffset)));
@ -394,7 +394,7 @@ struct LowerBoundsKernel
// necessarily true, but it is true for the current uses of this general
// function and I don't want to compete with STL if I don't have to.
typedef vtkm::cont::ArrayPortalToIterators<InputPortalType> InputIteratorsType;
using InputIteratorsType = vtkm::cont::ArrayPortalToIterators<InputPortalType>;
InputIteratorsType inputIterators(this->InputPortal);
typename InputIteratorsType::IteratorType resultPos = std::lower_bound(
inputIterators.GetBegin(), inputIterators.GetEnd(), this->ValuesPortal.Get(index));
@ -442,7 +442,7 @@ struct LowerBoundsComparisonKernel
// necessarily true, but it is true for the current uses of this general
// function and I don't want to compete with STL if I don't have to.
typedef vtkm::cont::ArrayPortalToIterators<InputPortalType> InputIteratorsType;
using InputIteratorsType = vtkm::cont::ArrayPortalToIterators<InputPortalType>;
InputIteratorsType inputIterators(this->InputPortal);
typename InputIteratorsType::IteratorType resultPos =
std::lower_bound(inputIterators.GetBegin(),
@ -462,7 +462,7 @@ struct LowerBoundsComparisonKernel
template <typename PortalType>
struct SetConstantKernel
{
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
PortalType Portal;
ValueType Value;
@ -500,7 +500,7 @@ struct BitonicSortMergeKernel : vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
vtkm::Id groupIndex = index % this->GroupSize;
vtkm::Id blockSize = 2 * this->GroupSize;
@ -543,7 +543,7 @@ struct BitonicSortCrossoverKernel : vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
vtkm::Id groupIndex = index % this->GroupSize;
vtkm::Id blockSize = 2 * this->GroupSize;
@ -568,7 +568,7 @@ struct BitonicSortCrossoverKernel : vtkm::exec::FunctorBase
template <class StencilPortalType, class OutputPortalType, class UnaryPredicate>
struct StencilToIndexFlagKernel
{
typedef typename StencilPortalType::ValueType StencilValueType;
using StencilValueType = typename StencilPortalType::ValueType;
StencilPortalType StencilPortal;
OutputPortalType OutputPortal;
UnaryPredicate Predicate;
@ -626,13 +626,13 @@ struct CopyIfKernel
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename StencilPortalType::ValueType StencilValueType;
using StencilValueType = typename StencilPortalType::ValueType;
StencilValueType stencilValue = this->StencilPortal.Get(index);
if (Predicate(stencilValue))
{
vtkm::Id outputIndex = this->IndexPortal.Get(index);
typedef typename OutputPortalType::ValueType OutputValueType;
using OutputValueType = typename OutputPortalType::ValueType;
OutputValueType value = this->InputPortal.Get(index);
this->OutputPortal.Set(outputIndex, value);
@ -660,7 +660,7 @@ struct ClassifyUniqueKernel
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename StencilPortalType::ValueType ValueType;
using ValueType = typename StencilPortalType::ValueType;
if (index == 0)
{
// Always copy first value.
@ -698,7 +698,7 @@ struct ClassifyUniqueComparisonKernel
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename StencilPortalType::ValueType ValueType;
using ValueType = typename StencilPortalType::ValueType;
if (index == 0)
{
// Always copy first value.
@ -746,7 +746,7 @@ struct UpperBoundsKernel
// necessarily true, but it is true for the current uses of this general
// function and I don't want to compete with STL if I don't have to.
typedef vtkm::cont::ArrayPortalToIterators<InputPortalType> InputIteratorsType;
using InputIteratorsType = vtkm::cont::ArrayPortalToIterators<InputPortalType>;
InputIteratorsType inputIterators(this->InputPortal);
typename InputIteratorsType::IteratorType resultPos = std::upper_bound(
inputIterators.GetBegin(), inputIterators.GetEnd(), this->ValuesPortal.Get(index));
@ -794,7 +794,7 @@ struct UpperBoundsKernelComparisonKernel
// necessarily true, but it is true for the current uses of this general
// function and I don't want to compete with STL if I don't have to.
typedef vtkm::cont::ArrayPortalToIterators<InputPortalType> InputIteratorsType;
using InputIteratorsType = vtkm::cont::ArrayPortalToIterators<InputPortalType>;
InputIteratorsType inputIterators(this->InputPortal);
typename InputIteratorsType::IteratorType resultPos =
std::upper_bound(inputIterators.GetBegin(),
@ -814,7 +814,7 @@ struct UpperBoundsKernelComparisonKernel
template <typename InPortalType, typename OutPortalType, typename BinaryFunctor>
struct InclusiveToExclusiveKernel : vtkm::exec::FunctorBase
{
typedef typename InPortalType::ValueType ValueType;
using ValueType = typename InPortalType::ValueType;
InPortalType InPortal;
OutPortalType OutPortal;
@ -870,7 +870,7 @@ struct ScanKernel : vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename PortalType::ValueType ValueType;
using ValueType = typename PortalType::ValueType;
vtkm::Id leftIndex = this->Offset + index * this->Stride;
vtkm::Id rightIndex = leftIndex + this->Distance;

6
vtkm/cont/serial/internal/ArrayManagerExecutionSerial.h
View File

@ -38,9 +38,9 @@ class ArrayManagerExecution<T, StorageTag, vtkm::cont::DeviceAdapterTagSerial>
{
public:
typedef vtkm::cont::internal::ArrayManagerExecutionShareWithControl<T, StorageTag> Superclass;
typedef typename Superclass::ValueType ValueType;
typedef typename Superclass::PortalType PortalType;
typedef typename Superclass::PortalConstType PortalConstType;
using ValueType = typename Superclass::ValueType;
using PortalType = typename Superclass::PortalType;
using PortalConstType = typename Superclass::PortalConstType;
VTKM_CONT
ArrayManagerExecution(typename Superclass::StorageType* storage)

11
vtkm/cont/serial/internal/DeviceAdapterAlgorithmSerial.h
View File

@ -47,7 +47,7 @@ struct DeviceAdapterAlgorithm<vtkm::cont::DeviceAdapterTagSerial>
vtkm::cont::DeviceAdapterTagSerial>
{
private:
typedef vtkm::cont::DeviceAdapterTagSerial Device;
using Device = vtkm::cont::DeviceAdapterTagSerial;
public:
template <typename T, typename U, class CIn>
@ -310,8 +310,9 @@ private:
//we than need to specify a custom compare function wrapper
//that only checks for key side of the pair, using the custom compare
//functor that the user passed in
typedef vtkm::cont::ArrayHandle<T, StorageT> KeyType;
typedef vtkm::cont::ArrayHandle<U, StorageU> ValueType;
using KeyType = vtkm::cont::ArrayHandle<T, StorageT>;
;
using ValueType = vtkm::cont::ArrayHandle<U, StorageU>;
typedef vtkm::cont::ArrayHandleZip<KeyType, ValueType> ZipHandleType;
ZipHandleType zipHandle = vtkm::cont::make_ArrayHandleZip(keys, values);
@ -336,8 +337,8 @@ public:
{
/// More efficient sort:
/// Move value indexes when sorting and reorder the value array at last
typedef vtkm::cont::ArrayHandle<U, StorageU> ValueType;
typedef vtkm::cont::ArrayHandle<vtkm::Id> IndexType;
using ValueType = vtkm::cont::ArrayHandle<U, StorageU>;
using IndexType = vtkm::cont::ArrayHandle<vtkm::Id>;
IndexType indexArray;
ValueType valuesScattered;

8
vtkm/cont/tbb/internal/ArrayManagerExecutionTBB.h
View File

@ -42,10 +42,10 @@ class ArrayManagerExecution<T, StorageTag, vtkm::cont::DeviceAdapterTagTBB>
{
public:
typedef vtkm::cont::internal::ArrayManagerExecutionShareWithControl<T, StorageTag> Superclass;
typedef typename Superclass::ValueType ValueType;
typedef typename Superclass::PortalType PortalType;
typedef typename Superclass::PortalConstType PortalConstType;
typedef typename Superclass::StorageType StorageType;
using ValueType = typename Superclass::ValueType;
using PortalType = typename Superclass::PortalType;
using PortalConstType = typename Superclass::PortalConstType;
using StorageType = typename Superclass::StorageType;
VTKM_CONT
ArrayManagerExecution(StorageType* storage)

11
vtkm/cont/tbb/internal/DeviceAdapterAlgorithmTBB.h
View File

@ -140,7 +140,7 @@ public:
vtkm::cont::DeviceAdapterTagTBB>::Portal PortalType;
PortalType arrayPortal = values.PrepareForInPlace(vtkm::cont::DeviceAdapterTagTBB());
typedef vtkm::cont::ArrayPortalToIterators<PortalType> IteratorsType;
using IteratorsType = vtkm::cont::ArrayPortalToIterators<PortalType>;
IteratorsType iterators(arrayPortal);
internal::WrappedBinaryOperator<bool, BinaryCompare> wrappedCompare(binary_compare);
@ -159,14 +159,15 @@ public:
vtkm::cont::ArrayHandle<U, StorageU>& values,
Compare comp)
{
typedef vtkm::cont::ArrayHandle<T, StorageT> KeyType;
using KeyType = vtkm::cont::ArrayHandle<T, StorageT>;
;
if (sizeof(U) > sizeof(vtkm::Id))
{
/// More efficient sort:
/// Move value indexes when sorting and reorder the value array at last
typedef vtkm::cont::ArrayHandle<U, StorageU> ValueType;
typedef vtkm::cont::ArrayHandle<vtkm::Id> IndexType;
using ValueType = vtkm::cont::ArrayHandle<U, StorageU>;
using IndexType = vtkm::cont::ArrayHandle<vtkm::Id>;
typedef vtkm::cont::ArrayHandleZip<KeyType, IndexType> ZipHandleType;
IndexType indexArray;
@ -186,7 +187,7 @@ public:
}
else
{
typedef vtkm::cont::ArrayHandle<U, StorageU> ValueType;
using ValueType = vtkm::cont::ArrayHandle<U, StorageU>;
typedef vtkm::cont::ArrayHandleZip<KeyType, ValueType> ZipHandleType;
ZipHandleType zipHandle = vtkm::cont::make_ArrayHandleZip(keys, values);

14
vtkm/cont/tbb/internal/FunctorsTBB.h
View File

@ -120,7 +120,7 @@ struct ReduceBody
VTKM_EXEC
void operator()(const ::tbb::blocked_range<vtkm::Id>& range)
{
typedef vtkm::cont::ArrayPortalToIterators<InputPortalType> InputIteratorsType;
using InputIteratorsType = vtkm::cont::ArrayPortalToIterators<InputPortalType>;
InputIteratorsType inputIterators(this->InputPortal);
//use temp, and iterators instead of member variable to reduce false sharing
@ -231,7 +231,7 @@ struct ScanInclusiveBody
VTKM_EXEC
void operator()(const ::tbb::blocked_range<vtkm::Id>& range, ::tbb::pre_scan_tag)
{
typedef vtkm::cont::ArrayPortalToIterators<InputPortalType> InputIteratorsType;
using InputIteratorsType = vtkm::cont::ArrayPortalToIterators<InputPortalType>;
InputIteratorsType inputIterators(this->InputPortal);
//use temp, and iterators instead of member variable to reduce false sharing
@ -250,8 +250,8 @@ struct ScanInclusiveBody
VTKM_EXEC
void operator()(const ::tbb::blocked_range<vtkm::Id>& range, ::tbb::final_scan_tag)
{
typedef vtkm::cont::ArrayPortalToIterators<InputPortalType> InputIteratorsType;
typedef vtkm::cont::ArrayPortalToIterators<OutputPortalType> OutputIteratorsType;
using InputIteratorsType = vtkm::cont::ArrayPortalToIterators<InputPortalType>;
using OutputIteratorsType = vtkm::cont::ArrayPortalToIterators<OutputPortalType>;
InputIteratorsType inputIterators(this->InputPortal);
OutputIteratorsType outputIterators(this->OutputPortal);
@ -321,7 +321,7 @@ struct ScanExclusiveBody
VTKM_EXEC
void operator()(const ::tbb::blocked_range<vtkm::Id>& range, ::tbb::pre_scan_tag)
{
typedef vtkm::cont::ArrayPortalToIterators<InputPortalType> InputIteratorsType;
using InputIteratorsType = vtkm::cont::ArrayPortalToIterators<InputPortalType>;
InputIteratorsType inputIterators(this->InputPortal);
//move the iterator to the first item
@ -346,8 +346,8 @@ struct ScanExclusiveBody
VTKM_EXEC
void operator()(const ::tbb::blocked_range<vtkm::Id>& range, ::tbb::final_scan_tag)
{
typedef vtkm::cont::ArrayPortalToIterators<InputPortalType> InputIteratorsType;
typedef vtkm::cont::ArrayPortalToIterators<OutputPortalType> OutputIteratorsType;
using InputIteratorsType = vtkm::cont::ArrayPortalToIterators<InputPortalType>;
using OutputIteratorsType = vtkm::cont::ArrayPortalToIterators<OutputPortalType>;
InputIteratorsType inputIterators(this->InputPortal);
OutputIteratorsType outputIterators(this->OutputPortal);

4
vtkm/cont/testing/CMakeLists.txt
View File

@ -39,14 +39,18 @@ set(unit_tests
UnitTestArrayHandleCompositeVector.cxx
UnitTestArrayHandleCounting.cxx
UnitTestArrayHandleDiscard.cxx
UnitTestArrayHandleExtractComponent.cxx
UnitTestArrayHandleImplicit.cxx
UnitTestArrayHandleIndex.cxx
UnitTestArrayHandleReverse.cxx
UnitTestArrayHandlePermutation.cxx
UnitTestArrayHandleSwizzle.cxx
UnitTestArrayHandleTransform.cxx
UnitTestArrayHandleUniformPointCoordinates.cxx
UnitTestArrayHandleConcatenate.cxx
UnitTestArrayPortalToIterators.cxx
UnitTestCellSetExplicit.cxx
UnitTestCellSetPermutation.cxx
UnitTestContTesting.cxx
UnitTestDataSetBuilderExplicit.cxx
UnitTestDataSetBuilderRectilinear.cxx

78
vtkm/cont/testing/MakeTestDataSet.h
View File

@ -69,6 +69,7 @@ public:
vtkm::cont::DataSet Make3DExplicitDataSet3();
vtkm::cont::DataSet Make3DExplicitDataSet4();
vtkm::cont::DataSet Make3DExplicitDataSet5();
vtkm::cont::DataSet Make3DExplicitDataSet6();
vtkm::cont::DataSet Make3DExplicitDataSetPolygonal();
vtkm::cont::DataSet Make3DExplicitDataSetCowNose();
};
@ -777,6 +778,83 @@ inline vtkm::cont::DataSet MakeTestDataSet::Make3DExplicitDataSet5()
return dataSet;
}
inline vtkm::cont::DataSet MakeTestDataSet::Make3DExplicitDataSet6()
{
vtkm::cont::DataSet dataSet;
vtkm::cont::DataSetBuilderExplicit dsb;
vtkm::cont::DataSetFieldAdd dsf;
// Coordinates
const int nVerts = 8;
const int nCells = 8;
typedef vtkm::Vec<vtkm::Float32, 3> CoordType;
std::vector<CoordType> coords = { { -0.707f, -0.354f, -0.354f }, { 0.000f, -0.854f, 0.146f },
{ 0.000f, -0.146f, 0.854f }, { -0.707f, 0.354f, 0.354f },
{ 10.0f, 10.0f, 10.0f }, { 5.0f, 5.0f, 5.0f },
{ 0.0f, 0.0f, 2.0f }, { 0.0f, 0.0f, -2.0f } };
// Connectivity
std::vector<vtkm::UInt8> shapes;
std::vector<vtkm::IdComponent> numindices;
std::vector<vtkm::Id> conn;
shapes.push_back(vtkm::CELL_SHAPE_LINE);
numindices.push_back(2);
conn.push_back(0);
conn.push_back(1);
shapes.push_back(vtkm::CELL_SHAPE_LINE);
numindices.push_back(2);
conn.push_back(2);
conn.push_back(3);
shapes.push_back(vtkm::CELL_SHAPE_VERTEX);
numindices.push_back(1);
conn.push_back(4);
shapes.push_back(vtkm::CELL_SHAPE_VERTEX);
numindices.push_back(1);
conn.push_back(5);
shapes.push_back(vtkm::CELL_SHAPE_TRIANGLE);
numindices.push_back(3);
conn.push_back(2);
conn.push_back(3);
conn.push_back(5);
shapes.push_back(vtkm::CELL_SHAPE_QUAD);
numindices.push_back(4);
conn.push_back(0);
conn.push_back(1);
conn.push_back(2);
conn.push_back(3);
shapes.push_back(vtkm::CELL_SHAPE_TETRA);
numindices.push_back(4);
conn.push_back(0);
conn.push_back(2);
conn.push_back(3);
conn.push_back(6);
shapes.push_back(vtkm::CELL_SHAPE_TETRA);
numindices.push_back(4);
conn.push_back(3);
conn.push_back(2);
conn.push_back(0);
conn.push_back(7);
dataSet = dsb.Create(coords, shapes, numindices, conn, "coordinates", "cells");
// Field data
vtkm::Float32 pointvar[nVerts] = { 100.0f, 78.0f, 49.0f, 17.0f, 94.0f, 71.0f, 47.0f, 57.0f };
vtkm::Float32 cellvar[nCells] = { 0.0f, 1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f };
dsf.AddPointField(dataSet, "pointvar", pointvar, nVerts);
dsf.AddCellField(dataSet, "cellvar", cellvar, nCells, "cells");
return dataSet;
}
inline vtkm::cont::DataSet MakeTestDataSet::Make3DExplicitDataSetPolygonal()
{
vtkm::cont::DataSet dataSet;

12
vtkm/cont/testing/TestingArrayHandles.h
View File

@ -121,7 +121,7 @@ struct TestingArrayHandles
private:
static const vtkm::Id ARRAY_SIZE = 100;
typedef vtkm::cont::DeviceAdapterAlgorithm<DeviceAdapterTag> Algorithm;
using Algorithm = vtkm::cont::DeviceAdapterAlgorithm<DeviceAdapterTag>;
typedef vtkm::worklet::DispatcherMapField<PassThrough, DeviceAdapterTag> DispatcherPassThrough;
struct VerifyEmptyArrays
@ -274,9 +274,8 @@ private:
VTKM_TEST_ASSERT(arrayHandle.GetNumberOfValues() == 0,
"ArrayHandle has wrong number of entries.");
{
typedef
typename vtkm::cont::ArrayHandle<T>::template ExecutionTypes<DeviceAdapterTag>::Portal
ExecutionPortalType;
using ExecutionPortalType =
typename vtkm::cont::ArrayHandle<T>::template ExecutionTypes<DeviceAdapterTag>::Portal;
ExecutionPortalType executionPortal =
arrayHandle.PrepareForOutput(ARRAY_SIZE * 2, DeviceAdapterTag());
@ -304,9 +303,8 @@ private:
std::cout << "Try in place operation." << std::endl;
{
typedef
typename vtkm::cont::ArrayHandle<T>::template ExecutionTypes<DeviceAdapterTag>::Portal
ExecutionPortalType;
using ExecutionPortalType =
typename vtkm::cont::ArrayHandle<T>::template ExecutionTypes<DeviceAdapterTag>::Portal;
ExecutionPortalType executionPortal = arrayHandle.PrepareForInPlace(DeviceAdapterTag());
//in place can't be done through the dispatcher

20
vtkm/cont/testing/TestingDeviceAdapter.h
View File

@ -70,7 +70,7 @@ template <class DeviceAdapterTag>
struct TestingDeviceAdapter
{
private:
typedef vtkm::cont::StorageTagBasic StorageTag;
using StorageTag = vtkm::cont::StorageTagBasic;
typedef vtkm::cont::ArrayHandle<vtkm::Id, StorageTag> IdArrayHandle;
@ -81,11 +81,11 @@ private:
typedef vtkm::cont::internal::Storage<vtkm::Id, StorageTag> IdStorage;
typedef typename IdArrayHandle::template ExecutionTypes<DeviceAdapterTag>::Portal IdPortalType;
typedef typename IdArrayHandle::template ExecutionTypes<DeviceAdapterTag>::PortalConst
IdPortalConstType;
using IdPortalType = typename IdArrayHandle::template ExecutionTypes<DeviceAdapterTag>::Portal;
using IdPortalConstType =
typename IdArrayHandle::template ExecutionTypes<DeviceAdapterTag>::PortalConst;
typedef vtkm::cont::DeviceAdapterAlgorithm<DeviceAdapterTag> Algorithm;
using Algorithm = vtkm::cont::DeviceAdapterAlgorithm<DeviceAdapterTag>;
public:
// Cuda kernels have to be public (in Cuda 4.0).
@ -362,8 +362,8 @@ private:
std::cout << "-------------------------------------------" << std::endl;
std::cout << "Testing device adapter tag" << std::endl;
typedef vtkm::cont::DeviceAdapterTraits<DeviceAdapterTag> Traits;
typedef vtkm::cont::DeviceAdapterTraits<vtkm::cont::DeviceAdapterTagError> ErrorTraits;
using Traits = vtkm::cont::DeviceAdapterTraits<DeviceAdapterTag>;
using ErrorTraits = vtkm::cont::DeviceAdapterTraits<vtkm::cont::DeviceAdapterTagError>;
VTKM_TEST_ASSERT(Traits::GetId() == Traits::GetId(),
"Device adapter Id does not equal itself.");
@ -1150,7 +1150,7 @@ private:
//and a custom reduce binary functor
const vtkm::Id indexLength = 30;
const vtkm::Id valuesLength = 10;
typedef vtkm::Float32 ValueType;
using ValueType = vtkm::Float32;
vtkm::Id indexs[indexLength] = { 0, 0, 0, 1, 1, 1, 2, 2, 2, 3, 3, 3, 4, 4, 4,
5, 5, 5, 1, 4, 9, 7, 7, 7, 8, 8, 8, 0, 1, 2 };
@ -1256,7 +1256,7 @@ private:
//and a custom reduce binary functor
const vtkm::Id inputLength = 30;
const vtkm::Id expectedLength = 10;
typedef vtkm::Float32 ValueType;
using ValueType = vtkm::Float32;
vtkm::Id inputKeys[inputLength] = { 0, 0, 0, 1, 1, 1, 2, 2, 2, 3, 3, 3, 4, 4, 4,
5, 5, 5, 6, 6, 6, 7, 7, 7, 8, 8, 8, 9, 9, 9 }; // input keys
ValueType inputValues1[inputLength] = {
@ -1274,7 +1274,7 @@ private:
IdArrayHandle keys = vtkm::cont::make_ArrayHandle(inputKeys, inputLength);
typedef vtkm::cont::ArrayHandle<ValueType, StorageTag> ValueArrayType;
ValueArrayType values1 = vtkm::cont::make_ArrayHandle(inputValues1, inputLength);
typedef vtkm::cont::ArrayHandleConstant<ValueType> ConstValueArrayType;
using ConstValueArrayType = vtkm::cont::ArrayHandleConstant<ValueType>;
ConstValueArrayType constOneArray(1.f, inputLength);
vtkm::cont::ArrayHandleZip<ValueArrayType, ConstValueArrayType> valuesZip;

2
vtkm/cont/testing/UnitTestArrayHandleCartesianProduct.cxx
View File

@ -32,7 +32,7 @@
namespace ArrayHandleCartesianProductNamespace
{
typedef vtkm::cont::DeviceAdapterAlgorithm<VTKM_DEFAULT_DEVICE_ADAPTER_TAG> DFA;
using DFA = vtkm::cont::DeviceAdapterAlgorithm<VTKM_DEFAULT_DEVICE_ADAPTER_TAG>;
template <typename T>
void ArrayHandleCPBasic(vtkm::cont::ArrayHandle<T> x,

12
vtkm/cont/testing/UnitTestArrayHandleCompositeVector.cxx
View File

@ -41,7 +41,7 @@ namespace
const vtkm::Id ARRAY_SIZE = 10;
typedef vtkm::cont::StorageTagBasic StorageTag;
using StorageTag = vtkm::cont::StorageTagBasic;
vtkm::FloatDefault TestValue3Ids(vtkm::Id index, vtkm::IdComponent inComponentIndex, int inArrayId)
{
@ -52,7 +52,7 @@ vtkm::FloatDefault TestValue3Ids(vtkm::Id index, vtkm::IdComponent inComponentIn
template <typename ValueType>
vtkm::cont::ArrayHandle<ValueType, StorageTag> MakeInputArray(int arrayId)
{
typedef vtkm::VecTraits<ValueType> VTraits;
using VTraits = vtkm::VecTraits<ValueType>;
// Create a buffer with valid test values.
ValueType buffer[ARRAY_SIZE];
@ -92,7 +92,7 @@ void CheckArray(const vtkm::cont::ArrayHandle<ValueType, C>& outArray,
vtkm::cont::DeviceAdapterAlgorithm<vtkm::cont::DeviceAdapterTagSerial>::Copy(outArray, arrayCopy);
typename ArrayHandleType::PortalConstControl portal = arrayCopy.GetPortalConstControl();
typedef vtkm::VecTraits<ValueType> VTraits;
using VTraits = vtkm::VecTraits<ValueType>;
for (vtkm::Id index = 0; index < ARRAY_SIZE; index++)
{
ValueType retreivedValue = portal.Get(index);
@ -118,7 +118,7 @@ void TryScalarArray()
int inArrayId = 0;
InArrayType inArray = MakeInputArray<InValueType>(inArrayId);
typedef typename vtkm::cont::ArrayHandleCompositeVectorType<InArrayType>::type OutArrayType;
using OutArrayType = typename vtkm::cont::ArrayHandleCompositeVectorType<InArrayType>::type;
for (vtkm::IdComponent inComponentIndex = 0; inComponentIndex < inComponents; inComponentIndex++)
{
OutArrayType outArray = vtkm::cont::make_ArrayHandleCompositeVector(inArray, inComponentIndex);
@ -253,10 +253,10 @@ void TrySpecialArrays()
{
std::cout << "Trying special arrays." << std::endl;
typedef vtkm::cont::ArrayHandleIndex ArrayType1;
using ArrayType1 = vtkm::cont::ArrayHandleIndex;
ArrayType1 array1(ARRAY_SIZE);
typedef vtkm::cont::ArrayHandleConstant<vtkm::Id> ArrayType2;
using ArrayType2 = vtkm::cont::ArrayHandleConstant<vtkm::Id>;
ArrayType2 array2(295, ARRAY_SIZE);
typedef vtkm::cont::ArrayHandleCompositeVectorType<ArrayType1, ArrayType2>::type

4
vtkm/cont/testing/UnitTestArrayHandleConcatenate.cxx
View File

@ -58,8 +58,8 @@ void TestConcatenateEmptyArray()
for (vtkm::Id i = 0; i < ARRAY_SIZE; i++)
vec.push_back(vtkm::Float64(i) * 1.5);
typedef vtkm::Float64 CoeffValueType;
typedef vtkm::cont::ArrayHandle<CoeffValueType> CoeffArrayTypeTmp;
using CoeffValueType = vtkm::Float64;
using CoeffArrayTypeTmp = vtkm::cont::ArrayHandle<CoeffValueType>;
typedef vtkm::cont::ArrayHandleConcatenate<CoeffArrayTypeTmp, CoeffArrayTypeTmp> ArrayConcat;
typedef vtkm::cont::ArrayHandleConcatenate<ArrayConcat, CoeffArrayTypeTmp> ArrayConcat2;

4
vtkm/cont/testing/UnitTestArrayHandleCounting.cxx
View File

@ -88,13 +88,13 @@ namespace UnitTestArrayHandleCountingNamespace
template <typename ValueType>
struct TemplatedTests
{
typedef vtkm::cont::ArrayHandleCounting<ValueType> ArrayHandleType;
using ArrayHandleType = vtkm::cont::ArrayHandleCounting<ValueType>;
typedef vtkm::cont::
ArrayHandle<ValueType, typename vtkm::cont::internal::ArrayHandleCountingTraits<ValueType>::Tag>
ArrayHandleType2;
typedef typename ArrayHandleType::PortalConstControl PortalType;
using PortalType = typename ArrayHandleType::PortalConstControl;
void operator()(const ValueType& startingValue, const ValueType& step)
{

247
vtkm/cont/testing/UnitTestArrayHandleExtractComponent.cxx Normal file
View File

@ -0,0 +1,247 @@
//=============================================================================
//
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
//
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2015 Sandia Corporation.
// Copyright 2015 UT-Battelle, LLC.
// Copyright 2015 Los Alamos National Security.
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
// Under the terms of Contract DE-AC52-06NA25396 with Los Alamos National
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//
//=============================================================================
#include <vtkm/cont/ArrayHandle.h>
#include <vtkm/cont/ArrayHandleCompositeVector.h>
#include <vtkm/cont/ArrayHandleCounting.h>
#include <vtkm/cont/ArrayHandleExtractComponent.h>
#include <vtkm/cont/DeviceAdapter.h>
#include <vtkm/cont/DeviceAdapterAlgorithm.h>
#include <vtkm/cont/testing/Testing.h>
namespace
{
template <typename ValueType>
struct ExtractComponentTests
{
using InputArray = vtkm::cont::ArrayHandle<vtkm::Vec<ValueType, 4>>;
template <vtkm::IdComponent Component>
using ExtractArray = vtkm::cont::ArrayHandleExtractComponent<InputArray, Component>;
using ReferenceComponentArray = vtkm::cont::ArrayHandleCounting<ValueType>;
using ReferenceCompositeArray =
typename vtkm::cont::ArrayHandleCompositeVectorType<ReferenceComponentArray,
ReferenceComponentArray,
ReferenceComponentArray,
ReferenceComponentArray>::type;
using DeviceTag = VTKM_DEFAULT_DEVICE_ADAPTER_TAG;
using Algo = vtkm::cont::DeviceAdapterAlgorithm<DeviceTag>;
// This is used to build a ArrayHandleExtractComponent's internal array.
ReferenceCompositeArray RefComposite;
void ConstructReferenceArray()
{
// Build the Ref array
const vtkm::Id numValues = 32;
ReferenceComponentArray c1 = vtkm::cont::make_ArrayHandleCounting<ValueType>(3, 2, numValues);
ReferenceComponentArray c2 = vtkm::cont::make_ArrayHandleCounting<ValueType>(2, 3, numValues);
ReferenceComponentArray c3 = vtkm::cont::make_ArrayHandleCounting<ValueType>(4, 4, numValues);
ReferenceComponentArray c4 = vtkm::cont::make_ArrayHandleCounting<ValueType>(1, 3, numValues);
this->RefComposite = vtkm::cont::make_ArrayHandleCompositeVector(c1, 0, c2, 0, c3, 0, c4, 0);
}
InputArray BuildInputArray() const
{
InputArray result;
Algo::Copy(this->RefComposite, result);
return result;
}
template <vtkm::IdComponent Component>
void SanityCheck() const
{
InputArray composite = this->BuildInputArray();
ExtractArray<Component> extract =
vtkm::cont::make_ArrayHandleExtractComponent<Component>(composite);
VTKM_TEST_ASSERT(composite.GetNumberOfValues() == extract.GetNumberOfValues(),
"Number of values in copied ExtractComponent array does not match input.");
}
template <vtkm::IdComponent Component>
void ReadTestComponentExtraction() const
{
// Test that the expected values are read from an ExtractComponent array.
InputArray composite = this->BuildInputArray();
ExtractArray<Component> extract =
vtkm::cont::make_ArrayHandleExtractComponent<Component>(composite);
// Test reading the data back in the control env:
this->ValidateReadTestArray<Component>(extract);
// Copy the extract array in the execution environment to test reading:
vtkm::cont::ArrayHandle<ValueType> execCopy;
Algo::Copy(extract, execCopy);
this->ValidateReadTestArray<Component>(execCopy);
}
template <vtkm::IdComponent Component, typename ArrayHandleType>
void ValidateReadTestArray(ArrayHandleType testArray) const
{
using RefVectorType = typename ReferenceCompositeArray::ValueType;
using Traits = vtkm::VecTraits<RefVectorType>;
auto testPortal = testArray.GetPortalConstControl();
auto refPortal = this->RefComposite.GetPortalConstControl();
VTKM_TEST_ASSERT(testPortal.GetNumberOfValues() == refPortal.GetNumberOfValues(),
"Number of values in read test output do not match input.");
for (vtkm::Id i = 0; i < testPortal.GetNumberOfValues(); ++i)
{
VTKM_TEST_ASSERT(
test_equal(testPortal.Get(i), Traits::GetComponent(refPortal.Get(i), Component), 0.),
"Value mismatch in read test.");
}
}
// Doubles the specified component (reading from RefVectorType).
template <typename PortalType, typename RefPortalType, vtkm::IdComponent Component>
struct WriteTestFunctor : vtkm::exec::FunctorBase
{
using RefVectorType = typename RefPortalType::ValueType;
using Traits = vtkm::VecTraits<RefVectorType>;
PortalType Portal;
RefPortalType RefPortal;
VTKM_CONT
WriteTestFunctor(const PortalType& portal, const RefPortalType& ref)
: Portal(portal)
, RefPortal(ref)
{
}
VTKM_EXEC_CONT
void operator()(vtkm::Id index) const
{
this->Portal.Set(index, Traits::GetComponent(this->RefPortal.Get(index), Component) * 2);
}
};
template <vtkm::IdComponent Component>
void WriteTestComponentExtraction() const
{
// Control test:
{
InputArray composite = this->BuildInputArray();
ExtractArray<Component> extract =
vtkm::cont::make_ArrayHandleExtractComponent<Component>(composite);
WriteTestFunctor<typename ExtractArray<Component>::PortalControl,
typename ReferenceCompositeArray::PortalConstControl,
Component>
functor(extract.GetPortalControl(), this->RefComposite.GetPortalConstControl());
for (vtkm::Id i = 0; i < extract.GetNumberOfValues(); ++i)
{
functor(i);
}
this->ValidateWriteTestArray<Component>(composite);
}
// Exec test:
{
InputArray composite = this->BuildInputArray();
ExtractArray<Component> extract =
vtkm::cont::make_ArrayHandleExtractComponent<Component>(composite);
using Portal = typename ExtractArray<Component>::template ExecutionTypes<DeviceTag>::Portal;
using RefPortal =
typename ReferenceCompositeArray::template ExecutionTypes<DeviceTag>::PortalConst;
WriteTestFunctor<Portal, RefPortal, Component> functor(
extract.PrepareForInPlace(DeviceTag()), this->RefComposite.PrepareForInput(DeviceTag()));
Algo::Schedule(functor, extract.GetNumberOfValues());
this->ValidateWriteTestArray<Component>(composite);
}
}
template <vtkm::IdComponent Component>
void ValidateWriteTestArray(InputArray testArray) const
{
using VectorType = typename ReferenceCompositeArray::ValueType;
using Traits = vtkm::VecTraits<VectorType>;
// Check that the indicated component is twice the reference value.
auto refPortal = this->RefComposite.GetPortalConstControl();
auto portal = testArray.GetPortalConstControl();
VTKM_TEST_ASSERT(portal.GetNumberOfValues() == refPortal.GetNumberOfValues(),
"Number of values in write test output do not match input.");
for (vtkm::Id i = 0; i < portal.GetNumberOfValues(); ++i)
{
auto value = portal.Get(i);
auto refValue = refPortal.Get(i);
Traits::SetComponent(refValue, Component, Traits::GetComponent(refValue, Component) * 2);
VTKM_TEST_ASSERT(test_equal(refValue, value, 0.), "Value mismatch in write test.");
}
}
template <vtkm::IdComponent Component>
void TestComponent() const
{
this->SanityCheck<Component>();
this->ReadTestComponentExtraction<Component>();
this->WriteTestComponentExtraction<Component>();
}
void operator()()
{
this->ConstructReferenceArray();
this->TestComponent<0>();
this->TestComponent<1>();
this->TestComponent<2>();
this->TestComponent<3>();
}
};
struct ArgToTemplateType
{
template <typename ValueType>
void operator()(ValueType) const
{
ExtractComponentTests<ValueType>()();
}
};
void TestArrayHandleExtractComponent()
{
using TestTypes = vtkm::ListTagBase<vtkm::Int32, vtkm::Int64, vtkm::Float32, vtkm::Float64>;
vtkm::testing::Testing::TryTypes(ArgToTemplateType(), TestTypes());
}
} // end anon namespace
int UnitTestArrayHandleExtractComponent(int, char* [])
{
return vtkm::cont::testing::Testing::Run(TestArrayHandleExtractComponent);
}

10
vtkm/cont/testing/UnitTestArrayHandleImplicit.cxx
View File

@ -39,7 +39,7 @@ struct IndexSquared
VTKM_EXEC_CONT
ValueType operator()(vtkm::Id i) const
{
typedef typename vtkm::VecTraits<ValueType>::ComponentType ComponentType;
using ComponentType = typename vtkm::VecTraits<ValueType>::ComponentType;
return ValueType(static_cast<ComponentType>(i * i));
}
};
@ -49,10 +49,10 @@ struct ImplicitTests
template <typename ValueType>
void operator()(const ValueType) const
{
typedef IndexSquared<ValueType> FunctorType;
using FunctorType = IndexSquared<ValueType>;
FunctorType functor;
typedef vtkm::cont::ArrayHandleImplicit<FunctorType> ImplicitHandle;
using ImplicitHandle = vtkm::cont::ArrayHandleImplicit<FunctorType>;
ImplicitHandle implict = vtkm::cont::make_ArrayHandleImplicit(functor, ARRAY_SIZE);
@ -65,8 +65,8 @@ struct ImplicitTests
}
//verify that the execution portal works
typedef vtkm::cont::DeviceAdapterTagSerial Device;
typedef typename ImplicitHandle::template ExecutionTypes<Device>::PortalConst CEPortal;
using Device = vtkm::cont::DeviceAdapterTagSerial;
using CEPortal = typename ImplicitHandle::template ExecutionTypes<Device>::PortalConst;
CEPortal execPortal = implict.PrepareForInput(Device());
for (int i = 0; i < ARRAY_SIZE; ++i)
{

30
vtkm/cont/testing/UnitTestArrayHandlePermutation.cxx
View File

@ -44,7 +44,7 @@ struct DoubleIndexFunctor
vtkm::Id operator()(vtkm::Id index) const { return 2 * index; }
};
typedef vtkm::cont::ArrayHandleImplicit<DoubleIndexFunctor> DoubleIndexArrayType;
using DoubleIndexArrayType = vtkm::cont::ArrayHandleImplicit<DoubleIndexFunctor>;
template <typename PermutedPortalType>
struct CheckPermutationFunctor : vtkm::exec::FunctorBase
@ -54,7 +54,7 @@ struct CheckPermutationFunctor : vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename PermutedPortalType::ValueType T;
using T = typename PermutedPortalType::ValueType;
T value = this->PermutedPortal.Get(index);
vtkm::Id permutedIndex = 2 * index;
@ -72,8 +72,8 @@ VTKM_CONT CheckPermutationFunctor<
typename PermutedArrayHandleType::template ExecutionTypes<Device>::PortalConst>
make_CheckPermutationFunctor(const PermutedArrayHandleType& permutedArray, Device)
{
typedef typename PermutedArrayHandleType::template ExecutionTypes<Device>::PortalConst
PermutedPortalType;
using PermutedPortalType =
typename PermutedArrayHandleType::template ExecutionTypes<Device>::PortalConst;
CheckPermutationFunctor<PermutedPortalType> functor;
functor.PermutedPortal = permutedArray.PrepareForInput(Device());
return functor;
@ -87,7 +87,7 @@ struct InPlacePermutationFunctor : vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename PermutedPortalType::ValueType T;
using T = typename PermutedPortalType::ValueType;
T value = this->PermutedPortal.Get(index);
value = value + T(1000);
@ -101,8 +101,8 @@ VTKM_CONT InPlacePermutationFunctor<
typename PermutedArrayHandleType::template ExecutionTypes<Device>::Portal>
make_InPlacePermutationFunctor(PermutedArrayHandleType& permutedArray, Device)
{
typedef
typename PermutedArrayHandleType::template ExecutionTypes<Device>::Portal PermutedPortalType;
using PermutedPortalType =
typename PermutedArrayHandleType::template ExecutionTypes<Device>::Portal;
InPlacePermutationFunctor<PermutedPortalType> functor;
functor.PermutedPortal = permutedArray.PrepareForInPlace(Device());
return functor;
@ -111,7 +111,7 @@ make_InPlacePermutationFunctor(PermutedArrayHandleType& permutedArray, Device)
template <typename PortalType>
VTKM_CONT void CheckInPlaceResult(PortalType portal)
{
typedef typename PortalType::ValueType T;
using T = typename PortalType::ValueType;
for (vtkm::Id permutedIndex = 0; permutedIndex < 2 * ARRAY_SIZE; permutedIndex++)
{
if (permutedIndex % 2 == 0)
@ -140,7 +140,7 @@ struct OutputPermutationFunctor : vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename PermutedPortalType::ValueType T;
using T = typename PermutedPortalType::ValueType;
this->PermutedPortal.Set(index, TestValue(static_cast<vtkm::Id>(index), T()));
}
};
@ -150,8 +150,8 @@ VTKM_CONT OutputPermutationFunctor<
typename PermutedArrayHandleType::template ExecutionTypes<Device>::Portal>
make_OutputPermutationFunctor(PermutedArrayHandleType& permutedArray, Device)
{
typedef
typename PermutedArrayHandleType::template ExecutionTypes<Device>::Portal PermutedPortalType;
using PermutedPortalType =
typename PermutedArrayHandleType::template ExecutionTypes<Device>::Portal;
OutputPermutationFunctor<PermutedPortalType> functor;
functor.PermutedPortal = permutedArray.PrepareForOutput(ARRAY_SIZE, Device());
return functor;
@ -160,7 +160,7 @@ make_OutputPermutationFunctor(PermutedArrayHandleType& permutedArray, Device)
template <typename PortalType>
VTKM_CONT void CheckOutputResult(PortalType portal)
{
typedef typename PortalType::ValueType T;
using T = typename PortalType::ValueType;
for (vtkm::IdComponent permutedIndex = 0; permutedIndex < 2 * ARRAY_SIZE; permutedIndex++)
{
if (permutedIndex % 2 == 0)
@ -185,12 +185,12 @@ VTKM_CONT void CheckOutputResult(PortalType portal)
template <typename ValueType>
struct PermutationTests
{
typedef vtkm::cont::ArrayHandleImplicit<DoubleIndexFunctor> IndexArrayType;
using IndexArrayType = vtkm::cont::ArrayHandleImplicit<DoubleIndexFunctor>;
typedef vtkm::cont::ArrayHandle<ValueType, vtkm::cont::StorageTagBasic> ValueArrayType;
typedef vtkm::cont::ArrayHandlePermutation<IndexArrayType, ValueArrayType> PermutationArrayType;
typedef VTKM_DEFAULT_DEVICE_ADAPTER_TAG Device;
typedef vtkm::cont::DeviceAdapterAlgorithm<Device> Algorithm;
using Device = VTKM_DEFAULT_DEVICE_ADAPTER_TAG;
using Algorithm = vtkm::cont::DeviceAdapterAlgorithm<Device>;
ValueArrayType MakeValueArray() const
{

2
vtkm/cont/testing/UnitTestArrayHandleReverse.cxx
View File

@ -82,7 +82,7 @@ void TestArrayHandleReverseScanInclusiveByKey()
vtkm::cont::ArrayHandleReverse<vtkm::cont::ArrayHandle<vtkm::Id>> reversed =
vtkm::cont::make_ArrayHandleReverse(output);
typedef vtkm::cont::DeviceAdapterAlgorithm<vtkm::cont::DeviceAdapterTagSerial> Algorithm;
using Algorithm = vtkm::cont::DeviceAdapterAlgorithm<vtkm::cont::DeviceAdapterTagSerial>;
Algorithm::ScanInclusiveByKey(keys, values, reversed);
vtkm::Id expected[] = { 0, 1, 3, 6, 4, 9, 6, 7, 15, 9 };

400
vtkm/cont/testing/UnitTestArrayHandleSwizzle.cxx Normal file
View File

@ -0,0 +1,400 @@
//=============================================================================
//
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
//
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2015 Sandia Corporation.
// Copyright 2015 UT-Battelle, LLC.
// Copyright 2015 Los Alamos National Security.
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
// Under the terms of Contract DE-AC52-06NA25396 with Los Alamos National
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//
//=============================================================================
#include <vtkm/cont/ArrayHandle.h>
#include <vtkm/cont/ArrayHandleCompositeVector.h>
#include <vtkm/cont/ArrayHandleCounting.h>
#include <vtkm/cont/ArrayHandleSwizzle.h>
#include <vtkm/cont/DeviceAdapter.h>
#include <vtkm/cont/DeviceAdapterAlgorithm.h>
#include <vtkm/cont/testing/Testing.h>
#include <type_traits>
namespace
{
template <typename ValueType>
struct SwizzleTests
{
using SwizzleInputArrayType = vtkm::cont::ArrayHandle<vtkm::Vec<ValueType, 4>>;
template <vtkm::IdComponent... ComponentMap>
using SwizzleArrayType = vtkm::cont::ArrayHandleSwizzle<SwizzleInputArrayType, ComponentMap...>;
using ReferenceComponentArrayType = vtkm::cont::ArrayHandleCounting<ValueType>;
using ReferenceArrayType =
typename vtkm::cont::ArrayHandleCompositeVectorType<ReferenceComponentArrayType,
ReferenceComponentArrayType,
ReferenceComponentArrayType,
ReferenceComponentArrayType>::type;
using DeviceTag = VTKM_DEFAULT_DEVICE_ADAPTER_TAG;
using Algo = vtkm::cont::DeviceAdapterAlgorithm<DeviceTag>;
// This is used to build a ArrayHandleSwizzle's internal array.
ReferenceArrayType RefArray;
void ConstructReferenceArray()
{
// Build the Ref array
const vtkm::Id numValues = 32;
ReferenceComponentArrayType c1 =
vtkm::cont::make_ArrayHandleCounting<ValueType>(3, 2, numValues);
ReferenceComponentArrayType c2 =
vtkm::cont::make_ArrayHandleCounting<ValueType>(2, 3, numValues);
ReferenceComponentArrayType c3 =
vtkm::cont::make_ArrayHandleCounting<ValueType>(4, 4, numValues);
ReferenceComponentArrayType c4 =
vtkm::cont::make_ArrayHandleCounting<ValueType>(1, 3, numValues);
this->RefArray = vtkm::cont::make_ArrayHandleCompositeVector(c1, 0, c2, 0, c3, 0, c4, 0);
}
SwizzleInputArrayType BuildSwizzleInputArray() const
{
SwizzleInputArrayType result;
Algo::Copy(this->RefArray, result);
return result;
}
template <vtkm::IdComponent... ComponentMap>
void SanityCheck() const
{
using Swizzle = SwizzleArrayType<ComponentMap...>;
using Traits = typename Swizzle::SwizzleTraits;
VTKM_TEST_ASSERT(Traits::COUNT == vtkm::VecTraits<typename Swizzle::ValueType>::NUM_COMPONENTS,
"Traits::COUNT invalid.");
VTKM_TEST_ASSERT(
VTKM_PASS_COMMAS(std::is_same<typename Traits::ComponentType, ValueType>::value),
"Traits::ComponentType invalid.");
VTKM_TEST_ASSERT(
VTKM_PASS_COMMAS(
std::is_same<
typename Traits::OutputType,
vtkm::Vec<ValueType, static_cast<vtkm::IdComponent>(sizeof...(ComponentMap))>>::value),
"Traits::OutputType invalid.");
SwizzleInputArrayType input = this->BuildSwizzleInputArray();
SwizzleArrayType<ComponentMap...> swizzle =
vtkm::cont::make_ArrayHandleSwizzle<ComponentMap...>(input);
VTKM_TEST_ASSERT(input.GetNumberOfValues() == swizzle.GetNumberOfValues(),
"Number of values in copied Swizzle array does not match input.");
}
template <vtkm::IdComponent... ComponentMap>
void ReadTest() const
{
using Traits = typename SwizzleArrayType<ComponentMap...>::SwizzleTraits;
// Test that the expected values are read from an Swizzle array.
SwizzleInputArrayType input = this->BuildSwizzleInputArray();
SwizzleArrayType<ComponentMap...> swizzle =
vtkm::cont::make_ArrayHandleSwizzle<ComponentMap...>(input);
// Test reading the data back in the control env:
this->ValidateReadTest<ComponentMap...>(swizzle);
// Copy the extract array in the execution environment to test reading:
vtkm::cont::ArrayHandle<typename Traits::OutputType> execCopy;
Algo::Copy(swizzle, execCopy);
this->ValidateReadTest<ComponentMap...>(execCopy);
}
template <vtkm::IdComponent... ComponentMap, typename ArrayHandleType>
void ValidateReadTest(ArrayHandleType testArray) const
{
using Traits = typename SwizzleArrayType<ComponentMap...>::SwizzleTraits;
using MapType = typename Traits::RuntimeComponentMapType;
const MapType map = Traits::GenerateRuntimeComponentMap();
using ReferenceVectorType = typename ReferenceArrayType::ValueType;
using SwizzleVectorType = typename Traits::OutputType;
VTKM_TEST_ASSERT(map.size() == vtkm::VecTraits<SwizzleVectorType>::NUM_COMPONENTS,
"Unexpected runtime component map size.");
VTKM_TEST_ASSERT(testArray.GetNumberOfValues() == this->RefArray.GetNumberOfValues(),
"Number of values incorrect in Read test.");
auto refPortal = this->RefArray.GetPortalConstControl();
auto testPortal = testArray.GetPortalConstControl();
for (vtkm::Id i = 0; i < testArray.GetNumberOfValues(); ++i)
{
// Manually swizzle the reference vector using the runtime map information:
ReferenceVectorType refVec = refPortal.Get(i);
SwizzleVectorType refVecSwizzle;
for (size_t j = 0; j < map.size(); ++j)
{
refVecSwizzle[static_cast<vtkm::IdComponent>(j)] = refVec[map[j]];
}
VTKM_TEST_ASSERT(test_equal(refVecSwizzle, testPortal.Get(i), 0.),
"Invalid value encountered in Read test.");
}
}
// Doubles everything in the input portal.
template <typename PortalType>
struct WriteTestFunctor : vtkm::exec::FunctorBase
{
PortalType Portal;
VTKM_CONT
WriteTestFunctor(const PortalType& portal)
: Portal(portal)
{
}
VTKM_EXEC_CONT
void operator()(vtkm::Id index) const { this->Portal.Set(index, this->Portal.Get(index) * 2.); }
};
template <vtkm::IdComponent... ComponentMap>
void WriteTest() const
{
// Control test:
{
SwizzleInputArrayType input = this->BuildSwizzleInputArray();
SwizzleArrayType<ComponentMap...> swizzle =
vtkm::cont::make_ArrayHandleSwizzle<ComponentMap...>(input);
WriteTestFunctor<typename SwizzleArrayType<ComponentMap...>::PortalControl> functor(
swizzle.GetPortalControl());
for (vtkm::Id i = 0; i < swizzle.GetNumberOfValues(); ++i)
{
functor(i);
}
this->ValidateWriteTestArray<ComponentMap...>(input);
}
// Exec test:
{
SwizzleInputArrayType input = this->BuildSwizzleInputArray();
SwizzleArrayType<ComponentMap...> swizzle =
vtkm::cont::make_ArrayHandleSwizzle<ComponentMap...>(input);
using Portal =
typename SwizzleArrayType<ComponentMap...>::template ExecutionTypes<DeviceTag>::Portal;
WriteTestFunctor<Portal> functor(swizzle.PrepareForInPlace(DeviceTag()));
Algo::Schedule(functor, swizzle.GetNumberOfValues());
this->ValidateWriteTestArray<ComponentMap...>(input);
}
}
// Check that the swizzled components are twice the reference value.
template <vtkm::IdComponent... ComponentMap>
void ValidateWriteTestArray(SwizzleInputArrayType testArray) const
{
using Traits = typename SwizzleArrayType<ComponentMap...>::SwizzleTraits;
using MapType = typename Traits::RuntimeComponentMapType;
const MapType map = Traits::GenerateRuntimeComponentMap();
auto refPortal = this->RefArray.GetPortalConstControl();
auto portal = testArray.GetPortalConstControl();
VTKM_TEST_ASSERT(portal.GetNumberOfValues() == refPortal.GetNumberOfValues(),
"Number of values in write test output do not match input.");
for (vtkm::Id i = 0; i < portal.GetNumberOfValues(); ++i)
{
auto value = portal.Get(i);
auto refValue = refPortal.Get(i);
// Double all of the components that appear in the map to replicate the
// test result:
for (size_t j = 0; j < map.size(); ++j)
{
refValue[map[j]] *= 2;
}
VTKM_TEST_ASSERT(test_equal(refValue, value, 0.), "Value mismatch in Write test.");
}
}
template <vtkm::IdComponent... ComponentMap>
void TestSwizzle() const
{
this->SanityCheck<ComponentMap...>();
this->ReadTest<ComponentMap...>();
this->WriteTest<ComponentMap...>();
}
void operator()()
{
this->ConstructReferenceArray();
// Enable for full test. We normally test a reduced set of component maps
// to keep compile times/sizes down:
#if 0
this->TestSwizzle<0, 1>();
this->TestSwizzle<0, 2>();
this->TestSwizzle<0, 3>();
this->TestSwizzle<1, 0>();
this->TestSwizzle<1, 2>();
this->TestSwizzle<1, 3>();
this->TestSwizzle<2, 0>();
this->TestSwizzle<2, 1>();
this->TestSwizzle<2, 3>();
this->TestSwizzle<3, 0>();
this->TestSwizzle<3, 1>();
this->TestSwizzle<3, 2>();
this->TestSwizzle<0, 1, 2>();
this->TestSwizzle<0, 1, 3>();
this->TestSwizzle<0, 2, 1>();
this->TestSwizzle<0, 2, 3>();
this->TestSwizzle<0, 3, 1>();
this->TestSwizzle<0, 3, 2>();
this->TestSwizzle<1, 0, 2>();
this->TestSwizzle<1, 0, 3>();
this->TestSwizzle<1, 2, 0>();
this->TestSwizzle<1, 2, 3>();
this->TestSwizzle<1, 3, 0>();
this->TestSwizzle<1, 3, 2>();
this->TestSwizzle<2, 0, 1>();
this->TestSwizzle<2, 0, 3>();
this->TestSwizzle<2, 1, 0>();
this->TestSwizzle<2, 1, 3>();
this->TestSwizzle<2, 3, 0>();
this->TestSwizzle<2, 3, 1>();
this->TestSwizzle<3, 0, 1>();
this->TestSwizzle<3, 0, 2>();
this->TestSwizzle<3, 1, 0>();
this->TestSwizzle<3, 1, 2>();
this->TestSwizzle<3, 2, 0>();
this->TestSwizzle<3, 2, 1>();
this->TestSwizzle<0, 1, 2, 3>();
this->TestSwizzle<0, 1, 3, 2>();
this->TestSwizzle<0, 2, 1, 3>();
this->TestSwizzle<0, 2, 3, 1>();
this->TestSwizzle<0, 3, 1, 2>();
this->TestSwizzle<0, 3, 2, 1>();
this->TestSwizzle<1, 0, 2, 3>();
this->TestSwizzle<1, 0, 3, 2>();
this->TestSwizzle<1, 2, 0, 3>();
this->TestSwizzle<1, 2, 3, 0>();
this->TestSwizzle<1, 3, 0, 2>();
this->TestSwizzle<1, 3, 2, 0>();
this->TestSwizzle<2, 0, 1, 3>();
this->TestSwizzle<2, 0, 3, 1>();
this->TestSwizzle<2, 1, 0, 3>();
this->TestSwizzle<2, 1, 3, 0>();
this->TestSwizzle<2, 3, 0, 1>();
this->TestSwizzle<2, 3, 1, 0>();
this->TestSwizzle<3, 0, 1, 2>();
this->TestSwizzle<3, 0, 2, 1>();
this->TestSwizzle<3, 1, 0, 2>();
this->TestSwizzle<3, 1, 2, 0>();
this->TestSwizzle<3, 2, 0, 1>();
this->TestSwizzle<3, 2, 1, 0>();
#else
this->TestSwizzle<0, 1>();
this->TestSwizzle<1, 0>();
this->TestSwizzle<2, 3>();
this->TestSwizzle<3, 2>();
this->TestSwizzle<0, 1, 2>();
this->TestSwizzle<0, 3, 1>();
this->TestSwizzle<2, 0, 3>();
this->TestSwizzle<3, 2, 1>();
this->TestSwizzle<0, 1, 2, 3>();
this->TestSwizzle<1, 3, 2, 0>();
this->TestSwizzle<2, 0, 1, 3>();
this->TestSwizzle<3, 1, 0, 2>();
this->TestSwizzle<3, 2, 1, 0>();
#endif
}
};
struct ArgToTemplateType
{
template <typename ValueType>
void operator()(ValueType) const
{
SwizzleTests<ValueType>()();
}
};
void TestArrayHandleSwizzle()
{
using TestTypes = vtkm::ListTagBase<vtkm::Int32, vtkm::Int64, vtkm::Float32, vtkm::Float64>;
vtkm::testing::Testing::TryTypes(ArgToTemplateType(), TestTypes());
}
template <vtkm::IdComponent InputSize, vtkm::IdComponent... ComponentMap>
using Validator = vtkm::cont::internal::ValidateComponentMap<InputSize, ComponentMap...>;
void TestComponentMapValidator()
{
using RepeatComps = Validator<5, 0, 3, 2, 3, 1, 4>;
VTKM_TEST_ASSERT(!RepeatComps::Valid, "Repeat components allowed.");
using NegativeComps = Validator<5, 0, 4, -3, 1, 2>;
VTKM_TEST_ASSERT(!NegativeComps::Valid, "Negative components allowed.");
using OutOfBoundsComps = Validator<5, 0, 2, 3, 5>;
VTKM_TEST_ASSERT(!OutOfBoundsComps::Valid, "Out-of-bounds components allowed.");
}
void TestRuntimeComponentMapGenerator()
{
// Dummy input vector type. Only concerned with the component map:
using Dummy = vtkm::Vec<char, 7>;
using Traits = vtkm::cont::ArrayHandleSwizzleTraits<Dummy, 3, 2, 4, 1, 6, 0>;
using MapType = Traits::RuntimeComponentMapType;
const MapType map = Traits::GenerateRuntimeComponentMap();
VTKM_TEST_ASSERT(map.size() == 6, "Invalid map size.");
VTKM_TEST_ASSERT(map[0] == 3, "Invalid map entry.");
VTKM_TEST_ASSERT(map[1] == 2, "Invalid map entry.");
VTKM_TEST_ASSERT(map[2] == 4, "Invalid map entry.");
VTKM_TEST_ASSERT(map[3] == 1, "Invalid map entry.");
VTKM_TEST_ASSERT(map[4] == 6, "Invalid map entry.");
VTKM_TEST_ASSERT(map[5] == 0, "Invalid map entry.");
}
} // end anon namespace
int UnitTestArrayHandleSwizzle(int, char* [])
{
try
{
TestComponentMapValidator();
TestRuntimeComponentMapGenerator();
}
catch (vtkm::cont::Error& e)
{
std::cerr << "Error: " << e.what() << "\n";
return EXIT_FAILURE;
}
return vtkm::cont::testing::Testing::Run(TestArrayHandleSwizzle);
}

26
vtkm/cont/testing/UnitTestArrayHandleTransform.cxx
View File

@ -56,7 +56,7 @@ struct CheckTransformFunctor : vtkm::exec::FunctorBase
VTKM_EXEC
void operator()(vtkm::Id index) const
{
typedef typename TransformedPortalType::ValueType T;
using T = typename TransformedPortalType::ValueType;
typename OriginalPortalType::ValueType original = this->OriginalPortal.Get(index);
T transformed = this->TransformedPortal.Get(index);
if (!test_equal(transformed, MySquare<T>()(original)))
@ -74,10 +74,10 @@ make_CheckTransformFunctor(const OriginalArrayHandleType& originalArray,
const TransformedArrayHandleType& transformedArray,
Device)
{
typedef typename OriginalArrayHandleType::template ExecutionTypes<Device>::PortalConst
OriginalPortalType;
typedef typename TransformedArrayHandleType::template ExecutionTypes<Device>::PortalConst
TransformedPortalType;
using OriginalPortalType =
typename OriginalArrayHandleType::template ExecutionTypes<Device>::PortalConst;
using TransformedPortalType =
typename TransformedArrayHandleType::template ExecutionTypes<Device>::PortalConst;
CheckTransformFunctor<OriginalPortalType, TransformedPortalType> functor;
functor.OriginalPortal = originalArray.PrepareForInput(Device());
functor.TransformedPortal = transformedArray.PrepareForInput(Device());
@ -90,8 +90,8 @@ VTKM_CONT void CheckControlPortals(const OriginalArrayHandleType& originalArray,
{
std::cout << " Verify that the control portal works" << std::endl;
typedef typename OriginalArrayHandleType::PortalConstControl OriginalPortalType;
typedef typename TransformedArrayHandleType::PortalConstControl TransformedPortalType;
using OriginalPortalType = typename OriginalArrayHandleType::PortalConstControl;
using TransformedPortalType = typename TransformedArrayHandleType::PortalConstControl;
VTKM_TEST_ASSERT(originalArray.GetNumberOfValues() == transformedArray.GetNumberOfValues(),
"Number of values in transformed array incorrect.");
@ -104,7 +104,7 @@ VTKM_CONT void CheckControlPortals(const OriginalArrayHandleType& originalArray,
for (vtkm::Id index = 0; index < originalArray.GetNumberOfValues(); index++)
{
typedef typename TransformedPortalType::ValueType T;
using T = typename TransformedPortalType::ValueType;
typename OriginalPortalType::ValueType original = originalPortal.Get(index);
T transformed = transformedPortal.Get(index);
VTKM_TEST_ASSERT(test_equal(transformed, MySquare<T>()(original)), "Bad transform value.");
@ -114,8 +114,8 @@ VTKM_CONT void CheckControlPortals(const OriginalArrayHandleType& originalArray,
template <typename InputValueType>
struct TransformTests
{
typedef typename vtkm::VecTraits<InputValueType>::ComponentType OutputValueType;
typedef MySquare<OutputValueType> FunctorType;
using OutputValueType = typename vtkm::VecTraits<InputValueType>::ComponentType;
using FunctorType = MySquare<OutputValueType>;
typedef vtkm::cont::ArrayHandleTransform<vtkm::cont::ArrayHandle<InputValueType>, FunctorType>
TransformHandle;
@ -124,8 +124,8 @@ struct TransformTests
FunctorType>
CountingTransformHandle;
typedef VTKM_DEFAULT_DEVICE_ADAPTER_TAG Device;
typedef vtkm::cont::DeviceAdapterAlgorithm<Device> Algorithm;
using Device = VTKM_DEFAULT_DEVICE_ADAPTER_TAG;
using Algorithm = vtkm::cont::DeviceAdapterAlgorithm<Device>;
void operator()() const
{
@ -149,7 +149,7 @@ struct TransformTests
vtkm::cont::ArrayHandle<InputValueType> input;
TransformHandle thandle(input, functor);
typedef typename vtkm::cont::ArrayHandle<InputValueType>::PortalControl Portal;
using Portal = typename vtkm::cont::ArrayHandle<InputValueType>::PortalControl;
input.Allocate(ARRAY_SIZE);
Portal portal = input.GetPortalControl();
for (vtkm::Id index = 0; index < ARRAY_SIZE; ++index)

16
vtkm/cont/testing/UnitTestArrayPortalToIterators.cxx
View File

@ -32,8 +32,8 @@ struct TemplatedTests
{
static const vtkm::Id ARRAY_SIZE = 10;
typedef T ValueType;
typedef typename vtkm::VecTraits<ValueType>::ComponentType ComponentType;
using ValueType = T;
using ComponentType = typename vtkm::VecTraits<ValueType>::ComponentType;
static ValueType ExpectedValue(vtkm::Id index, ComponentType value)
{
@ -43,7 +43,7 @@ struct TemplatedTests
class ReadOnlyArrayPortal
{
public:
typedef T ValueType;
using ValueType = T;
VTKM_CONT
ReadOnlyArrayPortal(ComponentType value)
@ -64,7 +64,7 @@ struct TemplatedTests
class WriteOnlyArrayPortal
{
public:
typedef T ValueType;
using ValueType = T;
VTKM_CONT
WriteOnlyArrayPortal(ComponentType value)
@ -122,8 +122,8 @@ struct TemplatedTests
void TestIteratorRead()
{
typedef ReadOnlyArrayPortal ArrayPortalType;
typedef vtkm::cont::ArrayPortalToIterators<ArrayPortalType> GetIteratorsType;
using ArrayPortalType = ReadOnlyArrayPortal;
using GetIteratorsType = vtkm::cont::ArrayPortalToIterators<ArrayPortalType>;
static const ComponentType READ_VALUE = 23;
ArrayPortalType portal(READ_VALUE);
@ -140,8 +140,8 @@ struct TemplatedTests
void TestIteratorWrite()
{
typedef WriteOnlyArrayPortal ArrayPortalType;
typedef vtkm::cont::ArrayPortalToIterators<ArrayPortalType> GetIteratorsType;
using ArrayPortalType = WriteOnlyArrayPortal;
using GetIteratorsType = vtkm::cont::ArrayPortalToIterators<ArrayPortalType>;
static const ComponentType WRITE_VALUE = 63;
ArrayPortalType portal(WRITE_VALUE);

161
vtkm/cont/testing/UnitTestCellSetExplicit.cxx Normal file
View File

@ -0,0 +1,161 @@
//============================================================================
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2017 Sandia Corporation.
// Copyright 2017 UT-Battelle, LLC.
// Copyright 2017 Los Alamos National Security.
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
//
// Under the terms of Contract DE-AC52-06NA25396 with Los Alamos National
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//============================================================================
#include <vtkm/cont/CellSetExplicit.h>
#include <vtkm/cont/ArrayHandle.h>
#include <vtkm/cont/testing/Testing.h>
#include <vtkm/worklet/DispatcherMapTopology.h>
#include <vtkm/worklet/WorkletMapTopology.h>
namespace
{
const vtkm::Id numberOfPoints = 11;
vtkm::UInt8 shapes[] = { static_cast<vtkm::UInt8>(vtkm::CELL_SHAPE_HEXAHEDRON),
static_cast<vtkm::UInt8>(vtkm::CELL_SHAPE_PYRAMID),
static_cast<vtkm::UInt8>(vtkm::CELL_SHAPE_TETRA),
static_cast<vtkm::UInt8>(vtkm::CELL_SHAPE_WEDGE) };
vtkm::IdComponent numIndices[] = { 8, 5, 4, 6 };
vtkm::Id indexOffset[] = { 0, 8, 13, 17 };
vtkm::Id connectivity[] = { 0, 1, 5, 4, 3, 2, 6, 7, 1, 5, 6, 2, 8, 5, 8, 10, 6, 4, 7, 9, 5, 6, 10 };
template <typename T, std::size_t Length>
vtkm::Id ArrayLength(const T (&)[Length])
{
return static_cast<vtkm::Id>(Length);
}
// all points are part of atleast 1 cell
vtkm::cont::CellSetExplicit<> MakeTestCellSet1()
{
vtkm::cont::CellSetExplicit<> cs;
cs.Fill(numberOfPoints,
vtkm::cont::make_ArrayHandle(shapes, 4),
vtkm::cont::make_ArrayHandle(numIndices, 4),
vtkm::cont::make_ArrayHandle(connectivity, ArrayLength(connectivity)),
vtkm::cont::make_ArrayHandle(indexOffset, 4));
return cs;
}
// some points are not part of any cell
vtkm::cont::CellSetExplicit<> MakeTestCellSet2()
{
vtkm::cont::CellSetExplicit<> cs;
cs.Fill(
numberOfPoints,
vtkm::cont::make_ArrayHandle(shapes + 1, 2),
vtkm::cont::make_ArrayHandle(numIndices + 1, 2),
vtkm::cont::make_ArrayHandle(connectivity + indexOffset[1], indexOffset[3] - indexOffset[1]));
return cs;
}
struct WorkletPointToCell : public vtkm::worklet::WorkletMapPointToCell
{
typedef void ControlSignature(CellSetIn cellset, FieldOutCell<IdType> numPoints);
typedef void ExecutionSignature(PointIndices, _2);
using InputDomain = _1;
template <typename PointIndicesType>
VTKM_EXEC void operator()(const PointIndicesType& pointIndices, vtkm::Id& numPoints) const
{
numPoints = pointIndices.GetNumberOfComponents();
}
};
struct WorkletCellToPoint : public vtkm::worklet::WorkletMapCellToPoint
{
typedef void ControlSignature(CellSetIn cellset, FieldOutPoint<IdType> numCells);
typedef void ExecutionSignature(CellIndices, _2);
using InputDomain = _1;
template <typename CellIndicesType>
VTKM_EXEC void operator()(const CellIndicesType& cellIndices, vtkm::Id& numCells) const
{
numCells = cellIndices.GetNumberOfComponents();
}
};
void TestCellSetExplicit()
{
vtkm::cont::CellSetExplicit<> cellset;
vtkm::cont::ArrayHandle<vtkm::Id> result;
std::cout << "----------------------------------------------------\n";
std::cout << "Testing Case 1 (all points are part of atleast 1 cell): \n";
cellset = MakeTestCellSet1();
std::cout << "\tTesting PointToCell\n";
vtkm::worklet::DispatcherMapTopology<WorkletPointToCell>().Invoke(cellset, result);
VTKM_TEST_ASSERT(result.GetNumberOfValues() == cellset.GetNumberOfCells(),
"result length not equal to number of cells");
for (vtkm::Id i = 0; i < result.GetNumberOfValues(); ++i)
{
VTKM_TEST_ASSERT(result.GetPortalConstControl().Get(i) == numIndices[i], "incorrect result");
}
std::cout << "\tTesting CellToPoint\n";
vtkm::worklet::DispatcherMapTopology<WorkletCellToPoint>().Invoke(cellset, result);
VTKM_TEST_ASSERT(result.GetNumberOfValues() == cellset.GetNumberOfPoints(),
"result length not equal to number of points");
vtkm::Id expected1[] = { 1, 2, 2, 1, 2, 4, 4, 2, 2, 1, 2 };
for (vtkm::Id i = 0; i < result.GetNumberOfValues(); ++i)
{
VTKM_TEST_ASSERT(result.GetPortalConstControl().Get(i) == expected1[i], "incorrect result");
}
std::cout << "----------------------------------------------------\n";
std::cout << "Testing Case 2 (some points are not part of any cell): \n";
cellset = MakeTestCellSet2();
std::cout << "\tTesting PointToCell\n";
vtkm::worklet::DispatcherMapTopology<WorkletPointToCell>().Invoke(cellset, result);
VTKM_TEST_ASSERT(result.GetNumberOfValues() == cellset.GetNumberOfCells(),
"result length not equal to number of cells");
VTKM_TEST_ASSERT(result.GetPortalConstControl().Get(0) == numIndices[1] &&
result.GetPortalConstControl().Get(1) == numIndices[2],
"incorrect result");
std::cout << "\tTesting CellToPoint\n";
vtkm::worklet::DispatcherMapTopology<WorkletCellToPoint>().Invoke(cellset, result);
VTKM_TEST_ASSERT(result.GetNumberOfValues() == cellset.GetNumberOfPoints(),
"result length not equal to number of points");
vtkm::Id expected2[] = { 0, 1, 1, 0, 0, 2, 2, 0, 2, 0, 1 };
for (vtkm::Id i = 0; i < result.GetNumberOfValues(); ++i)
{
VTKM_TEST_ASSERT(result.GetPortalConstControl().Get(i) == expected2[i], "incorrect result");
}
}
} // anonymous namespace
int UnitTestCellSetExplicit(int, char* [])
{
return vtkm::cont::testing::Testing::Run(TestCellSetExplicit);
}

195
vtkm/cont/testing/UnitTestCellSetPermutation.cxx Normal file
View File

@ -0,0 +1,195 @@
//============================================================================
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2017 Sandia Corporation.
// Copyright 2017 UT-Battelle, LLC.
// Copyright 2017 Los Alamos National Security.
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
//
// Under the terms of Contract DE-AC52-06NA25396 with Los Alamos National
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//============================================================================
#include <vtkm/cont/CellSetPermutation.h>
#include <vtkm/cont/ArrayHandle.h>
#include <vtkm/cont/ArrayHandleCounting.h>
#include <vtkm/cont/DeviceAdapterAlgorithm.h>
#include <vtkm/cont/testing/MakeTestDataSet.h>
#include <vtkm/cont/testing/Testing.h>
#include <vtkm/worklet/DispatcherMapTopology.h>
#include <vtkm/worklet/WorkletMapTopology.h>
namespace
{
struct WorkletPointToCell : public vtkm::worklet::WorkletMapPointToCell
{
typedef void ControlSignature(CellSetIn cellset, FieldOutCell<IdType> numPoints);
typedef void ExecutionSignature(PointIndices, _2);
using InputDomain = _1;
template <typename PointIndicesType>
VTKM_EXEC void operator()(const PointIndicesType& pointIndices, vtkm::Id& numPoints) const
{
numPoints = pointIndices.GetNumberOfComponents();
}
};
struct WorkletCellToPoint : public vtkm::worklet::WorkletMapCellToPoint
{
typedef void ControlSignature(CellSetIn cellset, FieldOutPoint<IdType> numCells);
typedef void ExecutionSignature(CellIndices, _2);
using InputDomain = _1;
template <typename CellIndicesType>
VTKM_EXEC void operator()(const CellIndicesType& cellIndices, vtkm::Id& numCells) const
{
numCells = cellIndices.GetNumberOfComponents();
}
};
struct CellsOfPoint : public vtkm::worklet::WorkletMapCellToPoint
{
typedef void ControlSignature(CellSetIn cellset,
FieldInPoint<IdType> offset,
WholeArrayOut<IdType> cellIds);
typedef void ExecutionSignature(CellIndices, _2, _3);
using InputDomain = _1;
template <typename CellIndicesType, typename CellIdsPortal>
VTKM_EXEC void operator()(const CellIndicesType& cellIndices,
vtkm::Id offset,
const CellIdsPortal& out) const
{
vtkm::IdComponent count = cellIndices.GetNumberOfComponents();
for (vtkm::IdComponent i = 0; i < count; ++i)
{
out.Set(offset++, cellIndices[i]);
}
}
};
template <typename CellSetType, typename PermutationArrayHandleType>
std::vector<vtkm::Id> ComputeCellToPointExpected(const CellSetType& cellset,
const PermutationArrayHandleType& permutation)
{
vtkm::cont::ArrayHandle<vtkm::Id> numIndices;
vtkm::worklet::DispatcherMapTopology<WorkletCellToPoint>().Invoke(cellset, numIndices);
std::cout << "\n";
vtkm::cont::ArrayHandle<vtkm::Id> indexOffsets;
vtkm::Id connectivityLength =
vtkm::cont::DeviceAdapterAlgorithm<VTKM_DEFAULT_DEVICE_ADAPTER_TAG>::ScanExclusive(
numIndices, indexOffsets);
vtkm::cont::ArrayHandle<vtkm::Id> connectivity;
connectivity.Allocate(connectivityLength);
vtkm::worklet::DispatcherMapTopology<CellsOfPoint>().Invoke(cellset, indexOffsets, connectivity);
std::vector<bool> permutationMask(static_cast<std::size_t>(cellset.GetNumberOfCells()), false);
for (vtkm::Id i = 0; i < permutation.GetNumberOfValues(); ++i)
{
permutationMask[static_cast<std::size_t>(permutation.GetPortalConstControl().Get(i))] = true;
}
vtkm::Id numberOfPoints = cellset.GetNumberOfPoints();
std::vector<vtkm::Id> expected(static_cast<std::size_t>(numberOfPoints), 0);
for (vtkm::Id i = 0; i < numberOfPoints; ++i)
{
vtkm::Id offset = indexOffsets.GetPortalConstControl().Get(i);
vtkm::Id count = numIndices.GetPortalConstControl().Get(i);
for (vtkm::Id j = 0; j < count; ++j)
{
vtkm::Id cellId = connectivity.GetPortalConstControl().Get(offset++);
if (permutationMask[static_cast<std::size_t>(cellId)])
{
++expected[static_cast<std::size_t>(i)];
}
}
}
return expected;
}
template <typename CellSetType>
vtkm::cont::CellSetPermutation<CellSetType, vtkm::cont::ArrayHandleCounting<vtkm::Id>> TestCellSet(
const CellSetType& cellset)
{
vtkm::Id numberOfCells = cellset.GetNumberOfCells() / 2;
vtkm::cont::ArrayHandleCounting<vtkm::Id> permutation(0, 2, numberOfCells);
auto cs = vtkm::cont::make_CellSetPermutation(permutation, cellset);
vtkm::cont::ArrayHandle<vtkm::Id> result;
std::cout << "\t\tTesting PointToCell\n";
vtkm::worklet::DispatcherMapTopology<WorkletPointToCell>().Invoke(cs, result);
VTKM_TEST_ASSERT(result.GetNumberOfValues() == numberOfCells,
"result length not equal to number of cells");
for (vtkm::Id i = 0; i < result.GetNumberOfValues(); ++i)
{
VTKM_TEST_ASSERT(result.GetPortalConstControl().Get(i) ==
cellset.GetNumberOfPointsInCell(permutation.GetPortalConstControl().Get(i)),
"incorrect result");
}
std::cout << "\t\tTesting CellToPoint\n";
vtkm::worklet::DispatcherMapTopology<WorkletCellToPoint>().Invoke(cs, result);
VTKM_TEST_ASSERT(result.GetNumberOfValues() == cellset.GetNumberOfPoints(),
"result length not equal to number of points");
auto expected = ComputeCellToPointExpected(cellset, permutation);
for (vtkm::Id i = 0; i < result.GetNumberOfValues(); ++i)
{
VTKM_TEST_ASSERT(result.GetPortalConstControl().Get(i) == expected[static_cast<std::size_t>(i)],
"incorrect result");
}
return cs;
}
template <typename CellSetType>
void RunTests(const CellSetType& cellset)
{
std::cout << "\tTesting CellSetPermutation:\n";
auto p1 = TestCellSet(cellset);
std::cout << "\tTesting CellSetPermutation of CellSetPermutation:\n";
TestCellSet(p1);
std::cout << "----------------------------------------------------------\n";
}
void TestCellSetPermutation()
{
vtkm::cont::DataSet dataset;
vtkm::cont::testing::MakeTestDataSet maker;
std::cout << "Testing CellSetStructured<2>\n";
dataset = maker.Make2DUniformDataSet1();
RunTests(dataset.GetCellSet().Cast<vtkm::cont::CellSetStructured<2>>());
std::cout << "Testing CellSetStructured<3>\n";
dataset = maker.Make3DUniformDataSet1();
RunTests(dataset.GetCellSet().Cast<vtkm::cont::CellSetStructured<3>>());
std::cout << "Testing CellSetExplicit\n";
dataset = maker.Make3DExplicitDataSetPolygonal();
RunTests(dataset.GetCellSet().Cast<vtkm::cont::CellSetExplicit<>>());
std::cout << "Testing CellSetSingleType\n";
dataset = maker.Make3DExplicitDataSetCowNose();
RunTests(dataset.GetCellSet().Cast<vtkm::cont::CellSetSingleType<>>());
}
} // anonymous namespace
int UnitTestCellSetPermutation(int, char* [])
{
return vtkm::cont::testing::Testing::Run(TestCellSetPermutation);
}

2
vtkm/cont/testing/UnitTestDataSetBuilderExplicit.cxx
View File

@ -31,7 +31,7 @@
namespace DataSetBuilderExplicitNamespace
{
typedef vtkm::cont::DeviceAdapterAlgorithm<VTKM_DEFAULT_DEVICE_ADAPTER_TAG> DFA;
using DFA = vtkm::cont::DeviceAdapterAlgorithm<VTKM_DEFAULT_DEVICE_ADAPTER_TAG>;
template <typename T>
vtkm::Bounds ComputeBounds(std::size_t numPoints, const T* coords)

4
vtkm/cont/testing/UnitTestDataSetPermutation.cxx
View File

@ -108,7 +108,7 @@ void TestDataSet_Explicit()
dataSet.GetCellSet(0).CopyTo(cellSet);
//verify that we can create a subset of a singlset
typedef vtkm::cont::CellSetPermutation<vtkm::cont::CellSetSingleType<>> SubsetType;
using SubsetType = vtkm::cont::CellSetPermutation<vtkm::cont::CellSetSingleType<>>;
SubsetType subset;
subset.Fill(validCellIds, cellSet);
@ -160,7 +160,7 @@ void TestDataSet_Structured2D()
subset.PrintSummary(std::cout);
//verify that we can call PrepareForInput on CellSetSingleType
typedef vtkm::cont::DeviceAdapterTagSerial DeviceAdapterTag;
using DeviceAdapterTag = vtkm::cont::DeviceAdapterTagSerial;
//verify that PrepareForInput exists
subset.PrepareForInput(

10
vtkm/cont/testing/UnitTestDeviceAdapterAlgorithmGeneral.cxx
View File

@ -48,9 +48,9 @@ struct DeviceAdapterAlgorithm<vtkm::cont::DeviceAdapterTagTestAlgorithmGeneral>
vtkm::cont::DeviceAdapterTagTestAlgorithmGeneral>
{
private:
typedef vtkm::cont::DeviceAdapterAlgorithm<vtkm::cont::DeviceAdapterTagSerial> Algorithm;
using Algorithm = vtkm::cont::DeviceAdapterAlgorithm<vtkm::cont::DeviceAdapterTagSerial>;
typedef vtkm::cont::DeviceAdapterTagTestAlgorithmGeneral DeviceAdapterTagTestAlgorithmGeneral;
using DeviceAdapterTagTestAlgorithmGeneral = vtkm::cont::DeviceAdapterTagTestAlgorithmGeneral;
public:
template <class Functor>
@ -82,9 +82,9 @@ public:
StorageTag,
vtkm::cont::DeviceAdapterTagSerial>
Superclass;
typedef typename Superclass::ValueType ValueType;
typedef typename Superclass::PortalType PortalType;
typedef typename Superclass::PortalConstType PortalConstType;
using ValueType = typename Superclass::ValueType;
using PortalType = typename Superclass::PortalType;
using PortalConstType = typename Superclass::PortalConstType;
ArrayManagerExecution(vtkm::cont::internal::Storage<T, StorageTag>* storage)
: Superclass(storage)

2
vtkm/cont/testing/UnitTestDynamicArrayHandle.cxx
View File

@ -71,7 +71,7 @@ struct TypeListTagString : vtkm::ListTagBase<std::string>
template <typename T>
struct UnusualPortal
{
typedef T ValueType;
using ValueType = T;
VTKM_EXEC_CONT
vtkm::Id GetNumberOfValues() const { return ARRAY_SIZE; }

8
vtkm/cont/testing/UnitTestRuntimeDeviceInformation.cxx
View File

@ -36,7 +36,7 @@ struct DoesExist;
template <typename DeviceAdapterTag>
void detect_if_exists(DeviceAdapterTag tag)
{
typedef vtkm::cont::DeviceAdapterTraits<DeviceAdapterTag> DeviceAdapterTraits;
using DeviceAdapterTraits = vtkm::cont::DeviceAdapterTraits<DeviceAdapterTag>;
DoesExist<DeviceAdapterTraits::Valid>::Exist(tag);
}
@ -68,9 +68,9 @@ struct DoesExist<true>
void Detection()
{
typedef ::vtkm::cont::DeviceAdapterTagSerial SerialTag;
typedef ::vtkm::cont::DeviceAdapterTagTBB TBBTag;
typedef ::vtkm::cont::DeviceAdapterTagCuda CudaTag;
using SerialTag = ::vtkm::cont::DeviceAdapterTagSerial;
using TBBTag = ::vtkm::cont::DeviceAdapterTagTBB;
using CudaTag = ::vtkm::cont::DeviceAdapterTagCuda;
//Verify that for each device adapter we compile code for, that it
//has valid runtime support.

10
vtkm/cont/testing/UnitTestStorageImplicit.cxx
View File

@ -36,7 +36,7 @@ namespace
template <typename T>
struct TestImplicitStorage
{
typedef T ValueType;
using ValueType = T;
ValueType Temp;
VTKM_EXEC_CONT
@ -57,12 +57,12 @@ const vtkm::Id ARRAY_SIZE = 1;
template <typename T>
struct TemplatedTests
{
typedef vtkm::cont::StorageTagImplicit<TestImplicitStorage<T>> StorageTagType;
using StorageTagType = vtkm::cont::StorageTagImplicit<TestImplicitStorage<T>>;
typedef vtkm::cont::internal::Storage<T, StorageTagType> StorageType;
typedef typename StorageType::ValueType ValueType;
typedef typename StorageType::PortalType PortalType;
typedef typename PortalType::IteratorType IteratorType;
using ValueType = typename StorageType::ValueType;
using PortalType = typename StorageType::PortalType;
using IteratorType = typename PortalType::IteratorType;
void BasicAllocation()
{

10
vtkm/cont/testing/UnitTestTryExecute.cxx
View File

@ -51,7 +51,7 @@ struct TryExecuteTestFunctor
template <typename Device>
VTKM_CONT bool operator()(Device)
{
typedef vtkm::cont::DeviceAdapterAlgorithm<Device> Algorithm;
using Algorithm = vtkm::cont::DeviceAdapterAlgorithm<Device>;
Algorithm::Copy(this->InArray, this->OutArray);
this->NumCalls++;
return true;
@ -85,11 +85,11 @@ void TryExecuteWithList(DeviceList, bool expectSuccess)
static void Run()
{
typedef vtkm::cont::DeviceAdapterTagSerial ValidDevice;
typedef vtkm::cont::DeviceAdapterTagError InvalidDevice;
using ValidDevice = vtkm::cont::DeviceAdapterTagSerial;
using InvalidDevice = vtkm::cont::DeviceAdapterTagError;
std::cout << "Try a list with a single entry." << std::endl;
typedef vtkm::ListTagBase<ValidDevice> SingleValidList;
using SingleValidList = vtkm::ListTagBase<ValidDevice>;
TryExecuteWithList(SingleValidList(), true);
std::cout << "Try a list with two valid devices." << std::endl;
@ -97,7 +97,7 @@ static void Run()
TryExecuteWithList(DoubleValidList(), true);
std::cout << "Try a list with only invalid device." << std::endl;
typedef vtkm::ListTagBase<InvalidDevice> SingleInvalidList;
using SingleInvalidList = vtkm::ListTagBase<InvalidDevice>;
TryExecuteWithList(SingleInvalidList(), false);
std::cout << "Try a list with an invalid and valid device." << std::endl;

2
vtkm/exec/AtomicArray.h
View File

@ -54,7 +54,7 @@ template <typename T, typename DeviceAdapterTag>
class AtomicArray : public vtkm::exec::ExecutionObjectBase
{
public:
typedef T ValueType;
using ValueType = T;
VTKM_CONT
AtomicArray()

Some files were not shown because too many files have changed in this diff Show More