From 22d60463a6a378e8baccfe6b15898c8bddc23660 Mon Sep 17 00:00:00 2001 From: tapplencourt Date: Fri, 31 Jan 2025 17:53:41 +0000 Subject: [PATCH] add clang-format on source/ --- .clang-format | 67 +++++++++++++++++-- .github/workflows/CI.yml | 24 +++++++ adoc/code/algorithms.cpp | 12 ++-- adoc/code/anatomy.cpp | 14 ++-- adoc/code/aspectTraitExample.cpp | 11 ++- adoc/code/basicParallelForItem.cpp | 2 +- adoc/code/basicparallelfor.cpp | 2 +- adoc/code/bundle-builtin-kernel.cpp | 6 +- adoc/code/bundle-kernel-introspection.cpp | 12 ++-- adoc/code/bundle-pre-compile.cpp | 4 +- adoc/code/bundle-spec-constants.cpp | 8 +-- adoc/code/deviceHas.cpp | 8 +-- adoc/code/explicitcopy.cpp | 6 +- adoc/code/lambdaNameExamples.cpp | 10 +-- adoc/code/largesample.cpp | 26 +++---- adoc/code/myfunctor.cpp | 9 ++- adoc/code/parallelForWithKernelHandler.cpp | 6 +- .../parallelForWorkGroupWithKernelHandler.cpp | 2 +- adoc/code/parallelfor.cpp | 6 +- adoc/code/parallelforworkgroup.cpp | 2 +- adoc/code/propertyExample.cpp | 9 ++- adoc/code/queueShortcuts.cpp | 6 +- adoc/code/reduction.cpp | 10 +-- adoc/code/specialization_id.cpp | 40 +++++------ adoc/code/subbuffer.cpp | 20 +++--- adoc/code/swizzle-example.cpp | 16 ++--- adoc/code/sycl-external.cpp | 3 +- adoc/code/twoOptionalFeatures.cpp | 4 +- adoc/code/usingSpecConstants.cpp | 6 +- adoc/code/usm_device.cpp | 4 +- adoc/code/usm_shared.cpp | 4 +- 31 files changed, 218 insertions(+), 141 deletions(-) diff --git a/.clang-format b/.clang-format index 87b344537..138ee121e 100644 --- a/.clang-format +++ b/.clang-format @@ -1,8 +1,63 @@ -Language: Cpp -BasedOnStyle: LLVM +--- +Language: Cpp +# BasedOnStyle: Google AccessModifierOffset: -1 -BreakInheritanceList: BeforeComma -BreakConstructorInitializers: BeforeComma -Cpp11BracedListStyle: false +ConstructorInitializerIndentWidth: 4 +AlignEscapedNewlinesLeft: true +AlignTrailingComments: true +AllowAllParametersOfDeclarationOnNextLine: true +AllowShortBlocksOnASingleLine: false +AllowShortCaseLabelsOnASingleLine: false +AllowShortIfStatementsOnASingleLine: true +AllowShortLoopsOnASingleLine: true +AllowShortFunctionsOnASingleLine: All +AlwaysBreakAfterDefinitionReturnType: false +AlwaysBreakTemplateDeclarations: true +AlwaysBreakBeforeMultilineStrings: true +BreakBeforeBinaryOperators: None +BreakBeforeTernaryOperators: true +BreakConstructorInitializersBeforeComma: false +BinPackParameters: true +BinPackArguments: true +ColumnLimit: 80 +ConstructorInitializerAllOnOneLineOrOnePerLine: true +# Do not adapt pointer and reference alignment to the current style. +# Just use PointerAlignment parameter unconditionally for consistency. +DerivePointerAlignment: false +ExperimentalAutoDetectBinPacking: false +IndentCaseLabels: true +IndentWrappedFunctionNames: false +IndentFunctionDeclarationAfterType: false +MaxEmptyLinesToKeep: 1 +KeepEmptyLinesAtTheStartOfBlocks: false +NamespaceIndentation: None +ObjCSpaceAfterProperty: false +ObjCSpaceBeforeProtocolList: false +PenaltyBreakBeforeFirstCallParameter: 1 +PenaltyBreakComment: 300 +PenaltyBreakString: 1000 +PenaltyBreakFirstLessLess: 120 +PenaltyExcessCharacter: 1000000 +PenaltyReturnTypeOnItsOwnLine: 200 PointerAlignment: Left -SpaceBeforeCpp11BracedList: true +SpacesBeforeTrailingComments: 2 +Cpp11BracedListStyle: true +Standard: c++17 +IndentWidth: 2 +TabWidth: 8 +UseTab: Never +BreakBeforeBraces: Attach +SpacesInParentheses: false +SpacesInSquareBrackets: false +SpacesInAngles: false +SpaceInEmptyParentheses: false +SpacesInCStyleCastParentheses: false +SpaceAfterCStyleCast: false +SpacesInContainerLiterals: true +SpaceBeforeAssignmentOperators: true +ContinuationIndentWidth: 4 +CommentPragmas: '^ IWYU pragma:' +ForEachMacros: [ foreach, Q_FOREACH, BOOST_FOREACH ] +SpaceBeforeParens: ControlStatements +DisableFormat: false +SortIncludes: true diff --git a/.github/workflows/CI.yml b/.github/workflows/CI.yml index 3923b6cb9..2c707695d 100644 --- a/.github/workflows/CI.yml +++ b/.github/workflows/CI.yml @@ -52,3 +52,27 @@ jobs: name: spec-outputs path: | /tmp/out + + check-clang-format: + if: github.event_name == 'pull_request' + runs-on: ubuntu-latest + steps: + - name: Checkout + uses: actions/checkout@v3 + with: + fetch-depth: 0 # Need to check out base branch as well + - name: Install clang-format + run: sudo apt update; sudo apt install -y clang-format + # Inspired by SYCL-CTS + - name: Run clang-format on changed files + run: | + set -o errexit -o pipefail -o noclobber -o nounset + DIFF=$( git diff -U0 --no-color -C ./adoc/code/ ${{ github.event.pull_request.base.sha }} | ./adoc/scripts/clang-format-diff.py -p1 ) + if [ ! -z "$DIFF" ]; then + echo 'The following changes are not formatted according to `clang-format`:' >> $GITHUB_STEP_SUMMARY + echo '```diff' >> $GITHUB_STEP_SUMMARY + echo "$DIFF" >> $GITHUB_STEP_SUMMARY + echo '```' >> $GITHUB_STEP_SUMMARY + echo "Not all files are formatted according to clang-format. See workflow summary for details." + exit 1 # Fail CI run + fi diff --git a/adoc/code/algorithms.cpp b/adoc/code/algorithms.cpp index 14843372e..0191ebf72 100644 --- a/adoc/code/algorithms.cpp +++ b/adoc/code/algorithms.cpp @@ -1,17 +1,17 @@ // Copyright (c) 2011-2024 The Khronos Group, Inc. // SPDX-License-Identifier: Apache-2.0 -buffer inputBuf { 1024 }; -buffer outputBuf { 2 }; +buffer inputBuf{1024}; +buffer outputBuf{2}; { // Initialize buffer on the host with 0, 1, 2, 3, ..., 1023 - host_accessor a { inputBuf }; + host_accessor a{inputBuf}; std::iota(a.begin(), a.end(), 0); } myQueue.submit([&](handler& cgh) { - accessor inputValues { inputBuf, cgh, read_only }; - accessor outputValues { outputBuf, cgh, write_only, no_init }; + accessor inputValues{inputBuf, cgh, read_only}; + accessor outputValues{outputBuf, cgh, write_only, no_init}; cgh.parallel_for(nd_range<1>(range<1>(16), range<1>(16)), [=](nd_item<1> it) { // Apply a group algorithm to any number of values, described by an iterator @@ -31,5 +31,5 @@ myQueue.submit([&](handler& cgh) { }); }); -host_accessor a { outputBuf }; +host_accessor a{outputBuf}; assert(a[0] == 523776 && a[1] == 120); diff --git a/adoc/code/anatomy.cpp b/adoc/code/anatomy.cpp index 29c8c5860..d3467c1f3 100644 --- a/adoc/code/anatomy.cpp +++ b/adoc/code/anatomy.cpp @@ -3,10 +3,10 @@ #include #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names int main() { - int data[1024]; // Allocate data to be worked on + int data[1024]; // Allocate data to be worked on // Create a default queue to enqueue work to the default device queue myQueue; @@ -16,20 +16,20 @@ int main() { // because the destructor of resultBuf will wait { // Wrap our data variable in a buffer - buffer resultBuf { data, range<1> { 1024 } }; + buffer resultBuf{data, range<1>{1024}}; // Create a command group to issue commands to the queue myQueue.submit([&](handler& cgh) { // Request write access to the buffer without initialization - accessor writeResult { resultBuf, cgh, write_only, no_init }; + accessor writeResult{resultBuf, cgh, write_only, no_init}; // Enqueue a parallel_for task with 1024 work-items cgh.parallel_for(1024, [=](id<1> idx) { // Initialize each buffer element with its own rank number starting at 0 writeResult[idx] = idx; - }); // End of the kernel function - }); // End of our commands for this queue - } // End of scope, so we wait for work producing resultBuf to complete + }); // End of the kernel function + }); // End of our commands for this queue + } // End of scope, so we wait for work producing resultBuf to complete // Print result for (int i = 0; i < 1024; i++) diff --git a/adoc/code/aspectTraitExample.cpp b/adoc/code/aspectTraitExample.cpp index 9c7a419f4..91df3e1f8 100644 --- a/adoc/code/aspectTraitExample.cpp +++ b/adoc/code/aspectTraitExample.cpp @@ -2,11 +2,12 @@ // SPDX-License-Identifier: MIT #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names constexpr int N = 512; -template class MyKernel { +template +class MyKernel { public: void operator()(id<1> i) { if constexpr (HasFp16) { @@ -22,11 +23,9 @@ int main() { myQueue.submit([&](handler& cgh) { device dev = myQueue.get_device(); if (dev.has(aspect::fp16)) { - cgh.parallel_for(range { N }, - MyKernel> {}); + cgh.parallel_for(range{N}, MyKernel>{}); } else { - cgh.parallel_for(range { N }, - MyKernel> {}); + cgh.parallel_for(range{N}, MyKernel>{}); } }); diff --git a/adoc/code/basicParallelForItem.cpp b/adoc/code/basicParallelForItem.cpp index b44e6004f..74a020740 100644 --- a/adoc/code/basicParallelForItem.cpp +++ b/adoc/code/basicParallelForItem.cpp @@ -2,7 +2,7 @@ // SPDX-License-Identifier: Apache-2.0 myQueue.submit([&](handler& cgh) { - accessor acc { myBuffer, cgh, write_only }; + accessor acc{myBuffer, cgh, write_only}; cgh.parallel_for(range<1>(numWorkItems), [=](item<1> item) { // kernel argument type is item diff --git a/adoc/code/basicparallelfor.cpp b/adoc/code/basicparallelfor.cpp index bcb826548..535da3ca4 100644 --- a/adoc/code/basicparallelfor.cpp +++ b/adoc/code/basicparallelfor.cpp @@ -2,7 +2,7 @@ // SPDX-License-Identifier: Apache-2.0 myQueue.submit([&](handler& cgh) { - accessor acc { myBuffer, cgh, write_only }; + accessor acc{myBuffer, cgh, write_only}; cgh.parallel_for(range<1>(numWorkItems), [=](id<1> index) { acc[index] = 42.0f; }); diff --git a/adoc/code/bundle-builtin-kernel.cpp b/adoc/code/bundle-builtin-kernel.cpp index b0e1dc883..d7e6931d6 100644 --- a/adoc/code/bundle-builtin-kernel.cpp +++ b/adoc/code/bundle-builtin-kernel.cpp @@ -2,7 +2,7 @@ // SPDX-License-Identifier: Apache-2.0 #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names int main() { queue myQueue; @@ -15,7 +15,7 @@ int main() { // Get an executable kernel_bundle containing all the built-in kernels // supported by the device. kernel_bundle myBundle = - get_kernel_bundle(myContext, { myDevice }, builtinKernelIds); + get_kernel_bundle(myContext, {myDevice}, builtinKernelIds); // Retrieve a kernel object that can be used to query for more information // about the built-in kernel or to submit it to a command group. We assume @@ -26,7 +26,7 @@ int main() { myQueue.submit([&](handler& cgh) { // Setting the arguments depends on the backend and the exact kernel used. cgh.set_args(...); - cgh.parallel_for(range { 1024 }, builtinKernel); + cgh.parallel_for(range{1024}, builtinKernel); }); myQueue.wait(); diff --git a/adoc/code/bundle-kernel-introspection.cpp b/adoc/code/bundle-kernel-introspection.cpp index f4f26ecd9..86f04cad3 100644 --- a/adoc/code/bundle-kernel-introspection.cpp +++ b/adoc/code/bundle-kernel-introspection.cpp @@ -2,9 +2,9 @@ // SPDX-License-Identifier: Apache-2.0 #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names -class MyKernel; // Forward declare the name of our kernel. +class MyKernel; // Forward declare the name of our kernel. int main() { size_t N = 1024; @@ -15,7 +15,7 @@ int main() { // Get an executable kernel bundle containing our kernel. kernel_id kernelId = get_kernel_id(); auto myBundle = - get_kernel_bundle(myContext, { kernelId }); + get_kernel_bundle(myContext, {kernelId}); // Get the kernel's maximum work-group size when running on our device. kernel myKernel = myBundle.get_kernel(kernelId); @@ -24,11 +24,11 @@ int main() { // Compute a good ND-range to use for iteration in the kernel // based on the maximum work-group size. - std::array divisors = { 1024, 512, 256, 128, 64, 32, - 16, 8, 4, 2, 1 }; + std::array divisors = {1024, 512, 256, 128, 64, 32, + 16, 8, 4, 2, 1}; size_t wgSize = *std::find_if(divisors.begin(), divisors.end(), [=](auto d) { return (d <= maxWgSize); }); - nd_range myRange { range { N }, range { wgSize } }; + nd_range myRange{range{N}, range{wgSize}}; myQueue.submit([&](handler& cgh) { // Use the kernel bundle we queried, so we are sure the queried work-group diff --git a/adoc/code/bundle-pre-compile.cpp b/adoc/code/bundle-pre-compile.cpp index 3115b5ee8..52a4cd644 100644 --- a/adoc/code/bundle-pre-compile.cpp +++ b/adoc/code/bundle-pre-compile.cpp @@ -2,7 +2,7 @@ // SPDX-License-Identifier: Apache-2.0 #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names int main() { queue myQueue; @@ -18,7 +18,7 @@ int main() { // pre-compiled kernel from "myBundle". cgh.use_kernel_bundle(myBundle); - cgh.parallel_for(range { 1024 }, ([=](item index) { + cgh.parallel_for(range{1024}, ([=](item index) { // kernel code })); }); diff --git a/adoc/code/bundle-spec-constants.cpp b/adoc/code/bundle-spec-constants.cpp index 5d42df944..94be8b1b3 100644 --- a/adoc/code/bundle-spec-constants.cpp +++ b/adoc/code/bundle-spec-constants.cpp @@ -2,7 +2,7 @@ // SPDX-License-Identifier: Apache-2.0 #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names // Forward declare names for our two kernels. class MyKernel1; @@ -21,7 +21,7 @@ int main() { // Get the identifiers for our kernels, then get an input kernel bundle that // contains our two kernels. - auto kernelIds = { get_kernel_id(), get_kernel_id() }; + auto kernelIds = {get_kernel_id(), get_kernel_id()}; auto inputBundle = get_kernel_bundle(myContext, kernelIds); @@ -37,7 +37,7 @@ int main() { // Use the kernel bundle we built in this command group. cgh.use_kernel_bundle(exeBundle); cgh.parallel_for( - range { 1024 }, ([=](item index, kernel_handler kh) { + range{1024}, ([=](item index, kernel_handler kh) { // Read the value of the specialization constant. int w = kh.get_specialization_constant(); // ... @@ -48,7 +48,7 @@ int main() { // This command group uses the same kernel bundle. cgh.use_kernel_bundle(exeBundle); cgh.parallel_for( - range { 1024 }, ([=](item index, kernel_handler kh) { + range{1024}, ([=](item index, kernel_handler kh) { int h = kh.get_specialization_constant(); // ... })); diff --git a/adoc/code/deviceHas.cpp b/adoc/code/deviceHas.cpp index e00bd01c1..36decb23b 100644 --- a/adoc/code/deviceHas.cpp +++ b/adoc/code/deviceHas.cpp @@ -10,13 +10,13 @@ class KernelFunctor { private: void foo() const { - half fp = 1.0; // No compiler diagnostic here + half fp = 1.0; // No compiler diagnostic here } void bar() const { sycl::atomic_ref longAtomic(longValue); - longAtomic.fetchAdd(1); // ERROR: Compiler issues diagnostic because - // "aspect::atomic64" missing from "device_has()" + longAtomic.fetchAdd(1); // ERROR: Compiler issues diagnostic because + // "aspect::atomic64" missing from "device_has()" } }; @@ -25,5 +25,5 @@ class KernelFunctor { // still check the device's aspects before submitting the kernel. if (myQueue.get_device().has(aspect::fp16)) { myQueue.submit( - [&](handler& h) { h.parallel_for(range { 16 }, KernelFunctor {}); }); + [&](handler& h) { h.parallel_for(range{16}, KernelFunctor{}); }); } diff --git a/adoc/code/explicitcopy.cpp b/adoc/code/explicitcopy.cpp index a68093753..47b5fa5b9 100644 --- a/adoc/code/explicitcopy.cpp +++ b/adoc/code/explicitcopy.cpp @@ -4,11 +4,11 @@ const size_t nElems = 10u; // Create a vector and fill it with values 0, 1, 2, 3, 4, 5, 6, 7, 8, 9 -std::vector v { nElems }; +std::vector v{nElems}; std::iota(std::begin(v), std::end(v), 0); // Create a buffer with no associated user storage -sycl::buffer b { range<1>(nElems) }; +sycl::buffer b{range<1>(nElems)}; // Create a queue queue myQueue; @@ -16,7 +16,7 @@ queue myQueue; myQueue.submit([&](handler& cgh) { // Retrieve a ranged write accessor to a global buffer with access to the // first half of the buffer - accessor acc { b, cgh, range<1>(nElems / 2), id<1>(0), write_only }; + accessor acc{b, cgh, range<1>(nElems / 2), id<1>(0), write_only}; // Copy the first five elements of the vector into the buffer associated with // the accessor cgh.copy(v.data(), acc); diff --git a/adoc/code/lambdaNameExamples.cpp b/adoc/code/lambdaNameExamples.cpp index cb8b3bb85..c315971b2 100644 --- a/adoc/code/lambdaNameExamples.cpp +++ b/adoc/code/lambdaNameExamples.cpp @@ -4,10 +4,12 @@ // Explicit kernel names can be optionally forward declared at namespace scope class MyForwardDeclName; -template class MyTemplatedKernelName; +template +class MyTemplatedKernelName; // Define and launch templated kernel -template void templatedFunction() { +template +void templatedFunction() { queue myQueue; // Launch A: No explicit kernel name @@ -47,9 +49,9 @@ int main() { }); }); - templatedFunction(); // OK + templatedFunction(); // OK - templatedFunction>(); // Launch A is OK, Launch B illegal + templatedFunction>(); // Launch A is OK, Launch B illegal // because std::complex is not forward declarable according to C++, and was // used in an explicit kernel name which must be forward declarable. } diff --git a/adoc/code/largesample.cpp b/adoc/code/largesample.cpp index dfa4d3389..379ec8c20 100644 --- a/adoc/code/largesample.cpp +++ b/adoc/code/largesample.cpp @@ -3,7 +3,7 @@ #include #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names // Size of the matrices constexpr size_t N = 2000; @@ -14,31 +14,31 @@ int main() { queue myQueue; // Create some 2D buffers of float for our matrices - buffer a { range<2> { N, M } }; - buffer b { range<2> { N, M } }; - buffer c { range<2> { N, M } }; + buffer a{range<2>{N, M}}; + buffer b{range<2>{N, M}}; + buffer c{range<2>{N, M}}; // Launch an asynchronous kernel to initialize a myQueue.submit([&](handler& cgh) { // The kernel writes a, so get a write accessor on it - accessor A { a, cgh, write_only }; + accessor A{a, cgh, write_only}; // Enqueue a parallel kernel iterating on a N*M 2D iteration space - cgh.parallel_for(range<2> { N, M }, + cgh.parallel_for(range<2>{N, M}, [=](id<2> index) { A[index] = index[0] * 2 + index[1]; }); }); // Launch an asynchronous kernel to initialize b myQueue.submit([&](handler& cgh) { // The kernel writes b, so get a write accessor on it - accessor B { b, cgh, write_only }; + accessor B{b, cgh, write_only}; // From the access pattern above, the SYCL runtime detects that this // command_group is independent from the first one and can be // scheduled independently // Enqueue a parallel kernel iterating on a N*M 2D iteration space - cgh.parallel_for(range<2> { N, M }, [=](id<2> index) { + cgh.parallel_for(range<2>{N, M}, [=](id<2> index) { B[index] = index[0] * 2014 + index[1] * 42; }); }); @@ -46,21 +46,21 @@ int main() { // Launch an asynchronous kernel to compute matrix addition c = a + b myQueue.submit([&](handler& cgh) { // In the kernel a and b are read, but c is written - accessor A { a, cgh, read_only }; - accessor B { b, cgh, read_only }; - accessor C { c, cgh, write_only }; + accessor A{a, cgh, read_only}; + accessor B{b, cgh, read_only}; + accessor C{c, cgh, write_only}; // From these accessors, the SYCL runtime will ensure that when // this kernel is run, the kernels computing a and b have completed // Enqueue a parallel kernel iterating on a N*M 2D iteration space - cgh.parallel_for(range<2> { N, M }, + cgh.parallel_for(range<2>{N, M}, [=](id<2> index) { C[index] = A[index] + B[index]; }); }); // Ask for an accessor to read c from application scope. The SYCL runtime // waits for c to be ready before returning from the constructor - host_accessor C { c, read_only }; + host_accessor C{c, read_only}; std::cout << std::endl << "Result:" << std::endl; for (size_t i = 0; i < N; i++) { for (size_t j = 0; j < M; j++) { diff --git a/adoc/code/myfunctor.cpp b/adoc/code/myfunctor.cpp index 76e9fbad7..faec5b9ea 100644 --- a/adoc/code/myfunctor.cpp +++ b/adoc/code/myfunctor.cpp @@ -3,10 +3,9 @@ class RandomFiller { public: - RandomFiller(accessor ptr) - : ptr_ { ptr } { + RandomFiller(accessor ptr) : ptr_{ptr} { std::random_device hwRand; - std::uniform_int_distribution<> r { 1, 100 }; + std::uniform_int_distribution<> r{1, 100}; randomNum_ = r(hwRand); } void operator()(item<1> item) const { ptr_[item.get_id()] = get_random(); } @@ -19,8 +18,8 @@ class RandomFiller { void workFunction(buffer& b, queue& q, const range<1> r) { q.submit([&](handler& cgh) { - accessor ptr { b, cgh }; - RandomFiller filler { ptr }; + accessor ptr{b, cgh}; + RandomFiller filler{ptr}; cgh.parallel_for(r, filler); }); diff --git a/adoc/code/parallelForWithKernelHandler.cpp b/adoc/code/parallelForWithKernelHandler.cpp index dc03e0862..525751bd1 100644 --- a/adoc/code/parallelForWithKernelHandler.cpp +++ b/adoc/code/parallelForWithKernelHandler.cpp @@ -2,7 +2,7 @@ // SPDX-License-Identifier: Apache-2.0 myQueue.submit([&](handler& cgh) { - cgh.parallel_for(range<3>(3, 3, 3), // global range + cgh.parallel_for(range<3>(3, 3, 3), // global range [=](item<3> it, kernel_handler kh) { //[kernel code] }); @@ -11,8 +11,8 @@ myQueue.submit([&](handler& cgh) { // This form of parallel_for with the "offset" parameter is deprecated in SYCL // 2020 myQueue.submit([&](handler& cgh) { - cgh.parallel_for(range<3>(3, 3, 3), // global range - id<3>(1, 1, 1), // offset + cgh.parallel_for(range<3>(3, 3, 3), // global range + id<3>(1, 1, 1), // offset [=](item<3> it, kernel_handler kh) { //[kernel code] }); diff --git a/adoc/code/parallelForWorkGroupWithKernelHandler.cpp b/adoc/code/parallelForWorkGroupWithKernelHandler.cpp index 1d391b836..de525405e 100644 --- a/adoc/code/parallelForWorkGroupWithKernelHandler.cpp +++ b/adoc/code/parallelForWorkGroupWithKernelHandler.cpp @@ -7,7 +7,7 @@ myQueue.submit([&](handler& cgh) { range<3>(2, 2, 2), range<3>(2, 2, 2), [=](group<3> myGroup, kernel_handler kh) { //[workgroup code] - int myLocal; // this variable is shared between workitems + int myLocal; // this variable is shared between workitems // this variable will be instantiated for each work-item separately private_memory myPrivate(myGroup); diff --git a/adoc/code/parallelfor.cpp b/adoc/code/parallelfor.cpp index 5004684f8..49f8df0f4 100644 --- a/adoc/code/parallelfor.cpp +++ b/adoc/code/parallelfor.cpp @@ -2,7 +2,7 @@ // SPDX-License-Identifier: Apache-2.0 myQueue.submit([&](handler& cgh) { - cgh.parallel_for(range<3>(3, 3, 3), // global range + cgh.parallel_for(range<3>(3, 3, 3), // global range [=](item<3> it) { //[kernel code] }); @@ -11,8 +11,8 @@ myQueue.submit([&](handler& cgh) { // This form of parallel_for with the "offset" parameter is deprecated in SYCL // 2020 myQueue.submit([&](handler& cgh) { - cgh.parallel_for(range<3>(3, 3, 3), // global range - id<3>(1, 1, 1), // offset + cgh.parallel_for(range<3>(3, 3, 3), // global range + id<3>(1, 1, 1), // offset [=](item<3> it) { //[kernel code] }); diff --git a/adoc/code/parallelforworkgroup.cpp b/adoc/code/parallelforworkgroup.cpp index 35041f77d..e70b4e09c 100644 --- a/adoc/code/parallelforworkgroup.cpp +++ b/adoc/code/parallelforworkgroup.cpp @@ -6,7 +6,7 @@ myQueue.submit([&](handler& cgh) { cgh.parallel_for_work_group( range<3>(2, 2, 2), range<3>(2, 2, 2), [=](group<3> myGroup) { //[workgroup code] - int myLocal; // this variable is shared between workitems + int myLocal; // this variable is shared between workitems // this variable will be instantiated for each work-item separately private_memory myPrivate(myGroup); diff --git a/adoc/code/propertyExample.cpp b/adoc/code/propertyExample.cpp index a915562c3..b12d0d499 100644 --- a/adoc/code/propertyExample.cpp +++ b/adoc/code/propertyExample.cpp @@ -4,11 +4,10 @@ { context myContext; - std::vector> bufferList { - buffer { ptr, rng }, - buffer { ptr, rng, property::use_host_ptr {} }, - buffer { ptr, rng, property::context_bound { myContext } } - }; + std::vector> bufferList{ + buffer{ptr, rng}, + buffer{ptr, rng, property::use_host_ptr{}}, + buffer{ptr, rng, property::context_bound{myContext}}}; for (auto& buf : bufferList) { if (buf.has_property()) { diff --git a/adoc/code/queueShortcuts.cpp b/adoc/code/queueShortcuts.cpp index 336e8a15a..e78bd2572 100644 --- a/adoc/code/queueShortcuts.cpp +++ b/adoc/code/queueShortcuts.cpp @@ -4,11 +4,11 @@ class MyKernel; queue myQueue; -auto usmPtr = malloc_device(1024, myQueue); // USM pointer +auto usmPtr = malloc_device(1024, myQueue); // USM pointer int* data = /* pointer to some data */; -buffer buf { data, 1024 }; -accessor acc { buf }; // Placeholder accessor +buffer buf{data, 1024}; +accessor acc{buf}; // Placeholder accessor // Queue shortcut for a kernel invocation myQueue.single_task([=] { diff --git a/adoc/code/reduction.cpp b/adoc/code/reduction.cpp index 63c55fb5c..30eb16037 100644 --- a/adoc/code/reduction.cpp +++ b/adoc/code/reduction.cpp @@ -1,18 +1,18 @@ // Copyright (c) 2011-2024 The Khronos Group, Inc. // SPDX-License-Identifier: Apache-2.0 -buffer valuesBuf { 1024 }; +buffer valuesBuf{1024}; { // Initialize buffer on the host with 0, 1, 2, 3, ..., 1023 - host_accessor a { valuesBuf }; + host_accessor a{valuesBuf}; std::iota(a.begin(), a.end(), 0); } // Buffers with just 1 element to get the reduction results int sumResult = 0; -buffer sumBuf { &sumResult, 1 }; +buffer sumBuf{&sumResult, 1}; int maxResult = 0; -buffer maxBuf { &maxResult, 1 }; +buffer maxBuf{&maxResult, 1}; myQueue.submit([&](handler& cgh) { // Input values to reductions are standard accessors @@ -26,7 +26,7 @@ myQueue.submit([&](handler& cgh) { // For each reduction variable, the implementation: // - Creates a corresponding reducer // - Passes a reference to the reducer to the lambda as a parameter - cgh.parallel_for(range<1> { 1024 }, sumReduction, maxReduction, + cgh.parallel_for(range<1>{1024}, sumReduction, maxReduction, [=](id<1> idx, auto& sum, auto& max) { // plus<>() corresponds to += operator, so sum can be // updated via += or combine() diff --git a/adoc/code/specialization_id.cpp b/adoc/code/specialization_id.cpp index 0ce798fd3..b78c32aa6 100644 --- a/adoc/code/specialization_id.cpp +++ b/adoc/code/specialization_id.cpp @@ -2,50 +2,50 @@ // SPDX-License-Identifier: Apache-2.0 #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names struct Compound { int i; float f; }; -constexpr specialization_id a { 1 }; // OK -constexpr specialization_id b { 2, 3.14 }; // OK -inline constexpr specialization_id c { 3 }; // OK -static constexpr specialization_id d { 4 }; // OK -specialization_id e { 5 }; // ILLEGAL: not constexpr +constexpr specialization_id a{1}; // OK +constexpr specialization_id b{2, 3.14}; // OK +inline constexpr specialization_id c{3}; // OK +static constexpr specialization_id d{4}; // OK +specialization_id e{5}; // ILLEGAL: not constexpr struct Bar { - static constexpr specialization_id f { 6 }; // OK + static constexpr specialization_id f{6}; // OK }; struct Baz { struct Inner { - static constexpr specialization_id g { 7 }; // OK + static constexpr specialization_id g{7}; // OK }; }; class Boo { - static constexpr specialization_id h { 8 }; // ILLEGAL: not public member + static constexpr specialization_id h{8}; // ILLEGAL: not public member }; void Func() { - static constexpr specialization_id i { 9 }; // ILLEGAL: not at namespace - // or class scope + static constexpr specialization_id i{9}; // ILLEGAL: not at namespace + // or class scope /* ... */ } -constexpr specialization_id same_name { 10 }; // OK +constexpr specialization_id same_name{10}; // OK namespace foo { -constexpr specialization_id same_name { 11 }; // OK +constexpr specialization_id same_name{11}; // OK } namespace { -constexpr specialization_id same_name { 12 }; // OK +constexpr specialization_id same_name{12}; // OK } inline namespace other { -int same_name; // ILLEGAL: shadows "specialization_id" variable with same name in - // enclosing namespace scope +int same_name; // ILLEGAL: shadows "specialization_id" variable with same name + // in enclosing namespace scope } inline namespace other2 { -namespace foo { // ILLEGAL: namespace name shadows "::foo" namespace which contains - // "specialization_id" variable. -} // namespace foo -} // namespace +namespace foo { // ILLEGAL: namespace name shadows "::foo" namespace which + // contains "specialization_id" variable. +} // namespace foo +} // namespace other2 diff --git a/adoc/code/subbuffer.cpp b/adoc/code/subbuffer.cpp index 3a262f406..e5b52c52b 100644 --- a/adoc/code/subbuffer.cpp +++ b/adoc/code/subbuffer.cpp @@ -1,19 +1,19 @@ // Copyright (c) 2011-2024 The Khronos Group, Inc. // SPDX-License-Identifier: Apache-2.0 -buffer parent_buffer { range<2> { - 8, 8 } }; // Create 2-d buffer with 8x8 ints +buffer parent_buffer{ + range<2>{8, 8}}; // Create 2-d buffer with 8x8 ints // OK: Contiguous region from middle of buffer -buffer sub_buf1 { parent_buffer, /*offset*/ range<2> { 2, 0 }, - /*size*/ range<2> { 2, 8 } }; +buffer sub_buf1{parent_buffer, /*offset*/ range<2>{2, 0}, + /*size*/ range<2>{2, 8}}; // invalid exception: Non-contiguous regions of 2-d buffer -buffer sub_buf2 { parent_buffer, /*offset*/ range<2> { 2, 0 }, - /*size*/ range<2> { 2, 2 } }; -buffer sub_buf3 { parent_buffer, /*offset*/ range<2> { 2, 2 }, - /*size*/ range<2> { 2, 6 } }; +buffer sub_buf2{parent_buffer, /*offset*/ range<2>{2, 0}, + /*size*/ range<2>{2, 2}}; +buffer sub_buf3{parent_buffer, /*offset*/ range<2>{2, 2}, + /*size*/ range<2>{2, 6}}; // invalid exception: Out-of-bounds size -buffer sub_buf4 { parent_buffer, /*offset*/ range<2> { 2, 2 }, - /*size*/ range<2> { 2, 8 } }; +buffer sub_buf4{parent_buffer, /*offset*/ range<2>{2, 2}, + /*size*/ range<2>{2, 8}}; diff --git a/adoc/code/swizzle-example.cpp b/adoc/code/swizzle-example.cpp index 8be04fe85..9e10e5be3 100644 --- a/adoc/code/swizzle-example.cpp +++ b/adoc/code/swizzle-example.cpp @@ -2,15 +2,15 @@ // SPDX-License-Identifier: Apache-2.0 #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names int main() { vec v{5, 6, 7, 8}; - vec slice = v.swizzle<2,1>(); // slice has the value {7,6} - slice = v.swizzle<2,2>(); // slice is now {7,7} - int i = v.swizzle<2,1>()[1]; // i has the value 6 - v.swizzle<2>() = 0; // v is now {5,6,0,8} - v.swizzle<1>()++; // v is now {5,7,0,8} - v.swizzle<2,3>(); // Has no effect because result of swizzle is - // neither read nor modified + vec slice = v.swizzle<2, 1>(); // slice has the value {7,6} + slice = v.swizzle<2, 2>(); // slice is now {7,7} + int i = v.swizzle<2, 1>()[1]; // i has the value 6 + v.swizzle<2>() = 0; // v is now {5,6,0,8} + v.swizzle<1>()++; // v is now {5,7,0,8} + v.swizzle<2, 3>(); // Has no effect because result of swizzle is + // neither read nor modified } diff --git a/adoc/code/sycl-external.cpp b/adoc/code/sycl-external.cpp index 86350d156..3bc00f67d 100644 --- a/adoc/code/sycl-external.cpp +++ b/adoc/code/sycl-external.cpp @@ -5,8 +5,7 @@ SYCL_EXTERNAL void Foo(); -SYCL_EXTERNAL void Bar() { /* ... */ -} +SYCL_EXTERNAL void Bar() { /* ... */ } SYCL_EXTERNAL extern void Baz(); diff --git a/adoc/code/twoOptionalFeatures.cpp b/adoc/code/twoOptionalFeatures.cpp index 8f8428a62..6a95059b9 100644 --- a/adoc/code/twoOptionalFeatures.cpp +++ b/adoc/code/twoOptionalFeatures.cpp @@ -4,7 +4,7 @@ queue q1(dev1); if (dev1.has(aspect::fp16)) { q1.submit([&](handler& cgh) { - cgh.parallel_for(range { N }, [=](id i) { + cgh.parallel_for(range{N}, [=](id i) { half fpShort = 1.0; /* ... */ }); @@ -14,7 +14,7 @@ if (dev1.has(aspect::fp16)) { queue q2(dev2); if (dev2.has(aspect::atomic64)) { q2.submit([&](handler& cgh) { - cgh.parallel_for(range { N }, [=](id i) { + cgh.parallel_for(range{N}, [=](id i) { /* ... */ sycl::atomic_ref longAtomic(longValue); longAtomic.fetch_add(1); diff --git a/adoc/code/usingSpecConstants.cpp b/adoc/code/usingSpecConstants.cpp index df1d13ed9..60e23b66b 100644 --- a/adoc/code/usingSpecConstants.cpp +++ b/adoc/code/usingSpecConstants.cpp @@ -2,7 +2,7 @@ // SPDX-License-Identifier: Apache-2.0 #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names using coeff_t = std::array, 3>; @@ -16,8 +16,8 @@ void do_conv(buffer in, buffer out) { queue myQueue; myQueue.submit([&](handler& cgh) { - accessor in_acc { in, cgh, read_only }; - accessor out_acc { out, cgh, write_only }; + accessor in_acc{in, cgh, read_only}; + accessor out_acc{out, cgh, write_only}; // Set the coefficient of the convolution as constant. // This will build a specific kernel the coefficient available as literals. diff --git a/adoc/code/usm_device.cpp b/adoc/code/usm_device.cpp index 8e0d88c8e..71dedd5b5 100644 --- a/adoc/code/usm_device.cpp +++ b/adoc/code/usm_device.cpp @@ -3,7 +3,7 @@ #include #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names int main() { // Create a default queue to enqueue work to the default device @@ -15,7 +15,7 @@ int main() { myQueue.parallel_for(1024, [=](id<1> idx) { // Initialize each buffer element with its own rank number starting at 0 data[idx] = idx; - }); // End of the kernel function + }); // End of the kernel function // Explicitly wait for kernel execution since there is no accessor involved myQueue.wait(); diff --git a/adoc/code/usm_shared.cpp b/adoc/code/usm_shared.cpp index 84a1c5483..7b4eb3aa8 100644 --- a/adoc/code/usm_shared.cpp +++ b/adoc/code/usm_shared.cpp @@ -3,7 +3,7 @@ #include #include -using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names +using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names int main() { // Create a default queue to enqueue work to the default device @@ -17,7 +17,7 @@ int main() { myQueue.parallel_for(1024, [=](id<1> idx) { // Initialize each buffer element with its own rank number starting at 0 data[idx] = idx; - }); // End of the kernel function + }); // End of the kernel function // Explicitly wait for kernel execution since there is no accessor involved myQueue.wait();