Skip to content

Commit

Permalink
Get rid of RawMemoryAllocationFailure::AllocationMechanism and deri…
Browse files Browse the repository at this point in the history 
…ved backend-specific exceptions (kokkos#7139)

* Get rid of RawMemoryAllocationFailure::AllocationMechanism and derived exceptions

* Fix format remove trailing empty line

* Fixup from CI
  • Loading branch information
@dalg24
dalg24 authored Jul 17, 2024
1 parent 7247c7f commit 0a64dfc
Show file tree
Hide file tree
Showing 9 changed files with 139 additions and 257 deletions.
40 changes: 28 additions & 12 deletions core/src/Cuda/Kokkos_CudaSpace.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -56,6 +56,29 @@ const std::unique_ptr<Kokkos::Cuda> &Kokkos::Impl::cuda_get_deep_copy_space(
return space;
}

namespace {

auto get_failure_mode(cudaError_t error_code) {
using FailureMode =
Kokkos::Experimental::RawMemoryAllocationFailure::FailureMode;
switch (error_code) {
case cudaErrorMemoryAllocation: return FailureMode::OutOfMemoryError;
case cudaErrorInvalidValue: return FailureMode::InvalidAllocationSize;
default: return FailureMode::Unknown;
}
}

void throw_cuda_allocation_failure(size_t alloc_size, cudaError_t error_code,
std::string msg) {
msg += " returned error code \"";
msg += cudaGetErrorName(error_code);
msg += "\"";
Kokkos::Impl::throw_bad_alloc(alloc_size, std::align_val_t{1},
get_failure_mode(error_code), std::move(msg));
}

} // namespace

namespace Kokkos {
namespace Impl {

Expand Down Expand Up @@ -198,10 +221,7 @@ void *impl_allocate_common(const int device_id,
// we should do here since we're turning it into an
// exception here
cudaGetLastError();
throw Experimental::CudaRawMemoryAllocationFailure(
arg_alloc_size, error_code,
Experimental::RawMemoryAllocationFailure::AllocationMechanism::
CudaMalloc);
throw_cuda_allocation_failure(arg_alloc_size, error_code, "cudaMalloc()");
}

if (Kokkos::Profiling::profileLibraryLoaded()) {
Expand Down Expand Up @@ -255,10 +275,8 @@ void *CudaUVMSpace::impl_allocate(
// we should do here since we're turning it into an
// exception here
cudaGetLastError();
throw Experimental::CudaRawMemoryAllocationFailure(
arg_alloc_size, error_code,
Experimental::RawMemoryAllocationFailure::AllocationMechanism::
CudaMallocManaged);
throw_cuda_allocation_failure(arg_alloc_size, error_code,
"cudaMallocManaged()");
}

#ifdef KOKKOS_IMPL_DEBUG_CUDA_PIN_UVM_TO_HOST
Expand Down Expand Up @@ -299,10 +317,8 @@ void *CudaHostPinnedSpace::impl_allocate(
// we should do here since we're turning it into an
// exception here
cudaGetLastError();
throw Experimental::CudaRawMemoryAllocationFailure(
arg_alloc_size, error_code,
Experimental::RawMemoryAllocationFailure::AllocationMechanism::
CudaHostAlloc);
throw_cuda_allocation_failure(arg_alloc_size, error_code,
"cudaHostMalloc()");
}
if (Kokkos::Profiling::profileLibraryLoaded()) {
const size_t reported_size =
Expand Down
47 changes: 0 additions & 47 deletions core/src/Cuda/Kokkos_Cuda_Error.hpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -22,7 +22,6 @@

#include <impl/Kokkos_Error.hpp>
#include <impl/Kokkos_Profiling.hpp>
#include <iosfwd>

namespace Kokkos {
namespace Impl {
Expand Down Expand Up @@ -69,52 +68,6 @@ inline void cuda_internal_safe_call(cudaError e, const char* name,
Kokkos::Impl::cuda_internal_safe_call(call, #call, __FILE__, __LINE__)

} // namespace Impl

namespace Experimental {

class CudaRawMemoryAllocationFailure : public RawMemoryAllocationFailure {
private:
using base_t = RawMemoryAllocationFailure;

cudaError_t m_error_code = cudaSuccess;

static FailureMode get_failure_mode(cudaError_t error_code) {
switch (error_code) {
case cudaErrorMemoryAllocation: return FailureMode::OutOfMemoryError;
case cudaErrorInvalidValue: return FailureMode::InvalidAllocationSize;
// TODO handle cudaErrorNotSupported for cudaMallocManaged
default: return FailureMode::Unknown;
}
}

public:
// using base_t::base_t;
// would trigger
//
// error: cannot determine the exception specification of the default
// constructor due to a circular dependency
//
// using NVCC 9.1 and gcc 7.4
CudaRawMemoryAllocationFailure(
size_t arg_attempted_size, size_t arg_attempted_alignment,
FailureMode arg_failure_mode = FailureMode::OutOfMemoryError,
AllocationMechanism arg_mechanism =
AllocationMechanism::StdMalloc) noexcept
: base_t(arg_attempted_size, arg_attempted_alignment, arg_failure_mode,
arg_mechanism) {}

CudaRawMemoryAllocationFailure(size_t arg_attempted_size,
cudaError_t arg_error_code,
AllocationMechanism arg_mechanism) noexcept
: base_t(arg_attempted_size, /* CudaSpace doesn't handle alignment? */ 1,
get_failure_mode(arg_error_code), arg_mechanism),
m_error_code(arg_error_code) {}

void append_additional_error_information(std::ostream& o) const override;
};

} // end namespace Experimental

} // namespace Kokkos

#endif // KOKKOS_ENABLE_CUDA
Expand Down
37 changes: 0 additions & 37 deletions core/src/HIP/Kokkos_HIP_Error.hpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -22,8 +22,6 @@

#include <hip/hip_runtime.h>

#include <ostream>

namespace Kokkos {
namespace Impl {

Expand All @@ -44,39 +42,4 @@ inline void hip_internal_safe_call(hipError_t e, const char* name,
#define KOKKOS_IMPL_HIP_SAFE_CALL(call) \
Kokkos::Impl::hip_internal_safe_call(call, #call, __FILE__, __LINE__)

namespace Kokkos {
namespace Experimental {

class HIPRawMemoryAllocationFailure : public RawMemoryAllocationFailure {
private:
hipError_t m_error_code = hipSuccess;

static FailureMode get_failure_mode(hipError_t error_code) {
switch (error_code) {
case hipErrorMemoryAllocation: return FailureMode::OutOfMemoryError;
case hipErrorInvalidValue: return FailureMode::InvalidAllocationSize;
default: return FailureMode::Unknown;
}
}

public:
HIPRawMemoryAllocationFailure(size_t arg_attempted_size,
hipError_t arg_error_code,
AllocationMechanism arg_mechanism) noexcept
: RawMemoryAllocationFailure(
arg_attempted_size, /* HIPSpace doesn't handle alignment? */ 1,
get_failure_mode(arg_error_code), arg_mechanism),
m_error_code(arg_error_code) {}

void append_additional_error_information(std::ostream& o) const override {
if (m_error_code != hipSuccess) {
o << " The HIP allocation returned the error code \""
<< hipGetErrorName(m_error_code) << "\".";
}
}
};

} // namespace Experimental
} // namespace Kokkos

#endif
35 changes: 23 additions & 12 deletions core/src/HIP/Kokkos_HIP_Space.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -43,6 +43,25 @@ namespace {

static std::atomic<bool> is_first_hip_managed_allocation(true);

auto get_failure_mode(hipError_t error_code) {
using FailureMode =
Kokkos::Experimental::RawMemoryAllocationFailure::FailureMode;
switch (error_code) {
case hipErrorMemoryAllocation: return FailureMode::OutOfMemoryError;
case hipErrorInvalidValue: return FailureMode::InvalidAllocationSize;
default: return FailureMode::Unknown;
}
}

void throw_hip_allocation_failure(size_t alloc_size, hipError_t error_code,
std::string msg) {
msg += "returned error code \"";
msg += hipGetErrorName(error_code);
msg += "\"";
Kokkos::Impl::throw_bad_alloc(alloc_size, std::align_val_t{1},
get_failure_mode(error_code), std::move(msg));
}

} // namespace

/*--------------------------------------------------------------------------*/
Expand Down Expand Up @@ -77,10 +96,7 @@ void* HIPSpace::impl_allocate(
// This is the only way to clear the last error, which we should do here
// since we're turning it into an exception here
(void)hipGetLastError();
throw Experimental::HIPRawMemoryAllocationFailure(
arg_alloc_size, error_code,
Experimental::RawMemoryAllocationFailure::AllocationMechanism::
HIPMalloc);
throw_hip_allocation_failure(arg_alloc_size, error_code, "hipMalloc()");
}
if (Kokkos::Profiling::profileLibraryLoaded()) {
const size_t reported_size =
Expand Down Expand Up @@ -111,10 +127,7 @@ void* HIPHostPinnedSpace::impl_allocate(
// This is the only way to clear the last error, which we should do here
// since we're turning it into an exception here
(void)hipGetLastError();
throw Experimental::HIPRawMemoryAllocationFailure(
arg_alloc_size, error_code,
Experimental::RawMemoryAllocationFailure::AllocationMechanism::
HIPHostMalloc);
throw_hip_allocation_failure(arg_alloc_size, error_code, "hipHostMalloc()");
}
if (Kokkos::Profiling::profileLibraryLoaded()) {
const size_t reported_size =
Expand Down Expand Up @@ -178,10 +191,8 @@ Kokkos::HIP::runtime WARNING: Kokkos did not find an environment variable 'HSA_X
// This is the only way to clear the last error, which we should do here
// since we're turning it into an exception here
(void)hipGetLastError();
throw Experimental::HIPRawMemoryAllocationFailure(
arg_alloc_size, error_code,
Experimental::RawMemoryAllocationFailure::AllocationMechanism::
HIPMallocManaged);
throw_hip_allocation_failure(arg_alloc_size, error_code,
"hipMallocManaged()");
}
KOKKOS_IMPL_HIP_SAFE_CALL(hipMemAdvise(
ptr, arg_alloc_size, hipMemAdviseSetCoarseGrain, m_device));
Expand Down
16 changes: 6 additions & 10 deletions core/src/OpenACC/Kokkos_OpenACCSpace.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -67,16 +67,12 @@ void *Kokkos::Experimental::OpenACCSpace::impl_allocate(
ptr = acc_malloc(arg_alloc_size);

if (!ptr) {
size_t alignment = 1; // OpenACC does not handle alignment
using Kokkos::Experimental::RawMemoryAllocationFailure;
auto failure_mode =
arg_alloc_size > 0
? RawMemoryAllocationFailure::FailureMode::OutOfMemoryError
: RawMemoryAllocationFailure::FailureMode::InvalidAllocationSize;
auto alloc_mechanism =
RawMemoryAllocationFailure::AllocationMechanism::OpenACCMalloc;
throw RawMemoryAllocationFailure(arg_alloc_size, alignment, failure_mode,
alloc_mechanism);
using FailureMode =
Kokkos::Experimental::RawMemoryAllocationFailure::FailureMode;
auto failure_mode = arg_alloc_size > 0 ? FailureMode::OutOfMemoryError
: FailureMode::InvalidAllocationSize;
Kokkos::Impl::throw_bad_alloc(arg_alloc_size, std::align_val_t{1},
failure_mode, "acc_malloc()");
}

if (Kokkos::Profiling::profileLibraryLoaded()) {
Expand Down
Loading

0 comments on commit 0a64dfc

Please sign in to comment.