Split out the recursive and non-recursive mutex. This split is needed
for platforms which may use differing types for the two mutex (e.g.
Win32 threads).
llvm-svn: 291145
It seems that the ARM buildbots do not include x86 support. However,
other x86 targets do not support the ARM target. Use a x86 triple and
require the registered target.
llvm-svn: 291142
Summary:
MSVC seems to use "__in" and "__out" for its own purposes, so we have to
pick different names in this macro.
Reviewers: tra
Subscribers: cfe-commits
Differential Revision: https://reviews.llvm.org/D28325
llvm-svn: 291138
Summary:
When compiling device code, we may still see host code with explicit
calling conventions. NVPTX needs to claim that it supports these CCs,
so that (a) we don't raise noisy warnings, and (b) we don't break
existing code which relies on the existence of these CCs when
specializing templates. (If a CC doesn't exist, clang ignores it, so
two template specializations which are different only insofar as one
specifies a CC are considered identical and therefore are an error if
that CC is not supported.)
Reviewers: tra
Subscribers: cfe-commits
Differential Revision: https://reviews.llvm.org/D28323
llvm-svn: 291136
Summary:
CUDA lets users share structs between the host and device, so for that
and other reasons, primitive types such as ptrdiff_t should be the same
on both sides of the compilation.
Our code to do this wasn't entirely successful. In particular, we did a
bunch of work during the NVPTXTargetInfo constructor, only to override
it in the NVPTX{32,64}TargetInfo constructors. It worked well enough on
Linux and Mac, but Windows is LLP64, which is different enough to break
it.
This patch removes the NVPTX{32,64}TargetInfo classes entirely and fixes
the bug described above.
Reviewers: tra
Subscribers: cfe-commits
Differential Revision: https://reviews.llvm.org/D28322
llvm-svn: 291135
Summary: CUDA attributes are spelled __declspec(__foo__) on Windows.
Reviewers: tra
Subscribers: cfe-commits, rnk
Differential Revision: https://reviews.llvm.org/D28321
llvm-svn: 291134
Summary:
Preheader instruction's operands will always be invariant w.r.t. the loop which its the preheader
for.
Memory aliases are handled in canSinkOrHoistInst.
Reviewers: danielcdh, davidxl
Subscribers: mzolotukhin, llvm-commits
Differential Revision: https://reviews.llvm.org/D28270
llvm-svn: 291132
Summary:
For the most part this is straightforward: Just add a CudaInstallation
object to the MSVC and MinGW toolchains.
CudaToolChain has to override computeMSVCVersion so that
Clang::constructJob passes the right version flag to cc1. We have to
modify IsWindowsMSVC and friends in Clang::constructJob to be true when
compiling CUDA device code on Windows for the same reason.
Depends on: D28319
Reviewers: tra
Subscribers: cfe-commits
Differential Revision: https://reviews.llvm.org/D28320
llvm-svn: 291131
Summary:
Previously it was taking the true target triple, which is not really
what it needs: The location of the CUDA installation depends on the host
OS.
Reviewers: tra
Subscribers: cfe-commits
Differential Revision: https://reviews.llvm.org/D28319
llvm-svn: 291130
Summary:
When Sema looks up an attribute name, it strips off leading and trailing
"__" if the attribute is GNU-style. That is, __attribute__((foo)) and
__attribute__((__foo__)) are equivalent.
This is only true for GNU-style attributes. In particular,
__declspec(__foo__) is not equivalent to __declspec(foo), and Sema
respects this difference.
This patch fixes TableGen to match Sema's behavior. The spelling
'GNU<"__foo__">' should be normalized to 'GNU<"foo">', but
'Declspec<"__foo__">' should not be changed.
This is necessary to make CUDA compilation work on Windows, because e.g.
the __device__ attribute is spelled __declspec(__device__).
Attr.td does not contain any Declspec spellings that start or end with
"__", so this change should not affect any other attributes.
Reviewers: rnk
Subscribers: cfe-commits, tra
Differential Revision: https://reviews.llvm.org/D28318
llvm-svn: 291129
Summary:
These duplicate declarations cause a problem for CUDA compiles on
Windows. All implicitly-defined functions are host+device, and this
applies to the declarations in Builtin.def. But then when we see the
declarations in intrin.h, they have no attributes, so are host-only
functions. This is an error.
(A better fix might be to make these builtins host-only, but that is a
much bigger change.)
Reviewers: rnk
Subscribers: cfe-commits, echristo
Differential Revision: https://reviews.llvm.org/D28317
llvm-svn: 291128
Summary:
This patch adds two fields to the offload entry descriptor. One field is meant to signal Ctors/Dtors and `link` global variables, and the other is reserved for runtime library use.
Currently, these fields are only filled with zeros in the current code generation, but that will change when `declare target` is added.
The reason, we are adding these fields now is to make the code generation consistent with the runtime library proposal under review in https://reviews.llvm.org/D14031.
Reviewers: ABataev, hfinkel, carlo.bertolli, kkwli0, arpith-jacob, Hahnfeld
Subscribers: cfe-commits, caomhin, jholewinski
Differential Revision: https://reviews.llvm.org/D28298
llvm-svn: 291124
inline assembly may use the `.include` directive to include other
content into the file. Without the integrated assembler, the `-I` group
gets passed to the assembler. Emulate this by collecting the header
search paths and passing them to the IAS.
Resolves PR24811!
llvm-svn: 291123
This patch includes updates for codegen of the target region for the NVPTX
device. It moves initializers from the compiler to the runtime and updates
the worker loop to assume parallel work is retrieved from the runtime. A
subsequent patch will update the codegen to retrieve the parallel work using
calls to the runtime. It includes the removal of the inline attribute
for the worker loop and disabling debug info in it.
This allows codegen for a target directive and serial execution on the
NVPTX device.
Reviewers: ABataev
Differential Revision: https://reviews.llvm.org/D28125
llvm-svn: 291121
Summary:
For instructions such as PSLLW/PSLLD/PSLLQ a variable shift amount may be passed in an XMM register.
The lower 64-bits of the register are evaluated to determine the shift amount.
This patch improves the construction of the vector containing the shift amount.
Reviewers: craig.topper, delena, RKSimon
Subscribers: llvm-commits
Differential Revision: https://reviews.llvm.org/D28353
llvm-svn: 291120
Should fix some more bot failures from r291108.
This should have been a DenseSet, since GUID is not a pointer type.
It caused some bots to fail, but for some reason I wasnt't getting a
build failure.
llvm-svn: 291115
Should fix bot failures due to r291108 which happened due to a
change required in ModuleSummaryIndexYAML.h which was just added in
r291069.
llvm-svn: 291111
David Blaikie