kpet requested changes to this revision.
kpet added a comment.
This revision now requires changes to proceed.
Very useful to have all of this documented! Thanks!
================
Comment at: docs/LanguageExtensions.rst:1527
+This functionality is built on top of OpenCL C v2.0 and C++17. Regular C++
+features can be used in OpenCL kernel code. All functionality from OpenCL C
+is inherited. This section describes minor differences to OpenCL C and any
----------------
I'm tempted to suggest you merge the first two sentences and tone down the
claim in the second. "This functionality [...] and C++17, enabling _some_
regular C++ features [...].
================
Comment at: docs/LanguageExtensions.rst:1527
+This functionality is built on top of OpenCL C v2.0 and C++17. Regular C++
+features can be used in OpenCL kernel code. All functionality from OpenCL C
+is inherited. This section describes minor differences to OpenCL C and any
----------------
kpet wrote:
> I'm tempted to suggest you merge the first two sentences and tone down the
> claim in the second. "This functionality [...] and C++17, enabling _some_
> regular C++ features [...].
s/All/Most/? Stating that all functionality from OpenCL C is available right
before stating that the differences will be described feels like a
contradiction.
================
Comment at: docs/LanguageExtensions.rst:1535
+
+The following features are not supported:
+
----------------
Probably worth adding exceptions to the list.
================
Comment at: docs/LanguageExtensions.rst:1541
+- Standard C++ libraries. Currently there is no solution for alternative C++
+ libraries provided. Future release will feature library support.
+
----------------
I would remove this last sentence. Only a very small portion of the libray can
be supported anyway.
================
Comment at: docs/LanguageExtensions.rst:1553
+extension ISO/IEC JTC1 SC22 WG14 N1021 s3.1. Note that since the address space
+behavior in C++ is not documented formally yet, Clang extends existing concept
+from C and OpenCL. For example conversion rules are extended from qualification
----------------
I would remove the "yet", unless we can refer to an on-going and visible effort.
================
Comment at: docs/LanguageExtensions.rst:1553
+extension ISO/IEC JTC1 SC22 WG14 N1021 s3.1. Note that since the address space
+behavior in C++ is not documented formally yet, Clang extends existing concept
+from C and OpenCL. For example conversion rules are extended from qualification
----------------
kpet wrote:
> I would remove the "yet", unless we can refer to an on-going and visible
> effort.
"Clang extends existing concept" feels like it should be re-worded.
================
Comment at: docs/LanguageExtensions.rst:1555
+from C and OpenCL. For example conversion rules are extended from qualification
+conversion but the compatibility is determined using sets and overlapping from
+Embedded C (ISO/IEC JTC1 SC22 WG14 N1021 s3.1.3). For OpenCL it means that
----------------
"using sets and overlapping" feels incomplete. Pretending I can't guess: sets
of what? What is overlapping what? :)
================
Comment at: docs/LanguageExtensions.rst:1557
+Embedded C (ISO/IEC JTC1 SC22 WG14 N1021 s3.1.3). For OpenCL it means that
+implicit conversions are allowed from named to ``__generic`` but not vice versa
+(OpenCL C v2.0 s6.5.5) except for ``__constant`` address space. Most of the
----------------
Why use `__generic` when `generic` is equally valid and easier on the eyes? The
same comment applies to other mentions of address space keywords.
================
Comment at: docs/LanguageExtensions.rst:1558
+implicit conversions are allowed from named to ``__generic`` but not vice versa
+(OpenCL C v2.0 s6.5.5) except for ``__constant`` address space. Most of the
+rules are built on top of this behavior.
----------------
This is ambiguous. One could read that conversions from `__generic` to
`__constant` are allowed.
================
Comment at: docs/LanguageExtensions.rst:1564
+C style cast will follow OpenCL C v2.0 rules (s6.5.5). All cast operators will
+permit implicit conversion to ``__generic``. However converting from named
+address spaces to ``__generic`` can only be done using ``addrspace_cast``. Note
----------------
Isn't a conversion explicit if a cast operator is used?
================
Comment at: docs/LanguageExtensions.rst:1565
+permit implicit conversion to ``__generic``. However converting from named
+address spaces to ``__generic`` can only be done using ``addrspace_cast``. Note
+that conversions between ``__constant`` and any other is still disallowed.
----------------
At the time of writing `addrspace_cast` doesn't seem to exist in trunk and
C-style casts seem to work.
================
Comment at: docs/LanguageExtensions.rst:1565
+permit implicit conversion to ``__generic``. However converting from named
+address spaces to ``__generic`` can only be done using ``addrspace_cast``. Note
+that conversions between ``__constant`` and any other is still disallowed.
----------------
kpet wrote:
> At the time of writing `addrspace_cast` doesn't seem to exist in trunk and
> C-style casts seem to work.
Didn't you mean "from `__generic` to named address spaces"?
================
Comment at: docs/LanguageExtensions.rst:1577
+- non-pointer/non-reference class members except for static data members that
are
+ deduced to ``__global`` address space.
+- non-pointer/non-reference alias declarations.
----------------
Makes me wonder whether const members couldn't be in the `constant` address
space.
================
Comment at: docs/LanguageExtensions.rst:1626
+``__generic`` address space first if the conversion is valid. Therefore
programs
+using objects in ``__constant`` address space won't be compiled unless address
+space is explicitly specified using address space method qualifiers
----------------
the address space?
================
Comment at: docs/LanguageExtensions.rst:1631
+be used in case conversion to ``__generic`` address space is undesirable (even
if
+it is legal), for example to take advantage of memory bank accesses. Note this
not
+only applies to regular methods but to constructors and destructors too.
----------------
I would remove the "for example to take advantage of memory bank accesses" or
make the example clear.
================
Comment at: docs/LanguageExtensions.rst:1641
+the same as qualifying methods with ``const`` or any other qualifiers. The
overloading
+resolution will select overload with most specific address space if multiple
candidates
+are provided. If there is no conversion to an address space among existing
overloads
----------------
the overload
================
Comment at: docs/LanguageExtensions.rst:1652-1657
+ __kernel void bar() {
+ __local C c1; // will resolve to the first foo
+ C c2; // will resolve to the second foo
+ __constant C c3; // error due to mismatching address spaces - can't convert
to
+ // __local or __generic
+ }
----------------
mantognini wrote:
> `foo()` isn't actually called here. You probably meant to write this:
>
>
> ```
> __local C c1;
> C c2;
> __constant C c3;
>
> c1.foo(); // will resolve to the first foo
> c2.foo(); // will resolve to the second foo
> c3.foo(); // error due to mismatching address spaces - can't convert to
> __local or __generic
> ```
+1, I was about to make the same comment.
================
Comment at: docs/LanguageExtensions.rst:1668
+ // Has the following implicit definition
+ // void () __generic;
+ // void (const __generic C &) __generic;
----------------
Shouldn't this be `C() __generic`? Similar comments can be made for the other
members.
================
Comment at: docs/LanguageExtensions.rst:1684
+and dependent types (see :ref:`Deduction <opencl_cpp_addrsp_deduction>`). The
address
+space of template parameter is deduced during the type deduction if it's not
explicitly
+provided in instantiation.
----------------
template parameterS
================
Comment at: docs/LanguageExtensions.rst:1702
+instantiation. If multiple different address spaces are specified in template
definition and
+instantiation compilation of such program will fail with a diagnostic.
+
----------------
"such a program" or "such programs"
================
Comment at: docs/LanguageExtensions.rst:1760
+objects is executed and destroyed just after the last kernel using the
+program objects is executed. In OpenCL v2.0 drivers there is no specific
+API for invoking global constructors. However, an easy workaround would be
----------------
I would say "must be constructed before the first kernel begins execution" and
"may be destroyed once the last kernel has completed its execution". These are
requirements for a runtime environment, not functionality provided by Clang.
The documentation should only describe the interface with the runtime and the
requirements on the runtime.
================
Comment at: docs/LanguageExtensions.rst:1761
+program objects is executed. In OpenCL v2.0 drivers there is no specific
+API for invoking global constructors. However, an easy workaround would be
+to enqueue constructor initialization kernel that has a name
----------------
OpenCL doesn't have an API for this per-se, regardless of the version. The only
mechanism currently specified is the SPIR-V `Initializer` and `Finalizer`
execution modes for entrypoints, either of which currently require an OpenCL
2.2 implementation.
================
Comment at: docs/LanguageExtensions.rst:1762
+API for invoking global constructors. However, an easy workaround would be
+to enqueue constructor initialization kernel that has a name
+``@_GLOBAL__sub_I_<compiled file name>``. This kernel is only present if there
----------------
I would say that "applications are required to run all of these initialization
kernels before any others if they exist".
================
Comment at: docs/LanguageExtensions.rst:1779
+
+Global destructors can not be invoked in OpenCL v2.0 drivers. However, all
+memory used for program scope objects is released on ``clReleaseProgram``.
----------------
This could be managed by the application with a scheme similar to the one
you're proposing for constructor functions.
================
Comment at: docs/UsersManual.rst:2400
-Clang currently supports OpenCL C language standards up to v2.0.
+Clang currently supports OpenCL C language standards up to v2.0. Starting from
Clang9
+C++ mode is available for OpenCL (see :ref:`C++ for OpenCL <opencl_cpp>`).
----------------
Shouldn't "Clang9" be "Clang 9"? You may even want to spell it out "version 9
of Clang".
================
Comment at: docs/UsersManual.rst:2401
+Clang currently supports OpenCL C language standards up to v2.0. Starting from
Clang9
+C++ mode is available for OpenCL (see :ref:`C++ for OpenCL <opencl_cpp>`).
----------------
"A C++ mode" or "A non-standard C++ language extension" to be more precise?
================
Comment at: docs/UsersManual.rst:2769
+<https://www.khronos.org/registry/OpenCL/specs/2.2/pdf/OpenCL_Cxx.pdf>`_ and
+there is no plan to support it in clang in any new releases in the near future.
+
----------------
No plans that we're aware of ;). For all we know, there could be a
company/individual working on this. I think conjectures on people's plans don't
belong in the documentation.
================
Comment at: docs/UsersManual.rst:2771
+
+There are only a few restrictions on allowed C++ features. For detailed
information
+please refer to documentation on Extensions (:doc:`LanguageExtensions`).
----------------
Everything is relative. I would keep this 100% factual: "there are
restrictions". This is not a sales pitch :).
================
Comment at: docs/UsersManual.rst:2779
+To enable the new mode pass the following command line option when compiling
``.cl``
+file ``-cl-std=c++`` or ``-std=c++``.
+
----------------
I find `-std=c++` confusing/misleading. Maybe we should change the way
`-cl-std` works so we can have `-std=clc++` and `-cl-std=c++`?
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D64418/new/
https://reviews.llvm.org/D64418
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
