[OpenCL] Disallow negative attribute arguments
Summary:
Negative arguments in kernel attributes are silently bitcast'ed to
unsigned, for example:
__attribute__((reqd_work_group_size(1, -1, 1)))
__kernel void k() {}
is a complete equivalent of:
__attribute__((reqd_work_group_size(1, 4294967294, 1)))
__kernel void k() {}
This is likely an error, so the patch forbids negative arguments in
several OpenCL attributes. Users who really want 4294967294 can still
use it as an unsigned representation.
Reviewers: Anastasia, yaxunl, bader
Reviewed By: Anastasia, yaxunl, bader
Subscribers: bader, cfe-commits
Differential Revision: https://reviews.llvm.org/D50259
llvm-svn: 341539
diff --git a/clang/lib/Sema/SemaDeclAttr.cpp b/clang/lib/Sema/SemaDeclAttr.cpp
index d72b8fb..488299a 100644
--- a/clang/lib/Sema/SemaDeclAttr.cpp
+++ b/clang/lib/Sema/SemaDeclAttr.cpp
@@ -227,9 +227,13 @@
/// If Expr is a valid integer constant, get the value of the integer
/// expression and return success or failure. May output an error.
+///
+/// Negative argument is implicitly converted to unsigned, unless
+/// \p StrictlyUnsigned is true.
template <typename AttrInfo>
static bool checkUInt32Argument(Sema &S, const AttrInfo &AI, const Expr *Expr,
- uint32_t &Val, unsigned Idx = UINT_MAX) {
+ uint32_t &Val, unsigned Idx = UINT_MAX,
+ bool StrictlyUnsigned = false) {
llvm::APSInt I(32);
if (Expr->isTypeDependent() || Expr->isValueDependent() ||
!Expr->isIntegerConstantExpr(I, S.Context)) {
@@ -249,6 +253,11 @@
return false;
}
+ if (StrictlyUnsigned && I.isSigned() && I.isNegative()) {
+ S.Diag(getAttrLoc(AI), diag::err_attribute_argument_negative) << AI;
+ return false;
+ }
+
Val = (uint32_t)I.getZExtValue();
return true;
}
@@ -2766,7 +2775,8 @@
uint32_t WGSize[3];
for (unsigned i = 0; i < 3; ++i) {
const Expr *E = AL.getArgAsExpr(i);
- if (!checkUInt32Argument(S, AL, E, WGSize[i], i))
+ if (!checkUInt32Argument(S, AL, E, WGSize[i], i,
+ /*StrictlyUnsigned=*/true))
return;
if (WGSize[i] == 0) {
S.Diag(AL.getLoc(), diag::err_attribute_argument_is_zero)