This is an archive of the discontinued LLVM Phabricator instance.

Differential D38326

[CUDA] Work around conflicting function definitions in CUDA-9 headers.
ClosedPublic

Authored by tra on Sep 27 2017, 12:01 PM.

Details

Reviewers
jlebar
Commits
rG93e33f8fb3db: [CUDA] Work around conflicting function definitions in CUDA-9 headers.
rC314334: [CUDA] Work around conflicting function definitions in CUDA-9 headers.
rL314334: [CUDA] Work around conflicting function definitions in CUDA-9 headers.

Diff Detail

Event Timeline

tra created this revision.Sep 27 2017, 12:01 PM
Herald added a subscriber: sanjoy. · View Herald TranscriptSep 27 2017, 12:01 PM
jlebar accepted this revision.Sep 27 2017, 12:07 PM
This revision is now accepted and ready to land.Sep 27 2017, 12:07 PM
Closed by commit rL314334: [CUDA] Work around conflicting function definitions in CUDA-9 headers. (authored by tra). · Explain WhySep 27 2017, 12:08 PM
This revision was automatically updated to reflect the committed changes.

Revision Contents

PathSize
clang/
lib/
Headers/
11 lines

Diff 116856

clang/lib/Headers/__clang_cuda_runtime_wrapper.h

Show First 20 Lines • Show All 167 Lines • ▼ Show 20 Lines
#endif #endif
// device_functions.hpp and math_functions*.hpp use 'static // device_functions.hpp and math_functions*.hpp use 'static
// __forceinline__' (with no __device__) for definitions of device // __forceinline__' (with no __device__) for definitions of device
// functions. Temporarily redefine __forceinline__ to include // functions. Temporarily redefine __forceinline__ to include
// __device__. // __device__.
#pragma push_macro("__forceinline__") #pragma push_macro("__forceinline__")
#define __forceinline__ __device__ __inline__ __attribute__((always_inline)) #define __forceinline__ __device__ __inline__ __attribute__((always_inline))
#pragma push_macro("__float2half_rn")
#if CUDA_VERSION >= 9000
// CUDA-9 has conflicting prototypes for __float2half_rn(float f) in
// cuda_fp16.h[pp] and device_functions.hpp. We need to get the one in
// device_functions.hpp out of the way.
#define __float2half_rn __float2half_rn_disabled
#endif
#include "device_functions.hpp" #include "device_functions.hpp"
#pragma pop_macro("__float2half_rn")
// math_function.hpp uses the __USE_FAST_MATH__ macro to determine whether we // math_function.hpp uses the __USE_FAST_MATH__ macro to determine whether we
// get the slow-but-accurate or fast-but-inaccurate versions of functions like // get the slow-but-accurate or fast-but-inaccurate versions of functions like
// sin and exp. This is controlled in clang by -fcuda-approx-transcendentals. // sin and exp. This is controlled in clang by -fcuda-approx-transcendentals.
// //
// device_functions.hpp uses __USE_FAST_MATH__ for a different purpose (fast vs. // device_functions.hpp uses __USE_FAST_MATH__ for a different purpose (fast vs.
// slow divides), so we need to scope our define carefully here. // slow divides), so we need to scope our define carefully here.
#pragma push_macro("__USE_FAST_MATH__") #pragma push_macro("__USE_FAST_MATH__")
▲ Show 20 Lines • Show All 180 Lines • Show Last 20 Lines