From 1bec1d7494bcf1f8623c64371f1bbff2d900b218 Mon Sep 17 00:00:00 2001 From: Soumi Manna Date: Tue, 27 Apr 2021 07:06:36 -0700 Subject: [PATCH 1/4] [SYCL][NFC] Update tests This patch 1. updates CodeGen tests for disable_loop_pipelining and initiation_interval attributes. 2. rewrites Sema and CodeGen tests for scheduler-target-fmax-mhz attribute according to FE test guidelines and also separates AST and diagnostic parts for scheduler-target-fmax-mhz attribute to avoid unrelated changes on https://github.com/intel/llvm/pull/3601. Signed-off-by: Soumi Manna --- .../CodeGenSYCL/disable_loop_pipelining.cpp | 6 +- .../test/CodeGenSYCL/initiation_interval.cpp | 8 +- .../CodeGenSYCL/scheduler-target-fmax-mhz.cpp | 65 +++++++--- clang/test/SemaSYCL/initiation_interval.cpp | 6 +- .../SemaSYCL/scheduler_target_fmax_mhz.cpp | 117 ++++++++++-------- .../scheduler_target_fmax_mhz_ast.cpp | 110 ++++++++++++++++ ...ice-scheduler_target_fmax_mhz-template.cpp | 95 -------------- 7 files changed, 237 insertions(+), 170 deletions(-) create mode 100644 clang/test/SemaSYCL/scheduler_target_fmax_mhz_ast.cpp delete mode 100644 clang/test/SemaSYCL/sycl-device-scheduler_target_fmax_mhz-template.cpp diff --git a/clang/test/CodeGenSYCL/disable_loop_pipelining.cpp b/clang/test/CodeGenSYCL/disable_loop_pipelining.cpp index 33f83f6b1961f..97c4c754896b5 100644 --- a/clang/test/CodeGenSYCL/disable_loop_pipelining.cpp +++ b/clang/test/CodeGenSYCL/disable_loop_pipelining.cpp @@ -29,8 +29,8 @@ int main() { return 0; } -// CHECK: define dso_local spir_kernel void @"{{.*}}test_kernel1"() #0 !kernel_arg_buffer_location ![[NUM4:[0-9]+]] !disable_loop_pipelining ![[NUM5:[0-9]+]] -// CHECK: define dso_local spir_kernel void @"{{.*}}test_kernel2"() #0 !kernel_arg_buffer_location ![[NUM4]] -// CHECK: define dso_local spir_kernel void @"{{.*}}test_kernel3"() #0 !kernel_arg_buffer_location ![[NUM4]] !disable_loop_pipelining ![[NUM5]] +// CHECK: define dso_local spir_kernel void @"{{.*}}test_kernel1"() #0 {{.*}} !disable_loop_pipelining ![[NUM5:[0-9]+]] +// CHECK: define dso_local spir_kernel void @"{{.*}}test_kernel2"() #0 {{.*}} ![[NUM4:[0-9]+]] +// CHECK: define dso_local spir_kernel void @"{{.*}}test_kernel3"() #0 {{.*}} !disable_loop_pipelining ![[NUM5]] // CHECK: ![[NUM4]] = !{} // CHECK: ![[NUM5]] = !{i32 1} diff --git a/clang/test/CodeGenSYCL/initiation_interval.cpp b/clang/test/CodeGenSYCL/initiation_interval.cpp index 95ef9ce4cde50..0fa5699b32e99 100644 --- a/clang/test/CodeGenSYCL/initiation_interval.cpp +++ b/clang/test/CodeGenSYCL/initiation_interval.cpp @@ -39,10 +39,10 @@ int main() { return 0; } -// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name1"() #0 !kernel_arg_buffer_location ![[NUM0:[0-9]+]] !initiation_interval ![[NUM1:[0-9]+]] -// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name2"() #0 !kernel_arg_buffer_location ![[NUM0]] !initiation_interval ![[NUM42:[0-9]+]] -// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name3"() #0 !kernel_arg_buffer_location ![[NUM0]] !initiation_interval ![[NUM2:[0-9]+]] -// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name4"() #0 !kernel_arg_buffer_location ![[NUM0]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !initiation_interval ![[NUM1:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !initiation_interval ![[NUM42:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !initiation_interval ![[NUM2:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} ![[NUM0:[0-9]+]] // CHECK: ![[NUM0]] = !{} // CHECK: ![[NUM1]] = !{i32 1} // CHECK: ![[NUM42]] = !{i32 42} diff --git a/clang/test/CodeGenSYCL/scheduler-target-fmax-mhz.cpp b/clang/test/CodeGenSYCL/scheduler-target-fmax-mhz.cpp index 21c981220528e..715e2dd204459 100644 --- a/clang/test/CodeGenSYCL/scheduler-target-fmax-mhz.cpp +++ b/clang/test/CodeGenSYCL/scheduler-target-fmax-mhz.cpp @@ -1,25 +1,58 @@ -// RUN: %clang_cc1 -fsycl-is-device -disable-llvm-passes -triple spir64-unknown-unknown-sycldevice -emit-llvm -o - %s | FileCheck %s +// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -sycl-std=2020 -emit-llvm -o - %s | FileCheck %s -#include "Inputs/sycl.hpp" -[[intel::scheduler_target_fmax_mhz(5)]] void -func() {} +#include "sycl.hpp" + +using namespace cl::sycl; +queue q; + +class Foo { +public: + [[intel::scheduler_target_fmax_mhz(5)]] void operator()() const {} +}; + +template +class Functor { +public: + [[intel::scheduler_target_fmax_mhz(N)]] void operator()() const {} +}; template [[intel::scheduler_target_fmax_mhz(N)]] void zoo() {} +[[intel::scheduler_target_fmax_mhz(2)]] void bar() {} + int main() { - cl::sycl::kernel_single_task( - []() [[intel::scheduler_target_fmax_mhz(2)]]{}); + q.submit([&](handler &h) { + // Test attribute argument size. + Foo boo; + h.single_task(boo); + + // Test attribute is applied on lambda. + h.single_task( + []() [[intel::scheduler_target_fmax_mhz(42)]]{}); - cl::sycl::kernel_single_task( - []() { func(); }); + // Test class template argument. + Functor<7> f; + h.single_task(f); - cl::sycl::kernel_single_task( - []() { zoo<75>(); }); + // Test attribute is propagated. + h.single_task( + []() { bar(); }); + + // Test function template argument. + h.single_task( + []() { zoo<75>(); }); + }); + return 0; } -// CHECK: define {{.*}}spir_kernel void @{{.*}}test_kernel1() {{.*}} !scheduler_target_fmax_mhz ![[PARAM1:[0-9]+]] -// CHECK: define {{.*}}spir_kernel void @{{.*}}test_kernel2() {{.*}} !scheduler_target_fmax_mhz ![[PARAM2:[0-9]+]] -// CHECK: define {{.*}}spir_kernel void @{{.*}}test_kernel3() {{.*}} !scheduler_target_fmax_mhz ![[PARAM3:[0-9]+]] -// CHECK: ![[PARAM1]] = !{i32 2} -// CHECK: ![[PARAM2]] = !{i32 5} -// CHECK: ![[PARAM3]] = !{i32 75} + +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !scheduler_target_fmax_mhz ![[NUM5:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !scheduler_target_fmax_mhz ![[NUM42:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !scheduler_target_fmax_mhz ![[NUM7:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !scheduler_target_fmax_mhz ![[NUM2:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name5"() #0 {{.*}} !scheduler_target_fmax_mhz ![[NUM75:[0-9]+]] +// CHECK: ![[NUM5]] = !{i32 5} +// CHECK: ![[NUM42]] = !{i32 42} +// CHECK: ![[NUM7]] = !{i32 7} +// CHECK: ![[NUM2]] = !{i32 2} +// CHECK: ![[NUM75]] = !{i32 75} diff --git a/clang/test/SemaSYCL/initiation_interval.cpp b/clang/test/SemaSYCL/initiation_interval.cpp index e70360828276d..9cca7b43b9686 100644 --- a/clang/test/SemaSYCL/initiation_interval.cpp +++ b/clang/test/SemaSYCL/initiation_interval.cpp @@ -1,8 +1,10 @@ // RUN: %clang_cc1 -fsycl-is-device -verify %s -// Test that checks disable_loop_pipelining attribute support on Function. +// Test that checks initiation_interval attribute support on function. // Tests for incorrect argument values for Intel FPGA initiation_interval function attribute. +[[intel::initiation_interval]] void one() {} // expected-error {{'initiation_interval' attribute takes one argument}} + [[intel::initiation_interval(5)]] int a; // expected-error{{'initiation_interval' attribute only applies to 'for', 'while', 'do' statements, and functions}} [[intel::initiation_interval("foo")]] void func() {} // expected-error{{integral constant expression must have integral or unscoped enumeration type, not 'const char [4]'}} @@ -27,7 +29,7 @@ [[intel::initiation_interval(1)]] void func6(); // expected-note {{previous attribute is here}} [[intel::initiation_interval(3)]] void func6(); // expected-warning {{attribute 'initiation_interval' is already applied with different arguments}} -// Tests for Intel FPGA loop fusion function attributes compatibility +// Tests for Intel FPGA initiation_interval and disable_loop_pipelining attributes compatibility checks. // expected-error@+2 {{'initiation_interval' and 'disable_loop_pipelining' attributes are not compatible}} // expected-note@+1 {{conflicting attribute is here}} [[intel::disable_loop_pipelining]] [[intel::initiation_interval(2)]] void func7(); diff --git a/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp b/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp index 984ceacd8b67e..636d36422b699 100644 --- a/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp +++ b/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp @@ -1,60 +1,77 @@ -// RUN: %clang_cc1 %s -fsyntax-only -ast-dump -fsycl-is-device -triple spir64 -Wno-sycl-2017-compat -verify | FileCheck %s +// RUN: %clang_cc1 -fsycl-is-device -verify %s -#include "Inputs/sycl.hpp" +// Test that checks scheduler_target_fmax_mhz attribute support on Function. + +// Test for deprecated spelling of scheduler_target_fmax_mhz attribute. // expected-warning@+2 {{attribute 'intelfpga::scheduler_target_fmax_mhz' is deprecated}} // expected-note@+1 {{did you mean to use 'intel::scheduler_target_fmax_mhz' instead?}} -[[intelfpga::scheduler_target_fmax_mhz(2)]] void -func() {} +[[intelfpga::scheduler_target_fmax_mhz(2)]] void deprecate() {} + +// Tests for incorrect argument values for Intel FPGA scheduler_target_fmax_mhz function attribute. +[[intel::scheduler_target_fmax_mhz(0)]] int Var = 0; // expected-error{{'scheduler_target_fmax_mhz' attribute only applies to functions}} + +[[intel::scheduler_target_fmax_mhz(1048577)]] void correct() {} // OK + +[[intel::scheduler_target_fmax_mhz("foo")]] void func() {} // expected-error{{integral constant expression must have integral or unscoped enumeration type, not 'const char [4]'}} + +[[intel::scheduler_target_fmax_mhz(-1)]] void func1() {} // expected-error{{'scheduler_target_fmax_mhz' attribute requires a non-negative integral compile time constant expression}} + +[[intel::scheduler_target_fmax_mhz(0, 1)]] void func2() {} // expected-error{{'scheduler_target_fmax_mhz' attribute takes one argument}} + +// Tests for Intel FPGA scheduler_target_fmax_mhz function attribute duplication. +// No diagnostic is emitted because the arguments match. Duplicate attribute is silently ignored. +[[intel::scheduler_target_fmax_mhz(2)]] +[[intel::scheduler_target_fmax_mhz(2)]] void func3() {} // No diagnostic is emitted because the arguments match. -[[intel::scheduler_target_fmax_mhz(12)]] void bar(); -[[intel::scheduler_target_fmax_mhz(12)]] void bar() {} // OK +[[intel::scheduler_target_fmax_mhz(4)]] void func4(); +[[intel::scheduler_target_fmax_mhz(4)]] void func4(); // OK // Diagnostic is emitted because the arguments mismatch. -[[intel::scheduler_target_fmax_mhz(12)]] void baz(); // expected-note {{previous attribute is here}} -[[intel::scheduler_target_fmax_mhz(100)]] void baz(); // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} +[[intel::scheduler_target_fmax_mhz(2)]] // expected-note {{previous attribute is here}} +[[intel::scheduler_target_fmax_mhz(4)]] void func5() {} // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} + +[[intel::scheduler_target_fmax_mhz(1)]] void func6(); // expected-note {{previous attribute is here}} +[[intel::scheduler_target_fmax_mhz(3)]] void func6(); // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} +// Tests that check template parameter support for Intel FPGA scheduler_target_fmax_mhz function attributes. template -[[intel::scheduler_target_fmax_mhz(N)]] void zoo() {} - -int main() { - // CHECK-LABEL: FunctionDecl {{.*}}test_kernel1 'void ()' - // CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} - // CHECK-NEXT: ConstantExpr {{.*}} 'int' - // CHECK-NEXT: value: Int 5 - // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 5 - // expected-warning@+3 {{attribute 'intelfpga::scheduler_target_fmax_mhz' is deprecated}} - // expected-note@+2 {{did you mean to use 'intel::scheduler_target_fmax_mhz' instead?}} - cl::sycl::kernel_single_task( - []() [[intelfpga::scheduler_target_fmax_mhz(5)]]{}); - - // CHECK-LABEL: FunctionDecl {{.*}}test_kernel2 'void ()' - // CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} - // CHECK-NEXT: ConstantExpr {{.*}} 'int' - // CHECK-NEXT: value: Int 2 - // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 2 - cl::sycl::kernel_single_task( - []() { func(); }); - - // CHECK-LABEL: FunctionDecl {{.*}}test_kernel3 'void ()' - // CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} - // CHECK-NEXT: ConstantExpr {{.*}} 'int' - // CHECK-NEXT: value: Int 75 - // CHECK-NEXT: SubstNonTypeTemplateParmExpr {{.*}} 'int' - // CHECK-NEXT: NonTypeTemplateParmDecl {{.*}} referenced 'int' depth 0 index 0 N - // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 75 - cl::sycl::kernel_single_task( - []() { zoo<75>(); }); - - [[intel::scheduler_target_fmax_mhz(0)]] int Var = 0; // expected-error{{'scheduler_target_fmax_mhz' attribute only applies to functions}} - - cl::sycl::kernel_single_task( - []() [[intel::scheduler_target_fmax_mhz(1048577)]]{}); // OK - - cl::sycl::kernel_single_task( - []() [[intel::scheduler_target_fmax_mhz(-4)]]{}); // expected-error{{'scheduler_target_fmax_mhz' attribute requires a non-negative integral compile time constant expression}} - - cl::sycl::kernel_single_task( - []() [[intel::scheduler_target_fmax_mhz(1), // expected-note {{previous attribute is here}} - intel::scheduler_target_fmax_mhz(2)]]{}); // expected-warning{{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} +[[intel::scheduler_target_fmax_mhz(N)]] void func7(); // expected-error {{'scheduler_target_fmax_mhz' attribute requires a non-negative integral compile time constant expression}} + +template +[[intel::scheduler_target_fmax_mhz(10)]] void func8(); // expected-note {{previous attribute is here}} +template +[[intel::scheduler_target_fmax_mhz(size)]] void func8() {} // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} + +void checkTemplates() { + func7<4>(); // OK + func7<-1>(); // expected-note {{in instantiation of function template specialization 'func7<-1>' requested here}} + func7<0>(); // OK + func8<20>(); // expected-note {{in instantiation of function template specialization 'func8<20>' requested here}} +} + +// Test that checks expression is not a constant expression. +// expected-note@+1{{declared here}} +int baz(); +// expected-error@+2{{expression is not an integral constant expression}} +// expected-note@+1{{non-constexpr function 'baz' cannot be used in a constant expression}} +[[intel::scheduler_target_fmax_mhz(baz() + 1)]] void func9(); + +// Test that checks expression is a constant expression. +constexpr int bar() { return 0; } +[[intel::scheduler_target_fmax_mhz(bar() + 2)]] void func10(); // OK + +// Test that checks wrong function template instantiation and ensures that the type +// is checked properly when instantiating from the template definition. +template +// expected-error@+2 {{integral constant expression must have integral or unscoped enumeration type, not 'S'}} +// expected-error@+1 {{integral constant expression must have integral or unscoped enumeration type, not 'float'}} +[[intel::scheduler_target_fmax_mhz(Ty{})]] void func11() {} + +struct S {}; +void test() { + //expected-note@+1{{in instantiation of function template specialization 'func11' requested here}} + func11(); + //expected-note@+1{{in instantiation of function template specialization 'func11' requested here}} + func11(); } diff --git a/clang/test/SemaSYCL/scheduler_target_fmax_mhz_ast.cpp b/clang/test/SemaSYCL/scheduler_target_fmax_mhz_ast.cpp new file mode 100644 index 0000000000000..8f13388e52097 --- /dev/null +++ b/clang/test/SemaSYCL/scheduler_target_fmax_mhz_ast.cpp @@ -0,0 +1,110 @@ +// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -sycl-std=2017 -ast-dump %s | FileCheck %s + +// Tests for AST of Intel FPGA scheduler_target_fmax_mhz function attribute. +#include "sycl.hpp" + +sycl::queue deviceQueue; + +// CHECK: FunctionDecl {{.*}} func1 'void ()' +// CHECK-NEXT: CompoundStmt +// CHECK-NEXT: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} +// CHECK-NEXT: ConstantExpr{{.*}}'int' +// CHECK-NEXT: value: Int 4 +// CHECK-NEXT: IntegerLiteral {{.*}} 'int' 4 +[[intel::scheduler_target_fmax_mhz(4)]] void func1() {} + +// Test that checks template parameter support on function. +// CHECK: FunctionTemplateDecl {{.*}} func2 +// CHECK: FunctionDecl {{.*}} func2 'void ()' +// CHECK-NEXT: CompoundStmt +// CHECK_NEXT: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} +// CHECK_NEXT: DeclRefExpr {{.*}} 'int' NonTypeTemplateParm {{.*}} 'N' 'int' +// CHECK: FunctionDecl {{.*}} func2 'void ()' +// CHECK-NEXT: TemplateArgument integral 6 +// CHECK-NEXT: CompoundStmt +// CHECK-NEXT: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} +// CHECK-NEXT: ConstantExpr{{.*}}'int' +// CHECK-NEXT: value: Int 6 +// CHECK-NEXT: SubstNonTypeTemplateParmExpr +// CHECK-NEXT: NonTypeTemplateParmDecl +// CHECK-NEXT: IntegerLiteral {{.*}} 'int' 6 +template +[[intel::scheduler_target_fmax_mhz(N)]] void func2() {} + +template +[[intel::scheduler_target_fmax_mhz(N)]] void func3() {} + +// No diagnostic is emitted because the arguments match. Duplicate attribute is silently ignored. +// CHECK: FunctionDecl {{.*}} {{.*}} func4 'void ()' +// CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} +// CHECK-NEXT: ConstantExpr {{.*}} 'int' +// CHECK-NEXT: value: Int 10 +// CHECK-NEXT: IntegerLiteral{{.*}}10{{$}} +[[intel::scheduler_target_fmax_mhz(10)]] +[[intel::scheduler_target_fmax_mhz(10)]] void func4() {} + +class KernelFunctor { +public: + void operator()() const { + func1(); + } +}; + +// Test that checks template parameter support on class member function. +template +class KernelFunctor2 { +public: + [[intel::scheduler_target_fmax_mhz(N)]] void operator()() const { + } +}; + +int main() { + deviceQueue.submit([&](sycl::handler &h) { + // CHECK-LABEL: FunctionDecl {{.*}}kernel_name_1 + // CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr + KernelFunctor f1; + h.single_task(f1); + + // CHECK-LABEL: FunctionDecl {{.*}}kernel_name_2 + // CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} + // CHECK-NEXT: ConstantExpr{{.*}}'int' + // CHECK-NEXT: value: Int 3 + // CHECK-NEXT: SubstNonTypeTemplateParmExpr + // CHECK-NEXT: NonTypeTemplateParmDecl + // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 3 + KernelFunctor2<3> f2; + h.single_task(f2); + + // CHECK-LABEL: FunctionDecl {{.*}}kernel_name_3 + // CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} + // CHECK-NEXT: ConstantExpr{{.*}}'int' + // CHECK-NEXT: value: Int 4 + // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 4 + h.single_task( + []() [[intel::scheduler_target_fmax_mhz(4)]]{}); + + // CHECK-LABEL: FunctionDecl {{.*}}kernel_name_4 + // CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} + // CHECK-NEXT: ConstantExpr {{.*}} 'int' + // CHECK-NEXT: value: Int 75 + // CHECK-NEXT: SubstNonTypeTemplateParmExpr {{.*}} 'int' + // CHECK-NEXT: NonTypeTemplateParmDecl {{.*}} referenced 'int' depth 0 index 0 N + // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 75 + h.single_task( + []() { func3<75>(); }); + + // Ignore duplicate attribute. + h.single_task( + // CHECK-LABEL: FunctionDecl {{.*}}kernel_name_5 + // CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} + // CHECK-NEXT: ConstantExpr {{.*}} 'int' + // CHECK-NEXT: value: Int 6 + // CHECK-NEXT: IntegerLiteral{{.*}}6{{$}} + []() [[intel::scheduler_target_fmax_mhz(6), + intel::scheduler_target_fmax_mhz(6)]]{}); + }); + + func2<6>(); + + return 0; +} diff --git a/clang/test/SemaSYCL/sycl-device-scheduler_target_fmax_mhz-template.cpp b/clang/test/SemaSYCL/sycl-device-scheduler_target_fmax_mhz-template.cpp deleted file mode 100644 index 407488f9bb0ae..0000000000000 --- a/clang/test/SemaSYCL/sycl-device-scheduler_target_fmax_mhz-template.cpp +++ /dev/null @@ -1,95 +0,0 @@ -// RUN: %clang_cc1 -fsycl-is-device -fsyntax-only -ast-dump -verify -pedantic %s | FileCheck %s - -// Test that checkes template parameter support for 'scheduler_target_fmax_mhz' attribute on sycl device. - -// Test that checks wrong function template instantiation and ensures that the type -// is checked properly when instantiating from the template definition. -template -// expected-error@+2 {{integral constant expression must have integral or unscoped enumeration type, not 'S'}} -// expected-error@+1 {{integral constant expression must have integral or unscoped enumeration type, not 'float'}} -[[intel::scheduler_target_fmax_mhz(Ty{})]] void func() {} - -struct S {}; -void test() { - //expected-note@+1{{in instantiation of function template specialization 'func' requested here}} - func(); - //expected-note@+1{{in instantiation of function template specialization 'func' requested here}} - func(); -} - -// Test that checks expression is not a constant expression. -// expected-note@+1{{declared here}} -int foo(); -// expected-error@+2{{expression is not an integral constant expression}} -// expected-note@+1{{non-constexpr function 'foo' cannot be used in a constant expression}} -[[intel::scheduler_target_fmax_mhz(foo() + 1)]] void func1(); - -// Test that checks expression is a constant expression. -constexpr int bar() { return 0; } -[[intel::scheduler_target_fmax_mhz(bar() + 2)]] void func2(); // OK - -// Test that checks template parameter support on member function of class template. -template -class KernelFunctor { -public: - // expected-error@+1{{'scheduler_target_fmax_mhz' attribute requires a non-negative integral compile time constant expression}} - [[intel::scheduler_target_fmax_mhz(SIZE)]] void operator()() {} -}; - -int main() { - //expected-note@+1{{in instantiation of template class 'KernelFunctor<-1>' requested here}} - KernelFunctor<-1>(); - // no error expected - KernelFunctor<2>(); -} - -// CHECK: ClassTemplateDecl {{.*}} {{.*}} KernelFunctor -// CHECK: ClassTemplateSpecializationDecl {{.*}} {{.*}} class KernelFunctor definition -// CHECK: TemplateArgument integral 2 -// CHECK: CXXRecordDecl {{.*}} {{.*}} implicit class KernelFunctor -// CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} -// CHECK-NEXT: ConstantExpr {{.*}} 'int' -// CHECK-NEXT: value: Int 2 -// CHECK-NEXT: SubstNonTypeTemplateParmExpr {{.*}} -// CHECK-NEXT: NonTypeTemplateParmDecl {{.*}} -// CHECK-NEXT: IntegerLiteral{{.*}}2{{$}} - -// Test that checks template parameter support on function. -template -// expected-error@+1{{'scheduler_target_fmax_mhz' attribute requires a non-negative integral compile time constant expression}} -[[intel::scheduler_target_fmax_mhz(N)]] void func3() {} - -template -[[intel::scheduler_target_fmax_mhz(4)]] void func4(); // expected-note {{previous attribute is here}} - -template -[[intel::scheduler_target_fmax_mhz(N)]] void func4() {} // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} - -int check() { - // no error expected - func3<3>(); - //expected-note@+1{{in instantiation of function template specialization 'func3<-1>' requested here}} - func3<-1>(); - //expected-note@+1 {{in instantiation of function template specialization 'func4<6>' requested here}} - func4<6>(); - return 0; -} - -// No diagnostic is emitted because the arguments match. Duplicate attribute is silently ignored. -[[intel::scheduler_target_fmax_mhz(8)]] -[[intel::scheduler_target_fmax_mhz(8)]] void func5() {} - -// CHECK: FunctionDecl {{.*}} {{.*}} func3 'void ()' -// CHECK: TemplateArgument integral 3 -// CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} -// CHECK-NEXT: ConstantExpr {{.*}} 'int' -// CHECK-NEXT: value: Int 3 -// CHECK-NEXT: SubstNonTypeTemplateParmExpr {{.*}} -// CHECK-NEXT: NonTypeTemplateParmDecl {{.*}} -// CHECK-NEXT: IntegerLiteral{{.*}}3{{$}} - -// CHECK: FunctionDecl {{.*}} {{.*}} func5 'void ()' -// CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} -// CHECK-NEXT: ConstantExpr {{.*}} 'int' -// CHECK-NEXT: value: Int 8 -// CHECK-NEXT: IntegerLiteral{{.*}}8{{$}} From 05fe9bbbd901a0e33bc76e0312377c8163db4f09 Mon Sep 17 00:00:00 2001 From: Soumi Manna Date: Tue, 27 Apr 2021 07:22:45 -0700 Subject: [PATCH 2/4] fix clang format issues Signed-off-by: Soumi Manna --- clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp | 12 +++++++----- .../test/SemaSYCL/scheduler_target_fmax_mhz_ast.cpp | 13 +++++++------ 2 files changed, 14 insertions(+), 11 deletions(-) diff --git a/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp b/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp index 636d36422b699..575749f1b5c98 100644 --- a/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp +++ b/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp @@ -10,7 +10,7 @@ // Tests for incorrect argument values for Intel FPGA scheduler_target_fmax_mhz function attribute. [[intel::scheduler_target_fmax_mhz(0)]] int Var = 0; // expected-error{{'scheduler_target_fmax_mhz' attribute only applies to functions}} -[[intel::scheduler_target_fmax_mhz(1048577)]] void correct() {} // OK +[[intel::scheduler_target_fmax_mhz(1048577)]] void correct() {} // OK [[intel::scheduler_target_fmax_mhz("foo")]] void func() {} // expected-error{{integral constant expression must have integral or unscoped enumeration type, not 'const char [4]'}} @@ -21,15 +21,17 @@ // Tests for Intel FPGA scheduler_target_fmax_mhz function attribute duplication. // No diagnostic is emitted because the arguments match. Duplicate attribute is silently ignored. [[intel::scheduler_target_fmax_mhz(2)]] -[[intel::scheduler_target_fmax_mhz(2)]] void func3() {} +[[intel::scheduler_target_fmax_mhz(2)]] void +func3() {} // No diagnostic is emitted because the arguments match. [[intel::scheduler_target_fmax_mhz(4)]] void func4(); [[intel::scheduler_target_fmax_mhz(4)]] void func4(); // OK // Diagnostic is emitted because the arguments mismatch. -[[intel::scheduler_target_fmax_mhz(2)]] // expected-note {{previous attribute is here}} -[[intel::scheduler_target_fmax_mhz(4)]] void func5() {} // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} +[[intel::scheduler_target_fmax_mhz(2)]] // expected-note {{previous attribute is here}} +[[intel::scheduler_target_fmax_mhz(4)]] void +func5() {} // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} [[intel::scheduler_target_fmax_mhz(1)]] void func6(); // expected-note {{previous attribute is here}} [[intel::scheduler_target_fmax_mhz(3)]] void func6(); // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} @@ -39,7 +41,7 @@ template [[intel::scheduler_target_fmax_mhz(N)]] void func7(); // expected-error {{'scheduler_target_fmax_mhz' attribute requires a non-negative integral compile time constant expression}} template -[[intel::scheduler_target_fmax_mhz(10)]] void func8(); // expected-note {{previous attribute is here}} +[[intel::scheduler_target_fmax_mhz(10)]] void func8(); // expected-note {{previous attribute is here}} template [[intel::scheduler_target_fmax_mhz(size)]] void func8() {} // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} diff --git a/clang/test/SemaSYCL/scheduler_target_fmax_mhz_ast.cpp b/clang/test/SemaSYCL/scheduler_target_fmax_mhz_ast.cpp index 8f13388e52097..eca66915670f3 100644 --- a/clang/test/SemaSYCL/scheduler_target_fmax_mhz_ast.cpp +++ b/clang/test/SemaSYCL/scheduler_target_fmax_mhz_ast.cpp @@ -41,7 +41,8 @@ template // CHECK-NEXT: value: Int 10 // CHECK-NEXT: IntegerLiteral{{.*}}10{{$}} [[intel::scheduler_target_fmax_mhz(10)]] -[[intel::scheduler_target_fmax_mhz(10)]] void func4() {} +[[intel::scheduler_target_fmax_mhz(10)]] void +func4() {} class KernelFunctor { public: @@ -95,11 +96,11 @@ int main() { // Ignore duplicate attribute. h.single_task( - // CHECK-LABEL: FunctionDecl {{.*}}kernel_name_5 - // CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} - // CHECK-NEXT: ConstantExpr {{.*}} 'int' - // CHECK-NEXT: value: Int 6 - // CHECK-NEXT: IntegerLiteral{{.*}}6{{$}} + // CHECK-LABEL: FunctionDecl {{.*}}kernel_name_5 + // CHECK: SYCLIntelSchedulerTargetFmaxMhzAttr {{.*}} + // CHECK-NEXT: ConstantExpr {{.*}} 'int' + // CHECK-NEXT: value: Int 6 + // CHECK-NEXT: IntegerLiteral{{.*}}6{{$}} []() [[intel::scheduler_target_fmax_mhz(6), intel::scheduler_target_fmax_mhz(6)]]{}); }); From 4f0d9e320bb73e2007f5146162d92fa6a6ea5b2f Mon Sep 17 00:00:00 2001 From: Soumi Manna Date: Tue, 27 Apr 2021 07:31:10 -0700 Subject: [PATCH 3/4] fix test Signed-off-by: Soumi Manna --- clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp | 3 +-- 1 file changed, 1 insertion(+), 2 deletions(-) diff --git a/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp b/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp index 575749f1b5c98..349958ae857be 100644 --- a/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp +++ b/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp @@ -30,8 +30,7 @@ func3() {} // Diagnostic is emitted because the arguments mismatch. [[intel::scheduler_target_fmax_mhz(2)]] // expected-note {{previous attribute is here}} -[[intel::scheduler_target_fmax_mhz(4)]] void -func5() {} // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} +[[intel::scheduler_target_fmax_mhz(4)]] void func5() {} // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} [[intel::scheduler_target_fmax_mhz(1)]] void func6(); // expected-note {{previous attribute is here}} [[intel::scheduler_target_fmax_mhz(3)]] void func6(); // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} From 609d517dca687868eb53f82d19b990924793a2bf Mon Sep 17 00:00:00 2001 From: Soumi Manna Date: Tue, 27 Apr 2021 07:35:48 -0700 Subject: [PATCH 4/4] fix test Signed-off-by: Soumi Manna --- clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp | 3 ++- 1 file changed, 2 insertions(+), 1 deletion(-) diff --git a/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp b/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp index 349958ae857be..79e1328c7db81 100644 --- a/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp +++ b/clang/test/SemaSYCL/scheduler_target_fmax_mhz.cpp @@ -30,7 +30,8 @@ func3() {} // Diagnostic is emitted because the arguments mismatch. [[intel::scheduler_target_fmax_mhz(2)]] // expected-note {{previous attribute is here}} -[[intel::scheduler_target_fmax_mhz(4)]] void func5() {} // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} +[[intel::scheduler_target_fmax_mhz(4)]] void +func5() {} // expected-warning@-1 {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}} [[intel::scheduler_target_fmax_mhz(1)]] void func6(); // expected-note {{previous attribute is here}} [[intel::scheduler_target_fmax_mhz(3)]] void func6(); // expected-warning {{attribute 'scheduler_target_fmax_mhz' is already applied with different arguments}}