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

CMake/VTKmCheckCopyright.cmake

@@ -234,7 +234,6 @@ foreach (glob_expression ${FILES_TO_CHECK})
     endforeach(exception)
 
     if (NOT skip)
-      message("Checking ${file}")
       check_copyright("${VTKm_SOURCE_DIR}/${file}")
     endif (NOT skip)
   endforeach (file)

CMake/VTKmCheckSourceInBuild.cmake

@@ -1,116 +0,0 @@
-##============================================================================
-##  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 2016 National Technology & Engineering Solutions of Sandia, LLC (NTESS).
-##  Copyright 2016 UT-Battelle, LLC.
-##  Copyright 2016 Los Alamos National Security.
-##
-##  Under the terms of Contract DE-NA0003525 with NTESS,
-##  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.
-##============================================================================
-
-## This CMake script checks to make sure that each source file is explicitly
-## listed in the CMakeLists.txt files. This helps ensure that all files that we
-## are using are appropriately listed in IDEs and installed as necessary. It
-## also helps identify dead files that should no longer be in the repository.
-## To run this script, execute CMake as follows:
-##
-## cmake -DVTKm_SOURCE_DIR=<VTKm_SOURCE_DIR> -P <VTKm_SOURCE_DIR>/CMake/VTKMCheckSourceInBuild.cmake
-##
-
-set(FILES_TO_CHECK
-  *.h
-  *.h.in
-  *.cxx
-  *.cu
-  )
-
-set(EXCEPTIONS
-  kxsort.h
-  )
-
-set(DIRECTORY_EXCEPTIONS
-  ${VTKm_SOURCE_DIR}/vtkm/thirdparty/taotuple/vtkmtaotuple
-  ${VTKm_SOURCE_DIR}/vtkm/thirdparty/diy/vtkmdiy
-  )
-
-if (NOT VTKm_SOURCE_DIR)
-  message(SEND_ERROR "VTKm_SOURCE_DIR not defined.")
-endif (NOT VTKm_SOURCE_DIR)
-
-function(check_directory directory parent_CMakeLists_contents)
-  foreach(exception IN LISTS DIRECTORY_EXCEPTIONS)
-    if(directory MATCHES "^${exception}$")
-      return()
-    endif()
-  endforeach(exception)
-
-  message("Checking directory ${directory}...")
-
-  get_filename_component(directory_name "${directory}" NAME)
-
-  if(EXISTS "${directory}/CMakeLists.txt")
-    file(READ "${directory}/CMakeLists.txt" CMakeLists_contents)
-  endif()
-
-  foreach (glob_expression ${FILES_TO_CHECK})
-    file(GLOB file_list
-      RELATIVE "${directory}"
-      "${directory}/${glob_expression}"
-      )
-
-    foreach (file ${file_list})
-      set(skip)
-      foreach(exception ${EXCEPTIONS})
-        if(file MATCHES "^${exception}(/.*)?$")
-          # This file is an exception
-          set(skip TRUE)
-        endif()
-      endforeach(exception)
-
-      if(NOT skip)
-        message("Checking ${file}")
-        # Remove .in suffix. These are generally configured files that generate
-        # new files that are actually used in the build.
-        string(REGEX REPLACE ".in$" "" file_check "${file}")
-        string(FIND "${CMakeLists_contents}" "${file_check}" position)
-        if(${position} LESS 0)
-          # Check the CMakeLists.txt of the parent directory. Some sources of
-          # internal directories are packaged into libraries in the parent
-          # directory.
-          string(FIND "${parent_CMakeLists_contents}"
-            "${directory_name}/${file_check}"
-            position
-            )
-          if(${position} LESS 0)
-            message(SEND_ERROR
-              "****************************************************************
-${file_check} is not found in ${directory}/CMakeLists.txt
-This indicates that the file is not part of the build system. Thus it might be missing build targets. All such files should be explicitly handled by CMake.")
-          endif() # Not in parent's CMakeLists.txt
-        endif() # Not in CMakeLists.txt
-      endif() # Not skipped
-    endforeach (file)
-  endforeach(glob_expression)
-
-  file(GLOB file_list
-    LIST_DIRECTORIES true
-    "${directory}/*")
-  foreach(file ${file_list})
-    if(IS_DIRECTORY "${file}")
-      check_directory("${file}" "${CMakeLists_contents}")
-    endif()
-  endforeach(file)
-endfunction(check_directory)
-
-check_directory("${VTKm_SOURCE_DIR}/vtkm" "")
-check_directory("${VTKm_SOURCE_DIR}/examples" "")

CMake/VTKmCheckSourceInInstall.cmake

@@ -0,0 +1,172 @@
+##============================================================================
+##  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 2016 National Technology & Engineering Solutions of Sandia, LLC (NTESS).
+##  Copyright 2016 UT-Battelle, LLC.
+##  Copyright 2016 Los Alamos National Security.
+##
+##  Under the terms of Contract DE-NA0003525 with NTESS,
+##  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.
+##============================================================================
+
+## This CMake script checks to make sure that each source file explicitly
+## inside a project is installed.
+## To run this script, execute CMake as follows:
+##
+## cmake -DMODE=[INSTALL|VERIFY|CLEANUP]
+#        -DVTKm_SOURCE_DIR=<VTKm_SOURCE_DIR>
+#        -DVTKm_BINARY_DIR=<VTKm_BINARY_DIR>
+#        -DVTKm_INSTALL_INCLUDE_DIR=<VTKm_INSTALL_INCLUDE_DIR>
+#        -DVTKm_ENABLE_RENDERING=<VTKm_ENABLE_RENDERING>
+#        -DVTKm_ENABLE_LOGGING=<VTKm_ENABLE_LOGGING>
+#        -P <VTKm_SOURCE_DIR>/CMake/VTKMCheckSourceInInstall.cmake
+##
+
+if (NOT DEFINED MODE)
+  message(FATAL_ERROR "Need to pass the MODE variable (INSTALL|VERIFY|CLEANUP) so the script knows what to do")
+endif ()
+if (NOT VTKm_SOURCE_DIR)
+  message(FATAL_ERROR "VTKm_SOURCE_DIR not defined.")
+endif ()
+if (NOT VTKm_BINARY_DIR)
+  message(FATAL_ERROR "VTKm_BINARY_DIR not defined.")
+endif ()
+if (NOT VTKm_INSTALL_INCLUDE_DIR)
+  message(FATAL_ERROR "VTKm_INSTALL_INCLUDE_DIR not defined.")
+endif ()
+if (NOT DEFINED VTKm_ENABLE_RENDERING)
+  message(FATAL_ERROR "VTKm_ENABLE_RENDERING not defined.")
+endif ()
+if (NOT DEFINED VTKm_ENABLE_LOGGING)
+  message(FATAL_ERROR "VTKm_ENABLE_LOGGING not defined.")
+endif ()
+
+
+include(CMakeParseArguments)
+# -----------------------------------------------------------------------------
+function(verify_install_per_dir src_directory build_dir)
+  set(options )
+  set(oneValueArgs )
+  set(multiValueArgs EXTENSIONS FILE_EXCEPTIONS DIR_EXCEPTIONS)
+  cmake_parse_arguments(verify
+    "${options}" "${oneValueArgs}" "${multiValueArgs}"
+    ${ARGN}
+    )
+
+  set(files_to_verify )
+  foreach(ext IN LISTS verify_EXTENSIONS)
+    file(GLOB_RECURSE listing
+         RELATIVE "${src_directory}"
+         "${src_directory}/${ext}"
+         )
+    list(APPEND files_to_verify ${listing})
+  endforeach()
+
+  #remove all files that are exempt
+  list(REMOVE_ITEM files_to_verify ${verify_FILE_EXCEPTIONS})
+
+  #remove all files inside directories that match
+  foreach(dir IN LISTS verify_DIR_EXCEPTIONS)
+    list(FILTER files_to_verify EXCLUDE REGEX "^${dir}")
+  endforeach()
+
+  set(to_fail FALSE) # error out after listing all missing headers
+  foreach(file IN LISTS files_to_verify)
+    if(NOT EXISTS ${build_dir}/${file})
+      message(STATUS "file: ${file} not installed \n\tWas expecting it to be at: ${build_dir}/${file}")
+      set(to_fail TRUE)
+    # else()
+    #   message(STATUS "file: ${file} installed")
+    endif()
+  endforeach()
+
+  if(to_fail)
+    message(FATAL_ERROR "unable to find all headers in the install tree")
+  endif()
+endfunction()
+
+# -----------------------------------------------------------------------------
+function(do_install root_dir prefix)
+  #Step 1. Setup up our new install prefix location
+  set(CMAKE_INSTALL_PREFIX  ${root_dir}/${prefix})
+
+  #Step 2. Gather all the cmake_install.cmake files
+  file(GLOB_RECURSE install_files
+     LIST_DIRECTORIES False
+     RELATIVE "${root_dir}"
+     "${root_dir}/*/cmake_install.cmake")
+
+  #Step 3. Execute all the install files
+  if(EXISTS "${root_dir}/cmake_install.cmake")
+    include(${root_dir}/cmake_install.cmake)
+  endif()
+  foreach(file ${install_files})
+    include("${file}")
+  endforeach()
+
+endfunction()
+
+# -----------------------------------------------------------------------------
+function(do_verify root_dir prefix)
+  #Step 1. Setup the extensions to check, and all file and directory
+  # extensions
+  set(files_extensions
+    *.hpp #needed for diy and taotuple
+    *.h
+    *.hxx
+    )
+
+  set(file_exceptions
+    cont/ColorTablePrivate.hxx
+    )
+
+  #by default every header in a testing directory doesn't need to be installed
+  set(directory_exceptions ".*/testing" )
+  # These exceptions should be based on the status of the associated
+  # cmake option
+  if(NOT VTKm_ENABLE_RENDERING)
+    list(APPEND directory_exceptions rendering)
+  endif()
+  if(NOT VTKm_ENABLE_LOGGING)
+    list(APPEND directory_exceptions thirdparty/loguru)
+  endif()
+
+  #Step 2. Verify the installed files match what headers are listed in each
+  # source directory
+  verify_install_per_dir("${VTKm_SOURCE_DIR}/vtkm"
+                         "${root_dir}/${prefix}/${VTKm_INSTALL_INCLUDE_DIR}/vtkm"
+                         EXTENSIONS ${files_extensions}
+                         FILE_EXCEPTIONS ${file_exceptions}
+                         DIR_EXCEPTIONS ${directory_exceptions}
+                         )
+endfunction()
+
+# -----------------------------------------------------------------------------
+function(do_cleanup root_dir prefix)
+  #Step 1. Remove temp directory
+  file(REMOVE_RECURSE "${root_dir}/${prefix}")
+endfunction()
+
+set(root_dir "${VTKm_BINARY_DIR}")
+set(prefix "/CMakeFiles/_tmp_install")
+
+message(STATUS "MODE: ${MODE}")
+if(MODE STREQUAL "INSTALL")
+  do_install(${root_dir} ${prefix})
+elseif(MODE STREQUAL "VERIFY")
+  do_verify(${root_dir} ${prefix})
+elseif(MODE STREQUAL "CLEANUP")
+  do_cleanup(${root_dir} ${prefix})
+endif()
+
+unset(prefix)
+unset(root_dir)

CMake/VTKmWrappers.cmake

@@ -149,16 +149,8 @@ endfunction(vtkm_install_headers)
 
 #-----------------------------------------------------------------------------
 function(vtkm_declare_headers)
-  set(options CUDA)
-  set(oneValueArgs)
-  set(multiValueArgs)
-  cmake_parse_arguments(VTKm_DH "${options}"
-    "${oneValueArgs}" "${multiValueArgs}"
-    ${ARGN}
-    )
-
   vtkm_get_kit_name(name dir_prefix)
-  vtkm_install_headers("${dir_prefix}" ${hfiles})
+  vtkm_install_headers("${dir_prefix}" ${ARGN})
 endfunction(vtkm_declare_headers)
 
 #-----------------------------------------------------------------------------
@@ -254,10 +246,9 @@ function(vtkm_library)
   #generate the export header and install it
   vtkm_generate_export_header(${lib_name})
 
-  #test and install the headers
+  #install the headers
   vtkm_declare_headers(${VTKm_LIB_HEADERS}
-                       EXCLUDE_FROM_TESTING ${VTKm_LIB_TEMPLATE_SOURCES}
+                       ${VTKm_LIB_TEMPLATE_SOURCES})
-                       )
 
   # When building libraries/tests that are part of the VTK-m repository inherit
   # the properties from vtkm_developer_flags. The flags are intended only for

CMakeLists.txt

@@ -152,9 +152,6 @@ include(VTKmWrappers)
 include(VTKmCompilerFlags)
 
 
-#-----------------------------------------------------------------------------
-
-
 #-----------------------------------------------------------------------------
 if (VTKm_ENABLE_TESTING)
   enable_testing()
@@ -171,24 +168,6 @@ if (VTKm_ENABLE_TESTING)
   configure_file(${VTKm_SOURCE_DIR}/CTestCustom.cmake.in
     ${VTKm_BINARY_DIR}/CTestCustom.cmake @ONLY)
 
-  #-----------------------------------------------------------------------------
-  # Add "meta" tests that check the state of the repository
-  # SystemInformation prints out information about the current configuration
-  # CopyrightStatement checks that the copyright statement is in all source files
-  # SourceInBuild checks that all source files are listed in the build
-  add_test(NAME SystemInformation
-    COMMAND ${CMAKE_COMMAND} "-DVTKm_SOURCE_DIR=${VTKm_SOURCE_DIR}" "-DVTKm_BINARY_DIR=${VTKm_BINARY_DIR}" -P "${VTKm_SOURCE_DIR}/CMake/VTKmSystemInformation.cmake"
-    )
-  add_test(NAME CopyrightStatement
-    COMMAND ${CMAKE_COMMAND} "-DVTKm_SOURCE_DIR=${VTKm_SOURCE_DIR}" -P "${VTKm_SOURCE_DIR}/CMake/VTKmCheckCopyright.cmake"
-    )
-  # increase timeout since on some machines CopyrightStatement test takes a long time.
-  set_tests_properties(CopyrightStatement PROPERTIES TIMEOUT 300)
-
-  add_test(NAME SourceInBuild
-    COMMAND ${CMAKE_COMMAND} "-DVTKm_SOURCE_DIR=${VTKm_SOURCE_DIR}" -P "${VTKm_SOURCE_DIR}/CMake/VTKmCheckSourceInBuild.cmake"
-    )
-
   #-----------------------------------------------------------------------------
   # Find the Python interpreter, which we will use during the build process
   find_package(PythonInterp QUIET)
@@ -325,3 +304,59 @@ endif()
 if(VTKm_ENABLE_EXAMPLES)
   add_subdirectory(examples)
 endif(VTKm_ENABLE_EXAMPLES)
+
+#-----------------------------------------------------------------------------
+if (VTKm_ENABLE_TESTING)
+
+  #-----------------------------------------------------------------------------
+  # Add "meta" tests that check the state of the repository
+  # SystemInformation prints out information about the current configuration
+  # CopyrightStatement checks that the copyright statement is in all source files
+  # SourceInBuild checks that all source files are listed in the build
+  # SourceInInstall checks that all source files are installed in the build
+  add_test(NAME SystemInformation
+    COMMAND ${CMAKE_COMMAND} "-DVTKm_SOURCE_DIR=${VTKm_SOURCE_DIR}" "-DVTKm_BINARY_DIR=${VTKm_BINARY_DIR}" -P "${VTKm_SOURCE_DIR}/CMake/VTKmSystemInformation.cmake"
+    )
+  add_test(NAME CopyrightStatement
+    COMMAND ${CMAKE_COMMAND} "-DVTKm_SOURCE_DIR=${VTKm_SOURCE_DIR}" -P "${VTKm_SOURCE_DIR}/CMake/VTKmCheckCopyright.cmake"
+    )
+  # increase timeout since on some machines CopyrightStatement test takes a long time.
+  set_tests_properties(CopyrightStatement PROPERTIES TIMEOUT 300)
+
+  if(NOT VTKm_INSTALL_ONLY_LIBRARIES)
+    set(command_args
+      "-DVTKm_SOURCE_DIR=${VTKm_SOURCE_DIR}"
+      "-DVTKm_BINARY_DIR=${VTKm_BINARY_DIR}"
+      "-DVTKm_INSTALL_INCLUDE_DIR=${VTKm_INSTALL_INCLUDE_DIR}"
+      "-DVTKm_ENABLE_RENDERING=${VTKm_ENABLE_RENDERING}"
+      "-DVTKm_ENABLE_LOGGING=${VTKm_ENABLE_LOGGING}"
+      )
+
+    #By having this as separate tests using fixtures, it will allow us in
+    #the future to write tests that build against the installed version
+    add_test(NAME TestInstallSetup
+      COMMAND ${CMAKE_COMMAND}
+        "-DMODE=INSTALL"
+        ${command_args}
+        -P "${VTKm_SOURCE_DIR}/CMake/VTKmCheckSourceInInstall.cmake"
+      )
+
+    add_test(NAME SourceInInstall
+      COMMAND ${CMAKE_COMMAND}
+        "-DMODE=VERIFY"
+        ${command_args}
+        -P "${VTKm_SOURCE_DIR}/CMake/VTKmCheckSourceInInstall.cmake"
+      )
+
+    add_test(NAME TestInstallCleanup
+      COMMAND ${CMAKE_COMMAND}
+       "-DMODE=CLEANUP"
+        ${command_args}
+        -P "${VTKm_SOURCE_DIR}/CMake/VTKmCheckSourceInInstall.cmake"
+      )
+
+    set_tests_properties(TestInstallSetup PROPERTIES FIXTURES_SETUP vtkm_installed)
+    set_tests_properties(SourceInInstall PROPERTIES FIXTURES_REQUIRED vtkm_installed)
+    set_tests_properties(TestInstallCleanup PROPERTIES FIXTURES_CLEANUP vtkm_installed)
+  endif()
+endif()

CTestCustom.cmake.in

@@ -37,6 +37,9 @@ list(APPEND CTEST_CUSTOM_WARNING_EXCEPTION
   "CONTRIBUTING.md.*warning"
   "CodingConventions.md.*warning"
 
+  # disable static/dynamic weak symbol warnings
+  ".*ld: warning: direct access in function.*"
+
 
   # disable PTX warning about recursive functions. These look like they can't be silenced
   # without disabling all PTX warnings, show hide them on the dashboard.

LICENSE.txt

@@ -77,6 +77,7 @@ vtkm/cont/tbb/internal/parallel_sort.h
 vtkm/cont/tbb/internal/parallel_radix_sort_tbb.h
 vtkm/cont/tbb/internal/kxsort.h
 vtkm/testing/OptionParser.h
+vtkm/thirdparty
 vtkm/internal/brigand.hpp
 version.txt
 

Utilities/DynamicAnalysis/lsan.supp

@@ -1 +1,4 @@
 leak:libX11*
+leak:loguru.hpp
+leak:/usr/lib/x86_64-linux-gnu/openmp*
+leak:/usr/lib/x86_64-linux-gnu/libopen-pal*

benchmarking/BenchmarkArrayTransfer.cxx

@@ -45,7 +45,7 @@ struct BenchmarkArrayTransfer
 {
   using Algo = vtkm::cont::DeviceAdapterAlgorithm<DeviceAdapter>;
   using StorageTag = vtkm::cont::StorageTagBasic;
-  using Timer = vtkm::cont::Timer<DeviceAdapter>;
+  using Timer = vtkm::cont::Timer;
 
   //------------- Functors for benchmarks --------------------------------------
 
@@ -171,7 +171,8 @@ struct BenchmarkArrayTransfer
       ArrayType array = vtkm::cont::make_ArrayHandle(vec);
 
       // Time the copy:
-      Timer timer;
+      Timer timer{ DeviceAdapter() };
+      timer.Start();
       ReadValues<PortalType> functor(array.PrepareForInput(DeviceAdapter()),
                                      ValueTypeTraits::ZeroInitialization());
       Algo::Schedule(functor, this->NumValues);
@@ -212,9 +213,11 @@ struct BenchmarkArrayTransfer
       ArrayType array;
 
       // Time the write:
-      Timer timer;
+      Timer timer{ DeviceAdapter() };
+      timer.Start();
       WriteValues<PortalType> functor(array.PrepareForOutput(this->NumValues, DeviceAdapter()));
       Algo::Schedule(functor, this->NumValues);
+
       return timer.GetElapsedTime();
     }
   };
@@ -254,7 +257,8 @@ struct BenchmarkArrayTransfer
       ArrayType array = vtkm::cont::make_ArrayHandle(vec);
 
       // Time the copy:
-      Timer timer;
+      Timer timer{ DeviceAdapter() };
+      timer.Start();
       ReadWriteValues<PortalType> functor(array.PrepareForInPlace(DeviceAdapter()));
       Algo::Schedule(functor, this->NumValues);
       return timer.GetElapsedTime();
@@ -301,7 +305,8 @@ struct BenchmarkArrayTransfer
       array.ReleaseResourcesExecution();
 
       // Time the copy:
-      Timer timer;
+      Timer timer{ DeviceAdapter() };
+      timer.Start();
 
       // Copy to device:
       ReadValues<PortalExecType> functor(array.PrepareForInput(DeviceAdapter()),
@@ -360,7 +365,8 @@ struct BenchmarkArrayTransfer
       array.ReleaseResourcesExecution();
 
       // Time the copy:
-      Timer timer;
+      Timer timer{ DeviceAdapter() };
+      timer.Start();
 
       // Do work on device:
       ReadWriteValues<PortalExecType> functor(array.PrepareForInPlace(DeviceAdapter()));
@@ -411,7 +417,8 @@ struct BenchmarkArrayTransfer
       ArrayType array;
 
       // Time the copy:
-      Timer timer;
+      Timer timer{ DeviceAdapter() };
+      timer.Start();
 
       // Allocate/write data on device
       WriteValues<PortalExecType> functor(array.PrepareForOutput(this->NumValues, DeviceAdapter()));
@@ -464,7 +471,8 @@ struct BenchmarkArrayTransfer
       ArrayType array;
 
       // Time the copy:
-      Timer timer;
+      Timer timer{ DeviceAdapter() };
+      timer.Start();
 
       // Allocate/write data on device
       WriteValues<PortalExecType> functor(array.PrepareForOutput(this->NumValues, DeviceAdapter()));
@@ -516,7 +524,8 @@ struct BenchmarkArrayTransfer
       ArrayType array;
 
       // Time the copy:
-      Timer timer;
+      Timer timer{ DeviceAdapter() };
+      timer.Start();
 
       // Allocate/write data on device
       WriteValues<PortalExecType> functor(array.PrepareForOutput(this->NumValues, DeviceAdapter()));

benchmarking/BenchmarkAtomicArray.cxx

@@ -62,7 +62,7 @@ class BenchmarkAtomicArray
 {
 public:
   using Algo = vtkm::cont::DeviceAdapterAlgorithm<Device>;
-  using Timer = vtkm::cont::Timer<Device>;
+  using Timer = vtkm::cont::Timer;
 
   // Benchmarks AtomicArray::Add such that each work index writes to adjacent
   // indices.
@@ -102,8 +102,10 @@ public:
       auto portal = array.PrepareForExecution(Device{});
       Worker<decltype(portal)> worker{ this->ArraySize, portal };
 
-      Timer timer;
+      Timer timer{ Device() };
+      timer.Start();
       Algo::Schedule(worker, NumWrites);
+
       return timer.GetElapsedTime();
     }
 
@@ -156,8 +158,10 @@ public:
       auto portal = this->Data.PrepareForOutput(this->ArraySize, Device{});
       Worker<decltype(portal)> worker{ this->ArraySize, portal };
 
-      Timer timer;
+      Timer timer{ Device() };
+      timer.Start();
       Algo::Schedule(worker, NumWrites);
+
       return timer.GetElapsedTime();
     }
 
@@ -217,8 +221,10 @@ public:
       auto portal = array.PrepareForExecution(Device{});
       Worker<decltype(portal)> worker{ this->ArraySize, this->Stride, portal };
 
-      Timer timer;
+      Timer timer{ Device() };
+      timer.Start();
       Algo::Schedule(worker, NumWrites);
+
       return timer.GetElapsedTime();
     }
 
@@ -276,8 +282,10 @@ public:
       auto portal = this->Data.PrepareForOutput(this->ArraySize, Device{});
       Worker<decltype(portal)> worker{ this->ArraySize, this->Stride, portal };
 
-      Timer timer;
+      Timer timer{ Device() };
+      timer.Start();
       Algo::Schedule(worker, NumWrites);
+
       return timer.GetElapsedTime();
     }
 
@@ -342,8 +350,10 @@ public:
       auto portal = array.PrepareForExecution(Device{});
       Worker<decltype(portal)> worker{ this->ArraySize, portal };
 
-      Timer timer;
+      Timer timer{ Device() };
+      timer.Start();
       Algo::Schedule(worker, NumWrites);
+
       return timer.GetElapsedTime();
     }
 
@@ -398,7 +408,8 @@ public:
       auto portal = this->Data.PrepareForOutput(this->ArraySize, Device{});
       Worker<decltype(portal)> worker{ this->ArraySize, portal };
 
-      Timer timer;
+      Timer timer{ Device() };
+      timer.Start();
       Algo::Schedule(worker, NumWrites);
       return timer.GetElapsedTime();
     }
@@ -468,8 +479,10 @@ public:
       auto portal = array.PrepareForExecution(Device{});
       Worker<decltype(portal)> worker{ this->ArraySize, this->Stride, portal };
 
-      Timer timer;
+      Timer timer{ Device() };
+      timer.Start();
       Algo::Schedule(worker, NumWrites);
+
       return timer.GetElapsedTime();
     }
 
@@ -529,8 +542,10 @@ public:
       auto portal = this->Data.PrepareForOutput(this->ArraySize, Device{});
       Worker<decltype(portal)> worker{ this->ArraySize, this->Stride, portal };
 
-      Timer timer;
+      Timer timer{ Device() };
+      timer.Start();
       Algo::Schedule(worker, NumWrites);
+
       return timer.GetElapsedTime();
     }
 

benchmarking/BenchmarkCopySpeeds.cxx

@@ -74,8 +74,10 @@ struct MeasureCopySpeed
 
   VTKM_CONT vtkm::Float64 operator()()
   {
-    vtkm::cont::Timer<DeviceAdapter> timer;
+    vtkm::cont::Timer timer{ DeviceAdapter() };
+    timer.Start();
     Algo::Copy(this->Source, this->Destination);
+
     return timer.GetElapsedTime();
   }
 

benchmarking/BenchmarkDeviceAdapter.cxx

@@ -177,7 +177,7 @@ class BenchmarkDeviceAdapter
 
   using Algorithm = vtkm::cont::DeviceAdapterAlgorithm<DeviceAdapterTag>;
 
-  using Timer = vtkm::cont::Timer<DeviceAdapterTag>;
+  using Timer = vtkm::cont::Timer;
 
 public:
   // Various kernels used by the different benchmarks to accelerate
@@ -284,8 +284,10 @@ private:
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       Algorithm::Copy(ValueHandle_src, ValueHandle_dst);
+
       return timer.GetElapsedTime();
     }
 
@@ -331,8 +333,10 @@ private:
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       Algorithm::CopyIf(ValueHandle, StencilHandle, OutHandle);
+
       return timer.GetElapsedTime();
     }
 
@@ -388,8 +392,10 @@ private:
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       Algorithm::LowerBounds(InputHandle, ValueHandle, OutHandle);
+
       return timer.GetElapsedTime();
     }
 
@@ -442,7 +448,8 @@ private:
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       Value tmp = Algorithm::Reduce(InputHandle, vtkm::TypeTraits<Value>::ZeroInitialization());
       vtkm::Float64 time = timer.GetElapsedTime();
       if (tmp != this->Result)
@@ -494,7 +501,8 @@ private:
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       Algorithm::ReduceByKey(KeyHandle, ValueHandle, KeysOut, ValuesOut, vtkm::Add());
       return timer.GetElapsedTime();
     }
@@ -542,7 +550,8 @@ private:
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       Algorithm::ScanInclusive(ValueHandle, OutHandle);
       return timer.GetElapsedTime();
     }
@@ -580,7 +589,8 @@ private:
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       Algorithm::ScanExclusive(ValueHandle, OutHandle);
       return timer.GetElapsedTime();
     }
@@ -624,7 +634,8 @@ private:
       ValueArrayHandle array;
       Algorithm::Copy(this->ValueHandle, array);
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       Algorithm::Sort(array);
       return timer.GetElapsedTime();
     }
@@ -679,7 +690,8 @@ private:
       Algorithm::Copy(this->KeyHandle, keys);
       Algorithm::Copy(this->ValueHandle, values);
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       Algorithm::SortByKey(keys, values);
       return timer.GetElapsedTime();
     }
@@ -737,7 +749,8 @@ private:
       vtkm::cont::ArrayHandle<vtkm::Id> indices;
       Algorithm::Copy(vtkm::cont::ArrayHandleIndex(arraySize), indices);
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       SSI::Sort(ValueHandle, indices);
       return timer.GetElapsedTime();
     }
@@ -786,7 +799,8 @@ private:
     {
       IndexArrayHandle indices;
       Algorithm::Copy(this->IndexHandle, indices);
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       SSI::Unique(this->ValueHandle, indices);
       return timer.GetElapsedTime();
     }
@@ -843,7 +857,8 @@ private:
       ValueArrayHandle array;
       Algorithm::Copy(this->ValueHandle, array);
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       Algorithm::Unique(array);
       return timer.GetElapsedTime();
     }
@@ -900,7 +915,8 @@ private:
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      vtkm::cont::Timer timer;
+      timer.Start();
       Algorithm::UpperBounds(InputHandle, ValueHandle, OutHandle);
       return timer.GetElapsedTime();
     }

benchmarking/BenchmarkFieldAlgorithms.cxx

@@ -316,7 +316,7 @@ class BenchmarkFieldAlgorithms
 {
   using StorageTag = vtkm::cont::StorageTagBasic;
 
-  using Timer = vtkm::cont::Timer<DeviceAdapterTag>;
+  using Timer = vtkm::cont::Timer;
 
   using ValueVariantHandle = vtkm::cont::VariantArrayHandleBase<ValueTypes>;
   using InterpVariantHandle = vtkm::cont::VariantArrayHandleBase<InterpValueTypes>;
@@ -366,7 +366,8 @@ private:
       const Value RISKFREE = 0.02f;
       const Value VOLATILITY = 0.30f;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       BlackScholes<Value> worklet(RISKFREE, VOLATILITY);
       vtkm::worklet::DispatcherMapField<BlackScholes<Value>> dispatcher(worklet);
       dispatcher.SetDevice(DeviceAdapterTag());
@@ -405,7 +406,8 @@ private:
       const Value RISKFREE = 0.02f;
       const Value VOLATILITY = 0.30f;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       BlackScholes<Value> worklet(RISKFREE, VOLATILITY);
       vtkm::worklet::DispatcherMapField<BlackScholes<Value>> dispatcher(worklet);
       dispatcher.SetDevice(DeviceAdapterTag());
@@ -448,7 +450,8 @@ private:
       vtkm::cont::ArrayHandle<Value> tempHandle1;
       vtkm::cont::ArrayHandle<Value> tempHandle2;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
 
       vtkm::worklet::Invoker invoke(DeviceAdapterTag{});
       invoke(Mag{}, this->InputHandle, tempHandle1);
@@ -487,7 +490,8 @@ private:
       ValueVariantHandle dtemp1(temp1);
       ValueVariantHandle dtemp2(temp2);
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
 
       vtkm::worklet::Invoker invoke(DeviceAdapterTag{});
       invoke(Mag{}, dinput, dtemp1);
@@ -530,10 +534,12 @@ private:
     {
       vtkm::cont::ArrayHandle<Value> result;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       vtkm::worklet::DispatcherMapField<FusedMath> dispatcher;
       dispatcher.SetDevice(DeviceAdapterTag());
       dispatcher.Invoke(this->InputHandle, result);
+
       return timer.GetElapsedTime();
     }
 
@@ -563,10 +569,12 @@ private:
 
       vtkm::cont::ArrayHandle<Value, StorageTag> result;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       vtkm::worklet::DispatcherMapField<FusedMath> dispatcher;
       dispatcher.SetDevice(DeviceAdapterTag());
       dispatcher.Invoke(dinput, result);
+
       return timer.GetElapsedTime();
     }
 
@@ -633,10 +641,12 @@ private:
     {
       vtkm::cont::ArrayHandle<Value> result;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       vtkm::worklet::DispatcherMapField<InterpolateField> dispatcher;
       dispatcher.SetDevice(DeviceAdapterTag());
       dispatcher.Invoke(this->EdgePairHandle, this->WeightHandle, this->FieldHandle, result);
+
       return timer.GetElapsedTime();
     }
 
@@ -666,10 +676,12 @@ private:
       EdgeIdVariantHandle dedges(this->EdgePairHandle);
       vtkm::cont::ArrayHandle<Value> result;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       vtkm::worklet::DispatcherMapField<InterpolateField> dispatcher;
       dispatcher.SetDevice(DeviceAdapterTag());
       dispatcher.Invoke(dedges, dweight, dfield, result);
+
       return timer.GetElapsedTime();
     }
 
@@ -731,10 +743,12 @@ private:
         static_cast<const vtkm::Sphere*>(handle.PrepareForExecution(DeviceAdapterTag()));
       EvalWorklet eval(function);
 
-      vtkm::cont::Timer<DeviceAdapterTag> timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       EvalDispatcher dispatcher(eval);
       dispatcher.SetDevice(DeviceAdapterTag());
       dispatcher.Invoke(this->Internal.Points, this->Internal.Result);
+
       return timer.GetElapsedTime();
     }
 
@@ -767,10 +781,12 @@ private:
       auto sphere = vtkm::cont::make_ImplicitFunctionHandle(Internal.Sphere1);
       EvalWorklet eval(sphere.PrepareForExecution(DeviceAdapterTag()));
 
-      vtkm::cont::Timer<DeviceAdapterTag> timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       EvalDispatcher dispatcher(eval);
       dispatcher.SetDevice(DeviceAdapterTag());
       dispatcher.Invoke(this->Internal.Points, this->Internal.Result);
+
       return timer.GetElapsedTime();
     }
 
@@ -806,10 +822,12 @@ private:
       auto f2 = static_cast<const vtkm::Sphere*>(h2.PrepareForExecution(DeviceAdapterTag()));
       EvalWorklet eval(f1, f2);
 
-      vtkm::cont::Timer<DeviceAdapterTag> timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       EvalDispatcher dispatcher(eval);
       dispatcher.SetDevice(DeviceAdapterTag());
       dispatcher.Invoke(this->Internal.Points, this->Internal.Result);
+
       return timer.GetElapsedTime();
     }
 
@@ -845,10 +863,12 @@ private:
       EvalWorklet eval(s1.PrepareForExecution(DeviceAdapterTag()),
                        s2.PrepareForExecution(DeviceAdapterTag()));
 
-      vtkm::cont::Timer<DeviceAdapterTag> timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       EvalDispatcher dispatcher(eval);
       dispatcher.SetDevice(DeviceAdapterTag());
       dispatcher.Invoke(this->Internal.Points, this->Internal.Result);
+
       return timer.GetElapsedTime();
     }
 

benchmarking/BenchmarkFilters.cxx

@@ -172,7 +172,7 @@ public:
 template <class DeviceAdapterTag>
 class BenchmarkFilters
 {
-  using Timer = vtkm::cont::Timer<DeviceAdapterTag>;
+  using Timer = vtkm::cont::Timer;
 
   enum GradOpts
   {
@@ -229,7 +229,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -316,7 +317,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -354,7 +356,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -379,7 +382,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -403,7 +407,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -430,7 +435,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -456,7 +462,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -498,7 +505,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -542,7 +550,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -576,7 +585,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -601,7 +611,8 @@ class BenchmarkFilters
     VTKM_CONT
     vtkm::Float64 operator()()
     {
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
       this->Filter.Execute(InputDataSet, BenchmarkFilterPolicy());
       return timer.GetElapsedTime();
     }
@@ -650,7 +661,8 @@ class BenchmarkFilters
                                      vtkm::TopologyElementTagPoint{});
         }
 
-        vtkm::cont::Timer<Device> timer;
+        Timer timer{ DeviceAdapterTag() };
+        timer.Start();
         cellSet.PrepareForInput(
           Device{}, vtkm::TopologyElementTagCell{}, vtkm::TopologyElementTagPoint{});
         this->Time = timer.GetElapsedTime();

benchmarking/BenchmarkRayTracing.cxx

@@ -114,7 +114,8 @@ struct BenchRayTracing
   VTKM_CONT
   vtkm::Float64 operator()()
   {
-    vtkm::cont::Timer<VTKM_DEFAULT_DEVICE_ADAPTER_TAG> timer;
+    vtkm::cont::Timer timer;
+    timer.Start();
 
     RayCamera.CreateRays(Rays, Coords.GetBounds());
     Tracer.Render(Rays);

benchmarking/BenchmarkTopologyAlgorithms.cxx

@@ -140,7 +140,7 @@ class BenchmarkTopologyAlgorithms
 {
   using StorageTag = vtkm::cont::StorageTagBasic;
 
-  using Timer = vtkm::cont::Timer<DeviceAdapterTag>;
+  using Timer = vtkm::cont::Timer;
 
   using ValueVariantHandle = vtkm::cont::VariantArrayHandleBase<ValueTypes>;
 
@@ -205,7 +205,8 @@ private:
       cellSet.SetPointDimensions(vtkm::Id3(CUBE_SIZE, CUBE_SIZE, CUBE_SIZE));
       vtkm::cont::ArrayHandle<Value, StorageTag> result;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
 
       vtkm::worklet::DispatcherMapTopology<AverageCellToPoint> dispatcher;
       dispatcher.SetDevice(DeviceAdapterTag());
@@ -241,7 +242,8 @@ private:
       ValueVariantHandle dinput(this->InputHandle);
       vtkm::cont::ArrayHandle<Value, StorageTag> result;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
 
       vtkm::worklet::DispatcherMapTopology<AverageCellToPoint> dispatcher;
       dispatcher.SetDevice(DeviceAdapterTag());
@@ -284,7 +286,8 @@ private:
       cellSet.SetPointDimensions(vtkm::Id3(CUBE_SIZE, CUBE_SIZE, CUBE_SIZE));
       vtkm::cont::ArrayHandle<Value, StorageTag> result;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
 
       vtkm::worklet::DispatcherMapTopology<AveragePointToCell> dispatcher;
       dispatcher.SetDevice(DeviceAdapterTag());
@@ -320,7 +323,8 @@ private:
       ValueVariantHandle dinput(this->InputHandle);
       vtkm::cont::ArrayHandle<Value, StorageTag> result;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
 
       vtkm::worklet::DispatcherMapTopology<AveragePointToCell> dispatcher;
       dispatcher.SetDevice(DeviceAdapterTag());
@@ -367,7 +371,8 @@ private:
 
       ValueVariantHandle dinput(this->InputHandle);
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
 
       Classification<Value> worklet(this->IsoValue);
       vtkm::worklet::DispatcherMapTopology<Classification<Value>> dispatcher(worklet);
@@ -401,13 +406,15 @@ private:
       cellSet.SetPointDimensions(vtkm::Id3(CUBE_SIZE, CUBE_SIZE, CUBE_SIZE));
       vtkm::cont::ArrayHandle<vtkm::IdComponent, StorageTag> result;
 
-      Timer timer;
+      Timer timer{ DeviceAdapterTag() };
+      timer.Start();
 
       Classification<Value> worklet(this->IsoValue);
       vtkm::worklet::DispatcherMapTopology<Classification<Value>> dispatcher(worklet);
       dispatcher.SetDevice(DeviceAdapterTag());
       dispatcher.Invoke(this->InputHandle, cellSet, result);
 
+      timer.Stop();
       return timer.GetElapsedTime();
     }
 

benchmarking/CMakeLists.txt

@@ -69,7 +69,7 @@ set(benchmarks
   )
 
 foreach (benchmark ${benchmarks})
-  add_benchmark(${benchmark} ${benchmark}.cxx vtkm_cont)
+  add_benchmark(${benchmark} ${benchmark}.cxx vtkm_filter)
 endforeach ()
 
 if(TARGET vtkm_rendering)

docs/changelog/asynchronize-device-independent-timer.md

@@ -0,0 +1,65 @@
+# Introduce asynchronous and device independent timer
+
+The timer class now is asynchronous and device independent. it's using an
+similiar API as vtkOpenGLRenderTimer with Start(), Stop(), Reset(), Ready(),
+and GetElapsedTime() function. For convenience and backward compability, Each
+Start() function call will call Reset() internally. GetElapsedTime() function
+can be used multiple times to time sequential operations and Stop() function
+can be helpful when you want to get the elapsed time latter.
+
+Bascially it can be used in two modes:
+
+* Create a Timer without any device info.
+  * It would enable the timer for all enabled devices on the machine. Users can get a
+specific elapsed time by passing a device id into the GetElapsedTime function.
+If no device is provided, it would pick the maximum of all timer results - the
+logic behind this decision is that if cuda is disabled, openmp, serial and tbb
+roughly give the same results; if cuda is enabled it's safe to return the
+maximum elapsed time since users are more interested in the device execution
+time rather than the kernal launch time. The Ready function can be handy here
+to query the status of the timer.
+
+``` Construct a generic timer
+// Assume CUDA is enabled on the machine
+vtkm::cont::Timer timer;
+timer.Start();
+// Run the algorithm
+
+auto timeHost = timer.GetElapsedTime(vtkm::cont::DeviceAdapterTagSerial());
+// To avoid the expensive device synchronization, we query is ready here.
+if (timer.IsReady())
+{
+  auto timeDevice = timer.GetElapsedTime(vtkm::cont::DeviceAdapterTagCuda());
+}
+// Force the synchronization. Ideally device execution time would be returned
+which takes longer time than ther kernal call
+auto timeGeneral = timer.GetElapsedTime();
+```
+
+* Create a Timer with a specific device.
+  * It works as the old timer that times for a specific device id.
+``` Construct a device specific timer
+// Assume TBB is enabled on the machine
+vtkm::cont::Timer timer{vtkm::cont::DeviceAdaptertagTBB()};
+timer.Start(); // t0
+// Run the algorithm
+
+// Timer would just return 0 and warn the user in the logger that an invalid
+// device is used to query elapsed time
+auto timeInvalid = timer.GetElapsedTime(vtkm::cont::DeviceAdapterTagSerial());
+if timer.IsReady()
+{
+  // Either will work and mark t1, return t1-t0
+  auto time1TBB = timer.GetElapsedTime(vtkm::cont::DeviceAdapterTagTBB());
+  auto time1General = timer.GetElapsedTime();
+}
+
+// Do something
+auto time2 = timer.GetElapsedTime(); // t2 will be marked and t2-t0 will be returned
+
+// Do something
+timer.Stop() // t3 marked
+
+// Do something then summarize latter
+auto timeFinal = timer.GetElapsedTime(); // t3-t0
+```

docs/changelog/verify-cmake-install.md

@@ -0,0 +1,9 @@
+# VTK-m now can verify that it installs itself correctly
+
+It was a fairly common occurrence of VTK-m to have a broken install
+tree as it had no easy way to verify that all headers would be installed.
+
+Now VTK-m offers a testing infrastructure that creates a temporary installed
+version and is able to run tests with that VTK-m installed version. Currently
+the only test is to verify that each header listed in VTK-m is also installed,
+but this can expand in the future to include compilation tests.

docs/changelog/vtkm-handles-busy-cuda-devices-better.md

@@ -0,0 +1,17 @@
+# VTK-m CUDA detection properly handles busy devices
+
+When an application that uses VTK-m is first launched it will
+do a check to see if CUDA is supported at runtime. If for
+some reason that CUDA card is not allowing kernel execution
+VTK-m would report the hardware doesn't have CUDA support.
+
+This was problematic as was over aggressive in disabling CUDA
+support for hardware that could support kernel execution in
+the future. With the fact that every VTK-m worklet is executed
+through a TryExecute it is no longer necessary to be so
+aggressive in disabling CUDA support.
+
+Now the behavior is that VTK-m considers a machine to have
+CUDA runtime support if it has 1+ GPU's of Kepler or
+higher hardware (SM_30+).
+

docs/changelog/vtkm-mangle-diy.md

@@ -0,0 +1,11 @@
+# VTK-m thirdparty diy now can coexist with external diy
+
+Previously VTK-m would leak macros that would cause an
+external diy to be incorrectly mangled breaking consumers
+of VTK-m that used diy.
+
+Going forward to use `diy` from VTK-m all calls must use the
+`vtkmdiy` namespace instead of the `diy` namespace. This
+allows for VTK-m to properly forward calls to either
+the external or internal version correctly.
+

examples/clipping/CMakeLists.txt

@@ -28,13 +28,13 @@ find_package(VTKm REQUIRED QUIET)
 add_executable(Clipping_SERIAL Clipping.cxx)
 target_compile_definitions(Clipping_SERIAL PRIVATE
                            "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_SERIAL")
-target_link_libraries(Clipping_SERIAL PRIVATE vtkm_cont)
+target_link_libraries(Clipping_SERIAL PRIVATE vtkm_filter)
 
 if(TARGET vtkm::tbb)
   add_executable(Clipping_TBB Clipping.cxx)
   target_compile_definitions(Clipping_TBB PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_TBB")
-  target_link_libraries(Clipping_TBB PRIVATE vtkm_cont)
+  target_link_libraries(Clipping_TBB PRIVATE vtkm_filter)
 endif()
 
 if(TARGET vtkm::cuda)
@@ -43,5 +43,5 @@ if(TARGET vtkm::cuda)
 
   target_compile_definitions(Clipping_CUDA PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_CUDA")
-  target_link_libraries(Clipping_CUDA PRIVATE vtkm_cont)
+  target_link_libraries(Clipping_CUDA PRIVATE vtkm_filter)
 endif()

examples/clipping/Clipping.cxx

@@ -85,8 +85,10 @@ int main(int argc, char* argv[])
   vtkm::Float32 clipValue = std::stof(argv[argc - 2]);
   vtkm::worklet::Clip clip;
 
-  vtkm::cont::Timer<> total;
+  vtkm::cont::Timer total;
-  vtkm::cont::Timer<> timer;
+  total.Start();
+  vtkm::cont::Timer timer;
+  timer.Start();
   bool invertClip = false;
   vtkm::cont::CellSetExplicit<> outputCellSet =
     clip.Run(input.GetCellSet(0),
@@ -100,12 +102,12 @@ int main(int argc, char* argv[])
 
 
   auto inCoords = input.GetCoordinateSystem(0).GetData();
-  timer.Reset();
+  timer.Start();
   auto outCoords = clip.ProcessCellField(inCoords);
   vtkm::Float64 processCoordinatesTime = timer.GetElapsedTime();
   output.AddCoordinateSystem(vtkm::cont::CoordinateSystem("coordinates", outCoords));
 
-  timer.Reset();
+  timer.Start();
   for (vtkm::Id i = 0; i < input.GetNumberOfFields(); ++i)
   {
     vtkm::cont::Field inField = input.GetField(i);

examples/contour_tree/CMakeLists.txt

@@ -26,10 +26,10 @@ project(ContourTree CXX)
 find_package(VTKm REQUIRED QUIET)
 
 add_executable(ContourTreeMesh2D_SERIAL ContourTreeMesh2D.cxx)
-target_link_libraries(ContourTreeMesh2D_SERIAL vtkm_cont)
+target_link_libraries(ContourTreeMesh2D_SERIAL vtkm_filter)
 
 add_executable(ContourTreeMesh3D_SERIAL ContourTreeMesh3D.cxx)
-target_link_libraries(ContourTreeMesh3D_SERIAL vtkm_cont)
+target_link_libraries(ContourTreeMesh3D_SERIAL vtkm_filter)
 
 target_compile_definitions(ContourTreeMesh2D_SERIAL PRIVATE
                            "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_SERIAL")
@@ -43,8 +43,8 @@ if(TARGET vtkm::cuda)
   vtkm_compile_as_cuda(cudaSource ContourTreeMesh3D.cxx)
   add_executable(ContourTreeMesh3D_CUDA ${cudaSource})
 
-  target_link_libraries(ContourTreeMesh2D_CUDA vtkm_cont)
+  target_link_libraries(ContourTreeMesh2D_CUDA vtkm_filter)
-  target_link_libraries(ContourTreeMesh3D_CUDA vtkm_cont)
+  target_link_libraries(ContourTreeMesh3D_CUDA vtkm_filter)
 
   target_compile_definitions(ContourTreeMesh2D_CUDA PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_CUDA")
@@ -56,8 +56,8 @@ if(TARGET vtkm::tbb)
   add_executable(ContourTreeMesh2D_TBB ContourTreeMesh2D.cxx)
   add_executable(ContourTreeMesh3D_TBB ContourTreeMesh3D.cxx)
 
-  target_link_libraries(ContourTreeMesh2D_TBB vtkm_cont)
+  target_link_libraries(ContourTreeMesh2D_TBB vtkm_filter)
-  target_link_libraries(ContourTreeMesh3D_TBB vtkm_cont)
+  target_link_libraries(ContourTreeMesh3D_TBB vtkm_filter)
 
   target_compile_definitions(ContourTreeMesh2D_TBB PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_TBB")

examples/contour_tree_augmented/CMakeLists.txt

@@ -77,7 +77,7 @@ find_package(VTKm REQUIRED QUIET)
 ####################################
 # Serial 2D / 3D / MC
 add_executable(ContourTree_PPP2_SERIAL ContourTreeApp.cxx)
-target_link_libraries(ContourTree_PPP2_SERIAL vtkm_cont)
+target_link_libraries(ContourTree_PPP2_SERIAL vtkm_filter)
 target_compile_definitions(ContourTree_PPP2_SERIAL PRIVATE
                            "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_SERIAL")
 
@@ -88,7 +88,7 @@ target_compile_definitions(ContourTree_PPP2_SERIAL PRIVATE
 
 # Debug Serial 2D / 3D / MC
 add_executable(ContourTree_PPP2_SERIAL_DEBUG ContourTreeApp.cxx)
-target_link_libraries(ContourTree_PPP2_SERIAL_DEBUG vtkm_cont)
+target_link_libraries(ContourTree_PPP2_SERIAL_DEBUG vtkm_filter)
 target_compile_definitions(ContourTree_PPP2_SERIAL_DEBUG PRIVATE
                            "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_SERIAL" "DEBUG_PRINT")
 
@@ -96,13 +96,13 @@ target_compile_definitions(ContourTree_PPP2_SERIAL_DEBUG PRIVATE
 if(TARGET vtkm::tbb)
   # TBB 2D/3D/MC
   add_executable(ContourTree_PPP2_TBB ContourTreeApp.cxx)
-  target_link_libraries(ContourTree_PPP2_TBB vtkm_cont)
+  target_link_libraries(ContourTree_PPP2_TBB vtkm_filter)
   target_compile_definitions(ContourTree_PPP2_TBB PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_TBB" "ENABLE_SET_NUM_THREADS")
 
   # TBB 2D/3D/MC DEBUG
   add_executable(ContourTree_PPP2_TBB_DEBUG ContourTreeApp.cxx)
-  target_link_libraries(ContourTree_PPP2_TBB_DEBUG vtkm_cont)
+  target_link_libraries(ContourTree_PPP2_TBB_DEBUG vtkm_filter)
   target_compile_definitions(ContourTree_PPP2_TBB_DEBUG PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_TBB" "ENABLE_SET_NUM_THREADS" "DEBUG_PRINT")
 
@@ -112,14 +112,14 @@ if(TARGET vtkm::cuda)
   # CUDA 2D/3D/MC
   vtkm_compile_as_cuda(cudaSource ContourTreeApp.cxx)
   add_executable(ContourTree_PPP2_CUDA ${cudaSource})
-  target_link_libraries(ContourTree_PPP2_CUDA vtkm_cont)
+  target_link_libraries(ContourTree_PPP2_CUDA vtkm_filter)
   target_compile_definitions(ContourTree_PPP2_CUDA PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_CUDA")
 
   # CUDA 2D/3D/MC
   # vtkm_compile_as_cuda(cudaSource ContourTreeApp.cxx)
   add_executable(ContourTree_PPP2_CUDA_DEBUG ${cudaSource})
-  target_link_libraries(ContourTree_PPP2_CUDA_DEBUG vtkm_cont)
+  target_link_libraries(ContourTree_PPP2_CUDA_DEBUG vtkm_filter)
   target_compile_definitions(ContourTree_PPP2_CUDA_DEBUG PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_CUDA" "DEBUG_PRINT")
 endif()

examples/contour_tree_augmented/ContourTreeApp.cxx

@@ -145,8 +145,8 @@ private:
 // Compute and render an isosurface for a uniform grid example
 int main(int argc, char* argv[])
 {
-  // TODO: Change timing to use logging in vtkm/cont/Logging.h
+  vtkm::cont::Timer totalTime;
-  vtkm::cont::Timer<> totalTime;
+  totalTime.Start();
   vtkm::Float64 prevTime = 0;
   vtkm::Float64 currTime = 0;
   std::cout << "ContourTreePPP2Mesh <options> <fileName>" << std::endl;
@@ -330,7 +330,8 @@ int main(int argc, char* argv[])
   if (computeBranchDecomposition)
   {
     // TODO: Change timing to use logging in vtkm/cont/Logging.h
-    vtkm::cont::Timer<> branchDecompTimer;
+    vtkm::cont::Timer branchDecompTimer;
+    branchDecompTimer.Start();
     // compute the volume for each hyperarc and superarc
     cppp2_ns::IdArrayType superarcIntrinsicWeight;
     cppp2_ns::IdArrayType superarcDependentWeight;
@@ -346,6 +347,7 @@ int main(int argc, char* argv[])
     std::cout << std::setw(42) << std::left << "Compute Volume Weights"
               << ": " << branchDecompTimer.GetElapsedTime() << " seconds" << std::endl;
     branchDecompTimer.Reset();
+    branchDecompTimer.Start();
 
     // compute the branch decomposition by volume
     cppp2_ns::IdArrayType whichBranch;

examples/cosmotools/CMakeLists.txt

@@ -36,5 +36,5 @@ else()
   add_executable(CosmoHaloFinder CosmoHaloFinder.cxx)
 endif()
 
-target_link_libraries(CosmoCenterFinder PRIVATE vtkm_cont)
+target_link_libraries(CosmoCenterFinder PRIVATE vtkm_filter)
-target_link_libraries(CosmoHaloFinder PRIVATE vtkm_cont)
+target_link_libraries(CosmoHaloFinder PRIVATE vtkm_filter)

examples/game_of_life/CMakeLists.txt

@@ -35,5 +35,5 @@ if(TARGET OpenGL::GL AND
     add_executable(GameOfLife GameOfLife.cxx LoadShaders.h)
   endif()
 
-  target_link_libraries(GameOfLife PRIVATE vtkm_cont OpenGL::GL GLEW::GLEW GLUT::GLUT)
+  target_link_libraries(GameOfLife PRIVATE vtkm_filter OpenGL::GL GLEW::GLEW GLUT::GLUT)
 endif()

examples/game_of_life/GameOfLife.cxx

@@ -250,7 +250,7 @@ struct RenderGameOfLife
   }
 };
 
-vtkm::cont::Timer<vtkm::cont::DeviceAdapterTagSerial> gTimer;
+vtkm::cont::Timer gTimer{ vtkm::cont::DeviceAdapterTagSerial() };
 vtkm::cont::DataSet* gData = nullptr;
 GameOfLife* gFilter = nullptr;
 RenderGameOfLife* gRenderer = nullptr;
@@ -363,6 +363,7 @@ int main(int argc, char** argv)
   gFilter = &filter;
   gRenderer = &renderer;
 
+  gTimer.Start();
   glutDisplayFunc([]() {
     const vtkm::Float32 c = static_cast<vtkm::Float32>(gTimer.GetElapsedTime());
 

examples/hello_world/CMakeLists.txt

@@ -37,6 +37,6 @@ if(TARGET OpenGL::GL AND
   else()
     add_executable(HelloWorld HelloWorld.cxx LoadShaders.h)
   endif()
-  target_link_libraries(HelloWorld PRIVATE vtkm_cont ${gl_libs})
+  target_link_libraries(HelloWorld PRIVATE vtkm_filter ${gl_libs})
 
 endif()

examples/hello_world/HelloWorld.cxx

@@ -59,7 +59,7 @@ struct HelloVTKMInterop
   vtkm::interop::BufferState VBOState;
   vtkm::interop::BufferState ColorState;
 
-  vtkm::cont::Timer<> Timer;
+  vtkm::cont::Timer Timer;
 
   std::vector<vtkm::Vec<T, 3>> InputData;
   vtkm::cont::ArrayHandle<vtkm::Vec<T, 3>> InHandle;
@@ -78,6 +78,7 @@ struct HelloVTKMInterop
     , OutCoords()
     , OutColors()
   {
+    Timer.Start();
     int dim = 256;
     this->InputData.reserve(static_cast<std::size_t>(dim * dim));
     for (int i = 0; i < dim; ++i)

examples/histogram/CMakeLists.txt

@@ -28,13 +28,13 @@ if (VTKm_ENABLE_MPI)
   add_executable(Histogram_SERIAL Histogram.cxx HistogramMPI.h HistogramMPI.hxx)
   target_compile_definitions(Histogram_SERIAL PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_SERIAL")
-  target_link_libraries(Histogram_SERIAL PRIVATE vtkm_cont)
+  target_link_libraries(Histogram_SERIAL PRIVATE vtkm_filter)
 
   if(TARGET vtkm::tbb)
     add_executable(Histogram_TBB Histogram.cxx HistogramMPI.h HistogramMPI.hxx)
     target_compile_definitions(Histogram_TBB PRIVATE
                                "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_TBB")
-    target_link_libraries(Histogram_TBB PRIVATE vtkm_cont)
+    target_link_libraries(Histogram_TBB PRIVATE vtkm_filter)
   endif()
 
   if(TARGET vtkm::cuda)
@@ -43,6 +43,6 @@ if (VTKm_ENABLE_MPI)
 
     target_compile_definitions(Histogram_CUDA PRIVATE
                                "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_CUDA")
-    target_link_libraries(Histogram_CUDA PRIVATE vtkm_cont)
+    target_link_libraries(Histogram_CUDA PRIVATE vtkm_filter)
   endif()
 endif()

examples/histogram/Histogram.cxx

@@ -31,11 +31,7 @@
 #include <vtkm/cont/DataSetFieldAdd.h>
 #include <vtkm/cont/EnvironmentTracker.h>
 
-// clang-format off
+#include <vtkm/thirdparty/diy/diy.h>
-VTKM_THIRDPARTY_PRE_INCLUDE
-#include VTKM_DIY(diy/mpi.hpp)
-VTKM_THIRDPARTY_POST_INCLUDE
-// clang-format on
 
 #include <mpi.h>
 
@@ -69,7 +65,7 @@ int main(int argc, char* argv[])
   MPI_Init(&argc, &argv);
 
   // tell VTK-m the communicator to use.
-  vtkm::cont::EnvironmentTracker::SetCommunicator(diy::mpi::communicator(MPI_COMM_WORLD));
+  vtkm::cont::EnvironmentTracker::SetCommunicator(vtkmdiy::mpi::communicator(MPI_COMM_WORLD));
 
   int rank, size;
   MPI_Comm_rank(MPI_COMM_WORLD, &rank);

examples/histogram/HistogramMPI.hxx

@@ -28,11 +28,7 @@
 #include <vtkm/cont/ErrorFilterExecution.h>
 #include <vtkm/cont/FieldRangeGlobalCompute.h>
 
-// clang-format off
+#include <vtkm/thirdparty/diy/diy.h>
-VTKM_THIRDPARTY_PRE_INCLUDE
-#include VTKM_DIY(diy/mpi.hpp)
-VTKM_THIRDPARTY_POST_INCLUDE
-// clang-format on
 
 namespace example
 {

examples/isosurface/CMakeLists.txt

@@ -34,16 +34,16 @@ if(TARGET OpenGL::GL AND
 
   set(gl_libs OpenGL::GL OpenGL::GLU GLEW::GLEW GLUT::GLUT)
   add_executable(IsosurfaceUniformGrid_SERIAL IsosurfaceUniformGrid.cxx quaternion.h)
-  target_link_libraries(IsosurfaceUniformGrid_SERIAL PRIVATE vtkm_cont ${gl_libs})
+  target_link_libraries(IsosurfaceUniformGrid_SERIAL PRIVATE vtkm_filter ${gl_libs})
 
   if(TARGET vtkm::cuda)
     add_executable(IsosurfaceUniformGrid_CUDA IsosurfaceUniformGrid.cu quaternion.h)
-    target_link_libraries(IsosurfaceUniformGrid_CUDA PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(IsosurfaceUniformGrid_CUDA PRIVATE vtkm_filter ${gl_libs})
   endif()
 
   if(TARGET vtkm::tbb)
     add_executable(IsosurfaceUniformGrid_TBB IsosurfaceUniformGridTBB.cxx quaternion.h)
-    target_link_libraries(IsosurfaceUniformGrid_TBB PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(IsosurfaceUniformGrid_TBB PRIVATE vtkm_filter ${gl_libs})
   endif()
 
 endif()

examples/lagrangian/CMakeLists.txt

@@ -25,5 +25,5 @@ find_package(VTKm REQUIRED QUIET)
 
 add_executable(Lagrangian_SERIAL lagrangian.cxx ABCfield.h)
 target_compile_definitions(Lagrangian_SERIAL PRIVATE "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_SERIAL")
-target_link_libraries(Lagrangian_SERIAL PRIVATE vtkm_cont)
+target_link_libraries(Lagrangian_SERIAL PRIVATE vtkm_filter)
 target_link_libraries(Lagrangian_SERIAL ${VTKm_LIBRARIES})

examples/multi_backend/CMakeLists.txt

@@ -47,4 +47,4 @@ if(TARGET vtkm::cuda)
 endif()
 
 add_executable(MultiBackend ${device_srcs} ${srcs} ${headers})
-target_link_libraries(MultiBackend PRIVATE vtkm_cont Threads::Threads)
+target_link_libraries(MultiBackend PRIVATE vtkm_filter Threads::Threads)

examples/oscillator/CMakeLists.txt

@@ -28,13 +28,13 @@ find_package(VTKm REQUIRED QUIET)
 add_executable(Oscillator_SERIAL Oscillator.cxx)
 target_compile_definitions(Oscillator_SERIAL PRIVATE
   "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_SERIAL")
-target_link_libraries(Oscillator_SERIAL PRIVATE vtkm_cont)
+target_link_libraries(Oscillator_SERIAL PRIVATE vtkm_filter)
 
 if (TARGET vtkm::tbb)
   add_executable(Oscillator_TBB Oscillator.cxx)
   target_compile_definitions(Oscillator_TBB PRIVATE
     "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_TBB")
-  target_link_libraries(Oscillator_TBB PRIVATE vtkm_cont)
+  target_link_libraries(Oscillator_TBB PRIVATE vtkm_filter)
 endif()
 
 if (TARGET vtkm::cuda)
@@ -42,5 +42,5 @@ if (TARGET vtkm::cuda)
   add_executable(Oscillator_CUDA ${oscillatorCudaSrc})
   target_compile_definitions(Oscillator_CUDA PRIVATE
     "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_CUDA")
-  target_link_libraries(Oscillator_CUDA PRIVATE vtkm_cont)
+  target_link_libraries(Oscillator_CUDA PRIVATE vtkm_filter)
 endif()

examples/particle_advection/CMakeLists.txt

@@ -26,14 +26,14 @@ project(ParticleAdvection CXX)
 find_package(VTKm REQUIRED QUIET)
 
 add_executable(Particle_Advection_SERIAL ParticleAdvection.cxx)
-target_link_libraries(Particle_Advection_SERIAL PRIVATE vtkm_cont)
+target_link_libraries(Particle_Advection_SERIAL PRIVATE vtkm_filter)
 
 if(TARGET vtkm::tbb)
   add_executable(Particle_Advection_TBB ParticleAdvectionTBB.cxx)
-  target_link_libraries(Particle_Advection_TBB PRIVATE vtkm_cont)
+  target_link_libraries(Particle_Advection_TBB PRIVATE vtkm_filter)
 endif()
 
 if(TARGET vtkm::cuda)
   add_executable(Particle_Advection_CUDA ParticleAdvection.cu)
-  target_link_libraries(Particle_Advection_CUDA PRIVATE vtkm_cont)
+  target_link_libraries(Particle_Advection_CUDA PRIVATE vtkm_filter)
 endif()

examples/redistribute_points/CMakeLists.txt

@@ -27,13 +27,13 @@ find_package(VTKm REQUIRED QUIET)
 add_executable(RedistributePoints_SERIAL RedistributePoints.cxx RedistributePoints.h)
 target_compile_definitions(RedistributePoints_SERIAL PRIVATE
                            "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_SERIAL")
-target_link_libraries(RedistributePoints_SERIAL PRIVATE vtkm_cont)
+target_link_libraries(RedistributePoints_SERIAL PRIVATE vtkm_filter)
 
 if(TARGET vtkm::tbb)
   add_executable(RedistributePoints_TBB RedistributePoints.cxx RedistributePoints.h)
   target_compile_definitions(RedistributePoints_TBB PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_TBB")
-  target_link_libraries(RedistributePoints_TBB PRIVATE vtkm_cont)
+  target_link_libraries(RedistributePoints_TBB PRIVATE vtkm_filter)
 endif()
 
 if(TARGET vtkm::cuda)
@@ -42,5 +42,5 @@ if(TARGET vtkm::cuda)
 
   target_compile_definitions(RedistributePoints_CUDA PRIVATE
                              "VTKM_DEVICE_ADAPTER=VTKM_DEVICE_ADAPTER_CUDA")
-  target_link_libraries(RedistributePoints_CUDA PRIVATE vtkm_cont)
+  target_link_libraries(RedistributePoints_CUDA PRIVATE vtkm_filter)
 endif()

examples/redistribute_points/RedistributePoints.cxx

@@ -23,10 +23,7 @@
 #include <vtkm/io/reader/VTKDataSetReader.h>
 #include <vtkm/io/writer/VTKDataSetWriter.h>
 
-// clang-format off
+#include <vtkm/thirdparty/diy/diy.h>
-VTKM_THIRDPARTY_PRE_INCLUDE
-#include VTKM_DIY(diy/mpi.hpp)
-VTKM_THIRDPARTY_POST_INCLUDE
 
 #include "RedistributePoints.h"
 
@@ -36,8 +33,8 @@ using std::endl;
 
 int main(int argc, char* argv[])
 {
-  diy::mpi::environment env(argc, argv);
+  vtkmdiy::mpi::environment env(argc, argv);
-  auto comm = diy::mpi::communicator(MPI_COMM_WORLD);
+  auto comm = vtkmdiy::mpi::communicator(MPI_COMM_WORLD);
   vtkm::cont::EnvironmentTracker::SetCommunicator(comm);
 
   if (argc != 3)

examples/redistribute_points/RedistributePoints.h

@@ -22,20 +22,12 @@
 #include <vtkm/cont/Algorithm.h>
 #include <vtkm/cont/AssignerMultiBlock.h>
 #include <vtkm/cont/BoundsGlobalCompute.h>
+#include <vtkm/cont/EnvironmentTracker.h>
 #include <vtkm/cont/Serialization.h>
 #include <vtkm/filter/ExtractPoints.h>
 #include <vtkm/filter/Filter.h>
 
-// clang-format off
+#include <vtkm/thirdparty/diy/diy.h>
-VTKM_THIRDPARTY_PRE_INCLUDE
-#include VTKM_DIY(diy/decomposition.hpp)
-#include VTKM_DIY(diy/link.hpp)
-#include VTKM_DIY(diy/master.hpp)
-#include VTKM_DIY(diy/proxy.hpp)
-#include VTKM_DIY(diy/reduce.hpp)
-#include VTKM_DIY(diy/reduce-operations.hpp)
-#include VTKM_DIY(diy/types.hpp)
-VTKM_THIRDPARTY_POST_INCLUDE
 
 namespace example
 {
@@ -43,9 +35,9 @@ namespace example
 namespace internal
 {
 
-static diy::ContinuousBounds convert(const vtkm::Bounds& bds)
+static vtkmdiy::ContinuousBounds convert(const vtkm::Bounds& bds)
 {
-  diy::ContinuousBounds result;
+  vtkmdiy::ContinuousBounds result;
   result.min[0] = static_cast<float>(bds.X.Min);
   result.min[1] = static_cast<float>(bds.Y.Min);
   result.min[2] = static_cast<float>(bds.Z.Min);
@@ -59,10 +51,11 @@ static diy::ContinuousBounds convert(const vtkm::Bounds& bds)
 template <typename DerivedPolicy>
 class Redistributor
 {
-  const diy::RegularDecomposer<diy::ContinuousBounds>& Decomposer;
+  const vtkmdiy::RegularDecomposer<vtkmdiy::ContinuousBounds>& Decomposer;
   const vtkm::filter::PolicyBase<DerivedPolicy>& Policy;
 
-  vtkm::cont::DataSet Extract(const vtkm::cont::DataSet& input, const diy::ContinuousBounds& bds) const
+  vtkm::cont::DataSet Extract(const vtkm::cont::DataSet& input,
+                              const vtkmdiy::ContinuousBounds& bds) const
   {
     // extract points
     vtkm::Box box(bds.min[0], bds.max[0], bds.min[1], bds.max[1], bds.min[2], bds.max[2]);
@@ -76,7 +69,11 @@ class Redistributor
   class ConcatenateFields
   {
   public:
-    explicit ConcatenateFields(vtkm::Id totalSize) : TotalSize(totalSize), CurrentIdx(0) {}
+    explicit ConcatenateFields(vtkm::Id totalSize)
+      : TotalSize(totalSize)
+      , CurrentIdx(0)
+    {
+    }
 
     void Append(const vtkm::cont::Field& field)
     {
@@ -97,10 +94,7 @@ class Redistributor
       this->CurrentIdx += field.GetData().GetNumberOfValues();
     }
 
-    const vtkm::cont::Field& GetResult() const
+    const vtkm::cont::Field& GetResult() const { return this->Field; }
-    {
-      return this->Field;
-    }
 
   private:
     struct Allocator
@@ -135,13 +129,14 @@ class Redistributor
   };
 
 public:
-  Redistributor(const diy::RegularDecomposer<diy::ContinuousBounds>& decomposer,
+  Redistributor(const vtkmdiy::RegularDecomposer<vtkmdiy::ContinuousBounds>& decomposer,
-      const vtkm::filter::PolicyBase<DerivedPolicy>& policy)
+                const vtkm::filter::PolicyBase<DerivedPolicy>& policy)
-    : Decomposer(decomposer), Policy(policy)
+    : Decomposer(decomposer)
+    , Policy(policy)
   {
   }
 
-  void operator()(vtkm::cont::DataSet* block, const diy::ReduceProxy& rp) const
+  void operator()(vtkm::cont::DataSet* block, const vtkmdiy::ReduceProxy& rp) const
   {
     if (rp.in_link().size() == 0)
     {
@@ -151,7 +146,7 @@ public:
         {
           auto target = rp.out_link().target(cc);
           // let's get the bounding box for the target block.
-          diy::ContinuousBounds bds;
+          vtkmdiy::ContinuousBounds bds;
           this->Decomposer.fill_bounds(bds, target.gid);
 
           auto extractedDS = this->Extract(*block, bds);
@@ -185,7 +180,7 @@ public:
       else if (receives.size() > 1)
       {
         ConcatenateFields concatCoords(numValues);
-        for (const auto& ds: receives)
+        for (const auto& ds : receives)
         {
           concatCoords.Append(ds.GetCoordinateSystem(0));
         }
@@ -195,7 +190,7 @@ public:
         for (vtkm::IdComponent i = 0; i < receives[0].GetNumberOfFields(); ++i)
         {
           ConcatenateFields concatField(numValues);
-          for (const auto& ds: receives)
+          for (const auto& ds : receives)
           {
             concatField.Append(ds.GetField(i));
           }
@@ -220,13 +215,12 @@ public:
 
   template <typename DerivedPolicy>
   VTKM_CONT vtkm::cont::MultiBlock PrepareForExecution(
-        const vtkm::cont::MultiBlock& input,
+    const vtkm::cont::MultiBlock& input,
-        const vtkm::filter::PolicyBase<DerivedPolicy>& policy);
+    const vtkm::filter::PolicyBase<DerivedPolicy>& policy);
 };
 
 template <typename DerivedPolicy>
-inline VTKM_CONT vtkm::cont::MultiBlock
+inline VTKM_CONT vtkm::cont::MultiBlock RedistributePoints::PrepareForExecution(
-RedistributePoints::PrepareForExecution(
   const vtkm::cont::MultiBlock& input,
   const vtkm::filter::PolicyBase<DerivedPolicy>& policy)
 {
@@ -236,31 +230,30 @@ RedistributePoints::PrepareForExecution(
   vtkm::Bounds gbounds = vtkm::cont::BoundsGlobalCompute(input);
 
   vtkm::cont::AssignerMultiBlock assigner(input.GetNumberOfBlocks());
-  diy::RegularDecomposer<diy::ContinuousBounds> decomposer(/*dim*/3, internal::convert(gbounds), assigner.nblocks());
+  vtkmdiy::RegularDecomposer<vtkmdiy::ContinuousBounds> decomposer(
+    /*dim*/ 3, internal::convert(gbounds), assigner.nblocks());
 
-  diy::Master master(comm,
+  vtkmdiy::Master master(comm,
-      /*threads*/ 1,
+                         /*threads*/ 1,
-      /*limit*/ -1,
+                         /*limit*/ -1,
-      []() -> void* { return new vtkm::cont::DataSet(); },
+                         []() -> void* { return new vtkm::cont::DataSet(); },
-      [](void*ptr) { delete static_cast<vtkm::cont::DataSet*>(ptr); });
+                         [](void* ptr) { delete static_cast<vtkm::cont::DataSet*>(ptr); });
   decomposer.decompose(comm.rank(), assigner, master);
 
   assert(static_cast<vtkm::Id>(master.size()) == input.GetNumberOfBlocks());
   // let's populate local blocks
-  master.foreach(
+  master.foreach ([&input](vtkm::cont::DataSet* ds, const vtkmdiy::Master::ProxyWithLink& proxy) {
-      [&input](vtkm::cont::DataSet* ds, const diy::Master::ProxyWithLink& proxy) {
+    auto lid = proxy.master()->lid(proxy.gid());
-      auto lid = proxy.master()->lid(proxy.gid());
+    *ds = input.GetBlock(lid);
-      *ds = input.GetBlock(lid);
+  });
-      });
 
   internal::Redistributor<DerivedPolicy> redistributor(decomposer, policy);
-  diy::all_to_all(master, assigner, redistributor, /*k=*/2);
+  vtkmdiy::all_to_all(master, assigner, redistributor, /*k=*/2);
 
   vtkm::cont::MultiBlock result;
-  master.foreach(
+  master.foreach ([&result](vtkm::cont::DataSet* ds, const vtkmdiy::Master::ProxyWithLink&) {
-      [&result](vtkm::cont::DataSet* ds, const diy::Master::ProxyWithLink&) {
+    result.AddBlock(*ds);
-      result.AddBlock(*ds);
+  });
-      });
 
   return result;
 }

examples/streamline/CMakeLists.txt

@@ -33,16 +33,16 @@ if(TARGET OpenGL::GL AND
   set(gl_libs OpenGL::GL OpenGL::GLU GLEW::GLEW GLUT::GLUT)
 
   add_executable(StreamLineUniformGrid_SERIAL StreamLineUniformGrid.cxx)
-  target_link_libraries(StreamLineUniformGrid_SERIAL PRIVATE vtkm_cont ${gl_libs})
+  target_link_libraries(StreamLineUniformGrid_SERIAL PRIVATE vtkm_filter ${gl_libs})
 
   if(TARGET vtkm::cont)
     add_executable(StreamLineUniformGrid_CUDA StreamLineUniformGrid.cu)
-    target_link_libraries(StreamLineUniformGrid_CUDA PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(StreamLineUniformGrid_CUDA PRIVATE vtkm_filter ${gl_libs})
   endif()
 
   if(TARGET vtkm::tbb)
     add_executable(StreamLineUniformGrid_TBB StreamLineUniformGridTBB.cxx)
-    target_link_libraries(StreamLineUniformGrid_TBB PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(StreamLineUniformGrid_TBB PRIVATE vtkm_filter ${gl_libs})
   endif()
 
 endif()

examples/temporal_advection/CMakeLists.txt

@@ -28,14 +28,14 @@ project(TemporalAdvection CXX)
 find_package(VTKm REQUIRED QUIET)
 
 add_executable(Temporal_Advection_SERIAL TemporalAdvection.cxx)
-target_link_libraries(Temporal_Advection_SERIAL PRIVATE vtkm_cont)
+target_link_libraries(Temporal_Advection_SERIAL PRIVATE vtkm_filter)
 
 if(TARGET vtkm::tbb)
   add_executable(Temporal_Advection_TBB TemporalAdvectionTBB.cxx)
-  target_link_libraries(Temporal_Advection_TBB PRIVATE vtkm_cont)
+  target_link_libraries(Temporal_Advection_TBB PRIVATE vtkm_filter)
 endif()
 
 if(TARGET vtkm::cuda)
   add_executable(Temporal_Advection_CUDA TemporalAdvection.cu)
-  target_link_libraries(Temporal_Advection_CUDA PRIVATE vtkm_cont)
+  target_link_libraries(Temporal_Advection_CUDA PRIVATE vtkm_filter)
 endif()

examples/tetrahedra/CMakeLists.txt

@@ -37,10 +37,10 @@ if(TARGET OpenGL::GL AND
   add_executable(TetrahedralizeUniformGrid_SERIAL TetrahedralizeUniformGrid.cxx)
   add_executable(TriangulateUniformGrid_SERIAL TriangulateUniformGrid.cxx)
 
-  target_link_libraries(TetrahedralizeExplicitGrid_SERIAL PRIVATE vtkm_cont ${gl_libs})
+  target_link_libraries(TetrahedralizeExplicitGrid_SERIAL PRIVATE vtkm_filter ${gl_libs})
-  target_link_libraries(TriangulateExplicitGrid_SERIAL PRIVATE vtkm_cont ${gl_libs})
+  target_link_libraries(TriangulateExplicitGrid_SERIAL PRIVATE vtkm_filter ${gl_libs})
-  target_link_libraries(TetrahedralizeUniformGrid_SERIAL PRIVATE vtkm_cont ${gl_libs})
+  target_link_libraries(TetrahedralizeUniformGrid_SERIAL PRIVATE vtkm_filter ${gl_libs})
-  target_link_libraries(TriangulateUniformGrid_SERIAL PRIVATE vtkm_cont ${gl_libs})
+  target_link_libraries(TriangulateUniformGrid_SERIAL PRIVATE vtkm_filter ${gl_libs})
 
   if(TARGET vtkm::cuda)
     add_executable(TetrahedralizeExplicitGrid_CUDA TetrahedralizeExplicitGrid.cu)
@@ -48,10 +48,10 @@ if(TARGET OpenGL::GL AND
     add_executable(TetrahedralizeUniformGrid_CUDA TetrahedralizeUniformGrid.cu)
     add_executable(TriangulateUniformGrid_CUDA TriangulateUniformGrid.cu)
 
-    target_link_libraries(TetrahedralizeExplicitGrid_CUDA PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(TetrahedralizeExplicitGrid_CUDA PRIVATE vtkm_filter ${gl_libs})
-    target_link_libraries(TriangulateExplicitGrid_CUDA PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(TriangulateExplicitGrid_CUDA PRIVATE vtkm_filter ${gl_libs})
-    target_link_libraries(TetrahedralizeUniformGrid_CUDA PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(TetrahedralizeUniformGrid_CUDA PRIVATE vtkm_filter ${gl_libs})
-    target_link_libraries(TriangulateUniformGrid_CUDA PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(TriangulateUniformGrid_CUDA PRIVATE vtkm_filter ${gl_libs})
   endif()
 
   if(TARGET vtkm::tbb)
@@ -60,9 +60,9 @@ if(TARGET OpenGL::GL AND
     add_executable(TetrahedralizeUniformGrid_TBB TetrahedralizeUniformGridTBB.cxx)
     add_executable(TriangulateUniformGrid_TBB TriangulateUniformGridTBB.cxx)
 
-    target_link_libraries(TetrahedralizeExplicitGrid_TBB PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(TetrahedralizeExplicitGrid_TBB PRIVATE vtkm_filter ${gl_libs})
-    target_link_libraries(TriangulateExplicitGrid_TBB PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(TriangulateExplicitGrid_TBB PRIVATE vtkm_filter ${gl_libs})
-    target_link_libraries(TetrahedralizeUniformGrid_TBB PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(TetrahedralizeUniformGrid_TBB PRIVATE vtkm_filter ${gl_libs})
-    target_link_libraries(TriangulateUniformGrid_TBB PRIVATE vtkm_cont ${gl_libs})
+    target_link_libraries(TriangulateUniformGrid_TBB PRIVATE vtkm_filter ${gl_libs})
   endif()
 endif()

examples/unified_memory/CMakeLists.txt

@@ -27,6 +27,6 @@ find_package(VTKm REQUIRED QUIET)
 
 if(TARGET vtkm::cuda)
   add_executable(UnifiedMemory_CUDA UnifiedMemory.cu)
-  target_link_libraries(UnifiedMemory_CUDA PRIVATE vtkm_cont)
+  target_link_libraries(UnifiedMemory_CUDA PRIVATE vtkm_filter)
 endif()
 

examples/unified_memory/UnifiedMemory.cu

@@ -84,7 +84,8 @@ int main(int argc, char* argv[])
     //run once to get the CUDA code warmed up
     dispatcher.Invoke(input, output);
 
-    vtkm::cont::Timer<DeviceTag> timer;
+    vtkm::cont::Timer timer{ DeviceTag() };
+    timer.Start();
 
     for (int i = 0; i < 3; ++i)
     {
@@ -117,7 +118,8 @@ int main(int argc, char* argv[])
   //run once to get the CUDA code warmed up
   dispatcher.Invoke(input, output);
 
-  vtkm::cont::Timer<DeviceTag> timer;
+  vtkm::cont::Timer timer{ DeviceTag() };
+  timer.Start();
 
   for (int i = 0; i < 3; ++i)
   {

vtkm/CMakeLists.txt

@@ -47,6 +47,7 @@ set(headers
   Pair.h
   Range.h
   RangeId.h
+  RangeId2.h
   RangeId3.h
   StaticAssert.h
   Swap.h

vtkm/Math.h

@@ -36,7 +36,7 @@
 #include <stdlib.h>
 #endif // !VTKM_CUDA
 
-#if !defined(__CUDA_ARCH__)
+#if !defined(VTKM_CUDA_DEVICE_PASS)
 #define VTKM_USE_STL
 #include <algorithm>
 #endif

vtkm/Math.h.in

@@ -48,7 +48,7 @@ $# Ignore the following comment. It is meant for the generated file.
 #include <stdlib.h>
 #endif // !VTKM_CUDA
 
-#if !defined(__CUDA_ARCH__)
+#if !defined(VTKM_CUDA_DEVICE_PASS)
 #define VTKM_USE_STL
 #include <algorithm>
 #endif

vtkm/RangeId2.h

@@ -0,0 +1,166 @@
+//============================================================================
+//  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 National Technology & Engineering Solutions of Sandia, LLC (NTESS).
+//  Copyright 2017 UT-Battelle, LLC.
+//  Copyright 2017 Los Alamos National Security.
+//
+//  Under the terms of Contract DE-NA0003525 with NTESS,
+//  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_RangeId2_h
+#define vtk_m_RangeId2_h
+
+#include <vtkm/RangeId.h>
+
+namespace vtkm
+{
+
+/// \brief Represent 2D integer range.
+///
+/// \c vtkm::RangeId2 is a helper class for representing a 2D range of integer
+/// values. The typical use of this class is to express a box of indices
+/// in the x, y, and z directions.
+///
+/// \c RangeId2 also contains several helper functions for computing and
+/// maintaining the range.
+///
+struct RangeId2
+{
+  vtkm::RangeId X;
+  vtkm::RangeId Y;
+
+  RangeId2() = default;
+
+  VTKM_EXEC_CONT
+  RangeId2(const vtkm::RangeId& xrange, const vtkm::RangeId& yrange)
+    : X(xrange)
+    , Y(yrange)
+  {
+  }
+
+  VTKM_EXEC_CONT
+  RangeId2(vtkm::Id minX, vtkm::Id maxX, vtkm::Id minY, vtkm::Id maxY)
+    : X(vtkm::RangeId(minX, maxX))
+    , Y(vtkm::RangeId(minY, maxY))
+  {
+  }
+
+  /// Initialize range with an array of 6 values in the order xmin, xmax,
+  /// ymin, ymax, zmin, zmax.
+  ///
+  VTKM_EXEC_CONT
+  explicit RangeId2(const vtkm::Id range[4])
+    : X(vtkm::RangeId(range[0], range[1]))
+    , Y(vtkm::RangeId(range[2], range[3]))
+  {
+  }
+
+  /// Initialize range with the minimum and the maximum corners
+  ///
+  VTKM_EXEC_CONT
+  RangeId2(const vtkm::Id2& min, const vtkm::Id2& max)
+    : X(vtkm::RangeId(min[0], max[0]))
+    , Y(vtkm::RangeId(min[1], max[1]))
+  {
+  }
+
+  /// \b Determine if the range is non-empty.
+  ///
+  /// \c IsNonEmpty returns true if the range is non-empty.
+  ///
+  VTKM_EXEC_CONT
+  bool IsNonEmpty() const { return (this->X.IsNonEmpty() && this->Y.IsNonEmpty()); }
+
+  /// \b Determines if an Id2 value is within the range.
+  ///
+  VTKM_EXEC_CONT
+  bool Contains(const vtkm::Id2& val) const
+  {
+    return (this->X.Contains(val[0]) && this->Y.Contains(val[1]));
+  }
+
+  /// \b Returns the center of the range.
+  ///
+  /// \c Center computes the middle of the range.
+  ///
+  VTKM_EXEC_CONT
+  vtkm::Id2 Center() const { return vtkm::Id2(this->X.Center(), this->Y.Center()); }
+
+  VTKM_EXEC_CONT
+  vtkm::Id2 Dimensions() const { return vtkm::Id2(this->X.Length(), this->Y.Length()); }
+
+  /// \b Expand range to include a value.
+  ///
+  /// This version of \c Include expands the range just enough to include the
+  /// given value. If the range already include this value, then
+  /// nothing is done.
+  ///
+  template <typename T>
+  VTKM_EXEC_CONT void Include(const vtkm::Vec<T, 2>& point)
+  {
+    this->X.Include(point[0]);
+    this->Y.Include(point[1]);
+  }
+
+  /// \b Expand range to include other range.
+  ///
+  /// This version of \c Include expands the range just enough to include
+  /// the other range. Essentially it is the union of the two ranges.
+  ///
+  VTKM_EXEC_CONT
+  void Include(const vtkm::RangeId2& range)
+  {
+    this->X.Include(range.X);
+    this->Y.Include(range.Y);
+  }
+
+  /// \b Return the union of this and another range.
+  ///
+  /// This is a nondestructive form of \c Include.
+  ///
+  VTKM_EXEC_CONT
+  vtkm::RangeId2 Union(const vtkm::RangeId2& other) const
+  {
+    vtkm::RangeId2 unionRangeId2(*this);
+    unionRangeId2.Include(other);
+    return unionRangeId2;
+  }
+
+  /// \b Operator for union
+  ///
+  VTKM_EXEC_CONT
+  vtkm::RangeId2 operator+(const vtkm::RangeId2& other) const { return this->Union(other); }
+
+  VTKM_EXEC_CONT
+  bool operator==(const vtkm::RangeId2& range) const
+  {
+    return ((this->X == range.X) && (this->Y == range.Y));
+  }
+
+  VTKM_EXEC_CONT
+  bool operator!=(const vtkm::RangeId2& range) const
+  {
+    return ((this->X != range.X) || (this->Y != range.Y));
+  }
+};
+
+} // namespace vtkm
+
+/// Helper function for printing range during testing
+///
+static inline VTKM_CONT std::ostream& operator<<(std::ostream& stream, const vtkm::RangeId2& range)
+{
+  return stream << "{ X:" << range.X << ", Y:" << range.Y << " }";
+}
+
+#endif //vtk_m_RangeId2_h

vtkm/cont/ArrayHandle.h

@@ -696,14 +696,14 @@ namespace internal
 {
 
 template <typename T, typename S>
-void VTKM_CONT ArrayHandleDefaultSerialization(diy::BinaryBuffer& bb,
+void VTKM_CONT ArrayHandleDefaultSerialization(vtkmdiy::BinaryBuffer& bb,
                                                const vtkm::cont::ArrayHandle<T, S>& obj);
 
 } // internal
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename T>

vtkm/cont/ArrayHandle.hxx

@@ -357,36 +357,36 @@ namespace internal
 namespace detail
 {
 template <typename ArrayHandle>
-inline void VTKM_CONT StorageSerialization(diy::BinaryBuffer& bb,
+inline void VTKM_CONT StorageSerialization(vtkmdiy::BinaryBuffer& bb,
                                            const ArrayHandle& obj,
                                            std::false_type)
 {
   vtkm::Id count = obj.GetNumberOfValues();
-  diy::save(bb, count);
+  vtkmdiy::save(bb, count);
 
-  diy::save(bb, vtkm::Id(0)); //not a basic storage
+  vtkmdiy::save(bb, vtkm::Id(0)); //not a basic storage
   auto portal = obj.GetPortalConstControl();
   for (vtkm::Id i = 0; i < count; ++i)
   {
-    diy::save(bb, portal.Get(i));
+    vtkmdiy::save(bb, portal.Get(i));
   }
 }
 
 template <typename ArrayHandle>
-inline void VTKM_CONT StorageSerialization(diy::BinaryBuffer& bb,
+inline void VTKM_CONT StorageSerialization(vtkmdiy::BinaryBuffer& bb,
                                            const ArrayHandle& obj,
                                            std::true_type)
 {
   vtkm::Id count = obj.GetNumberOfValues();
-  diy::save(bb, count);
+  vtkmdiy::save(bb, count);
 
-  diy::save(bb, vtkm::Id(1)); //is basic storage
+  vtkmdiy::save(bb, vtkm::Id(1)); //is basic storage
-  diy::save(bb, obj.GetStorage().GetArray(), static_cast<std::size_t>(count));
+  vtkmdiy::save(bb, obj.GetStorage().GetArray(), static_cast<std::size_t>(count));
 }
 }
 
 template <typename T, typename S>
-inline void VTKM_CONT ArrayHandleDefaultSerialization(diy::BinaryBuffer& bb,
+inline void VTKM_CONT ArrayHandleDefaultSerialization(vtkmdiy::BinaryBuffer& bb,
                                                       const vtkm::cont::ArrayHandle<T, S>& obj)
 {
   using is_basic = typename std::is_same<S, vtkm::cont::StorageTagBasic>::type;
@@ -396,7 +396,7 @@ inline void VTKM_CONT ArrayHandleDefaultSerialization(diy::BinaryBuffer& bb,
 }
 } // vtkm::cont::internal
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename T>
@@ -404,14 +404,14 @@ VTKM_CONT void Serialization<vtkm::cont::ArrayHandle<T>>::load(BinaryBuffer& bb,
                                                                vtkm::cont::ArrayHandle<T>& obj)
 {
   vtkm::Id count = 0;
-  diy::load(bb, count);
+  vtkmdiy::load(bb, count);
   obj.Allocate(count);
 
   vtkm::Id input_was_basic_storage = 0;
-  diy::load(bb, input_was_basic_storage);
+  vtkmdiy::load(bb, input_was_basic_storage);
   if (input_was_basic_storage)
   {
-    diy::load(bb, obj.GetStorage().GetArray(), static_cast<std::size_t>(count));
+    vtkmdiy::load(bb, obj.GetStorage().GetArray(), static_cast<std::size_t>(count));
   }
   else
   {
@@ -419,7 +419,7 @@ VTKM_CONT void Serialization<vtkm::cont::ArrayHandle<T>>::load(BinaryBuffer& bb,
     for (vtkm::Id i = 0; i < count; ++i)
     {
       T val{};
-      diy::load(bb, val);
+      vtkmdiy::load(bb, val);
       portal.Set(i, val);
     }
   }

vtkm/cont/ArrayHandleCartesianProduct.h

@@ -458,7 +458,7 @@ struct TypeString<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename AH1, typename AH2, typename AH3>
@@ -472,9 +472,9 @@ public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
     auto storage = obj.GetStorage();
-    diy::save(bb, storage.GetFirstArray());
+    vtkmdiy::save(bb, storage.GetFirstArray());
-    diy::save(bb, storage.GetSecondArray());
+    vtkmdiy::save(bb, storage.GetSecondArray());
-    diy::save(bb, storage.GetThirdArray());
+    vtkmdiy::save(bb, storage.GetThirdArray());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
@@ -483,9 +483,9 @@ public:
     AH2 array2;
     AH3 array3;
 
-    diy::load(bb, array1);
+    vtkmdiy::load(bb, array1);
-    diy::load(bb, array2);
+    vtkmdiy::load(bb, array2);
-    diy::load(bb, array3);
+    vtkmdiy::load(bb, array3);
 
     obj = vtkm::cont::make_ArrayHandleCartesianProduct(array1, array2, array3);
   }

vtkm/cont/ArrayHandleCast.h

@@ -129,7 +129,7 @@ struct TypeString<vtkm::cont::ArrayHandleCast<T, AH>>
 }
 } // namespace vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename T1, typename T2>

vtkm/cont/ArrayHandleCompositeVector.h

@@ -752,7 +752,7 @@ struct TypeString<vtkm::cont::ArrayHandle<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 namespace internal
@@ -804,7 +804,7 @@ private:
     template <typename AH>
     void operator()(const AH& ah, BinaryBuffer& bb) const
     {
-      diy::save(bb, ah);
+      vtkmdiy::save(bb, ah);
     }
   };
 
@@ -813,7 +813,7 @@ private:
     template <typename AH>
     void operator()(AH& ah, BinaryBuffer& bb) const
     {
-      diy::load(bb, ah);
+      vtkmdiy::load(bb, ah);
     }
   };
 

vtkm/cont/ArrayHandleConcatenate.h

@@ -351,7 +351,7 @@ struct TypeString<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename AH1, typename AH2>
@@ -365,8 +365,8 @@ public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
     auto storage = obj.GetStorage();
-    diy::save(bb, storage.GetArray1());
+    vtkmdiy::save(bb, storage.GetArray1());
-    diy::save(bb, storage.GetArray2());
+    vtkmdiy::save(bb, storage.GetArray2());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
@@ -374,8 +374,8 @@ public:
     AH1 array1;
     AH2 array2;
 
-    diy::load(bb, array1);
+    vtkmdiy::load(bb, array1);
-    diy::load(bb, array2);
+    vtkmdiy::load(bb, array2);
 
     obj = vtkm::cont::make_ArrayHandleConcatenate(array1, array2);
   }

vtkm/cont/ArrayHandleConstant.h

@@ -110,7 +110,7 @@ struct TypeString<vtkm::cont::ArrayHandleConstant<T>>
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename T>

vtkm/cont/ArrayHandleCounting.h

@@ -174,7 +174,7 @@ struct TypeString<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename T>
@@ -188,9 +188,9 @@ public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
     auto portal = obj.GetPortalConstControl();
-    diy::save(bb, portal.GetStart());
+    vtkmdiy::save(bb, portal.GetStart());
-    diy::save(bb, portal.GetStep());
+    vtkmdiy::save(bb, portal.GetStep());
-    diy::save(bb, portal.GetNumberOfValues());
+    vtkmdiy::save(bb, portal.GetNumberOfValues());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
@@ -198,9 +198,9 @@ public:
     T start{}, step{};
     vtkm::Id count = 0;
 
-    diy::load(bb, start);
+    vtkmdiy::load(bb, start);
-    diy::load(bb, step);
+    vtkmdiy::load(bb, step);
-    diy::load(bb, count);
+    vtkmdiy::load(bb, count);
 
     obj = vtkm::cont::make_ArrayHandleCounting(start, step, count);
   }

vtkm/cont/ArrayHandleExtractComponent.h

@@ -337,7 +337,7 @@ struct TypeString<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename AH>
@@ -351,16 +351,16 @@ public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
     auto storage = obj.GetStorage();
-    diy::save(bb, storage.GetComponent());
+    vtkmdiy::save(bb, storage.GetComponent());
-    diy::save(bb, storage.GetArray());
+    vtkmdiy::save(bb, storage.GetArray());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
   {
     vtkm::IdComponent component = 0;
     AH array;
-    diy::load(bb, component);
+    vtkmdiy::load(bb, component);
-    diy::load(bb, array);
+    vtkmdiy::load(bb, array);
 
     obj = vtkm::cont::make_ArrayHandleExtractComponent(array, component);
   }

vtkm/cont/ArrayHandleGroupVec.h

@@ -393,7 +393,7 @@ struct TypeString<vtkm::cont::ArrayHandle<vtkm::Vec<typename AH::ValueType, NUM_
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename AH, vtkm::IdComponent NUM_COMPS>
@@ -406,13 +406,13 @@ private:
 public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
-    diy::save(bb, obj.GetStorage().GetSourceArray());
+    vtkmdiy::save(bb, obj.GetStorage().GetSourceArray());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
   {
     AH array;
-    diy::load(bb, array);
+    vtkmdiy::load(bb, array);
 
     obj = vtkm::cont::make_ArrayHandleGroupVec<NUM_COMPS>(array);
   }

vtkm/cont/ArrayHandleGroupVecVariable.h

@@ -534,7 +534,7 @@ struct TypeString<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename SAH, typename OAH>
@@ -547,8 +547,8 @@ private:
 public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
-    diy::save(bb, obj.GetStorage().GetSourceArray());
+    vtkmdiy::save(bb, obj.GetStorage().GetSourceArray());
-    diy::save(bb, obj.GetStorage().GetOffsetsArray());
+    vtkmdiy::save(bb, obj.GetStorage().GetOffsetsArray());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
@@ -556,8 +556,8 @@ public:
     SAH src;
     OAH off;
 
-    diy::load(bb, src);
+    vtkmdiy::load(bb, src);
-    diy::load(bb, off);
+    vtkmdiy::load(bb, off);
 
     obj = vtkm::cont::make_ArrayHandleGroupVecVariable(src, off);
   }

vtkm/cont/ArrayHandleImplicit.h

@@ -173,7 +173,7 @@ struct TypeString<vtkm::cont::ArrayHandle<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename Functor>
@@ -186,17 +186,17 @@ private:
 public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
-    diy::save(bb, obj.GetNumberOfValues());
+    vtkmdiy::save(bb, obj.GetNumberOfValues());
-    diy::save(bb, obj.GetPortalConstControl().GetFunctor());
+    vtkmdiy::save(bb, obj.GetPortalConstControl().GetFunctor());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
   {
     vtkm::Id count = 0;
-    diy::load(bb, count);
+    vtkmdiy::load(bb, count);
 
     Functor functor;
-    diy::load(bb, functor);
+    vtkmdiy::load(bb, functor);
 
     obj = vtkm::cont::make_ArrayHandleImplicit(functor, count);
   }

vtkm/cont/ArrayHandleIndex.h

@@ -83,7 +83,7 @@ struct TypeString<vtkm::cont::ArrayHandleIndex>
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <>

vtkm/cont/ArrayHandlePermutation.h

@@ -400,7 +400,7 @@ struct TypeString<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename IdxAH, typename ValAH>
@@ -414,8 +414,8 @@ public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
     auto storage = obj.GetStorage();
-    diy::save(bb, storage.GetIndexArray());
+    vtkmdiy::save(bb, storage.GetIndexArray());
-    diy::save(bb, storage.GetValueArray());
+    vtkmdiy::save(bb, storage.GetValueArray());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
@@ -423,8 +423,8 @@ public:
     IdxAH indices;
     ValAH values;
 
-    diy::load(bb, indices);
+    vtkmdiy::load(bb, indices);
-    diy::load(bb, values);
+    vtkmdiy::load(bb, values);
 
     obj = vtkm::cont::make_ArrayHandlePermutation(indices, values);
   }

vtkm/cont/ArrayHandleReverse.h

@@ -269,7 +269,7 @@ struct TypeString<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename AH>
@@ -282,13 +282,13 @@ private:
 public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
-    diy::save(bb, obj.GetStorage().GetArray());
+    vtkmdiy::save(bb, obj.GetStorage().GetArray());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
   {
     AH array;
-    diy::load(bb, array);
+    vtkmdiy::load(bb, array);
     obj = vtkm::cont::make_ArrayHandleReverse(array);
   }
 };

vtkm/cont/ArrayHandleSwizzle.h

@@ -426,7 +426,7 @@ struct TypeString<vtkm::cont::ArrayHandle<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename AH, vtkm::IdComponent NComps>
@@ -440,16 +440,16 @@ public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
     auto storage = obj.GetStorage();
-    diy::save(bb, storage.GetArray());
+    vtkmdiy::save(bb, storage.GetArray());
-    diy::save(bb, storage.GetMap());
+    vtkmdiy::save(bb, storage.GetMap());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
   {
     AH array;
-    diy::load(bb, array);
+    vtkmdiy::load(bb, array);
     vtkm::Vec<vtkm::IdComponent, NComps> map;
-    diy::load(bb, map);
+    vtkmdiy::load(bb, map);
     obj = vtkm::cont::make_ArrayHandleSwizzle(array, map);
   }
 };

vtkm/cont/ArrayHandleTransform.h

@@ -778,7 +778,7 @@ struct TypeString<vtkm::cont::ArrayHandle<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename AH, typename Functor>
@@ -792,16 +792,16 @@ public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
     auto storage = obj.GetStorage();
-    diy::save(bb, storage.GetArray());
+    vtkmdiy::save(bb, storage.GetArray());
-    diy::save(bb, storage.GetFunctor());
+    vtkmdiy::save(bb, storage.GetFunctor());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
   {
     AH array;
-    diy::load(bb, array);
+    vtkmdiy::load(bb, array);
     Functor functor;
-    diy::load(bb, functor);
+    vtkmdiy::load(bb, functor);
     obj = vtkm::cont::make_ArrayHandleTransform(array, functor);
   }
 };
@@ -817,19 +817,19 @@ public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
     auto storage = obj.GetStorage();
-    diy::save(bb, storage.GetArray());
+    vtkmdiy::save(bb, storage.GetArray());
-    diy::save(bb, storage.GetFunctor());
+    vtkmdiy::save(bb, storage.GetFunctor());
-    diy::save(bb, storage.GetInverseFunctor());
+    vtkmdiy::save(bb, storage.GetInverseFunctor());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
   {
     AH array;
-    diy::load(bb, array);
+    vtkmdiy::load(bb, array);
     Functor functor;
-    diy::load(bb, functor);
+    vtkmdiy::load(bb, functor);
     InvFunctor invFunctor;
-    diy::load(bb, invFunctor);
+    vtkmdiy::load(bb, invFunctor);
     obj = vtkm::cont::make_ArrayHandleTransform(array, functor, invFunctor);
   }
 };

vtkm/cont/ArrayHandleUniformPointCoordinates.h

@@ -85,7 +85,7 @@ struct TypeString<vtkm::cont::ArrayHandle<
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <>
@@ -99,9 +99,9 @@ public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
     auto portal = obj.GetPortalConstControl();
-    diy::save(bb, portal.GetDimensions());
+    vtkmdiy::save(bb, portal.GetDimensions());
-    diy::save(bb, portal.GetOrigin());
+    vtkmdiy::save(bb, portal.GetOrigin());
-    diy::save(bb, portal.GetSpacing());
+    vtkmdiy::save(bb, portal.GetSpacing());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
@@ -109,9 +109,9 @@ public:
     vtkm::Id3 dims;
     typename BaseType::ValueType origin, spacing;
 
-    diy::load(bb, dims);
+    vtkmdiy::load(bb, dims);
-    diy::load(bb, origin);
+    vtkmdiy::load(bb, origin);
-    diy::load(bb, spacing);
+    vtkmdiy::load(bb, spacing);
 
     obj = vtkm::cont::ArrayHandleUniformPointCoordinates(dims, origin, spacing);
   }

vtkm/cont/ArrayHandleVirtual.hxx

@@ -53,14 +53,15 @@ ArrayHandleType inline ArrayHandle<T, StorageTagVirtual>::CastToType(
 
 //=============================================================================
 // Specializations of serialization related classes
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename T>
 struct Serialization<vtkm::cont::ArrayHandleVirtual<T>>
 {
 
-  static VTKM_CONT void save(diy::BinaryBuffer& bb, const vtkm::cont::ArrayHandleVirtual<T>& obj)
+  static VTKM_CONT void save(vtkmdiy::BinaryBuffer& bb,
+                             const vtkm::cont::ArrayHandleVirtual<T>& obj)
   {
     vtkm::cont::internal::ArrayHandleDefaultSerialization(bb, obj);
   }
@@ -68,7 +69,7 @@ struct Serialization<vtkm::cont::ArrayHandleVirtual<T>>
   static VTKM_CONT void load(BinaryBuffer& bb, vtkm::cont::ArrayHandleVirtual<T>& obj)
   {
     vtkm::cont::ArrayHandle<T> array;
-    diy::load(bb, array);
+    vtkmdiy::load(bb, array);
     obj = std::move(vtkm::cont::ArrayHandleVirtual<T>{ array });
   }
 };
@@ -80,29 +81,30 @@ struct IntAnySerializer
   using ConstantType = vtkm::cont::ArrayHandleConstant<T>;
   using BasicType = vtkm::cont::ArrayHandle<T>;
 
-  static VTKM_CONT void save(diy::BinaryBuffer& bb, const vtkm::cont::ArrayHandleVirtual<T>& obj)
+  static VTKM_CONT void save(vtkmdiy::BinaryBuffer& bb,
+                             const vtkm::cont::ArrayHandleVirtual<T>& obj)
   {
     if (obj.template IsType<CountingType>())
     {
-      diy::save(bb, vtkm::cont::TypeString<CountingType>::Get());
+      vtkmdiy::save(bb, vtkm::cont::TypeString<CountingType>::Get());
 
       using S = typename CountingType::StorageTag;
       const vtkm::cont::StorageVirtual* storage = obj.GetStorage();
       auto* any = storage->Cast<vtkm::cont::StorageAny<T, S>>();
-      diy::save(bb, any->GetHandle());
+      vtkmdiy::save(bb, any->GetHandle());
     }
     else if (obj.template IsType<ConstantType>())
     {
-      diy::save(bb, vtkm::cont::TypeString<ConstantType>::Get());
+      vtkmdiy::save(bb, vtkm::cont::TypeString<ConstantType>::Get());
 
       using S = typename ConstantType::StorageTag;
       const vtkm::cont::StorageVirtual* storage = obj.GetStorage();
       auto* any = storage->Cast<vtkm::cont::StorageAny<T, S>>();
-      diy::save(bb, any->GetHandle());
+      vtkmdiy::save(bb, any->GetHandle());
     }
     else
     {
-      diy::save(bb, vtkm::cont::TypeString<BasicType>::Get());
+      vtkmdiy::save(bb, vtkm::cont::TypeString<BasicType>::Get());
       vtkm::cont::internal::ArrayHandleDefaultSerialization(bb, obj);
     }
   }
@@ -110,24 +112,24 @@ struct IntAnySerializer
   static VTKM_CONT void load(BinaryBuffer& bb, vtkm::cont::ArrayHandleVirtual<T>& obj)
   {
     std::string typeString;
-    diy::load(bb, typeString);
+    vtkmdiy::load(bb, typeString);
 
     if (typeString == vtkm::cont::TypeString<CountingType>::Get())
     {
       CountingType array;
-      diy::load(bb, array);
+      vtkmdiy::load(bb, array);
       obj = std::move(vtkm::cont::ArrayHandleVirtual<T>{ array });
     }
     else if (typeString == vtkm::cont::TypeString<ConstantType>::Get())
     {
       ConstantType array;
-      diy::load(bb, array);
+      vtkmdiy::load(bb, array);
       obj = std::move(vtkm::cont::ArrayHandleVirtual<T>{ array });
     }
     else
     {
       vtkm::cont::ArrayHandle<T> array;
-      diy::load(bb, array);
+      vtkmdiy::load(bb, array);
       obj = std::move(vtkm::cont::ArrayHandleVirtual<T>{ array });
     }
   }

vtkm/cont/ArrayHandleVirtualCoordinates.h

@@ -106,7 +106,7 @@ struct TypeString<vtkm::cont::ArrayHandleVirtualCoordinates>
 
 //=============================================================================
 // Specializations of serialization related classes
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <>
@@ -132,8 +132,8 @@ public:
       using T = typename HandleType::ValueType;
       using S = typename HandleType::StorageTag;
       auto array = storage->Cast<vtkm::cont::StorageAny<T, S>>();
-      diy::save(bb, vtkm::cont::TypeString<HandleType>::Get());
+      vtkmdiy::save(bb, vtkm::cont::TypeString<HandleType>::Get());
-      diy::save(bb, array->GetHandle());
+      vtkmdiy::save(bb, array->GetHandle());
     }
     else if (obj.IsType<RectilinearCoordsArrayType>())
     {
@@ -141,12 +141,12 @@ public:
       using T = typename HandleType::ValueType;
       using S = typename HandleType::StorageTag;
       auto array = storage->Cast<vtkm::cont::StorageAny<T, S>>();
-      diy::save(bb, vtkm::cont::TypeString<HandleType>::Get());
+      vtkmdiy::save(bb, vtkm::cont::TypeString<HandleType>::Get());
-      diy::save(bb, array->GetHandle());
+      vtkmdiy::save(bb, array->GetHandle());
     }
     else
     {
-      diy::save(bb, vtkm::cont::TypeString<BasicCoordsType>::Get());
+      vtkmdiy::save(bb, vtkm::cont::TypeString<BasicCoordsType>::Get());
       vtkm::cont::internal::ArrayHandleDefaultSerialization(bb, obj);
     }
   }
@@ -154,24 +154,24 @@ public:
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
   {
     std::string typeString;
-    diy::load(bb, typeString);
+    vtkmdiy::load(bb, typeString);
 
     if (typeString == vtkm::cont::TypeString<vtkm::cont::ArrayHandleUniformPointCoordinates>::Get())
     {
       vtkm::cont::ArrayHandleUniformPointCoordinates array;
-      diy::load(bb, array);
+      vtkmdiy::load(bb, array);
       obj = vtkm::cont::ArrayHandleVirtualCoordinates(array);
     }
     else if (typeString == vtkm::cont::TypeString<RectilinearCoordsArrayType>::Get())
     {
       RectilinearCoordsArrayType array;
-      diy::load(bb, array);
+      vtkmdiy::load(bb, array);
       obj = vtkm::cont::ArrayHandleVirtualCoordinates(array);
     }
     else if (typeString == vtkm::cont::TypeString<BasicCoordsType>::Get())
     {
       BasicCoordsType array;
-      diy::load(bb, array);
+      vtkmdiy::load(bb, array);
       obj = vtkm::cont::ArrayHandleVirtualCoordinates(array);
     }
     else

vtkm/cont/ArrayHandleZip.h

@@ -412,7 +412,7 @@ struct TypeString<
 }
 } // namespace vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename AH1, typename AH2>
@@ -426,8 +426,8 @@ public:
   static VTKM_CONT void save(BinaryBuffer& bb, const BaseType& obj)
   {
     auto storage = obj.GetStorage();
-    diy::save(bb, storage.GetFirstArray());
+    vtkmdiy::save(bb, storage.GetFirstArray());
-    diy::save(bb, storage.GetSecondArray());
+    vtkmdiy::save(bb, storage.GetSecondArray());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, BaseType& obj)
@@ -435,8 +435,8 @@ public:
     AH1 a1;
     AH2 a2;
 
-    diy::load(bb, a1);
+    vtkmdiy::load(bb, a1);
-    diy::load(bb, a2);
+    vtkmdiy::load(bb, a2);
 
     obj = vtkm::cont::make_ArrayHandleZip(a1, a2);
   }

vtkm/cont/AssignerMultiBlock.cxx

@@ -22,11 +22,8 @@
 #include <vtkm/cont/EnvironmentTracker.h>
 #include <vtkm/cont/MultiBlock.h>
 
-// clang-format off
+
-VTKM_THIRDPARTY_PRE_INCLUDE
+#include <vtkm/thirdparty/diy/diy.h>
-#include VTKM_DIY(diy/mpi.hpp)
-VTKM_THIRDPARTY_POST_INCLUDE
-// clang-format on
 
 #include <algorithm> // std::lower_bound
 #include <numeric>   // std::iota
@@ -44,15 +41,15 @@ AssignerMultiBlock::AssignerMultiBlock(const vtkm::cont::MultiBlock& mb)
 
 VTKM_CONT
 AssignerMultiBlock::AssignerMultiBlock(vtkm::Id num_blocks)
-  : diy::StaticAssigner(vtkm::cont::EnvironmentTracker::GetCommunicator().size(), 1)
+  : vtkmdiy::StaticAssigner(vtkm::cont::EnvironmentTracker::GetCommunicator().size(), 1)
   , IScanBlockCounts()
 {
   auto comm = vtkm::cont::EnvironmentTracker::GetCommunicator();
   if (comm.size() > 1)
   {
     vtkm::Id iscan;
-    diy::mpi::scan(comm, num_blocks, iscan, std::plus<vtkm::Id>());
+    vtkmdiy::mpi::scan(comm, num_blocks, iscan, std::plus<vtkm::Id>());
-    diy::mpi::all_gather(comm, iscan, this->IScanBlockCounts);
+    vtkmdiy::mpi::all_gather(comm, iscan, this->IScanBlockCounts);
   }
   else
   {

vtkm/cont/AssignerMultiBlock.h

@@ -28,11 +28,7 @@
 
 #include <vector>
 
-// clang-format off
+#include <vtkm/thirdparty/diy/diy.h>
-VTKM_THIRDPARTY_PRE_INCLUDE
-#include VTKM_DIY(diy/assigner.hpp)
-VTKM_THIRDPARTY_POST_INCLUDE
-// clang-format on
 
 #ifdef VTKM_MSVC
 #pragma warning(push)
@@ -49,7 +45,7 @@ class MultiBlock;
 
 /// \brief Assigner for `MultiBlock` blocks.
 ///
-/// `AssignerMultiBlock` is a `diy::StaticAssigner` implementation that uses
+/// `AssignerMultiBlock` is a `vtkmdiy::StaticAssigner` implementation that uses
 /// `MultiBlock`'s block distribution to build global-id/rank associations
 /// needed for several `diy` operations.
 /// It uses a contiguous assignment strategy to map blocks to global ids i.e.
@@ -59,7 +55,7 @@ class MultiBlock;
 /// essential it gets created on all ranks irrespective of whether the rank has
 /// any blocks.
 ///
-class VTKM_CONT_EXPORT AssignerMultiBlock : public diy::StaticAssigner
+class VTKM_CONT_EXPORT AssignerMultiBlock : public vtkmdiy::StaticAssigner
 {
 public:
   /// Initialize the assigner using a multiblock dataset.
@@ -72,7 +68,7 @@ public:
   AssignerMultiBlock(vtkm::Id num_blocks);
 
   ///@{
-  /// diy::Assigner API implementation.
+  /// vtkmdiy::Assigner API implementation.
   VTKM_CONT
   void local_gids(int rank, std::vector<int>& gids) const override;
 

vtkm/cont/CMakeLists.txt

@@ -125,6 +125,7 @@ set(template_sources
   StorageAny.hxx
   StorageBasic.hxx
   StorageVirtual.hxx
+  VirtualObjectHandle.hxx
   )
 
 set(sources
@@ -170,6 +171,7 @@ set(device_sources
   ArrayRangeCompute.cxx
   CellSetExplicit.cxx
   CoordinateSystem.cxx
+  Timer.cxx
   )
 
 #-----------------------------------------------------------------------------
@@ -199,7 +201,9 @@ if(TARGET vtkm::openmp)
   list(APPEND backends vtkm::openmp)
 endif()
 if (VTKm_ENABLE_LOGGING)
-  list(APPEND DL_LIBS ${CMAKE_DL_LIBS}) # dladdr function
+  set(THREADS_PREFER_PTHREAD_FLAG ON)
+  find_package(Threads REQUIRED)
+  list(APPEND DL_LIBS ${CMAKE_DL_LIBS} Threads::Threads) # dladdr function
 endif()
 target_link_libraries(vtkm_cont PUBLIC vtkm_compiler_flags ${backends} ${DL_LIBS})
 if(TARGET vtkm_diy)

vtkm/cont/CellSetExplicit.h

@@ -366,7 +366,7 @@ struct TypeString<vtkm::cont::CellSetExplicit<ShapeST, CountST, ConnectivityST,
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename ShapeST, typename CountST, typename ConnectivityST, typename OffsetST>
@@ -378,32 +378,32 @@ private:
 public:
   static VTKM_CONT void save(BinaryBuffer& bb, const Type& cs)
   {
-    diy::save(bb, cs.GetName());
+    vtkmdiy::save(bb, cs.GetName());
-    diy::save(bb, cs.GetNumberOfPoints());
+    vtkmdiy::save(bb, cs.GetNumberOfPoints());
-    diy::save(bb,
+    vtkmdiy::save(
-              cs.GetShapesArray(vtkm::TopologyElementTagPoint{}, vtkm::TopologyElementTagCell{}));
+      bb, cs.GetShapesArray(vtkm::TopologyElementTagPoint{}, vtkm::TopologyElementTagCell{}));
-    diy::save(
+    vtkmdiy::save(
       bb, cs.GetNumIndicesArray(vtkm::TopologyElementTagPoint{}, vtkm::TopologyElementTagCell{}));
-    diy::save(
+    vtkmdiy::save(
       bb, cs.GetConnectivityArray(vtkm::TopologyElementTagPoint{}, vtkm::TopologyElementTagCell{}));
-    diy::save(
+    vtkmdiy::save(
       bb, cs.GetIndexOffsetArray(vtkm::TopologyElementTagPoint{}, vtkm::TopologyElementTagCell{}));
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, Type& cs)
   {
     std::string name;
-    diy::load(bb, name);
+    vtkmdiy::load(bb, name);
     vtkm::Id numberOfPoints = 0;
-    diy::load(bb, numberOfPoints);
+    vtkmdiy::load(bb, numberOfPoints);
     vtkm::cont::ArrayHandle<vtkm::UInt8, ShapeST> shapes;
-    diy::load(bb, shapes);
+    vtkmdiy::load(bb, shapes);
     vtkm::cont::ArrayHandle<vtkm::IdComponent, CountST> counts;
-    diy::load(bb, counts);
+    vtkmdiy::load(bb, counts);
     vtkm::cont::ArrayHandle<vtkm::Id, ConnectivityST> connectivity;
-    diy::load(bb, connectivity);
+    vtkmdiy::load(bb, connectivity);
     vtkm::cont::ArrayHandle<vtkm::Id, OffsetST> offsets;
-    diy::load(bb, offsets);
+    vtkmdiy::load(bb, offsets);
 
     cs = Type(name);
     cs.Fill(numberOfPoints, shapes, counts, connectivity, offsets);

vtkm/cont/CellSetPermutation.h

@@ -414,7 +414,7 @@ struct TypeString<vtkm::cont::CellSetPermutation<CSType, AHValidCellIds>>
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename CSType, typename AHValidCellIds>
@@ -426,19 +426,19 @@ private:
 public:
   static VTKM_CONT void save(BinaryBuffer& bb, const Type& cs)
   {
-    diy::save(bb, cs.GetName());
+    vtkmdiy::save(bb, cs.GetName());
-    diy::save(bb, cs.GetFullCellSet());
+    vtkmdiy::save(bb, cs.GetFullCellSet());
-    diy::save(bb, cs.GetValidCellIds());
+    vtkmdiy::save(bb, cs.GetValidCellIds());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, Type& cs)
   {
     std::string name;
-    diy::load(bb, name);
+    vtkmdiy::load(bb, name);
     CSType fullCS;
-    diy::load(bb, fullCS);
+    vtkmdiy::load(bb, fullCS);
     AHValidCellIds validCellIds;
-    diy::load(bb, validCellIds);
+    vtkmdiy::load(bb, validCellIds);
 
     cs = make_CellSetPermutation(validCellIds, fullCS, name);
   }

vtkm/cont/CellSetSingleType.h

@@ -282,7 +282,7 @@ struct TypeString<vtkm::cont::CellSetSingleType<ConnectivityST>>
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename ConnectivityST>
@@ -294,26 +294,26 @@ private:
 public:
   static VTKM_CONT void save(BinaryBuffer& bb, const Type& cs)
   {
-    diy::save(bb, cs.GetName());
+    vtkmdiy::save(bb, cs.GetName());
-    diy::save(bb, cs.GetNumberOfPoints());
+    vtkmdiy::save(bb, cs.GetNumberOfPoints());
-    diy::save(bb, cs.GetCellShape(0));
+    vtkmdiy::save(bb, cs.GetCellShape(0));
-    diy::save(bb, cs.GetNumberOfPointsInCell(0));
+    vtkmdiy::save(bb, cs.GetNumberOfPointsInCell(0));
-    diy::save(
+    vtkmdiy::save(
       bb, cs.GetConnectivityArray(vtkm::TopologyElementTagPoint{}, vtkm::TopologyElementTagCell{}));
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, Type& cs)
   {
     std::string name;
-    diy::load(bb, name);
+    vtkmdiy::load(bb, name);
     vtkm::Id numberOfPoints = 0;
-    diy::load(bb, numberOfPoints);
+    vtkmdiy::load(bb, numberOfPoints);
     vtkm::UInt8 shape;
-    diy::load(bb, shape);
+    vtkmdiy::load(bb, shape);
     vtkm::IdComponent count;
-    diy::load(bb, count);
+    vtkmdiy::load(bb, count);
     vtkm::cont::ArrayHandle<vtkm::Id, ConnectivityST> connectivity;
-    diy::load(bb, connectivity);
+    vtkmdiy::load(bb, connectivity);
 
     cs = Type(name);
     cs.Fill(numberOfPoints, shape, count, connectivity);

vtkm/cont/CellSetStructured.h

@@ -143,7 +143,7 @@ struct TypeString<vtkm::cont::CellSetStructured<DIMENSION>>
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <vtkm::IdComponent DIMENSION>
@@ -155,18 +155,18 @@ private:
 public:
   static VTKM_CONT void save(BinaryBuffer& bb, const Type& cs)
   {
-    diy::save(bb, cs.GetName());
+    vtkmdiy::save(bb, cs.GetName());
-    diy::save(bb, cs.GetPointDimensions());
+    vtkmdiy::save(bb, cs.GetPointDimensions());
-    diy::save(bb, cs.GetGlobalPointIndexStart());
+    vtkmdiy::save(bb, cs.GetGlobalPointIndexStart());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, Type& cs)
   {
     std::string name;
-    diy::load(bb, name);
+    vtkmdiy::load(bb, name);
     typename Type::SchedulingRangeType dims, start;
-    diy::load(bb, dims);
+    vtkmdiy::load(bb, dims);
-    diy::load(bb, start);
+    vtkmdiy::load(bb, start);
 
     cs = Type(name);
     cs.SetPointDimensions(dims);

vtkm/cont/CoordinateSystem.h

@@ -143,7 +143,7 @@ struct DynamicTransformTraits<vtkm::cont::CoordinateSystem>
 
 //=============================================================================
 // Specializations of serialization related classes
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <>
@@ -151,16 +151,16 @@ struct Serialization<vtkm::cont::CoordinateSystem>
 {
   static VTKM_CONT void save(BinaryBuffer& bb, const vtkm::cont::CoordinateSystem& cs)
   {
-    diy::save(bb, cs.GetName());
+    vtkmdiy::save(bb, cs.GetName());
-    diy::save(bb, cs.GetData());
+    vtkmdiy::save(bb, cs.GetData());
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, vtkm::cont::CoordinateSystem& cs)
   {
     std::string name;
-    diy::load(bb, name);
+    vtkmdiy::load(bb, name);
     vtkm::cont::ArrayHandleVirtualCoordinates array;
-    diy::load(bb, array);
+    vtkmdiy::load(bb, array);
     cs = vtkm::cont::CoordinateSystem(name, array);
   }
 };

vtkm/cont/DataSet.h

@@ -208,7 +208,7 @@ struct SerializableDataSet
 }
 } // vtkm::cont
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename FieldTypeList, typename CellSetTypesList>
@@ -223,24 +223,24 @@ public:
     const auto& dataset = serializable.DataSet;
 
     vtkm::IdComponent numberOfCoordinateSystems = dataset.GetNumberOfCoordinateSystems();
-    diy::save(bb, numberOfCoordinateSystems);
+    vtkmdiy::save(bb, numberOfCoordinateSystems);
     for (vtkm::IdComponent i = 0; i < numberOfCoordinateSystems; ++i)
     {
-      diy::save(bb, dataset.GetCoordinateSystem(i));
+      vtkmdiy::save(bb, dataset.GetCoordinateSystem(i));
     }
 
     vtkm::IdComponent numberOfCellSets = dataset.GetNumberOfCellSets();
-    diy::save(bb, numberOfCellSets);
+    vtkmdiy::save(bb, numberOfCellSets);
     for (vtkm::IdComponent i = 0; i < numberOfCellSets; ++i)
     {
-      diy::save(bb, dataset.GetCellSet(i).ResetCellSetList(CellSetTypesList{}));
+      vtkmdiy::save(bb, dataset.GetCellSet(i).ResetCellSetList(CellSetTypesList{}));
     }
 
     vtkm::IdComponent numberOfFields = dataset.GetNumberOfFields();
-    diy::save(bb, numberOfFields);
+    vtkmdiy::save(bb, numberOfFields);
     for (vtkm::IdComponent i = 0; i < numberOfFields; ++i)
     {
-      diy::save(bb, vtkm::cont::SerializableField<FieldTypeList>(dataset.GetField(i)));
+      vtkmdiy::save(bb, vtkm::cont::SerializableField<FieldTypeList>(dataset.GetField(i)));
     }
   }
 
@@ -250,29 +250,29 @@ public:
     dataset = {}; // clear
 
     vtkm::IdComponent numberOfCoordinateSystems = 0;
-    diy::load(bb, numberOfCoordinateSystems);
+    vtkmdiy::load(bb, numberOfCoordinateSystems);
     for (vtkm::IdComponent i = 0; i < numberOfCoordinateSystems; ++i)
     {
       vtkm::cont::CoordinateSystem coords;
-      diy::load(bb, coords);
+      vtkmdiy::load(bb, coords);
       dataset.AddCoordinateSystem(coords);
     }
 
     vtkm::IdComponent numberOfCellSets = 0;
-    diy::load(bb, numberOfCellSets);
+    vtkmdiy::load(bb, numberOfCellSets);
     for (vtkm::IdComponent i = 0; i < numberOfCellSets; ++i)
     {
       vtkm::cont::DynamicCellSetBase<CellSetTypesList> cells;
-      diy::load(bb, cells);
+      vtkmdiy::load(bb, cells);
       dataset.AddCellSet(vtkm::cont::DynamicCellSet(cells));
     }
 
     vtkm::IdComponent numberOfFields = 0;
-    diy::load(bb, numberOfFields);
+    vtkmdiy::load(bb, numberOfFields);
     for (vtkm::IdComponent i = 0; i < numberOfFields; ++i)
     {
       vtkm::cont::SerializableField<FieldTypeList> field;
-      diy::load(bb, field);
+      vtkmdiy::load(bb, field);
       dataset.AddField(field.Field);
     }
   }

vtkm/cont/DeviceAdapterAlgorithm.h

@@ -22,6 +22,7 @@
 
 #include <vtkm/Types.h>
 
+#include <vtkm/cont/Logging.h>
 #include <vtkm/cont/internal/ArrayManagerExecution.h>
 #include <vtkm/cont/internal/DeviceAdapterTag.h>
 
@@ -539,6 +540,11 @@ template <class DeviceAdapterTag>
 class DeviceAdapterTimerImplementation
 {
 public:
+  struct TimeStamp
+  {
+    vtkm::Int64 Seconds;
+    vtkm::Int64 Microseconds;
+  };
   /// When a timer is constructed, all threads are synchronized and the
   /// current time is marked so that GetElapsedTime returns the number of
   /// seconds elapsed since the construction.
@@ -548,7 +554,30 @@ public:
   /// number of seconds elapsed since the call to this. This method
   /// synchronizes all asynchronous operations.
   ///
-  VTKM_CONT void Reset() { this->StartTime = this->GetCurrentTime(); }
+  VTKM_CONT void Reset()
+  {
+    this->StartReady = false;
+    this->StopReady = false;
+  }
+
+  VTKM_CONT void Start()
+  {
+    this->Reset();
+    this->StartTime = this->GetCurrentTime();
+    this->StartReady = true;
+  }
+
+  VTKM_CONT void Stop()
+  {
+    this->StopTime = this->GetCurrentTime();
+    this->StopReady = true;
+  }
+
+  VTKM_CONT bool Started() { return this->StartReady; }
+
+  VTKM_CONT bool Stopped() { return this->StopReady; }
+
+  VTKM_CONT bool Ready() { return true; }
 
   /// Returns the elapsed time in seconds between the construction of this
   /// class or the last call to Reset and the time this function is called. The
@@ -558,21 +587,29 @@ public:
   ///
   VTKM_CONT vtkm::Float64 GetElapsedTime()
   {
-    TimeStamp currentTime = this->GetCurrentTime();
+    assert(this->StartReady);
+    if (!this->StartReady)
+    {
+      VTKM_LOG_S(vtkm::cont::LogLevel::Error,
+                 "Start() function should be called first then trying to call GetElapsedTime().");
+      return 0;
+    }
 
+    bool manualStop = true;
+    if (!this->StopReady)
+    {
+      manualStop = false;
+      this->Stop();
+    }
     vtkm::Float64 elapsedTime;
-    elapsedTime = vtkm::Float64(currentTime.Seconds - this->StartTime.Seconds);
+    elapsedTime = vtkm::Float64(this->StopTime.Seconds - this->StartTime.Seconds);
-    elapsedTime += (vtkm::Float64(currentTime.Microseconds - this->StartTime.Microseconds) /
+    elapsedTime += (vtkm::Float64(this->StopTime.Microseconds - this->StartTime.Microseconds) /
                     vtkm::Float64(1000000));
+    // Reset StopReady flag to its original state
+    this->StopReady = manualStop;
 
     return elapsedTime;
   }
-  struct TimeStamp
-  {
-    vtkm::Int64 Seconds;
-    vtkm::Int64 Microseconds;
-  };
-  TimeStamp StartTime;
 
   VTKM_CONT TimeStamp GetCurrentTime()
   {
@@ -592,6 +629,11 @@ public:
 #endif
     return retval;
   }
+
+  bool StartReady;
+  bool StopReady;
+  TimeStamp StartTime;
+  TimeStamp StopTime;
 };
 
 /// \brief Class providing a device-specific runtime support detector.

vtkm/cont/DynamicCellSet.h

@@ -364,7 +364,7 @@ struct DynamicCellSetCheck<vtkm::cont::DynamicCellSetBase<CellSetList>>
 
 //=============================================================================
 // Specializations of serialization related classes
-namespace diy
+namespace mangled_diy_namespace
 {
 
 namespace internal
@@ -375,8 +375,8 @@ struct DynamicCellSetSerializeFunctor
   template <typename CellSetType>
   void operator()(const CellSetType& cs, BinaryBuffer& bb) const
   {
-    diy::save(bb, vtkm::cont::TypeString<CellSetType>::Get());
+    vtkmdiy::save(bb, vtkm::cont::TypeString<CellSetType>::Get());
-    diy::save(bb, cs);
+    vtkmdiy::save(bb, cs);
   }
 };
 
@@ -393,7 +393,7 @@ struct DynamicCellSetDeserializeFunctor
     if (!success && (typeString == vtkm::cont::TypeString<CellSetType>::Get()))
     {
       CellSetType cs;
-      diy::load(bb, cs);
+      vtkmdiy::load(bb, cs);
       dh = vtkm::cont::DynamicCellSetBase<CellSetTypes>(cs);
       success = true;
     }
@@ -417,7 +417,7 @@ public:
   static VTKM_CONT void load(BinaryBuffer& bb, Type& obj)
   {
     std::string typeString;
-    diy::load(bb, typeString);
+    vtkmdiy::load(bb, typeString);
 
     bool success = false;
     vtkm::ListForEach(internal::DynamicCellSetDeserializeFunctor<CellSetTypes>{},

vtkm/cont/EnvironmentTracker.cxx

@@ -19,11 +19,7 @@
 //============================================================================
 #include <vtkm/cont/EnvironmentTracker.h>
 
-// clang-format off
+#include <vtkm/thirdparty/diy/diy.h>
-VTKM_THIRDPARTY_PRE_INCLUDE
-#include VTKM_DIY(diy/mpi.hpp)
-VTKM_THIRDPARTY_POST_INCLUDE
-// clang-format on
 
 namespace vtkm
 {
@@ -31,17 +27,17 @@ namespace cont
 {
 namespace internal
 {
-static diy::mpi::communicator GlobalCommuncator(MPI_COMM_NULL);
+static vtkmdiy::mpi::communicator GlobalCommuncator(MPI_COMM_NULL);
 }
 
-void EnvironmentTracker::SetCommunicator(const diy::mpi::communicator& comm)
+void EnvironmentTracker::SetCommunicator(const vtkmdiy::mpi::communicator& comm)
 {
   vtkm::cont::internal::GlobalCommuncator = comm;
 }
 
-const diy::mpi::communicator& EnvironmentTracker::GetCommunicator()
+const vtkmdiy::mpi::communicator& EnvironmentTracker::GetCommunicator()
 {
-#ifndef DIY_NO_MPI
+#ifndef VTKM_DIY_NO_MPI
   int flag;
   MPI_Initialized(&flag);
   if (!flag)
@@ -49,7 +45,7 @@ const diy::mpi::communicator& EnvironmentTracker::GetCommunicator()
     int argc = 0;
     char** argv = nullptr;
     MPI_Init(&argc, &argv);
-    internal::GlobalCommuncator = diy::mpi::communicator(MPI_COMM_WORLD);
+    internal::GlobalCommuncator = vtkmdiy::mpi::communicator(MPI_COMM_WORLD);
   }
 #endif
   return vtkm::cont::internal::GlobalCommuncator;

vtkm/cont/EnvironmentTracker.h

@@ -23,15 +23,8 @@
 #include <vtkm/Types.h>
 #include <vtkm/cont/vtkm_cont_export.h>
 #include <vtkm/internal/ExportMacros.h>
-#include <vtkm/thirdparty/diy/Configure.h>
 
-namespace diy
+#include <vtkm/thirdparty/diy/diy.h>
-{
-namespace mpi
-{
-class communicator;
-}
-}
 
 namespace vtkm
 {
@@ -46,10 +39,10 @@ class VTKM_CONT_EXPORT EnvironmentTracker
 {
 public:
   VTKM_CONT
-  static void SetCommunicator(const diy::mpi::communicator& comm);
+  static void SetCommunicator(const vtkmdiy::mpi::communicator& comm);
 
   VTKM_CONT
-  static const diy::mpi::communicator& GetCommunicator();
+  static const vtkmdiy::mpi::communicator& GetCommunicator();
 };
 }
 }

vtkm/cont/ErrorBadType.h

@@ -36,7 +36,7 @@ class VTKM_ALWAYS_EXPORT ErrorBadType : public Error
 {
 public:
   ErrorBadType(const std::string& message)
-    : Error(message)
+    : Error(message, true)
   {
   }
 };

vtkm/cont/Field.h

@@ -321,7 +321,7 @@ struct SerializableField
 } // namespace cont
 } // namespace vtkm
 
-namespace diy
+namespace mangled_diy_namespace
 {
 
 template <typename TypeList>
@@ -335,17 +335,17 @@ public:
   {
     const auto& field = serializable.Field;
 
-    diy::save(bb, field.GetName());
+    vtkmdiy::save(bb, field.GetName());
-    diy::save(bb, static_cast<int>(field.GetAssociation()));
+    vtkmdiy::save(bb, static_cast<int>(field.GetAssociation()));
     if (field.GetAssociation() == vtkm::cont::Field::Association::CELL_SET)
     {
-      diy::save(bb, field.GetAssocCellSet());
+      vtkmdiy::save(bb, field.GetAssocCellSet());
     }
     else if (field.GetAssociation() == vtkm::cont::Field::Association::LOGICAL_DIM)
     {
-      diy::save(bb, field.GetAssocLogicalDim());
+      vtkmdiy::save(bb, field.GetAssocLogicalDim());
     }
-    diy::save(bb, field.GetData().ResetTypes(TypeList{}));
+    vtkmdiy::save(bb, field.GetData().ResetTypes(TypeList{}));
   }
 
   static VTKM_CONT void load(BinaryBuffer& bb, Type& serializable)
@@ -353,30 +353,30 @@ public:
     auto& field = serializable.Field;
 
     std::string name;
-    diy::load(bb, name);
+    vtkmdiy::load(bb, name);
     int assocVal = 0;
-    diy::load(bb, assocVal);
+    vtkmdiy::load(bb, assocVal);
 
     auto assoc = static_cast<vtkm::cont::Field::Association>(assocVal);
     vtkm::cont::VariantArrayHandleBase<TypeList> data;
     if (assoc == vtkm::cont::Field::Association::CELL_SET)
     {
       std::string assocCellSetName;
-      diy::load(bb, assocCellSetName);
+      vtkmdiy::load(bb, assocCellSetName);
-      diy::load(bb, data);
+      vtkmdiy::load(bb, data);
       field =
         vtkm::cont::Field(name, assoc, assocCellSetName, vtkm::cont::VariantArrayHandle(data));
     }
     else if (assoc == vtkm::cont::Field::Association::LOGICAL_DIM)
     {
       vtkm::IdComponent assocLogicalDim;
-      diy::load(bb, assocLogicalDim);
+      vtkmdiy::load(bb, assocLogicalDim);
-      diy::load(bb, data);
+      vtkmdiy::load(bb, data);
       field = vtkm::cont::Field(name, assoc, assocLogicalDim, vtkm::cont::VariantArrayHandle(data));
     }
     else
     {
-      diy::load(bb, data);
+      vtkmdiy::load(bb, data);
       field = vtkm::cont::Field(name, assoc, vtkm::cont::VariantArrayHandle(data));
     }
   }

vtkm/cont/FieldRangeGlobalCompute.cxx

@@ -21,15 +21,7 @@
 
 #include <vtkm/cont/EnvironmentTracker.h>
 
-// clang-format off
+#include <vtkm/thirdparty/diy/diy.h>
-VTKM_THIRDPARTY_PRE_INCLUDE
-#include <vtkm/thirdparty/diy/Configure.h>
-#include VTKM_DIY(diy/decomposition.hpp)
-#include VTKM_DIY(diy/master.hpp)
-#include VTKM_DIY(diy/partners/all-reduce.hpp)
-#include VTKM_DIY(diy/reduce.hpp)
-VTKM_THIRDPARTY_POST_INCLUDE
-// clang-format on
 
 #include <algorithm>
 #include <functional>
@@ -78,56 +70,57 @@ vtkm::cont::ArrayHandle<vtkm::Range> MergeRangesGlobal(
 
   using VectorOfRangesT = std::vector<vtkm::Range>;
 
-  diy::Master master(comm,
+  vtkmdiy::Master master(comm,
-                     1,
+                         1,
-                     -1,
+                         -1,
-                     []() -> void* { return new VectorOfRangesT(); },
+                         []() -> void* { return new VectorOfRangesT(); },
-                     [](void* ptr) { delete static_cast<VectorOfRangesT*>(ptr); });
+                         [](void* ptr) { delete static_cast<VectorOfRangesT*>(ptr); });
 
-  diy::ContiguousAssigner assigner(/*num ranks*/ comm.size(), /*global-num-blocks*/ comm.size());
+  vtkmdiy::ContiguousAssigner assigner(/*num ranks*/ comm.size(),
-  diy::RegularDecomposer<diy::DiscreteBounds> decomposer(
+                                       /*global-num-blocks*/ comm.size());
-    /*dim*/ 1, diy::interval(0, comm.size() - 1), comm.size());
+  vtkmdiy::RegularDecomposer<vtkmdiy::DiscreteBounds> decomposer(
+    /*dim*/ 1, vtkmdiy::interval(0, comm.size() - 1), comm.size());
   decomposer.decompose(comm.rank(), assigner, master);
   assert(master.size() == 1); // each rank will have exactly 1 block.
   *master.block<VectorOfRangesT>(0) = v_ranges;
 
-  diy::RegularAllReducePartners all_reduce_partners(decomposer, /*k*/ 2);
+  vtkmdiy::RegularAllReducePartners all_reduce_partners(decomposer, /*k*/ 2);
 
-  auto callback =
+  auto callback = [](
-    [](VectorOfRangesT* data, const diy::ReduceProxy& srp, const diy::RegularMergePartners&) {
+    VectorOfRangesT* data, const vtkmdiy::ReduceProxy& srp, const vtkmdiy::RegularMergePartners&) {
-      const auto selfid = srp.gid();
+    const auto selfid = srp.gid();
-      // 1. dequeue.
+    // 1. dequeue.
-      std::vector<int> incoming;
+    std::vector<int> incoming;
-      srp.incoming(incoming);
+    srp.incoming(incoming);
-      for (const int gid : incoming)
+    for (const int gid : incoming)
+    {
+      if (gid != selfid)
       {
-        if (gid != selfid)
+        VectorOfRangesT message;
-        {
+        srp.dequeue(gid, message);
-          VectorOfRangesT message;
-          srp.dequeue(gid, message);
 
-          // if the number of components we've seen so far is less than those
+        // if the number of components we've seen so far is less than those
-          // in the received message, resize so we can accommodate all components
+        // in the received message, resize so we can accommodate all components
-          // in the message. If the message has fewer components, it has no
+        // in the message. If the message has fewer components, it has no
-          // effect.
+        // effect.
-          data->resize(std::max(data->size(), message.size()));
+        data->resize(std::max(data->size(), message.size()));
 
-          std::transform(
+        std::transform(
-            message.begin(), message.end(), data->begin(), data->begin(), std::plus<vtkm::Range>());
+          message.begin(), message.end(), data->begin(), data->begin(), std::plus<vtkm::Range>());
-        }
       }
-      // 2. enqueue
+    }
-      for (int cc = 0; cc < srp.out_link().size(); ++cc)
+    // 2. enqueue
+    for (int cc = 0; cc < srp.out_link().size(); ++cc)
+    {
+      auto target = srp.out_link().target(cc);
+      if (target.gid != selfid)
       {
-        auto target = srp.out_link().target(cc);
+        srp.enqueue(target, *data);
-        if (target.gid != selfid)
-        {
-          srp.enqueue(target, *data);
-        }
       }
-    };
+    }
+  };
 
-  diy::reduce(master, assigner, all_reduce_partners, callback);
+  vtkmdiy::reduce(master, assigner, all_reduce_partners, callback);
   assert(master.size() == 1); // each rank will have exactly 1 block.
 
   return vtkm::cont::make_ArrayHandle(*master.block<VectorOfRangesT>(0), vtkm::CopyFlag::On);

vtkm/cont/Serialization.h

@@ -22,10 +22,6 @@
 
 #include <vtkm/cont/TypeString.h>
 
-// clang-format off
+#include <vtkm/thirdparty/diy/serialization.h>
-VTKM_THIRDPARTY_PRE_INCLUDE
-#include <vtkm/thirdparty/diy/Configure.h>
-#include VTKM_DIY(diy/serialization.hpp)
-VTKM_THIRDPARTY_POST_INCLUDE
 
 #endif // vtk_m_cont_Serialization_h

vtkm/cont/Timer.cxx

@@ -0,0 +1,254 @@
+//============================================================================
+//  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 2016 National Technology & Engineering Solutions of Sandia, LLC (NTESS).
+//  Copyright 2016 UT-Battelle, LLC.
+//  Copyright 2016 Los Alamos National Security.
+//
+//  Under the terms of Contract DE-NA0003525 with NTESS,
+//  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/Logging.h>
+#include <vtkm/cont/RuntimeDeviceTracker.h>
+#include <vtkm/cont/Timer.h>
+
+#include <vtkm/internal/brigand.hpp>
+
+namespace vtkm
+{
+namespace cont
+{
+
+namespace detail
+{
+template <typename State, typename T>
+struct RemoveDisabledDevice
+{
+  using type = typename std::conditional<T::IsEnabled, brigand::push_back<State, T>, State>::type;
+};
+
+/// TMP code to generate enabled device timer container
+using AllDeviceList = DeviceAdapterListTagCommon::list;
+using EnabledDeviceList =
+  brigand::fold<AllDeviceList,
+                brigand::list<>,
+                detail::RemoveDisabledDevice<brigand::_state, brigand::_element>>;
+struct EnabledDeviceListTag : vtkm::ListTagBase<>
+{
+  using list = EnabledDeviceList;
+};
+using EnabledTimerImpls =
+  brigand::transform<EnabledDeviceList,
+                     brigand::bind<DeviceAdapterTimerImplementation, brigand::_1>>;
+using EnabledTimerImplTuple = brigand::as_tuple<EnabledTimerImpls>;
+}
+
+enum class TimerDispatchTag : int
+{
+  Reset,
+  Start,
+  Stop,
+  Started,
+  Stopped,
+  Ready,
+  GetElapsedTime
+};
+
+class EnabledDeviceTimerImpls
+{
+public:
+  EnabledDeviceTimerImpls() {}
+  ~EnabledDeviceTimerImpls() {}
+  // A tuple of enabled timer implementations
+  detail::EnabledTimerImplTuple timerImplTuple;
+};
+
+namespace detail
+{
+// C++11 does not support get tuple element by type. C++14 does support that.
+// Get the index of a type in tuple elements
+template <class T, class Tuple>
+struct Index;
+
+template <class T, template <typename...> class Container, class... Types>
+struct Index<T, Container<T, Types...>>
+{
+  static const std::size_t value = 0;
+};
+
+template <class T, class U, template <typename...> class Container, class... Types>
+struct Index<T, Container<U, Types...>>
+{
+  static const std::size_t value = 1 + Index<T, Container<Types...>>::value;
+};
+
+struct TimerFunctor
+{
+  TimerFunctor()
+    : elapsedTime(0)
+    , value(true)
+  {
+  }
+
+  template <typename DeviceAdapter>
+  VTKM_CONT void operator()(DeviceAdapter device, Timer* timer, TimerDispatchTag tag)
+  {
+    if (timer->Device == device || timer->Device == DeviceAdapterTagAny())
+    {
+      auto& timerImpl = std::get<Index<DeviceAdapter, detail::EnabledDeviceList>::value>(
+        timer->Internal->timerImplTuple);
+      switch (tag)
+      {
+        case TimerDispatchTag::Reset:
+          timerImpl.Reset();
+          break;
+        case TimerDispatchTag::Start:
+          timerImpl.Start();
+          break;
+        case TimerDispatchTag::Stop:
+          timerImpl.Stop();
+          break;
+        case TimerDispatchTag::Started:
+          value &= timerImpl.Started();
+          break;
+        case TimerDispatchTag::Stopped:
+          value &= timerImpl.Stopped();
+          break;
+        case TimerDispatchTag::Ready:
+          value &= timerImpl.Ready();
+          break;
+        case TimerDispatchTag::GetElapsedTime:
+        {
+          if (timer->Device == DeviceAdapterTagAny() &&
+              timer->DeviceForQuery == DeviceAdapterTagAny())
+          { // Just want to do timing jobs
+            elapsedTime = std::max(elapsedTime, timerImpl.GetElapsedTime());
+            break;
+          }
+          else if (timer->Device == DeviceAdapterTagAny() && timer->DeviceForQuery == device)
+          { // Given a generic timer, querying for a specific device time
+            elapsedTime = timerImpl.GetElapsedTime();
+            break;
+          }
+          else if (timer->Device == device && (timer->DeviceForQuery == DeviceAdapterTagAny() ||
+                                               timer->Device == timer->DeviceForQuery))
+          { // Given a specific timer, querying its elapsed time
+            elapsedTime = timerImpl.GetElapsedTime();
+            break;
+          }
+          break;
+        }
+      }
+    }
+  }
+
+  vtkm::Float64 elapsedTime;
+  bool value;
+};
+}
+
+
+Timer::Timer()
+  : Device(vtkm::cont::DeviceAdapterTagAny())
+  , DeviceForQuery(vtkm::cont::DeviceAdapterTagAny())
+  , Internal(nullptr)
+{
+  this->Init();
+}
+
+
+Timer::~Timer()
+{
+  delete this->Internal;
+}
+
+void Timer::Init()
+{
+  if (!this->Internal)
+  {
+    this->Internal = new EnabledDeviceTimerImpls();
+  }
+}
+
+void Timer::Reset()
+{
+  detail::TimerFunctor functor;
+  vtkm::ListForEach(functor, detail::EnabledDeviceListTag(), this, TimerDispatchTag::Reset);
+}
+
+void Timer::Start()
+{
+
+  detail::TimerFunctor functor;
+  vtkm::ListForEach(functor, detail::EnabledDeviceListTag(), this, TimerDispatchTag::Start);
+}
+
+void Timer::Stop()
+{
+  detail::TimerFunctor functor;
+  vtkm::ListForEach(functor, detail::EnabledDeviceListTag(), this, TimerDispatchTag::Stop);
+}
+
+bool Timer::Started()
+{
+  detail::TimerFunctor functor;
+  vtkm::ListForEach(functor, detail::EnabledDeviceListTag(), this, TimerDispatchTag::Started);
+  return functor.value;
+}
+
+bool Timer::Stopped()
+{
+  detail::TimerFunctor functor;
+  vtkm::ListForEach(functor, detail::EnabledDeviceListTag(), this, TimerDispatchTag::Stopped);
+  return functor.value;
+}
+
+bool Timer::Ready()
+{
+  detail::TimerFunctor functor;
+  vtkm::ListForEach(functor, detail::EnabledDeviceListTag(), this, TimerDispatchTag::Ready);
+  return functor.value;
+}
+
+vtkm::Float64 Timer::GetElapsedTime(DeviceAdapterId id)
+{
+  // Timer is constructed with a specific device. Only querying this device is allowed.
+  if (this->Device != DeviceAdapterTagAny() && (id != DeviceAdapterTagAny() && this->Device != id))
+  {
+    VTKM_LOG_S(vtkm::cont::LogLevel::Error,
+               "Device '" << id.GetName() << "' is not supported for current timer"
+                          << "("
+                          << this->Device.GetName()
+                          << ")");
+    return 0.0;
+  }
+
+  // Timer is constructed with any device. Only querying enabled device is allowed.
+  vtkm::cont::RuntimeDeviceTracker tracker;
+  if (this->Device == DeviceAdapterTagAny() &&
+      (id != DeviceAdapterTagAny() && !tracker.CanRunOn(id)))
+  {
+    VTKM_LOG_S(vtkm::cont::LogLevel::Error,
+               "Device '" << id.GetName() << "' can not run on current Device."
+                                             "Thus timer is not usable");
+    return 0.0;
+  }
+
+  this->DeviceForQuery = id;
+  detail::TimerFunctor functor;
+  vtkm::ListForEach(
+    functor, detail::EnabledDeviceListTag(), this, TimerDispatchTag::GetElapsedTime);
+
+  return functor.elapsedTime;
+}
+}
+} // namespace vtkm::cont

vtkm/cont/Timer.h

@@ -20,56 +20,85 @@
 #ifndef vtk_m_cont_Timer_h
 #define vtk_m_cont_Timer_h
 
+#include <vtkm/ListTag.h>
 #include <vtkm/cont/DeviceAdapter.h>
+#include <vtkm/cont/DeviceAdapterListTag.h>
+
+#include <vtkm/cont/vtkm_cont_export.h>
 
 namespace vtkm
 {
 namespace cont
 {
+namespace detail
+{
+struct TimerFunctor;
+}
+class EnabledDeviceTimerImpls;
 
-/// A class that can be used to time operations in VTK-m that might be occurring
+/// A class that can be used to time operations in VTK-m that might be occuring
-/// in parallel. You should make sure that the device adapter for the timer
+/// in parallel. Users are recommended to provide a device adapter at construction
-/// matches that being used to execute algorithms to ensure that the thread
+/// time which matches the one being used to execute algorithms to ensure that thread
-/// synchronization is correct.
+/// synchronization is correct and accurate.
+/// If no device adapter is provided at construction time, the maximum
+/// elapsed time of all enabled deivces will be returned. Normally cuda is expected to
+/// have the longest execution time if enabled.
+/// Per device adapter time query is also supported. It's useful when users want to reuse
+/// the same timer to measure the cuda kernal call as well as the cuda device execution.
 ///
 /// The there is no guaranteed resolution of the time but should generally be
 /// good to about a millisecond.
 ///
-template <class Device = VTKM_DEFAULT_DEVICE_ADAPTER_TAG>
+class VTKM_CONT_EXPORT Timer
-class Timer
 {
+  friend struct detail::TimerFunctor;
+
 public:
-  /// When a timer is constructed, all threads are synchronized and the
-  /// current time is marked so that GetElapsedTime returns the number of
-  /// seconds elapsed since the construction.
   VTKM_CONT
-  Timer()
+  Timer();
-    : TimerImplementation()
+
+  template <typename DeviceAdapter>
+  VTKM_CONT Timer(DeviceAdapter id)
+    : Device(id)
+    , DeviceForQuery(DeviceAdapterTagAny())
+    , Internal(nullptr)
   {
+    VTKM_IS_DEVICE_ADAPTER_TAG(DeviceAdapter);
+    static_assert(DeviceAdapter::IsEnabled, "A disabled device is passed to the Timer");
+    this->Init();
   }
 
-  /// Resets the timer. All further calls to GetElapsedTime will report the
+  VTKM_CONT ~Timer();
-  /// number of seconds elapsed since the call to this. This method
-  /// synchronizes all asynchronous operations.
-  ///
-  VTKM_CONT
-  void Reset() { this->TimerImplementation.Reset(); }
 
-  /// Returns the elapsed time in seconds between the construction of this
+  /// Resets the timer.
-  /// class or the last call to Reset and the time this function is called. The
+  VTKM_CONT void Reset();
-  /// time returned is measured in wall time. GetElapsedTime may be called any
+
-  /// number of times to get the progressive time. This method synchronizes all
+  /// Start would call Reset function before starting the timer for convenience
-  /// asynchronous operations.
+  VTKM_CONT void Start();
-  ///
+
+  VTKM_CONT void Stop();
+
+  VTKM_CONT bool Started();
+
+  VTKM_CONT bool Stopped();
+
+  /// Used to check if Timer has finished the synchronization to get the result from the device.
+  VTKM_CONT bool Ready();
+
+  /// Get the elapsed time measured by the given device adapter. If no device is
+  /// specified, the max time of all device measurements will be returned.
   VTKM_CONT
-  vtkm::Float64 GetElapsedTime() { return this->TimerImplementation.GetElapsedTime(); }
+  vtkm::Float64 GetElapsedTime(DeviceAdapterId id = DeviceAdapterTagAny());
 
 private:
+  VTKM_CONT void Init();
   /// Some timers are ill-defined when copied, so disallow that for all timers.
-  VTKM_CONT Timer(const Timer<Device>&) = delete;
+  VTKM_CONT Timer(const Timer&) = delete;
-  VTKM_CONT void operator=(const Timer<Device>&) = delete;
+  VTKM_CONT void operator=(const Timer&) = delete;
 
-  vtkm::cont::DeviceAdapterTimerImplementation<Device> TimerImplementation;
+  DeviceAdapterId Device;
+  DeviceAdapterId DeviceForQuery;
+  EnabledDeviceTimerImpls* Internal;
 };
 }
 } // namespace vtkm::cont

vtkm/cont/VariantArrayHandle.h

@@ -434,7 +434,7 @@ struct DynamicTransformTraits<vtkm::cont::VariantArrayHandleBase<TypeList>>
 
 //=============================================================================
 // Specializations of serialization related classes
-namespace diy
+namespace mangled_diy_namespace
 {
 
 namespace internal
@@ -445,8 +445,8 @@ struct VariantArrayHandleSerializeFunctor
   template <typename ArrayHandleType>
   void operator()(const ArrayHandleType& ah, BinaryBuffer& bb) const
   {
-    diy::save(bb, vtkm::cont::TypeString<ArrayHandleType>::Get());
+    vtkmdiy::save(bb, vtkm::cont::TypeString<ArrayHandleType>::Get());
-    diy::save(bb, ah);
+    vtkmdiy::save(bb, ah);
   }
 };
 
@@ -464,7 +464,7 @@ struct VariantArrayHandleDeserializeFunctor
     if (!success && (typeString == vtkm::cont::TypeString<ArrayHandleType>::Get()))
     {
       ArrayHandleType ah;
-      diy::load(bb, ah);
+      vtkmdiy::load(bb, ah);
       dh = vtkm::cont::VariantArrayHandleBase<TypeList>(ah);
       success = true;
     }
@@ -488,7 +488,7 @@ public:
   static VTKM_CONT void load(BinaryBuffer& bb, Type& obj)
   {
     std::string typeString;
-    diy::load(bb, typeString);
+    vtkmdiy::load(bb, typeString);
 
     bool success = false;
     vtkm::ListForEach(

vtkm/cont/cuda/CMakeLists.txt

@@ -26,7 +26,7 @@ set(headers
 
 #-----------------------------------------------------------------------------
 add_subdirectory(internal)
-vtkm_declare_headers(CUDA ${headers})
+vtkm_declare_headers(${headers})
 
 #-----------------------------------------------------------------------------
 if (TARGET vtkm::cuda)

vtkm/cont/cuda/internal/CMakeLists.txt

@@ -33,7 +33,7 @@ set(headers
   VirtualObjectTransferCuda.h
   )
 
-vtkm_declare_headers(CUDA ${headers})
+vtkm_declare_headers(${headers})
 
 
 if (TARGET vtkm::cuda)

vtkm/cont/cuda/internal/DeviceAdapterAtomicArrayImplementationCuda.h

@@ -61,7 +61,7 @@ public:
 // We work around this by calling the __device__ function inside of a
 // __CUDA_ARCH__ guard, as nvcc is smart enough to recognize that this is a
 // safe usage of a __device__ function in a __host__ __device__ context.
-#ifdef __CUDA_ARCH__
+#ifdef VTKM_CUDA_DEVICE_PASS
     T* lockedValue = ::thrust::raw_pointer_cast(this->Portal.GetIteratorBegin() + index);
     return this->vtkmAtomicAdd(lockedValue, value);
 #else
@@ -85,7 +85,7 @@ public:
 // We work around this by calling the __device__ function inside of a
 // __CUDA_ARCH__ guard, as nvcc is smart enough to recognize that this is a
 // safe usage of a __device__ function in a __host__ __device__ context.
-#ifdef __CUDA_ARCH__
+#ifdef VTKM_CUDA_DEVICE_PASS
     T* lockedValue = ::thrust::raw_pointer_cast(this->Portal.GetIteratorBegin() + index);
     return this->vtkmCompareAndSwap(lockedValue, newValue, oldValue);
 #else

vtkm/cont/cuda/internal/DeviceAdapterRuntimeDetectorCuda.cxx

@@ -27,24 +27,6 @@
 #include <vtkm/Math.h>
 #include <vtkm/cont/cuda/ErrorCuda.h>
 
-namespace vtkm
-{
-namespace cont
-{
-namespace cuda
-{
-namespace internal
-{
-
-static __global__ void DetermineIfValidCudaDevice()
-{
-  //used only to see if we can launch kernels. It is possible to have a
-  //CUDA capable device, but still fail to have CUDA support.
-}
-}
-}
-}
-}
 namespace
 {
 static std::once_flag deviceQueryFlag;
@@ -64,28 +46,11 @@ void queryNumberOfDevicesandHighestArchSupported(vtkm::Int32& nod, vtkm::Int32&
     for (vtkm::Int32 i = 0; i < numDevices; i++)
     {
       cudaDeviceProp prop;
-      VTKM_CUDA_CALL(cudaGetDeviceProperties(&prop, i));
+      res = cudaGetDeviceProperties(&prop, i);
-      const vtkm::Int32 arch = (prop.major * 10) + prop.minor;
+      if (res == cudaSuccess)
-      archVersion = vtkm::Max(arch, archVersion);
-    }
-
-    //Make sure we can actually launch a kernel. This could fail for any
-    //of the following reasons:
-    //
-    // 1. cudaErrorInsufficientDriver, caused by out of data drives
-    // 2. cudaErrorDevicesUnavailable, caused by another process locking the
-    //    device or somebody disabling cuda support on the device
-    // 3. cudaErrorNoKernelImageForDevice we built for a compute version
-    //    greater than the device we are running on
-    // Most likely others that I don't even know about
-    if (numDevices > 0)
-    {
-      vtkm::cont::cuda::internal::DetermineIfValidCudaDevice<<<1, 1, 0, cudaStreamPerThread>>>();
-      cudaStreamSynchronize(cudaStreamPerThread);
-      if (cudaSuccess != cudaGetLastError())
       {
-        numDevices = 0;
+        const vtkm::Int32 arch = (prop.major * 10) + prop.minor;
-        archVersion = 0;
+        archVersion = vtkm::Max(arch, archVersion);
       }
     }
   });
@@ -112,7 +77,7 @@ DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagCuda>::DeviceAdapterRun
 
 bool DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagCuda>::Exists() const
 {
-  return this->NumberOfDevices > 0 && this->HighestArchSupported >= 20;
+  return this->NumberOfDevices > 0 && this->HighestArchSupported >= 30;
 }
 }
 } // namespace vtkm::cont

vtkm/cont/cuda/internal/DeviceAdapterTimerImplementationCuda.cu

@@ -33,7 +33,7 @@ DeviceAdapterTimerImplementation<
   vtkm::cont::DeviceAdapterTagCuda>::DeviceAdapterTimerImplementation()
 {
   VTKM_CUDA_CALL(cudaEventCreate(&this->StartEvent));
-  VTKM_CUDA_CALL(cudaEventCreate(&this->EndEvent));
+  VTKM_CUDA_CALL(cudaEventCreate(&this->StopEvent));
   this->Reset();
 }
 
@@ -45,21 +45,72 @@ DeviceAdapterTimerImplementation<
   // VTKM_CUDA_CHECK_ASYNCHRONOUS_ERROR catching any issues from these calls
   // later.
   cudaEventDestroy(this->StartEvent);
-  cudaEventDestroy(this->EndEvent);
+  cudaEventDestroy(this->StopEvent);
 }
 
 void DeviceAdapterTimerImplementation<vtkm::cont::DeviceAdapterTagCuda>::Reset()
+{
+  this->StartReady = false;
+  this->StopReady = false;
+}
+
+void DeviceAdapterTimerImplementation<vtkm::cont::DeviceAdapterTagCuda>::Start()
 {
   VTKM_CUDA_CALL(cudaEventRecord(this->StartEvent, cudaStreamPerThread));
   VTKM_CUDA_CALL(cudaEventSynchronize(this->StartEvent));
+  this->StartReady = true;
 }
 
+void DeviceAdapterTimerImplementation<vtkm::cont::DeviceAdapterTagCuda>::Stop()
+{
+  VTKM_CUDA_CALL(cudaEventRecord(this->StopEvent, cudaStreamPerThread));
+  this->StopReady = true;
+}
+
+bool DeviceAdapterTimerImplementation<vtkm::cont::DeviceAdapterTagCuda>::Started()
+{
+  return this->StartReady;
+}
+
+bool DeviceAdapterTimerImplementation<vtkm::cont::DeviceAdapterTagCuda>::Stopped()
+{
+  return this->StopReady;
+}
+
+// Callbacks without a mandated order(in independent streams) execute in undefined
+// order and maybe serialized. So Instead CudaEventQuery is used here.
+// Ref link: https://docs.nvidia.com/cuda/cuda-driver-api/group__CUDA__STREAM.html
+bool DeviceAdapterTimerImplementation<vtkm::cont::DeviceAdapterTagCuda>::Ready()
+{
+  if (cudaEventQuery(this->StopEvent) == cudaSuccess)
+  {
+    return true;
+  }
+  return false;
+}
+
+
 vtkm::Float64 DeviceAdapterTimerImplementation<vtkm::cont::DeviceAdapterTagCuda>::GetElapsedTime()
 {
-  VTKM_CUDA_CALL(cudaEventRecord(this->EndEvent, cudaStreamPerThread));
+  assert(this->StartReady);
-  VTKM_CUDA_CALL(cudaEventSynchronize(this->EndEvent));
+  if (!this->StartReady)
+  {
+    VTKM_LOG_F(vtkm::cont::LogLevel::Error,
+               "Start() function should be called first then trying to call GetElapsedTime().");
+    return 0;
+  }
+  bool manualStop = true;
+  if (!this->StopReady)
+  {
+    manualStop = false;
+    this->Stop();
+  }
+
+  VTKM_CUDA_CALL(cudaEventSynchronize(this->StopEvent));
   float elapsedTimeMilliseconds;
-  VTKM_CUDA_CALL(cudaEventElapsedTime(&elapsedTimeMilliseconds, this->StartEvent, this->EndEvent));
+  VTKM_CUDA_CALL(cudaEventElapsedTime(&elapsedTimeMilliseconds, this->StartEvent, this->StopEvent));
+  // Reset Stop flag to its original state
+  this->StopReady = manualStop;
   return static_cast<vtkm::Float64>(0.001f * elapsedTimeMilliseconds);
 }
 }

vtkm/cont/cuda/internal/DeviceAdapterTimerImplementationCuda.h

@@ -50,6 +50,16 @@ public:
 
   VTKM_CONT void Reset();
 
+  VTKM_CONT void Start();
+
+  VTKM_CONT void Stop();
+
+  VTKM_CONT bool Started();
+
+  VTKM_CONT bool Stopped();
+
+  VTKM_CONT bool Ready();
+
   VTKM_CONT vtkm::Float64 GetElapsedTime();
 
 private:
@@ -59,8 +69,10 @@ private:
   void operator=(const DeviceAdapterTimerImplementation<vtkm::cont::DeviceAdapterTagCuda>&) =
     delete;
 
+  bool StartReady;
+  bool StopReady;
   cudaEvent_t StartEvent;
-  cudaEvent_t EndEvent;
+  cudaEvent_t StopEvent;
 };
 }
 } // namespace vtkm::cont

vtkm/cont/cuda/internal/testing/CMakeLists.txt

@@ -23,4 +23,4 @@ set(headers
    Testing.h
    )
 
-vtkm_declare_headers(CUDA ${headers})
+vtkm_declare_headers(${headers})

vtkm/cont/cuda/testing/UnitTestCudaMath.cu

@@ -48,7 +48,7 @@ struct TriggerICE : public vtkm::worklet::WorkletMapField
   using ControlSignature = void(FieldIn, FieldIn, FieldOut);
   using ExecutionSignature = _3(_1, _2, WorkIndex);
 
-#if __CUDA_ARCH__
+#ifdef VTKM_CUDA_DEVICE_PASS
   template <class ValueType>
   __device__ ValueType operator()(const ValueType& bad,
                                   const ValueType& sane,