Home
last modified time | relevance | path

Searched full:kernel (Results 1 – 25 of 1621) sorted by relevance

12345678910>>...65

/llvm-project/clang-tools-extra/test/clang-tidy/checkers/altera/
H A Dkernel-name-restriction.cpp1 // RUN: %check_clang_tidy %s altera-kernel-name-restriction %t -- -- -I%S/Inputs/kernel-name-restri…
2 …check_clang_tidy -check-suffix=UPPERCASE %s altera-kernel-name-restriction %t -- -- -I%S/Inputs/ke…
6 #include "KERNEL.cl"
7 …KERNEL.cl' may cause additional compilation errors due to the name of the kernel source file; cons…
9 …ion errors due to the name of the kernel source file; consider renaming the included kernel source…
11 …ion errors due to the name of the kernel source file; consider renaming the included kernel source…
13 // These are the banned kernel filenames, and should trigger warnings
14 #include "kernel.cl"
15 …kernel.cl' may cause additional compilation errors due to the name of the kernel source file; cons…
17 …ion errors due to the name of the kernel source file; consider renaming the included kernel source…
[all …]
/llvm-project/clang/test/SemaOpenCLCXX/
H A Dinvalid-kernel.clcpp9 kernel void m(); //expected-error{{kernel functions cannot be class members}}
13 //expected-error@+1{{'T' cannot be used as the type of a kernel parameter}}
14 kernel void templ(T par) { //expected-error{{kernel functions cannot be used in a template declarat…
18 kernel void bar(int par) { //expected-error{{kernel functions cannot be used in a template declarat…
21 kernel void foo(int); //expected-note{{previous declaration is here}}
23 kernel void foo(float); //expected-error{{conflicting types for 'foo'}}
25 kernel void int_v(int in);
26 kernel void int_p(__global int *in);
27 kernel void int_r(__global int &in);
28 kernel void int_p_p(__global int *__global *in);
[all …]
/llvm-project/llvm/test/CodeGen/X86/
H A Dx86-64-intrcc-uintr.ll4 ; RUN: llc -code-model=kernel < %s | FileCheck %s -check-prefixes=CHECK-KERNEL
5 ; RUN: llc -O0 -code-model=kernel < %s | FileCheck %s -check-prefixes=CHECK0-KERNEL
33 ; CHECK-KERNEL-LABEL: test_uintr_isr_cc_empty:
34 ; CHECK-KERNEL: # %bb.0: # %entry
35 ; CHECK-KERNEL-NEXT: pushq %rax
36 ; CHECK-KERNEL-NEXT: addq $16, %rsp
37 ; CHECK-KERNEL-NEXT: iretq
39 ; CHECK0-KERNEL-LABEL: test_uintr_isr_cc_empty:
40 ; CHECK0-KERNEL: # %bb.0: # %entry
41 ; CHECK0-KERNEL-NEXT: pushq %rax
[all …]
H A Dstack-protector.ll3 ; RUN: llc -code-model=kernel -mtriple=x86_64-pc-linux-gnu < %s -o - | FileCheck --check-prefix=LIN…
4 ; RUN: llc -code-model=kernel -mtriple=x86_64-unknown-freebsd < %s -o - | FileCheck --check-prefix=…
40 ; LINUX-KERNEL-X64-LABEL: test1a:
41 ; LINUX-KERNEL-X64-NOT: callq __stack_chk_fail
42 ; LINUX-KERNEL-X64: .cfi_endproc
79 ; FREEBSD-KERNEL-X64-LABEL: test1b:
80 ; FREEBSD-KERNEL-X64-NOT: mov{{l|q}} __stack_chk_guard@GOTPCREL
81 ; FREEBSD-KERNEL-X64: callq __stack_chk_fail
83 ; LINUX-KERNEL-X64-LABEL: test1b:
84 ; LINUX-KERNEL-X64: mov{{l|q}} %gs:
[all …]
/llvm-project/clang/test/SemaOpenCL/
H A Dinvalid-kernel-parameters.cl4 kernel void half_arg(half x) { } // expected-error{{declaring function parameter of type '__private…
9 // expected-error@+4{{kernel parameter cannot be declared as a pointer to a pointer}}
10 // expected-error@+4{{kernel parameter cannot be declared as a pointer to a pointer}}
11 // expected-error@+4{{kernel parameter cannot be declared as a pointer to a pointer}}
13 kernel void no_ptrptr(global int * global *i) { }
14 kernel void no_lptrcptr(constant int * local *i) { }
15 kernel void no_ptrptrptr(global int * global * global *i) { }
17 // expected-error@+1{{pointer arguments to kernel functions must reside in '__global', '__constant'…
20 // expected-error@+1{{pointer arguments to kernel functions must reside in '__global', '__constant'…
23 // expected-error@+1{{pointer arguments to kernel functions must reside in '__global', '__constant'…
[all …]
H A Damdgpu-attrs.cl3 …64 { // expected-error {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
7 …s_per_eu_2 { // expected-error {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
11 …per_eu_2_4 { // expected-error {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
15 …ct_num_sgpr_32 { // expected-error {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
19 …ct_num_vgpr_64 { // expected-error {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
24 …) {} // expected-error {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
25 …er_eu_2() {} // expected-error {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
26 …_eu_2_4() {} // expected-error {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
27 …num_sgpr_32() {} // expected-error {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
28 …num_vgpr_64() {} // expected-error {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
[all …]
H A Daccess-qualifier.cl43 kernel void k1(img1d_wo img) {
52 kernel void k2(img1d_ro img) {
61 kernel void k3(img1d_wo img) {
66 kernel void k4(img1d_rw img) {
71 kernel void k5(img1d_ro_default img) {
80 kernel void k6(img1d_ro img) {
84 kernel void k7(read_only img1d_wo img){} // expected-error {{multiple access qualifiers}}
86 kernel void k8(write_only img1d_ro_default img){} // expected-error {{multiple access qualifiers}}
88 kernel void k9(read_only int i){} // expected-error{{access qualifier can only be used for pipe and image type}}
90 kernel voi
[all...]
H A Dinvalid-kernel-attrs.cl3 kernel __attribute__((vec_type_hint)) void kernel1() {} //expected-error{{'vec_type_hint' attribute…
5 kernel __attribute__((vec_type_hint(not_type))) void kernel2() {} //expected-error{{unknown type na…
7 kernel __attribute__((vec_type_hint(void))) void kernel3() {} //expected-error{{a non-vector or non…
9 kernel __attribute__((vec_type_hint(bool))) void kernel4() {} //expected-error{{a non-vector or non…
11 kernel __attribute__((vec_type_hint(int))) __attribute__((vec_type_hint(float))) void kernel5() {} …
13 kernel __attribute__((work_group_size_hint(8,16,32,4))) void kernel6() {} //expected-error{{'work_g…
15 kernel __attribute__((work_group_size_hint(1,2,3))) __attribute__((work_group_size_hint(3,2,1))) vo…
17 …(){} // expected-error {{attribute 'reqd_work_group_size' can only be applied to an OpenCL kernel}}
19 …(){} // expected-error {{attribute 'work_group_size_hint' can only be applied to an OpenCL kernel}}
21 …ernel10(){} // expected-error {{attribute 'vec_type_hint' can only be applied to an OpenCL kernel}}
[all …]
/llvm-project/flang/test/Semantics/
H A Dreduce.cuf12 !$cuf kernel do <<<*,*>>> reduce (+:mr,ar)
14 !ERROR: !$CUF KERNEL DO REDUCE operation is not acceptable for a variable with type LOGICAL(4)
15 !$cuf kernel do <<<*,*>>> reduce (+:lr)
17 !$cuf kernel do <<<*,*>>> reduce (*:mr,ar)
19 !ERROR: !$CUF KERNEL DO REDUCE operation is not acceptable for a variable with type LOGICAL(4)
20 !$cuf kernel do <<<*,*>>> reduce (*:lr)
22 !$cuf kernel do <<<*,*>>> reduce (max:mr,ar)
24 !ERROR: !$CUF KERNEL DO REDUCE operation is not acceptable for a variable with type LOGICAL(4)
25 !$cuf kernel do <<<*,*>>> reduce (max:lr)
27 !$cuf kernel d
[all...]
H A Dcuf09.cuf9 !$cuf kernel do <<< 1, 2 >>>
92 !$cuf kernel do
94 !ERROR: Statement may not appear in cuf kernel code
98 !$cuf kernel do
101 !ERROR: Statement may not appear in cuf kernel code
105 !ERROR: Statement may not appear in cuf kernel code
115 !$cuf kernel do <<< *, * >>> ! ok
118 !$cuf kernel do <<< (*), (*) >>> ! ok
121 !$cuf kernel do <<< (1,*), (2,*) >>> ! ok
124 !ERROR: !$CUF KERNEL D
[all...]
/llvm-project/llvm/lib/Support/
H A DAMDGPUMetadata.cpp22 LLVM_YAML_IS_SEQUENCE_VECTOR(Kernel::Arg::Metadata)
23 LLVM_YAML_IS_SEQUENCE_VECTOR(Kernel::Metadata)
93 struct MappingTraits<Kernel::Attrs::Metadata> {
94 static void mapping(IO &YIO, Kernel::Attrs::Metadata &MD) { in mapping()
95 YIO.mapOptional(Kernel::Attrs::Key::ReqdWorkGroupSize, in mapping()
97 YIO.mapOptional(Kernel::Attrs::Key::WorkGroupSizeHint, in mapping()
99 YIO.mapOptional(Kernel::Attrs::Key::VecTypeHint, in mapping()
101 YIO.mapOptional(Kernel::Attrs::Key::RuntimeHandle, MD.mRuntimeHandle, in mapping()
107 struct MappingTraits<Kernel::Arg::Metadata> {
108 static void mapping(IO &YIO, Kernel::Arg::Metadata &MD) { in mapping()
[all …]
/llvm-project/polly/test/ScheduleOptimizer/
H A Dpattern-matching-based-opts_3.ll16 …rint<polly-ast>' -disable-output < %s 2>&1 | FileCheck %s --check-prefix=EXTRACTION-OF-MACRO-KERNEL
79 ; EXTRACTION-OF-MACRO-KERNEL-LABEL: :: isl ast :: kernel_gemm :: %bb8---%bb32
80 ; EXTRACTION-OF-MACRO-KERNEL: {
81 ; EXTRACTION-OF-MACRO-KERNEL-NEXT: // 1st level tiling - Tiles
82 ; EXTRACTION-OF-MACRO-KERNEL-NEXT: for (int c0 = 0; c0 <= 32; c0 += 1)
83 ; EXTRACTION-OF-MACRO-KERNEL-NEXT: for (int c1 = 0; c1 <= 32; c1 += 1) {
84 ; EXTRACTION-OF-MACRO-KERNEL-NEXT: // 1st level tiling - Points
85 ; EXTRACTION-OF-MACRO-KERNEL-NEXT: for (int c2 = 0; c2 <= 31; c2 += 1)
86 ; EXTRACTION-OF-MACRO-KERNEL-NEXT: for (int c3 = 0; c3 <= 31; c3 += 1)
87 ; EXTRACTION-OF-MACRO-KERNEL-NEXT: Stmt_bb9(32 * c0 + c2, 32 * c1 + c3);
[all …]
/llvm-project/clang/test/Index/
H A Dopencl-types.cl10 void kernel testFloatTypes() {
28 void kernel OCLImage1dROTest(read_only image1d_t scalarOCLImage1dRO);
29 void kernel OCLImage1dArrayROTest(read_only image1d_array_t scalarOCLImage1dArrayRO);
30 void kernel OCLImage1dBufferROTest(read_only image1d_buffer_t scalarOCLImage1dBufferRO);
31 void kernel OCLImage2dROTest(read_only image2d_t scalarOCLImage2dRO);
32 void kernel OCLImage2dArrayROTest(read_only image2d_array_t scalarOCLImage2dArrayRO);
33 void kernel OCLImage2dDepthROTest(read_only image2d_depth_t scalarOCLImage2dDepthRO);
34 void kernel OCLImage2dArrayDepthROTest(read_only image2d_array_depth_t scalarOCLImage2dArrayDepthRO…
35 void kernel OCLImage2dMSAAROTest(read_only image2d_msaa_t scalarOCLImage2dMSAARO);
36 void kernel OCLImage2dArrayMSAAROTest(read_only image2d_array_msaa_t scalarOCLImage2dArrayMSAARO);
[all …]
/llvm-project/llvm/test/MC/AMDGPU/
H A Dsym_kernel_scope.s3 .byte .kernel.sgpr_count
5 .byte .kernel.vgpr_count
9 .byte .kernel.sgpr_count
11 .byte .kernel.vgpr_count
16 .byte .kernel.sgpr_count
18 .byte .kernel.vgpr_count
22 .byte .kernel.sgpr_count
24 .byte .kernel.vgpr_count
28 .byte .kernel.sgpr_count
30 .byte .kernel.vgpr_count
[all …]
H A Dsym_kernel_scope_agpr.s5 .byte .kernel.agpr_count
7 .byte .kernel.vgpr_count
13 .byte .kernel.agpr_count
16 .byte .kernel.vgpr_count
22 .byte .kernel.agpr_count
24 .byte .kernel.vgpr_count
28 .byte .kernel.agpr_count
31 .byte .kernel.vgpr_count
36 .byte .kernel.agpr_count
38 .byte .kernel.vgpr_count
[all …]
/llvm-project/bolt/docs/
H A DOptimizingLinux.md1 # Optimizing Linux Kernel with BOLT
6 Many Linux applications spend a significant amount of their execution time in the kernel. Thus, when we consider code optimization for system performance, it is essential to improve the CPU utilization not only in the user-space applications and libraries but also in the kernel. BOLT has demonstrated double-digit gains while being applied to user-space programs. This guide shows how to apply BOLT to the x86-64 Linux kernel and enhance your system's performance. In our experiments, BOLT boosted database TPS by 2 percent when applied to the kernel compiled with the highest level optimizations, including PGO and LTO. The database spent ~40% of the time in the kernel and was quite sensitive to kernel performance.
10 While BOLT optimizations are not specific to the Linux kernel, certain quirks distinguish the kernel from user-level applications.
12 BOLT has been successfully applied to and tested with several flavors of the x86-64 Linux kernel
[all...]
/llvm-project/clang/test/Driver/
H A Dcl-zc.cpp27 // RUN: %clang_cl /c -### /kernel -- %s 2>&1 | FileCheck -check-prefixes=KERNEL-NO-RTTI,KERNEL-NO-EXCEPTIONS %s
28 // KERNEL-NO-RTTI: "-fno-rtti"
29 // KERNEL-NO-EXCEPTIONS-NOT: "-fcxx-exceptions" "-fexceptions"
31 // RUN: not %clang_cl /c -### --target=i686-pc-windows-msvc /kernel /arch:SSE -- %s 2>&1 | FileCheck -check-prefixes=KERNEL-SSE %s
32 // RUN: not %clang_cl /c -### --target=i686-pc-windows-msvc /kernel /arch:SSE2 -- %s 2>&1 | FileCheck -check-prefixes=KERNEL-SSE2 %s
33 // RUN: not %clang_cl /c -### --target=i686-pc-windows-msvc /kernel /arc
[all...]
/llvm-project/llvm/include/llvm/Support/
H A DAMDGPUMetadata.h118 // Kernel Metadata.
120 namespace Kernel {
123 // Kernel Attributes Metadata.
128 /// Key for Kernel::Attr::Metadata::mReqdWorkGroupSize.
130 /// Key for Kernel::Attr::Metadata::mWorkGroupSizeHint.
132 /// Key for Kernel::Attr::Metadata::mVecTypeHint.
134 /// Key for Kernel::Attr::Metadata::mRuntimeHandle.
138 /// In-memory representation of kernel attributes metadata.
146 /// External symbol created by runtime to store the kernel address
153 /// \returns True if kernel attributes metadata is empty, false otherwise.
[all …]
/llvm-project/llvm/test/CodeGen/AMDGPU/
H A Dlower-kernel-lds-constexpr.ll7 ; CHECK: %llvm.amdgcn.kernel.k0.lds.t = type { [2 x i8] }
8 ; CHECK: %llvm.amdgcn.kernel.k1.lds.t = type { [2 x i8] }
9 ; CHECK: %llvm.amdgcn.kernel.k2.lds.t = type { i32 }
10 ; CHECK: %llvm.amdgcn.kernel.k3.lds.t = type { [32 x i8] }
11 ; CHECK: %llvm.amdgcn.kernel.k4.lds.t = type { [2 x i8] }
12 ; CHECK: %llvm.amdgcn.kernel.k5.lds.t = type { [505 x i32] }
13 ; CHECK: %llvm.amdgcn.kernel.k6.lds.t = type { [4 x i32] }
17 ; CHECK: @llvm.amdgcn.kernel.k0.lds = internal addrspace(3) global %llvm.amdgcn.kernel.k0.lds.t poi…
18 ; CHECK: @llvm.amdgcn.kernel.k1.lds = internal addrspace(3) global %llvm.amdgcn.kernel.k1.lds.t poi…
19 ; CHECK: @llvm.amdgcn.kernel.k2.lds = internal addrspace(3) global %llvm.amdgcn.kernel.k2.lds.t poi…
[all …]
H A Dlower-kernel-and-module-lds.ll11 ; CHECK: %llvm.amdgcn.kernel.k0.lds.t = type { [16 x i8], [4 x i8], [2 x i8], [1 x i8] }
12 ; CHECK: %llvm.amdgcn.kernel.k1.lds.t = type { [16 x i8], [4 x i8], [2 x i8] }
13 ; CHECK: %llvm.amdgcn.kernel.k2.lds.t = type { [2 x i8] }
14 ; CHECK: %llvm.amdgcn.kernel.k3.lds.t = type { [4 x i8] }
19 ; CHECK: @llvm.amdgcn.kernel.k0.lds = internal addrspace(3) global %llvm.amdgcn.kernel.k0.lds.t poi…
20 ; CHECK: @llvm.amdgcn.kernel.k1.lds = internal addrspace(3) global %llvm.amdgcn.kernel.k1.lds.t poi…
21 ; CHECK: @llvm.amdgcn.kernel.k2.lds = internal addrspace(3) global %llvm.amdgcn.kernel.k2.lds.t poi…
22 ; CHECK: @llvm.amdgcn.kernel.k3.lds = internal addrspace(3) global %llvm.amdgcn.kernel.k3.lds.t poi…
26 …addrspace(3) getelementptr inbounds (%llvm.amdgcn.kernel.k0.lds.t, ptr addrspace(3) @llvm.amdgcn.k…
27 …addrspace(3) getelementptr inbounds (%llvm.amdgcn.kernel.k0.lds.t, ptr addrspace(3) @llvm.amdgcn.k…
[all …]
/llvm-project/llvm/test/Instrumentation/AddressSanitizer/
H A Dasan-optimize-callbacks.ll2 ; RUN: opt < %s -passes=asan -asan-instrumentation-with-call-threshold=0 -asan-optimize-callbacks --asan-kernel -S | FileCheck %s --check-prefixes=LOAD-KERNEL,STORE-KERNEL
24 ; LOAD-KERNEL: call void @llvm.asan.check.memaccess(ptr %p1, i32 1)
25 ; LOAD-KERNEL-NEXT: %n1 = load i8, ptr %p1, align 1
26 ; LOAD-KERNEL-NEXT: call void @llvm.asan.check.memaccess(ptr %p2, i32 3)
27 ; LOAD-KERNEL-NEXT: %n2 = load i16, ptr %p2, align 2
28 ; LOAD-KERNEL-NEXT: call void @llvm.asan.check.memaccess(ptr %p4, i32 5)
29 ; LOAD-KERNEL-NEXT: %n4 = load i32, ptr %p4, align 4
30 ; LOAD-KERNEL
[all...]
/llvm-project/clang/test/CodeGenOpenCL/
H A Damdgpu-attrs.cl6 kernel void flat_work_group_size_0_0() {}
8 kernel void waves_per_eu_0() {}
10 kernel void waves_per_eu_0_0() {}
12 kernel void num_sgpr0() {}
14 kernel void num_vgpr0() {}
17 kernel void flat_work_group_size_0_0_waves_per_eu_0() {}
19 kernel void flat_work_group_size_0_0_waves_per_eu_0_0() {}
21 kernel void flat_work_group_size_0_0_num_sgpr_0() {}
23 kernel void flat_work_group_size_0_0_num_vgpr_0() {}
25 kernel void waves_per_eu_0_num_sgpr_0() {}
[all …]
/llvm-project/mlir/test/Conversion/GPUToSPIRV/
H A Dreductions.mlir5 spirv.target_env = #spirv.target_env<#spirv.vce<v1.3, [Kernel, Addresses, Groups, GroupNonUniformArithmetic, GroupUniformArithmeticKHR], []>, #spirv.resource_limits<>>
11 gpu.func @test(%arg : f32) kernel
25 spirv.target_env = #spirv.target_env<#spirv.vce<v1.3, [Kernel, Addresses, Groups, GroupNonUniformArithmetic, GroupUniformArithmeticKHR], []>, #spirv.resource_limits<>>
31 gpu.func @test(%arg : f32) kernel
45 spirv.target_env = #spirv.target_env<#spirv.vce<v1.3, [Kernel, Addresses, Groups, GroupNonUniformArithmetic, GroupUniformArithmeticKHR], []>, #spirv.resource_limits<>>
51 gpu.func @test(%arg : i32) kernel
65 spirv.target_env = #spirv.target_env<#spirv.vce<v1.3, [Kernel, Addresses, Groups, GroupNonUniformArithmetic, GroupUniformArithmeticKHR], []>, #spirv.resource_limits<>>
71 gpu.func @test(%arg : i32) kernel
85 spirv.target_env = #spirv.target_env<#spirv.vce<v1.3, [Kernel, Addresses, Groups, GroupNonUniformArithmetic, GroupUniformArithmeticKHR], []>, #spirv.resource_limits<>>
91 gpu.func @test(%arg : f32) kernel
[all...]
/llvm-project/clang/test/SemaSYCL/
H A Dunique_stable_name.cpp12 // passed to a kernel. Test that passing the given function to the unique
13 // stable name builtin and then to the kernel throws an error because the
28 // then an empty lambda is passed to kernel.
30 // then passing a different lambda to the kernel still throws an error because
31 // the calling context is part of naming the kernel. Even though the given
32 // function (F2) is not passed to the kernel, its mangling changes due to
33 // kernel call with the unrelated lambda.
55 // because the kernel uses do not interfere with each other or invalidate in kernel3_4func()
79 // using the lambda in a way that does not contribute to the kernel name in main()
86 [=]() { l5(); }); // Used in the kernel, but not the kernel name itself in main()
[all …]
/llvm-project/clang/test/SemaCUDA/
H A Dlambda.cu13 __global__ void kernel(F f) { f(); } in kernel() function
14 // dev-note@-1 3{{called by 'kernel<(lambda}}
15 // warn-note@-2 5{{called by 'kernel<(lambda}}
27 kernel<<<1,1>>>([](){ hd(0); }); in test()
29 kernel<<<1,1>>>([=](){ hd(b); }); in test()
32 kernel<<<1,1>>>([&](){ hd(b); }); in test()
35 kernel<<<1,1>>>([&] __device__ (){ hd(b); }); in test()
38 kernel<<<1,1>>>([&](){ in test()
46 kernel<<<1,1>>>(lambda1); in test()
52 // com-error@-1 {{kernel function 'operator()' must be a free function or static member function}} in main()
[all …]

12345678910>>...65