<html>
<head>
<meta http-equiv="Content-Type" content="text/html; charset=iso-8859-1">
<style type="text/css" style="display:none;"><!-- P {margin-top:0;margin-bottom:0;} --></style>
</head>
<body dir="ltr">
<div id="divtagdefaultwrapper" style="font-size: 12pt; color: rgb(0, 0, 0); font-family: Calibri,Helvetica,sans-serif,"EmojiFont","Apple Color Emoji","Segoe UI Emoji",NotoColorEmoji,"Segoe UI Symbol","Android Emoji",EmojiSymbols;" dir="ltr">
<p style="margin-top:0;margin-bottom:0">Hi Alexey,</p>
<p style="margin-top:0;margin-bottom:0"><br>
</p>
<p style="margin-top:0;margin-bottom:0">I understand that avoiding adding extra diagnostics is desirable, but I am not sure it's a good idea to accept and generate code that is not going to work. Variadic functions are currently not expected to be supported
by all OpenCL targets hence we disallow this in the spec explicitly. <br>
</p>
<p style="margin-top:0;margin-bottom:0"><br>
</p>
<p style="margin-top:0;margin-bottom:0">I would be inclined to explicitly reject the code by default, unless there is a valid way to handle the generated bitcast of variadic prototype to non-variadic one? Then we could look at updating the spec and explain
that this is the difference in implicit function prototypes for OpenCL.</p>
<p style="margin-top:0;margin-bottom:0"><br>
</p>
<p style="margin-top:0;margin-bottom:0"></p>
<div>One compromise could be to try to set "<font size="2"><span style="font-size:11pt;">-Werror=implicit-function-declaration</span></font>" by default in the OpenCL mode? Not sure if it's easier/better than an extra diagnostic though.</div>
<br>
<p style="margin-top:0;margin-bottom:0">Cheers,</p>
<p style="margin-top:0;margin-bottom:0">Anastasia<br>
</p>
<br>
<br>
<div style="color: rgb(0, 0, 0);">
<hr style="display:inline-block;width:98%" tabindex="-1">
<div id="divRplyFwdMsg" dir="ltr"><font style="font-size:11pt" color="#000000" face="Calibri, sans-serif"><b>From:</b> aleksey.bader@gmail.com <aleksey.bader@gmail.com><br>
<b>Sent:</b> 21 August 2018 16:11<br>
<b>To:</b> Anastasia Stulova<br>
<b>Cc:</b> Richard Smith; cfe-commits; nd<br>
<b>Subject:</b> Re: r314872 - We allow implicit function declarations as an extension in all C dialects. Remove OpenCL special case.</font>
<div> </div>
</div>
<div class="BodyFragment"><font size="2"><span style="font-size:11pt;">
<div class="PlainText">Hi Anastasia,<br>
<br>
Richard and I discussed the way to enforce the error w/o customizing<br>
the diagnostic for OpenCL off-list and Richard proposed to use<br>
"-Werror=implicit-function-declaration" command line option if error<br>
is required. Will it work for you?<br>
<br>
Thanks,<br>
Alexey<br>
On Tue, Aug 21, 2018 at 5:41 PM Anastasia Stulova via cfe-commits<br>
<cfe-commits@lists.llvm.org> wrote:<br>
><br>
> If there are no objections I would like to revert this old commit that coverts error about implicit function declaration into a warning.<br>
><br>
><br>
> We have decided to generate an error for this <a href="https://reviews.llvm.org/D31745">
https://reviews.llvm.org/D31745</a> because for OpenCL variadic prototypes are disallowed (section 6.9.e,
<a href="https://www.khronos.org/registry/OpenCL/specs/opencl-2.0-openclc.pdf">https://www.khronos.org/registry/OpenCL/specs/opencl-2.0-openclc.pdf</a>) and the implicit prototype requires variadic support. As most vendors that support OpenCL don't support
variadic functions it was decided to restrict this explicitly in the spec (section s6.9.u). There is a little bit of more history in
<a href="https://reviews.llvm.org/D17438">https://reviews.llvm.org/D17438</a>.<br>
><br>
> Currently the code that can't run correctly on most OpenCL targets compiles successfully. The problem can't be easily seen by the OpenCL developers since it's not very common to retrieve the compilation warning log during online compilation. Also generated
IR doesn't seem to be correct if I compare with the similar code in C.<br>
><br>
> Example:<br>
> 1 typedef long long16 __attribute__((ext_vector_type(16)));<br>
> 2 void test_somefunc( __global int *d, __global void *s )<br>
> 3 {<br>
> 4 int i = get_global_id(0);<br>
> 5 d[i] = somefunc((( __global long16 *)s)[i]);<br>
> 6 }<br>
><br>
> Is generated to:<br>
><br>
> %call1 = call i32 (<16 x i64>*, ...) bitcast (i32 ()* @somefunc to i32 (<16 x i64>*, ...)*)(<16 x i64>* byval nonnull align 128 %indirect-arg-temp) #2<br>
> ...<br>
><br>
> declare i32 @somefunc() local_unnamed_addr #1<br>
><br>
> Equivalent C code at least generates variadic function prototype correctly:<br>
><br>
> %call1 = call i32 (<16 x i64>*, ...) bitcast (i32 (...)* @somefunc to i32 (<16 x i64>*, ...)*)(<16 x i64>* byval align 128 %indirect-arg-temp)<br>
> ...<br>
> declare i32 @somefunc(...)<br>
><br>
> Anastasia<br>
> ________________________________<br>
> From: cfe-commits <cfe-commits-bounces@lists.llvm.org> on behalf of Richard Smith via cfe-commits <cfe-commits@lists.llvm.org><br>
> Sent: 04 October 2017 02:58<br>
> To: cfe-commits@lists.llvm.org<br>
> Subject: r314872 - We allow implicit function declarations as an extension in all C dialects. Remove OpenCL special case.<br>
><br>
> Author: rsmith<br>
> Date: Tue Oct 3 18:58:22 2017<br>
> New Revision: 314872<br>
><br>
> URL: <a href="http://llvm.org/viewvc/llvm-project?rev=314872&view=rev">http://llvm.org/viewvc/llvm-project?rev=314872&view=rev</a><br>
> Log:<br>
> We allow implicit function declarations as an extension in all C dialects. Remove OpenCL special case.<br>
><br>
> Modified:<br>
> cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td<br>
> cfe/trunk/lib/Sema/SemaDecl.cpp<br>
> cfe/trunk/test/SemaOpenCL/clang-builtin-version.cl<br>
> cfe/trunk/test/SemaOpenCL/to_addr_builtin.cl<br>
><br>
> Modified: cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td<br>
> URL: <a href="http://llvm.org/viewvc/llvm-project/cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td?rev=314872&r1=314871&r2=314872&view=diff">
http://llvm.org/viewvc/llvm-project/cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td?rev=314872&r1=314871&r2=314872&view=diff</a><br>
> ==============================================================================<br>
> --- cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td (original)<br>
> +++ cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td Tue Oct 3 18:58:22 2017<br>
> @@ -355,7 +355,7 @@ def warn_implicit_function_decl : Warnin<br>
> "implicit declaration of function %0">,<br>
> InGroup<ImplicitFunctionDeclare>, DefaultIgnore;<br>
> def ext_implicit_function_decl : ExtWarn<<br>
> - "implicit declaration of function %0 is invalid in C99">,<br>
> + "implicit declaration of function %0 is invalid in %select{C99|OpenCL}1">,<br>
> InGroup<ImplicitFunctionDeclare>;<br>
> def note_function_suggestion : Note<"did you mean %0?">;<br>
><br>
> @@ -8449,8 +8449,6 @@ def err_opencl_scalar_type_rank_greater_<br>
> "element. (%0 and %1)">;<br>
> def err_bad_kernel_param_type : Error<<br>
> "%0 cannot be used as the type of a kernel parameter">;<br>
> -def err_opencl_implicit_function_decl : Error<<br>
> - "implicit declaration of function %0 is invalid in OpenCL">;<br>
> def err_record_with_pointers_kernel_param : Error<<br>
> "%select{struct|union}0 kernel parameters may not contain pointers">;<br>
> def note_within_field_of_type : Note<<br>
><br>
> Modified: cfe/trunk/lib/Sema/SemaDecl.cpp<br>
> URL: <a href="http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Sema/SemaDecl.cpp?rev=314872&r1=314871&r2=314872&view=diff">
http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Sema/SemaDecl.cpp?rev=314872&r1=314871&r2=314872&view=diff</a><br>
> ==============================================================================<br>
> --- cfe/trunk/lib/Sema/SemaDecl.cpp (original)<br>
> +++ cfe/trunk/lib/Sema/SemaDecl.cpp Tue Oct 3 18:58:22 2017<br>
> @@ -12642,17 +12642,15 @@ NamedDecl *Sema::ImplicitlyDefineFunctio<br>
> }<br>
><br>
> // Extension in C99. Legal in C90, but warn about it.<br>
> + // OpenCL v2.0 s6.9.u - Implicit function declaration is not supported.<br>
> unsigned diag_id;<br>
> if (II.getName().startswith("__builtin_"))<br>
> diag_id = diag::warn_builtin_unknown;<br>
> - // OpenCL v2.0 s6.9.u - Implicit function declaration is not supported.<br>
> - else if (getLangOpts().OpenCL)<br>
> - diag_id = diag::err_opencl_implicit_function_decl;<br>
> - else if (getLangOpts().C99)<br>
> + else if (getLangOpts().C99 || getLangOpts().OpenCL)<br>
> diag_id = diag::ext_implicit_function_decl;<br>
> else<br>
> diag_id = diag::warn_implicit_function_decl;<br>
> - Diag(Loc, diag_id) << &II;<br>
> + Diag(Loc, diag_id) << &II << getLangOpts().OpenCL;<br>
><br>
> // If we found a prior declaration of this function, don't bother building<br>
> // another one. We've already pushed that one into scope, so there's nothing<br>
><br>
> Modified: cfe/trunk/test/SemaOpenCL/clang-builtin-version.cl<br>
> URL: <a href="http://llvm.org/viewvc/llvm-project/cfe/trunk/test/SemaOpenCL/clang-builtin-version.cl?rev=314872&r1=314871&r2=314872&view=diff">
http://llvm.org/viewvc/llvm-project/cfe/trunk/test/SemaOpenCL/clang-builtin-version.cl?rev=314872&r1=314871&r2=314872&view=diff</a><br>
> ==============================================================================<br>
> --- cfe/trunk/test/SemaOpenCL/clang-builtin-version.cl (original)<br>
> +++ cfe/trunk/test/SemaOpenCL/clang-builtin-version.cl Tue Oct 3 18:58:22 2017<br>
> @@ -1,4 +1,4 @@<br>
> -// RUN: %clang_cc1 %s -fblocks -verify -pedantic -fsyntax-only -ferror-limit 100<br>
> +// RUN: %clang_cc1 %s -fblocks -verify -pedantic-errors -fsyntax-only -ferror-limit 100<br>
><br>
> // Confirm CL2.0 Clang builtins are not available in earlier versions<br>
><br>
><br>
> Modified: cfe/trunk/test/SemaOpenCL/to_addr_builtin.cl<br>
> URL: <a href="http://llvm.org/viewvc/llvm-project/cfe/trunk/test/SemaOpenCL/to_addr_builtin.cl?rev=314872&r1=314871&r2=314872&view=diff">
http://llvm.org/viewvc/llvm-project/cfe/trunk/test/SemaOpenCL/to_addr_builtin.cl?rev=314872&r1=314871&r2=314872&view=diff</a><br>
> ==============================================================================<br>
> --- cfe/trunk/test/SemaOpenCL/to_addr_builtin.cl (original)<br>
> +++ cfe/trunk/test/SemaOpenCL/to_addr_builtin.cl Tue Oct 3 18:58:22 2017<br>
> @@ -10,7 +10,7 @@ void test(void) {<br>
><br>
> glob = to_global(glob, loc);<br>
> #if __OPENCL_C_VERSION__ < CL_VERSION_2_0<br>
> - // expected-error@-2{{implicit declaration of function 'to_global' is invalid in OpenCL}}<br>
> + // expected-warning@-2{{implicit declaration of function 'to_global' is invalid in OpenCL}}<br>
> // expected-warning@-3{{incompatible integer to pointer conversion assigning to '__global int *' from 'int'}}<br>
> #else<br>
> // expected-error@-5{{invalid number of arguments to function: 'to_global'}}<br>
><br>
><br>
> _______________________________________________<br>
> cfe-commits mailing list<br>
> cfe-commits@lists.llvm.org<br>
> <a href="http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits">http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits</a><br>
> _______________________________________________<br>
> cfe-commits mailing list<br>
> cfe-commits@lists.llvm.org<br>
> <a href="http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits">http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits</a><br>
</div>
</span></font></div>
</div>
</div>
</body>
</html>