Skip to content

Commit f376e43

Browse files
smanna12smanna12
committed
Update comments
1 parent 1137c2d commit f376e43

File tree

1 file changed

+32
-96
lines changed

1 file changed

+32
-96
lines changed

‎clang/lib/Sema/SemaDeclAttr.cpp

+32-96
Original file line numberDiff line numberDiff line change
@@ -3417,8 +3417,17 @@ static bool InvalidWorkGroupSizeAttrs(const Expr *MGValue, const Expr *XDim,
34173417
//
34183418
// If the 'reqd_work_group_size' attribute is specified on a declaration along
34193419
// with 'max_work_group_size' attribute, check to see if values of
3420-
// 'reqd_work_group_size' attribute arguments are equal and less than values of
3421-
// 'max_work_group_size' attribute arguments.
3420+
// 'reqd_work_group_size' attribute arguments are equal to or less than values
3421+
// of 'max_work_group_size' attribute arguments.
3422+
//
3423+
// The arguments to reqd_work_group_size are ordered based on which index
3424+
// increments the fastest. In OpenCL, the first argument is the index that
3425+
// increments the fastest, and in SYCL, the last argument is the index that
3426+
// increments the fastest.
3427+
//
3428+
// __attribute__((reqd_work_group_size)) follows the OpenCL rules in OpenCL
3429+
// mode. All spellings of reqd_work_group_size attribute (regardless of
3430+
// syntax used) follow the SYCL rules when in SYCL mode.
34223431
static bool checkMaxAllowedWorkGroupSize(
34233432
Sema &S, const Expr *RWGSXDim, const Expr *RWGSYDim, const Expr *RWGSZDim,
34243433
const Expr *MWGSXDim, const Expr *MWGSYDim, const Expr *MWGSZDim) {
@@ -3491,21 +3500,11 @@ void Sema::AddSYCLIntelMaxWorkGroupSizeAttr(Decl *D,
34913500

34923501
// If the 'max_work_group_size' attribute is specified on a declaration along
34933502
// with 'reqd_work_group_size' attribute, check to see if values of
3494-
// 'reqd_work_group_size' attribute arguments are equal or less than values
3503+
// 'reqd_work_group_size' attribute arguments are equal to or less than values
34953504
// of 'max_work_group_size' attribute arguments.
34963505
//
34973506
// We emit diagnostic if values of 'reqd_work_group_size' attribute arguments
34983507
// are greater than values of 'max_work_group_size' attribute arguments.
3499-
//
3500-
// The arguments to reqd_work_group_size are ordered based on which index
3501-
// increments the fastest. In OpenCL, the first argument is the index that
3502-
// increments the fastest, and in SYCL, the last argument is the index that
3503-
// increments the fastest.
3504-
//
3505-
// [[sycl::reqd_work_group_size]] and [[cl::reqd_work_group_size]] are
3506-
// available in SYCL modes and follow the SYCL rules.
3507-
// __attribute__((reqd_work_group_size)) is only available in OpenCL mode
3508-
// and follows the OpenCL rules.
35093508
if (const auto *DeclAttr = D->getAttr<ReqdWorkGroupSizeAttr>()) {
35103509
if (checkMaxAllowedWorkGroupSize(*this, DeclAttr->getXDim(),
35113510
DeclAttr->getYDim(), DeclAttr->getZDim(),
@@ -3518,7 +3517,7 @@ void Sema::AddSYCLIntelMaxWorkGroupSizeAttr(Decl *D,
35183517
}
35193518

35203519
// If the declaration has a SYCLIntelMaxWorkGroupSizeAttr, check to see if
3521-
// the attribute holds equal values to (1, 1, 1) in case the value of
3520+
// the attribute holds values equal to (1, 1, 1) in case the value of
35223521
// SYCLIntelMaxGlobalWorkDimAttr equals to 0.
35233522
if (const auto *DeclAttr = D->getAttr<SYCLIntelMaxGlobalWorkDimAttr>()) {
35243523
if (InvalidWorkGroupSizeAttrs(DeclAttr->getValue(), XDim, YDim, ZDim)) {
@@ -3579,21 +3578,11 @@ SYCLIntelMaxWorkGroupSizeAttr *Sema::MergeSYCLIntelMaxWorkGroupSizeAttr(
35793578

35803579
// If the 'max_work_group_size' attribute is specified on a declaration along
35813580
// with 'reqd_work_group_size' attribute, check to see if values of
3582-
// 'reqd_work_group_size' attribute arguments are equal or less than values
3581+
// 'reqd_work_group_size' attribute arguments are equal to or less than values
35833582
// of 'max_work_group_size' attribute arguments.
35843583
//
35853584
// We emit diagnostic if values of 'reqd_work_group_size' attribute arguments
35863585
// are greater than values of 'max_work_group_size' attribute arguments.
3587-
//
3588-
// The arguments to reqd_work_group_size are ordered based on which index
3589-
// increments the fastest. In OpenCL, the first argument is the index that
3590-
// increments the fastest, and in SYCL, the last argument is the index that
3591-
// increments the fastest.
3592-
//
3593-
// [[sycl::reqd_work_group_size]] and [[cl::reqd_work_group_size]] are
3594-
// available in SYCL modes and follow the SYCL rules.
3595-
// __attribute__((reqd_work_group_size)) is only available in OpenCL mode
3596-
// and follows the OpenCL rules.
35973586
if (const auto *DeclAttr = D->getAttr<ReqdWorkGroupSizeAttr>()) {
35983587
if (checkMaxAllowedWorkGroupSize(*this, DeclAttr->getXDim(),
35993588
DeclAttr->getYDim(), DeclAttr->getZDim(),
@@ -3605,10 +3594,9 @@ SYCLIntelMaxWorkGroupSizeAttr *Sema::MergeSYCLIntelMaxWorkGroupSizeAttr(
36053594
}
36063595
}
36073596

3608-
// If the declaration has a SYCLIntelMaxWorkGroupSizeAttr,
3609-
// check to see if the attribute holds equal values to
3610-
// (1, 1, 1) in case the value of SYCLIntelMaxGlobalWorkDimAttr
3611-
// equals to 0.
3597+
// If the declaration has a SYCLIntelMaxWorkGroupSizeAttr, check to see if
3598+
// the attribute holds values equal to (1, 1, 1) in case the value of
3599+
// SYCLIntelMaxGlobalWorkDimAttr equals to 0.
36123600
if (const auto *DeclAttr = D->getAttr<SYCLIntelMaxGlobalWorkDimAttr>()) {
36133601
if (InvalidWorkGroupSizeAttrs(DeclAttr->getValue(), A.getXDim(),
36143602
A.getYDim(), A.getZDim())) {
@@ -3639,6 +3627,10 @@ static void handleSYCLIntelMaxWorkGroupSize(Sema &S, Decl *D,
36393627
// increments the fastest. In OpenCL, the first argument is the index that
36403628
// increments the fastest, and in SYCL, the last argument is the index that
36413629
// increments the fastest.
3630+
//
3631+
// __attribute__((reqd_work_group_size)) follows the OpenCL rules in OpenCL
3632+
// mode. All spellings of reqd_work_group_size attribute (regardless of
3633+
// syntax used) follow the SYCL rules when in SYCL mode.
36423634
static bool CheckWorkGroupSize(Sema &S, const Expr *NSWIValue,
36433635
const Expr *RWGSXDim, const Expr *RWGSZDim) {
36443636
// If any of the operand is still value dependent, we can't test anything.
@@ -3657,7 +3649,7 @@ static bool CheckWorkGroupSize(Sema &S, const Expr *NSWIValue,
36573649
: (RWGSZDimExpr->getResultAsAPSInt()).getZExtValue();
36583650

36593651
// Check if the required work group size specified by 'num_simd_work_items'
3660-
// attribute must evenly divide the index that increments fastest in the
3652+
// attribute evenly divides the index that increments fastest in the
36613653
// 'reqd_work_group_size' attribute.
36623654
return WorkGroupSize % NSWIValueExpr->getResultAsAPSInt().getZExtValue() != 0;
36633655
}
@@ -3694,7 +3686,7 @@ void Sema::AddReqdWorkGroupSizeAttr(Decl *D, const AttributeCommonInfo &CI,
36943686
return;
36953687

36963688
// If the declaration has a ReqdWorkGroupSizeAttr, check to see if
3697-
// the attribute holds equal values to (1, 1, 1) in case the value of
3689+
// the attribute holds values equal to (1, 1, 1) in case the value of
36983690
// SYCLIntelMaxGlobalWorkDimAttr equals to 0.
36993691
if (const auto *DeclAttr = D->getAttr<SYCLIntelMaxGlobalWorkDimAttr>()) {
37003692
if (InvalidWorkGroupSizeAttrs(DeclAttr->getValue(), XDim, YDim, ZDim)) {
@@ -3705,20 +3697,11 @@ void Sema::AddReqdWorkGroupSizeAttr(Decl *D, const AttributeCommonInfo &CI,
37053697

37063698
// If the 'max_work_group_size' attribute is specified on a declaration along
37073699
// with 'reqd_work_group_size' attribute, check to see if values of
3708-
// 'reqd_work_group_size' attribute arguments are equal or less than values
3700+
// 'reqd_work_group_size' attribute arguments are equal to or less than values
37093701
// of 'max_work_group_size' attribute arguments.
37103702
//
37113703
// We emit diagnostic if values of 'reqd_work_group_size' attribute arguments
37123704
// are greater than values of 'max_work_group_size' attribute arguments.
3713-
//
3714-
// The arguments to reqd_work_group_size are ordered based on which index
3715-
// increments the fastest. In OpenCL, the first argument is the index that
3716-
// increments the fastest, and in SYCL, the last argument is the index that
3717-
// increments the fastest.
3718-
//
3719-
// __attribute__((reqd_work_group_size)) follows the OpenCL rules in OpenCL
3720-
// mode. All spellings of reqd_work_group_size attribute (regardless of
3721-
// syntax used) follow the SYCL rules when in SYCL mode.
37223705
if (const auto *DeclAttr = D->getAttr<SYCLIntelMaxWorkGroupSizeAttr>()) {
37233706
if (checkMaxAllowedWorkGroupSize(*this, XDim, YDim, ZDim,
37243707
DeclAttr->getXDim(), DeclAttr->getYDim(),
@@ -3734,15 +3717,6 @@ void Sema::AddReqdWorkGroupSizeAttr(Decl *D, const AttributeCommonInfo &CI,
37343717
// along with 'num_simd_work_items' attribute, the required work group size
37353718
// specified by 'num_simd_work_items' attribute must evenly divide the index
37363719
// that increments fastest in the 'reqd_work_group_size' attribute.
3737-
//
3738-
// The arguments to reqd_work_group_size are ordered based on which index
3739-
// increments the fastest. In OpenCL, the first argument is the index that
3740-
// increments the fastest, and in SYCL, the last argument is the index that
3741-
// increments the fastest.
3742-
//
3743-
// __attribute__((reqd_work_group_size)) follows the OpenCL rules in OpenCL
3744-
// mode. All spellings of reqd_work_group_size attribute
3745-
// (regardless of syntax used) follow the SYCL rules when in SYCL mode.
37463720
if (const auto *DeclAttr = D->getAttr<SYCLIntelNumSimdWorkItemsAttr>()) {
37473721
if (CheckWorkGroupSize(*this, DeclAttr->getValue(), XDim, ZDim)) {
37483722
Diag(DeclAttr->getLoc(), diag::err_sycl_num_kernel_wrong_reqd_wg_size)
@@ -3780,10 +3754,9 @@ void Sema::AddReqdWorkGroupSizeAttr(Decl *D, const AttributeCommonInfo &CI,
37803754

37813755
ReqdWorkGroupSizeAttr *
37823756
Sema::MergeReqdWorkGroupSizeAttr(Decl *D, const ReqdWorkGroupSizeAttr &A) {
3783-
// If the declaration has a ReqdWorkGroupSizeAttr,
3784-
// check to see if the attribute holds equal values to
3785-
// (1, 1, 1) in case the value of SYCLIntelMaxGlobalWorkDimAttr
3786-
// equals to 0.
3757+
// If the declaration has a ReqdWorkGroupSizeAttr, check to see if the
3758+
// attribute holds values equal to (1, 1, 1) in case the value of
3759+
// SYCLIntelMaxGlobalWorkDimAttr equals to 0.
37873760
if (const auto *DeclAttr = D->getAttr<SYCLIntelMaxGlobalWorkDimAttr>()) {
37883761
if (InvalidWorkGroupSizeAttrs(DeclAttr->getValue(), A.getXDim(),
37893762
A.getYDim(), A.getZDim())) {
@@ -3800,15 +3773,6 @@ Sema::MergeReqdWorkGroupSizeAttr(Decl *D, const ReqdWorkGroupSizeAttr &A) {
38003773
//
38013774
// We emit diagnostic if values of 'reqd_work_group_size' attribute arguments
38023775
// are greater than values of 'max_work_group_size' attribute arguments.
3803-
//
3804-
// The arguments to reqd_work_group_size are ordered based on which index
3805-
// increments the fastest. In OpenCL, the first argument is the index that
3806-
// increments the fastest, and in SYCL, the last argument is the index that
3807-
// increments the fastest.
3808-
//
3809-
// __attribute__((reqd_work_group_size)) follows the OpenCL rules in OpenCL
3810-
// mode. All spellings of reqd_work_group_size attribute (regardless of
3811-
// syntax used) follow the SYCL rules when in SYCL mode.
38123776
if (const auto *DeclAttr = D->getAttr<SYCLIntelMaxWorkGroupSizeAttr>()) {
38133777
if (checkMaxAllowedWorkGroupSize(
38143778
*this, A.getXDim(), A.getYDim(), A.getZDim(), DeclAttr->getXDim(),
@@ -3824,15 +3788,6 @@ Sema::MergeReqdWorkGroupSizeAttr(Decl *D, const ReqdWorkGroupSizeAttr &A) {
38243788
// along with 'num_simd_work_items' attribute, the required work group size
38253789
// specified by 'num_simd_work_items' attribute must evenly divide the index
38263790
// that increments fastest in the 'reqd_work_group_size' attribute.
3827-
//
3828-
// The arguments to reqd_work_group_size are ordered based on which index
3829-
// increments the fastest. In OpenCL, the first argument is the index that
3830-
// increments the fastest, and in SYCL, the last argument is the index that
3831-
// increments the fastest.
3832-
//
3833-
// __attribute__((reqd_work_group_size)) follows the OpenCL rules in OpenCL
3834-
// mode. All spellings of reqd_work_group_size attribute
3835-
// (regardless of syntax used) follow the SYCL rules when in SYCL mode.
38363791
if (const auto *DeclAttr = D->getAttr<SYCLIntelNumSimdWorkItemsAttr>()) {
38373792
if (CheckWorkGroupSize(*this, DeclAttr->getValue(), A.getXDim(),
38383793
A.getZDim())) {
@@ -4048,15 +4003,6 @@ void Sema::AddSYCLIntelNumSimdWorkItemsAttr(Decl *D,
40484003
// along with 'num_simd_work_items' attribute, the required work group size
40494004
// specified by 'num_simd_work_items' attribute must evenly divide the index
40504005
// that increments fastest in the 'reqd_work_group_size' attribute.
4051-
//
4052-
// The arguments to reqd_work_group_size are ordered based on which index
4053-
// increments the fastest. In OpenCL, the first argument is the index that
4054-
// increments the fastest, and in SYCL, the last argument is the index that
4055-
// increments the fastest.
4056-
//
4057-
// __attribute__((reqd_work_group_size)) follows the OpenCL rules in OpenCL
4058-
// mode. All spellings of reqd_work_group_size attribute
4059-
// (regardless of syntax used) follow the SYCL rules when in SYCL mode.
40604006
if (const auto *DeclAttr = D->getAttr<ReqdWorkGroupSizeAttr>()) {
40614007
if (CheckWorkGroupSize(*this, E, DeclAttr->getXDim(),
40624008
DeclAttr->getZDim())) {
@@ -4092,15 +4038,6 @@ SYCLIntelNumSimdWorkItemsAttr *Sema::MergeSYCLIntelNumSimdWorkItemsAttr(
40924038
// along with 'num_simd_work_items' attribute, the required work group size
40934039
// specified by 'num_simd_work_items' attribute must evenly divide the index
40944040
// that increments fastest in the 'reqd_work_group_size' attribute.
4095-
//
4096-
// The arguments to reqd_work_group_size are ordered based on which index
4097-
// increments the fastest. In OpenCL, the first argument is the index that
4098-
// increments the fastest, and in SYCL, the last argument is the index that
4099-
// increments the fastest.
4100-
//
4101-
// __attribute__((reqd_work_group_size)) follows the OpenCL rules in OpenCL
4102-
// mode. All spellings of reqd_work_group_size attribute
4103-
// (regardless of syntax used) follow the SYCL rules when in SYCL mode.
41044041
if (const auto *DeclAttr = D->getAttr<ReqdWorkGroupSizeAttr>()) {
41054042
if (CheckWorkGroupSize(*this, A.getValue(), DeclAttr->getXDim(),
41064043
DeclAttr->getZDim())) {
@@ -4344,9 +4281,9 @@ void Sema::AddSYCLIntelMaxGlobalWorkDimAttr(Decl *D,
43444281
}
43454282

43464283
// If the declaration has a SYCLIntelMaxWorkGroupSizeAttr or
4347-
// ReqdWorkGroupSizeAttr, check to see if they hold equal values
4348-
// (1, 1, 1) in case the value of SYCLIntelMaxGlobalWorkDimAttr
4349-
// equals to 0.
4284+
// ReqdWorkGroupSizeAttr, check to see if the attribute holds values equal
4285+
// to (1, 1, 1) in case the value of SYCLIntelMaxGlobalWorkDimAttr equals
4286+
// to 0.
43504287
if (ArgVal == 0) {
43514288
if (checkWorkGroupSizeAttrExpr<SYCLIntelMaxWorkGroupSizeAttr>(*this, D,
43524289
CI) ||
@@ -4376,9 +4313,8 @@ SYCLIntelMaxGlobalWorkDimAttr *Sema::MergeSYCLIntelMaxGlobalWorkDimAttr(
43764313
}
43774314

43784315
// If the declaration has a SYCLIntelMaxWorkGroupSizeAttr or
4379-
// ReqdWorkGroupSizeAttr, check to see if they hold equal values
4380-
// (1, 1, 1) in case the value of SYCLIntelMaxGlobalWorkDimAttr
4381-
// equals to 0.
4316+
// ReqdWorkGroupSizeAttr, check to see if the attribute holds values equal to
4317+
// (1, 1, 1) in case the value of SYCLIntelMaxGlobalWorkDimAttr equals to 0.
43824318
const auto *MergeExpr = dyn_cast<ConstantExpr>(A.getValue());
43834319
if (MergeExpr->getResultAsAPSInt() == 0) {
43844320
if (checkWorkGroupSizeAttrExpr<SYCLIntelMaxWorkGroupSizeAttr>(*this, D,

0 commit comments

Comments
 (0)