cuda_stream_view.hpp

前往此文件的文档。

/*
 * Copyright (c) 2020-2025, NVIDIA CORPORATION.
 *
 * Licensed under the Apache License, Version 2.0 (the "License");
 * you may not use this file except in compliance with the License.
 * You may obtain a copy of the License at
 *
 * https://apache.ac.cn/licenses/LICENSE-2.0
 *
 * Unless required by applicable law or agreed to in writing, software
 * distributed under the License is distributed on an "AS IS" BASIS,
 * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 * See the License for the specific language governing permissions and
 * limitations under the License.
 */

#pragma once

#include <rmm/detail/error.hpp>
#include <rmm/detail/export.hpp>

#include <cuda/stream_ref>
#include <cuda_runtime_api.h>

#include <cstddef>

namespace RMM_NAMESPACE {
class cuda_stream_view {
 public
 constexpr cuda_stream_view() = default;
 ~cuda_stream_view() = default;
 constexpr cuda_stream_view(cuda_stream_view const&) = default;
 constexpr cuda_stream_view(cuda_stream_view&&) = default;
 constexpr cuda_stream_view& operator=(cuda_stream_view const&) =
 default;
 constexpr cuda_stream_view& operator=(cuda_stream_view&&) =
 default;

 // Disable construction from literal 0
 constexpr cuda_stream_view(int) = delete; //< Prevent cast from 0
 constexpr cuda_stream_view(std::nullptr_t) = delete; //< Prevent cast from nullptr

 constexpr cuda_stream_view(cudaStream_t stream) noexcept : stream_{stream} {}

 constexpr cuda_stream_view(cuda::stream_ref stream) noexcept : stream_{stream.get()} {}

 [[nodiscard]] constexpr cudaStream_t value() const noexcept { return stream_; }

 constexpr operator cudaStream_t() const noexcept { return value(); }

 constexpr operator cuda::stream_ref() const noexcept { return value(); }

 [[nodiscard]] inline bool is_per_thread_default() const noexcept;

 [[nodiscard]] inline bool is_default() const noexcept;

 void synchronize() const { RMM_CUDA_TRY(cudaStreamSynchronize(stream_)); }

 void synchronize_no_throw() const noexcept
 {
 RMM_ASSERT_CUDA_SUCCESS(cudaStreamSynchronize(stream_));
 }

 private
 cudaStream_t stream_{};
};

static constexpr cuda_stream_view cuda_stream_default{};

static const cuda_stream_view cuda_stream_legacy{
 cudaStreamLegacy // NOLINT(cppcoreguidelines-pro-type-cstyle-cast)
};

static const cuda_stream_view cuda_stream_per_thread{
 cudaStreamPerThread // NOLINT(cppcoreguidelines-pro-type-cstyle-cast)
};

// Need to avoid putting is_per_thread_default and is_default into the group twice. // end of group

[[nodiscard]] inline bool cuda_stream_view::is_per_thread_default() const noexcept
{
#ifdef CUDA_API_PER_THREAD_DEFAULT_STREAM
 return value() == cuda_stream_per_thread || value() == nullptr;
#else
 return value() == cuda_stream_per_thread;
#endif
}

[[nodiscard]] inline bool cuda_stream_view::is_default() const noexcept
{
#ifdef CUDA_API_PER_THREAD_DEFAULT_STREAM
 return value() == cuda_stream_legacy;
#else
 return value() == cuda_stream_legacy || value() == nullptr;
#endif
}

inline bool operator==(cuda_stream_view lhs, cuda_stream_view rhs)
{
 return lhs.value() == rhs.value();
}

inline bool operator!=(cuda_stream_view lhs, cuda_stream_view rhs) { return not(lhs == rhs); }

inline std::ostream& operator<<(std::ostream& os, cuda_stream_view stream)
{
 os << stream.value();
 return os;
}
 // end of group
} // namespace RMM_NAMESPACE