[SYCL] Support intel::reqd_work_group_size#1328
Conversation
clang/lib/Sema/SemaDeclAttr.cpp
Outdated
There was a problem hiding this comment.
getAttributeSpellingListIndex exists for this purpose.
There was a problem hiding this comment.
But in this situation this statement sifts not just ReqdWorkGroupSize spellings, but also different WorkGroupAttr.
There was a problem hiding this comment.
I think, that getNormalizedFullName() more applicable here.
There was a problem hiding this comment.
We have 2 attributes with the same spelling!?! That is completely unacceptable.
There was a problem hiding this comment.
No, but getAttributeSpellingListIndex() returns unsigned and an attribute spelling index is a just unsigned in fact. And in result we get the same outcome (true) with diff spellings indices of diff attributes.
There was a problem hiding this comment.
Ah... I think I see the problem in TableGen. I'm going to push a fix to LLOrg in the next hour or so to make it so that this enum still gets emitted, since it is useful here.
There was a problem hiding this comment.
Suggested a review, but my change is here: https://reviews.llvm.org/D76289
There was a problem hiding this comment.
There, got it in! rG661c950630fb. Once that comes down, this line should be AL.getKind == ?? && AL.getSpelling == ReqdWorkGroupSize::CXX11_intel_reqd_work_group_size.
There was a problem hiding this comment.
Should I wait until the merge?
There was a problem hiding this comment.
Any examples with 2? What happens when I pass a negative to these? How about when things conflict with '1'?
There was a problem hiding this comment.
Since the defaults are 1, why set them so here (and around here)?
Since the advantage to this form of attribute is the 'default' values, perhaps we should have more tests evaluating that.
There was a problem hiding this comment.
What about GNU spelling?
There was a problem hiding this comment.
I don't sure, I will ask about it.
There was a problem hiding this comment.
The answer of John Pennycook is following:
"I hadn't thought about this before, but I'm going to say "No" – at least for now. If we're not using the C++11 attribute specifier syntax we can't have a namespace, and I don't want to set a precedent for converting between intel:: and intel_ that might accidentally introduce name conflicts."
There was a problem hiding this comment.
So that a GNU spelling doesn't need here.
There was a problem hiding this comment.
@Pennycook, please note that this is documentation for the compiler itself, not for the particular language. It is used to generated the following documentation: https://clang.llvm.org/docs/AttributeReference.html
So, this attribute is accepted by the compiler in GNU spelling and therefore, must be documented here as well if you adding such documentation. In SYCL-specific documents we can omit GNU spelling, but not here.
BTW, looking at the web-page I referenced above, it would be okay for me to either mention all declared spellings or just leave reqd_work_group_size generic string
There was a problem hiding this comment.
@fadeeval, this comment is the only one I have. I would like header to be aligned with the actual documentation, which describes all three spellings
erichkeane
left a comment
There was a problem hiding this comment.
As a side note, please give meaningful names to your commits/updates. I have no idea what 10.2 means.
clang/lib/Sema/SemaDeclAttr.cpp
Outdated
There was a problem hiding this comment.
There is an enum class that gets created in the attribute type itself that has these lists. Something like ReqdWorkGroupSizeAttr::CXX11_intel_reqd_work_group_size
This comment was marked as resolved.
This comment was marked as resolved.
Sorry, something went wrong.
There was a problem hiding this comment.
I cannot see the exact spelling because I don't have your patch installed, so i sort of guessed based on the spelling. Look in your build directory at:
whatever/build/tools/clang/include/clang/Sema/Attrs.inc
You'll see something like this:
class SectionAttr : public InheritableAttr {
unsigned nameLength;
char *name;
public:
enum Spelling {
GNU_section = 0,
CXX11_gnu_section = 1,
Declspec_allocate = 2,
SpellingNotCalculated = 15
};
...
See the Spelling enum which should have the entry for your new spelling.
There was a problem hiding this comment.
I watched in this file and I didn't find enum Spelling in class ReqdWorkGroupSize. It is very strange.
clang/lib/Sema/SemaDeclAttr.cpp
Outdated
There was a problem hiding this comment.
Note: Every comparison with this gives me a -Wsign-compare warning. XDim, YDim, and ZDim are all 'int' in type. This should be so as well likely.
There was a problem hiding this comment.
The problem is TableGen has no unsigned type.
There was a problem hiding this comment.
Okey, this problem I solved, it seems.
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
erichkeane
left a comment
There was a problem hiding this comment.
I'm OK with this, but @AlexeySachkov has a number of comments that I think are unaddressed. Once he's OK, I am as well.
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
ca62510 to
971eaa4
Compare
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
|
clang-format-check doesn't pass, because I don't like the option, it suggests to me. |
There was a problem hiding this comment.
@Pennycook, please note that this is documentation for the compiler itself, not for the particular language. It is used to generated the following documentation: https://clang.llvm.org/docs/AttributeReference.html
So, this attribute is accepted by the compiler in GNU spelling and therefore, must be documented here as well if you adding such documentation. In SYCL-specific documents we can omit GNU spelling, but not here.
BTW, looking at the web-page I referenced above, it would be okay for me to either mention all declared spellings or just leave reqd_work_group_size generic string
There is nothing wrong with suggested formatting, I suggest to apply it. Tagging @bader to comment |
Absolutely agree with @AlexeySachkov. |
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
| let Heading = "GNU<reqd_work_group_size>, CXX11<cl::reqd_work_group_size, | ||
| CXX11<intel::reqd_work_group_size>"; |
There was a problem hiding this comment.
| let Heading = "GNU<reqd_work_group_size>, CXX11<cl::reqd_work_group_size, | |
| CXX11<intel::reqd_work_group_size>"; | |
| let Heading = "__attribute__((reqd_work_group_size)), [[cl::reqd_work_group_size]], | |
| [[intel::reqd_work_group_size]]"; |
There was a problem hiding this comment.
buildbot/Build_PR_Win didn't pass because Heading, and writes:
"D:/buildbot/Product_worker_intel/Build_PR_With_Lit_Win/llvm.src/clang/include\clang/Basic/AttrDocs.td:2008:17: error: Unknown token when parsing a value
let Heading = "attribute((reqd_work_group_size)), [[cl::reqd_work_group_size]],"
There was a problem hiding this comment.
Maybe just "reqd_work_group_size" will be better?
There was a problem hiding this comment.
Just "reqd_work_group_size" works for me
unknown token when parsing a value
let Heading = "attribute((reqd_work_group_size)), [[cl::reqd_work_group_size]],"
Probably this is some kind of keyword which is being parsed regardless of the fact that it is within a string literal
| and allows to specify exact *local_work_size* which must be used as | ||
| argument to **clEnqueueNDRangeKernel** (in OpenCL) or to | ||
| **parallel_for** in SYCL. This allows the compiler to optimize the | ||
| generated code appropriately for this kernel. |
There was a problem hiding this comment.
I'm just a dumb purple bee, so
This allows the compiler to optimize the
generated code appropriately for this kernel.
Which kernel?
There was a problem hiding this comment.
The kernel to which attribute is applied. Maybe "the kernel" will be more correct then "this kernel", I think.
There was a problem hiding this comment.
Or better to write: "for the kernel to which attribute is applied"?
There was a problem hiding this comment.
Both variants looks good to me. I just copied this from OpenCL spec, but apparently it has more context so "this" is applicable there
| spelling is a bit different: | ||
|
|
||
| SYCL 1.2.1 describes ``[[cl::reqd_work_group_size(X, Y, Z)]]`` spelling: this | ||
| attribute is legal on device functions and their specification is propagated |
There was a problem hiding this comment.
Do you mean specification of the attribute is propagated somewhere?
There was a problem hiding this comment.
No, it says about function specification, that the attribute setup.
| down to any caller of those device functions, such that the kernel attributes | ||
| are the sum of all the kernel attributes of all device functions called. | ||
| See section 6.7 Attributes for more details. |
There was a problem hiding this comment.
I see that this phrase is actually copy-pasted from the spec, but it looks ugly. Maybe we could ask someone with good enough English to rewrite it...
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
| As Intel extension, ``[[intel::reqd_work_group_size(X, Y, Z)]]`` spelling is allowed | ||
| which features optional arguments `Y` and `Z`, which simplifies its usage if | ||
| only 1- or 2-dimensional ND-range is assumed by a programmer. These arguments | ||
| defaults to ``1``. |
There was a problem hiding this comment.
This is from the spec too? which after which feels ugly too. :)
| (``__attribute__((reqd_work_group_size(X, Y, Z)))``), see, for example, section | ||
| 6.7.2 Optional Attribute Qualifiers of OpenCL 1.2 specification |
There was a problem hiding this comment.
| (``__attribute__((reqd_work_group_size(X, Y, Z)))``), see, for example, section | |
| 6.7.2 Optional Attribute Qualifiers of OpenCL 1.2 specification | |
| (``__attribute__((reqd_work_group_size(X, Y, Z)))``), see section | |
| 6.7.2 Optional Attribute Qualifiers of OpenCL 1.2 specification for details. |
Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
…hinx * upstream/sycl: (357 commits) [Support] Implement a simple tabular data management library (intel#1358) [Support] Implement a property set I/O library (intel#1357) [SYCL] Fix buffer constructor using iterators (intel#1386) [SYCL][FPGA] Enable a set of loop attributes (intel#1312) [Driver][SYCL][FPGA] Proper dependency output location when given /Fo<dir> (intel#1346) [SPIR-V] Enabling SPIR-V builtin lookup in device SYCL mode (intel#1384) [SYCL][NFC] Unify setting kernel arguments (intel#1379) [SYCL][Doc] First revision of standard layout relaxation extension (intel#1344) [SYCL] Fixed sub-buffer alloca search (intel#1385) [SYCL][FPGA] Emit multiple IR variants for the IVDep attribute (intel#1383) [SYCL] Add experimental flag to enable front-end optimizations (intel#1376) [SYCL] Remove unexpected double in complex SPIR-V for float support (intel#1381) [SYCL] Default work-group sizes based on max (intel#952) [SYCL][CUDA] Fix usage of multiple backends in the same program (intel#1252) [SPIR-V] Add SPIR-V builtin definitions to the builtin lookup. [SPIR-V] Add macro definition when -fdeclare-spirv-builtins is activated [SYCL] Fix sycl_generic printing [SYCL] Support intel::reqd_work_group_size (intel#1328) [SYCL][NFC] Make the RT::PiPlugin object private (intel#1375) [SPIRV] Add convergent attribute to SPIR-V built-ins (intel#1373) ...
Implementing intel::reqd_work_group_size.
Signed-off-by: Aleksander Fadeev aleksander.fadeev@intel.com