Commit Graph

11 Commits

Author SHA1 Message Date
Justin Lebar 11d5116904 [CUDA] Don't define functions that the CUDA headers themselves define on Windows.
Reviewers: tra

Subscribers: cfe-commits

Differential Revision: https://reviews.llvm.org/D28324

llvm-svn: 291137
2017-01-05 16:53:55 +00:00
Justin Lebar 50fe985349 [CUDA] Wrapper header changes necessary to support MacOS.
Reviewers: tra

Subscribers: cfe-commits

Differential Revision: https://reviews.llvm.org/D26780

llvm-svn: 287288
2016-11-18 00:41:35 +00:00
Justin Lebar e9eb792a0f [CUDA] Declare our __device__ math functions in the same inline namespace as our standard library.
Summary:
Currently we declare our inline __device__ math functions in namespace
std.  But libstdc++ and libc++ declare these functions in an inline
namespace inside namespace std.  We need to match this because, in a
later patch, we want to get e.g. <complex> to use our device overloads,
and it only will if those overloads are in the right inline namespace.

Reviewers: tra

Subscribers: cfe-commits, jhen

Differential Revision: https://reviews.llvm.org/D24977

llvm-svn: 283678
2016-10-08 22:16:03 +00:00
Justin Lebar e3612a039f [CUDA] Make __clang_cuda_cmath.h compatible with libc++.
Summary:
We need to add a bunch more "using"s, which weren't necessary with
libstdc++.

Once this is in I can check in a test to the test-suite.

Reviewers: tra

Subscribers: cfe-commits

Differential Revision: https://reviews.llvm.org/D24588

llvm-svn: 281544
2016-09-14 21:50:14 +00:00
Justin Lebar cb20a09f54 [CUDA] Improve handling of math functions.
Summary:
A bunch of related changes here to our CUDA math headers.

- The second arg to nexttoward is a double (well, technically, long
  double, but we don't have that), not a float.

- Add a forward-declare of llround(float), which is defined in the CUDA
  headers.  We need this for the same reason we need most of the other
  forward-declares: To prevent a constexpr function in our standard
  library from becoming host+device.

- Add nexttowardf implementation.

- Pull "foobarf" functions defined by the CUDA headers in the global
  namespace into namespace std.  This lets you do e.g. std::sinf.

- Add overloads for math functions accepting integer types.  This lets
  you do e.g. std::sin(0) without having an ambiguity between the
  overload that takes a float and the one that takes a double.

With these changes, we pass testcases derived from libc++ for cmath and
math.h.  We can check these testcases in to the test-suite once support
for CUDA lands there.

Reviewers: tra

Subscribers: cfe-commits

Differential Revision: https://reviews.llvm.org/D23627

llvm-svn: 279140
2016-08-18 20:43:13 +00:00
Justin Lebar 25c36fd61b [CUDA] Tweak math forward declares so we're compatible with libstdc++4.9.
Summary:
See comments in patch; we were assuming that some stdlib math functions
would be defined in namespace std, when in fact the spec says they
should be defined in the global namespace.  libstdc++4.9 became more
conforming and broke us.

This new implementation seems to cover the known knowns.

Reviewers: rsmith

Subscribers: cfe-commits, tra

Differential Revision: http://reviews.llvm.org/D18882

llvm-svn: 265751
2016-04-07 23:55:53 +00:00
Justin Lebar 50e5f184d8 [CUDA] Add missing #undef __DEVICE__ to CUDA shim header.
llvm-svn: 264742
2016-03-29 16:24:23 +00:00
Artem Belevich 7f522b7876 Added missing '__'.
llvm-svn: 260719
2016-02-12 20:26:43 +00:00
Eric Christopher 39a84d0b9b Update functions in clang supplied headers to use the compiler reserved
namespace for arguments.

llvm-svn: 260647
2016-02-12 02:22:53 +00:00
Artem Belevich 2aad2b3500 [CUDA] Bug 26497 : Remove wrappers for variants provided by CUDA headers.
... and pull global-scope ones into std namespace with using-declaration.

Differential Revision: http://reviews.llvm.org/D16932

llvm-svn: 259944
2016-02-05 22:54:05 +00:00
Artem Belevich c5f41a34e5 [CUDA] Implemented device-side support functions in <cmath>.
CUDA expects math functions in std:: namespace to work on device side.
In order to make it work with clang without allowing device-side code
generation for functions w/o appropriate target attributes, this patch
provides device-side implementations for <cmath> functions. Most of
them call global-scope math functions provided by CUDA headers. In few
cases we use clang builtins.

Tested out-of tree by compiling and running thrust's unit_tests.
https://github.com/thrust/thrust/tree/master/testing

Differential Revision: http://reviews.llvm.org/D16593

llvm-svn: 258880
2016-01-26 23:37:29 +00:00