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

Re-implement DeviceAdapterRuntimeDetector to avoid ODR violations.

Browse Source 
The previous implementation of DeviceAdapterRuntimeDetector caused
multiple differing definitions of the same class to exist and
was causing the runtime device tracker to report CUDA as disabled
when it actually was enabled.

The ODR was caused by having a default implementation for
DeviceAdapterRuntimeDetector and a specific specialization for
CUDA. If a library had both CUDA and C++ sources it would pick up
both implementations and would have undefined behavior. In general
it would think the CUDA backend was disabled.

To avoid this kind of situation in the future I have reworked VTK-m
so that each device adapter must implement DeviceAdapterRuntimeDetector
for that device.
This commit is contained in:
Robert Maynard 2018-05-14 16:21:30 -04:00
parent 7cf0926172
commit e28244f345
19 changed files with 361 additions and 101 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

19
vtkm/cont/DeviceAdapterAlgorithm.h
View File

@ -607,17 +607,14 @@ template <class DeviceAdapterTag>
class DeviceAdapterRuntimeDetector
{
public:
/// Returns true if the given device adapter is supported on the current
/// machine.
///
/// The default implementation is to return the value of
/// vtkm::cont::DeviceAdapterTraits<DeviceAdapterTag>::Valid
///
VTKM_CONT bool Exists() const
{
using DeviceAdapterTraits = vtkm::cont::DeviceAdapterTraits<DeviceAdapterTag>;
return DeviceAdapterTraits::Valid;
}
/// Returns true if the given device adapter is supported on the current
/// machine.
///
/// No default implementation is provided as it could possible cause
/// ODR violations when headers are included in differing order.
#ifdef VTKM_DOXYGEN_ONLY
VTKM_CONT bool Exists() const;
#endif
};
/// \brief Class providing a device-specific support for atomic operations.

8
vtkm/cont/RuntimeDeviceInformation.h
View File

@ -22,6 +22,12 @@
#include <vtkm/cont/DeviceAdapter.h>
//Bring in each device adapters runtime class
#include <vtkm/cont/cuda/internal/DeviceAdapterRuntimeDetectorCuda.h>
#include <vtkm/cont/internal/DeviceAdapterError.h>
#include <vtkm/cont/serial/internal/DeviceAdapterRuntimeDetectorSerial.h>
#include <vtkm/cont/tbb/internal/DeviceAdapterRuntimeDetectorTBB.h>
namespace vtkm
{
namespace cont
@ -33,7 +39,7 @@ namespace cont
/// or a Accelerator Card.
///
///
template <class Device = VTKM_DEFAULT_DEVICE_ADAPTER_TAG>
template <class Device>
class RuntimeDeviceInformation
{
public:

4
vtkm/cont/RuntimeDeviceTracker.h
View File

@ -20,11 +20,15 @@
#ifndef vtk_m_cont_RuntimeDeviceTracker_h
#define vtk_m_cont_RuntimeDeviceTracker_h
#include <vtkm/cont/vtkm_cont_export.h>
#include <vtkm/cont/DeviceAdapterAlgorithm.h>
#include <vtkm/cont/ErrorBadAllocation.h>
#include <vtkm/cont/ErrorBadDevice.h>
#include <vtkm/cont/RuntimeDeviceInformation.h>
#include <memory>
namespace vtkm
{
namespace cont

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

@ -36,8 +36,13 @@ set(headers
vtkm_declare_headers(CUDA ${headers} TESTABLE ${VTKm_ENABLE_CUDA})
#-----------------------------------------------------------------------------
if (NOT VTKm_ENABLE_CUDA)
#build the file with cpp compiler if cuda is disabled
target_sources(vtkm_cont PRIVATE
${CMAKE_CURRENT_SOURCE_DIR}/DeviceAdapterRuntimeDetectorCuda.cxx
)
return()
endif()

82
vtkm/cont/cuda/internal/DeviceAdapterRuntimeDetectorCuda.cu
View File

@ -17,84 +17,4 @@
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//============================================================================
#include <vtkm/cont/cuda/internal/DeviceAdapterRuntimeDetectorCuda.h>
#include <vtkm/Math.h>
#include <vtkm/cont/cuda/ErrorCuda.h>
#include <cuda.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 vtkm
{
namespace cont
{
DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagCuda>::DeviceAdapterRuntimeDetector()
: NumberOfDevices(0)
, HighestArchSupported(0)
{
static bool deviceQueryInit = false;
static int numDevices = 0;
static int archVersion = 0;
if (!deviceQueryInit)
{
deviceQueryInit = true;
//first query for the number of devices
VTKM_CUDA_CALL(cudaGetDeviceCount(&numDevices));
for (vtkm::Int32 i = 0; i < numDevices; i++)
{
cudaDeviceProp prop;
VTKM_CUDA_CALL(cudaGetDeviceProperties(&prop, i));
const vtkm::Int32 arch = (prop.major * 10) + prop.minor;
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
vtkm::cont::cuda::internal::DetermineIfValidCudaDevice<<<1, 1, 0, cudaStreamPerThread>>>();
if (cudaSuccess != cudaGetLastError())
{
numDevices = 0;
archVersion = 0;
}
}
this->NumberOfDevices = numDevices;
this->HighestArchSupported = archVersion;
}
bool DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagCuda>::Exists() const
{
return this->NumberOfDevices > 0 && this->HighestArchSupported >= 20;
}
}
} // namespace vtkm::cont
#include "DeviceAdapterRuntimeDetectorCuda.cxx"

106
vtkm/cont/cuda/internal/DeviceAdapterRuntimeDetectorCuda.cxx Normal file
View File

@ -0,0 +1,106 @@
//============================================================================
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2014 National Technology & Engineering Solutions of Sandia, LLC (NTESS).
// Copyright 2014 UT-Battelle, LLC.
// Copyright 2014 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/cuda/internal/DeviceAdapterRuntimeDetectorCuda.h>
#ifdef VTKM_CUDA
#include <cuda.h>
#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.
}
}
}
}
}
#endif
namespace vtkm
{
namespace cont
{
DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagCuda>::DeviceAdapterRuntimeDetector()
: NumberOfDevices(0)
, HighestArchSupported(0)
{
#ifdef VTKM_CUDA
static bool deviceQueryInit = false;
static int numDevices = 0;
static int archVersion = 0;
if (!deviceQueryInit)
{
deviceQueryInit = true;
//first query for the number of devices
VTKM_CUDA_CALL(cudaGetDeviceCount(&numDevices));
for (vtkm::Int32 i = 0; i < numDevices; i++)
{
cudaDeviceProp prop;
VTKM_CUDA_CALL(cudaGetDeviceProperties(&prop, i));
const vtkm::Int32 arch = (prop.major * 10) + prop.minor;
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
vtkm::cont::cuda::internal::DetermineIfValidCudaDevice<<<1, 1, 0, cudaStreamPerThread>>>();
cudaStreamSynchronize(cudaStreamPerThread);
if (cudaSuccess != cudaGetLastError())
{
numDevices = 0;
archVersion = 0;
}
}
this->NumberOfDevices = numDevices;
this->HighestArchSupported = archVersion;
#endif
}
bool DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagCuda>::Exists() const
{
return this->NumberOfDevices > 0 && this->HighestArchSupported >= 20;
}
}
} // namespace vtkm::cont

6
vtkm/cont/cuda/internal/DeviceAdapterRuntimeDetectorCuda.h
View File

@ -24,15 +24,17 @@
#include <vtkm/Types.h>
#include <vtkm/cont/DeviceAdapterAlgorithm.h>
#include <vtkm/cont/cuda/internal/DeviceAdapterTagCuda.h>
namespace vtkm
{
namespace cont
{
template <class DeviceAdapterTag>
class DeviceAdapterRuntimeDetector;
/// \brief Class providing a CUDA runtime support detector.
///
/// The class provide the actual implementation used by

4
vtkm/cont/cuda/internal/DeviceAdapterTimerImplementationCuda.h
View File

@ -17,8 +17,8 @@
// Laboratory (LANL), the U.S. Government retains certain rights in
// this software.
//============================================================================
#ifndef vtk_m_cont_cuda_internal_DeviceAdapterRuntimeDetectorCuda_h
#define vtk_m_cont_cuda_internal_DeviceAdapterRuntimeDetectorCuda_h
#ifndef vtk_m_cont_cuda_internal_DeviceAdapterTimerImplementationCuda_h
#define vtk_m_cont_cuda_internal_DeviceAdapterTimerImplementationCuda_h
#include <vtkm/cont/vtkm_cont_export.h>

26
vtkm/cont/internal/DeviceAdapterError.h
View File

@ -29,4 +29,30 @@
///
VTKM_INVALID_DEVICE_ADAPTER(Error, VTKM_DEVICE_ADAPTER_ERROR);
namespace vtkm
{
namespace cont
{
/// \brief Class providing a Error runtime support detector.
///
/// The class provide the actual implementation used by
/// vtkm::cont::RuntimeDeviceInformation for the Error backend.
///
/// We will always state that the current machine doesn't support
/// the error backend.
///
template <class DeviceAdapterTag>
class DeviceAdapterRuntimeDetector;
template <>
class DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagError>
{
public:
/// Returns false as the Error Device can never be run on.
VTKM_CONT bool Exists() const { return false; }
};
}
}
#endif //vtk_m_cont_internal_DeviceAdapterError_h

1
vtkm/cont/serial/DeviceAdapterSerial.h
View File

@ -23,6 +23,7 @@
// Keep headers in this order.
// clang-format off
#include <vtkm/cont/serial/internal/DeviceAdapterTagSerial.h>
#include <vtkm/cont/serial/internal/DeviceAdapterRuntimeDetectorSerial.h>
#include <vtkm/cont/serial/internal/ArrayManagerExecutionSerial.h>
#include <vtkm/cont/serial/internal/DeviceAdapterAlgorithmSerial.h>
#include <vtkm/cont/serial/internal/VirtualObjectTransferSerial.h>

2
vtkm/cont/serial/internal/CMakeLists.txt
View File

@ -21,6 +21,7 @@
set(headers
ArrayManagerExecutionSerial.h
DeviceAdapterAlgorithmSerial.h
DeviceAdapterRuntimeDetectorSerial.h
DeviceAdapterTagSerial.h
ExecutionArrayInterfaceBasicSerial.h
VirtualObjectTransferSerial.h
@ -30,5 +31,6 @@ vtkm_declare_headers(${headers})
target_sources(vtkm_cont PRIVATE
${CMAKE_CURRENT_SOURCE_DIR}/ArrayManagerExecutionSerial.cxx
${CMAKE_CURRENT_SOURCE_DIR}/DeviceAdapterAlgorithmSerial.cxx
${CMAKE_CURRENT_SOURCE_DIR}/DeviceAdapterRuntimeDetectorSerial.cxx
${CMAKE_CURRENT_SOURCE_DIR}/ExecutionArrayInterfaceBasicSerial.cxx
)

32
vtkm/cont/serial/internal/DeviceAdapterRuntimeDetectorSerial.cxx Normal file
View File

@ -0,0 +1,32 @@
//============================================================================
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2014 National Technology & Engineering Solutions of Sandia, LLC (NTESS).
// Copyright 2014 UT-Battelle, LLC.
// Copyright 2014 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/serial/internal/DeviceAdapterRuntimeDetectorSerial.h>
namespace vtkm
{
namespace cont
{
VTKM_CONT bool DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagSerial>::Exists() const
{
using DeviceAdapterTraits = vtkm::cont::DeviceAdapterTraits<vtkm::cont::DeviceAdapterTagSerial>;
return DeviceAdapterTraits::Valid;
}
}
}

47
vtkm/cont/serial/internal/DeviceAdapterRuntimeDetectorSerial.h Normal file
View File

@ -0,0 +1,47 @@
//============================================================================
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2014 National Technology & Engineering Solutions of Sandia, LLC (NTESS).
// Copyright 2014 UT-Battelle, LLC.
// Copyright 2014 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_cont_serial_internal_DeviceAdapterRuntimeDetector_h
#define vtk_m_cont_serial_internal_DeviceAdapterRuntimeDetector_h
#include <vtkm/cont/serial/internal/DeviceAdapterTagSerial.h>
#include <vtkm/cont/vtkm_cont_export.h>
namespace vtkm
{
namespace cont
{
template <class DeviceAdapterTag>
class DeviceAdapterRuntimeDetector;
/// Determine if this machine supports Serial backend
///
template <>
class VTKM_CONT_EXPORT DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagSerial>
{
public:
/// Returns true if the given device adapter is supported on the current
/// machine.
VTKM_CONT bool Exists() const;
};
}
}
#endif

1
vtkm/cont/tbb/DeviceAdapterTBB.h
View File

@ -20,6 +20,7 @@
#ifndef vtk_m_cont_tbb_DeviceAdapterTBB_h
#define vtk_m_cont_tbb_DeviceAdapterTBB_h
#include <vtkm/cont/tbb/internal/DeviceAdapterRuntimeDetectorTBB.h>
#include <vtkm/cont/tbb/internal/DeviceAdapterTagTBB.h>
#ifdef VTKM_ENABLE_TBB

6
vtkm/cont/tbb/internal/CMakeLists.txt
View File

@ -21,6 +21,7 @@
set(headers
ArrayManagerExecutionTBB.h
DeviceAdapterAlgorithmTBB.h
DeviceAdapterRuntimeDetectorTBB.h
DeviceAdapterTagTBB.h
ExecutionArrayInterfaceBasicTBB.h
FunctorsTBB.h
@ -41,6 +42,11 @@ vtkm_declare_headers(ParallelSortTBB.hxx
vtkm_declare_headers(${headers} TESTABLE ${VTKm_ENABLE_TBB})
#These sources need to always be built
target_sources(vtkm_cont PRIVATE
${CMAKE_CURRENT_SOURCE_DIR}/DeviceAdapterRuntimeDetectorTBB.cxx
)
#-----------------------------------------------------------------------------
if (NOT VTKm_ENABLE_TBB)
return()

31
vtkm/cont/tbb/internal/DeviceAdapterRuntimeDetectorTBB.cxx Normal file
View File

@ -0,0 +1,31 @@
//============================================================================
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2014 National Technology & Engineering Solutions of Sandia, LLC (NTESS).
// Copyright 2014 UT-Battelle, LLC.
// Copyright 2014 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/tbb/internal/DeviceAdapterRuntimeDetectorTBB.h>
namespace vtkm
{
namespace cont
{
VTKM_CONT bool DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagTBB>::Exists() const
{
using DeviceAdapterTraits = vtkm::cont::DeviceAdapterTraits<vtkm::cont::DeviceAdapterTagTBB>;
return DeviceAdapterTraits::Valid;
}
}
}

47
vtkm/cont/tbb/internal/DeviceAdapterRuntimeDetectorTBB.h Normal file
View File

@ -0,0 +1,47 @@
//============================================================================
// Copyright (c) Kitware, Inc.
// All rights reserved.
// See LICENSE.txt for details.
// This software is distributed WITHOUT ANY WARRANTY; without even
// the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR
// PURPOSE. See the above copyright notice for more information.
//
// Copyright 2014 National Technology & Engineering Solutions of Sandia, LLC (NTESS).
// Copyright 2014 UT-Battelle, LLC.
// Copyright 2014 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_cont_tbb_internal_DeviceAdapterRuntimeDetector_h
#define vtk_m_cont_tbb_internal_DeviceAdapterRuntimeDetector_h
#include <vtkm/cont/tbb/internal/DeviceAdapterTagTBB.h>
#include <vtkm/cont/vtkm_cont_export.h>
namespace vtkm
{
namespace cont
{
template <class DeviceAdapterTag>
class DeviceAdapterRuntimeDetector;
/// Determine if this machine supports Serial backend
///
template <>
class VTKM_CONT_EXPORT DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagTBB>
{
public:
/// Returns true if the given device adapter is supported on the current
/// machine.
VTKM_CONT bool Exists() const;
};
}
}
#endif

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

@ -68,6 +68,15 @@ public:
VTKM_CONT static void Synchronize() { Algorithm::Synchronize(); }
};
template <>
class DeviceAdapterRuntimeDetector<vtkm::cont::DeviceAdapterTagTestAlgorithmGeneral>
{
public:
/// Returns true as the General Algorithm Device can always be used.
VTKM_CONT bool Exists() const { return true; }
};
namespace internal
{

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

@ -37,27 +37,45 @@ template <typename DeviceAdapterTag>
void detect_if_exists(DeviceAdapterTag tag)
{
using DeviceAdapterTraits = vtkm::cont::DeviceAdapterTraits<DeviceAdapterTag>;
DoesExist<DeviceAdapterTraits::Valid>::Exist(tag);
std::cout << "testing runtime support for " << DeviceAdapterTraits::GetName() << std::endl;
DoesExist<DeviceAdapterTraits::Valid> exist;
exist.Exist(tag);
}
template <>
struct DoesExist<false>
{
template <typename DeviceAdapterTag>
static void Exist(DeviceAdapterTag)
void Exist(DeviceAdapterTag) const
{
//runtime information for this device should return false
vtkm::cont::RuntimeDeviceInformation<DeviceAdapterTag> runtime;
VTKM_TEST_ASSERT(runtime.Exists() == false,
"A backend with zero compile time support, can't have runtime support");
}
void Exist(vtkm::cont::DeviceAdapterTagCuda) const
{
//Since we are in a C++ compilation unit the Device Adapter
//trait should be false. But CUDA could still be enabled.
//That is why we check VTKM_ENABLE_CUDA.
vtkm::cont::RuntimeDeviceInformation<vtkm::cont::DeviceAdapterTagCuda> runtime;
#ifdef VTKM_ENABLE_CUDA
VTKM_TEST_ASSERT(runtime.Exists() == true,
"with cuda backend enabled, runtime support should be enabled");
#else
VTKM_TEST_ASSERT(runtime.Exists() == false,
"with cuda backend disabled, runtime support should be disabled");
#endif
}
};
template <>
struct DoesExist<true>
{
template <typename DeviceAdapterTag>
static void Exist(DeviceAdapterTag)
void Exist(DeviceAdapterTag) const
{
//runtime information for this device should return true
vtkm::cont::RuntimeDeviceInformation<DeviceAdapterTag> runtime;