| /*===---- complex - CUDA wrapper for <algorithm> ----------------------------=== |
| * |
| * Permission is hereby granted, free of charge, to any person obtaining a copy |
| * of this software and associated documentation files (the "Software"), to deal |
| * in the Software without restriction, including without limitation the rights |
| * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell |
| * copies of the Software, and to permit persons to whom the Software is |
| * furnished to do so, subject to the following conditions: |
| * |
| * The above copyright notice and this permission notice shall be included in |
| * all copies or substantial portions of the Software. |
| * |
| * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR |
| * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, |
| * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE |
| * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER |
| * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, |
| * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN |
| * THE SOFTWARE. |
| * |
| *===-----------------------------------------------------------------------=== |
| */ |
| |
| #ifndef __CLANG_CUDA_WRAPPERS_ALGORITHM |
| #define __CLANG_CUDA_WRAPPERS_ALGORITHM |
| |
| // This header defines __device__ overloads of std::min/max, but only if we're |
| // <= C++11. In C++14, these functions are constexpr, and so are implicitly |
| // __host__ __device__. |
| // |
| // We don't support the initializer_list overloads because |
| // initializer_list::begin() and end() are not __host__ __device__ functions. |
| // |
| // When compiling in C++14 mode, we could force std::min/max to have different |
| // implementations for host and device, by declaring the device overloads |
| // before the constexpr overloads appear. We choose not to do this because |
| |
| // a) why write our own implementation when we can use one from the standard |
| // library? and |
| // b) libstdc++ is evil and declares min/max inside a header that is included |
| // *before* we include <algorithm>. So we'd have to unconditionally |
| // declare our __device__ overloads of min/max, but that would pollute |
| // things for people who choose not to include <algorithm>. |
| |
| #include_next <algorithm> |
| |
| #if __cplusplus <= 201103L |
| |
| // We need to define these overloads in exactly the namespace our standard |
| // library uses (including the right inline namespace), otherwise they won't be |
| // picked up by other functions in the standard library (e.g. functions in |
| // <complex>). Thus the ugliness below. |
| #ifdef _LIBCPP_BEGIN_NAMESPACE_STD |
| _LIBCPP_BEGIN_NAMESPACE_STD |
| #else |
| namespace std { |
| #ifdef _GLIBCXX_BEGIN_NAMESPACE_VERSION |
| _GLIBCXX_BEGIN_NAMESPACE_VERSION |
| #endif |
| #endif |
| |
| template <class __T, class __Cmp> |
| inline __device__ const __T & |
| max(const __T &__a, const __T &__b, __Cmp __cmp) { |
| return __cmp(__a, __b) ? __b : __a; |
| } |
| |
| template <class __T> |
| inline __device__ const __T & |
| max(const __T &__a, const __T &__b) { |
| return __a < __b ? __b : __a; |
| } |
| |
| template <class __T, class __Cmp> |
| inline __device__ const __T & |
| min(const __T &__a, const __T &__b, __Cmp __cmp) { |
| return __cmp(__b, __a) ? __b : __a; |
| } |
| |
| template <class __T> |
| inline __device__ const __T & |
| min(const __T &__a, const __T &__b) { |
| return __a < __b ? __b : __a; |
| } |
| |
| #ifdef _LIBCPP_END_NAMESPACE_STD |
| _LIBCPP_END_NAMESPACE_STD |
| #else |
| #ifdef _GLIBCXX_BEGIN_NAMESPACE_VERSION |
| _GLIBCXX_END_NAMESPACE_VERSION |
| #endif |
| } // namespace std |
| #endif |
| |
| #endif // __cplusplus <= 201103L |
| #endif // __CLANG_CUDA_WRAPPERS_ALGORITHM |