This is an archive of the discontinued LLVM Phabricator instance.

Allow implicit conversions between incompatible pointer types in overload resolution in C.
ClosedPublic

Authored by george.burgess.iv on Aug 31 2016, 5:34 PM.

Details

Summary

In C, we allow pointer conversions like:

void foo(int *a);
long b;
void runFoo() { foo(&b); }

...But not when overloading:

int bar(char *a) __attribute__((overloadable));
int bar(struct{}) __attribute__((overloadable));
unsigned char d;
void runBar() { bar(&d); } // error: no matching function for call to 'bar'

This patch makes it so bar(&d) isn't an error.

A few notes:

  • We'll still emit a warning about the conversion, assuming the user hasn't opted out of said warnings.
  • Every such conversion is ranked as the worst possible thing. So, if there's any ambiguity at all (e.g. if bar had a void(signed char*) overload, as well), we'll still emit an error.
  • We consider conversions that drop qualifiers to be equally as bad as converting between incompatible pointer types. I'm happy to make conversions that only drop qualifiers preferred over incompatible pointer types, if that would be better.

Diff Detail

Repository
rL LLVM

Event Timeline

george.burgess.iv retitled this revision from to Allow implicit conversions between incompatible pointer types in overload resolution in C..
george.burgess.iv updated this object.
george.burgess.iv added a reviewer: rsmith.
george.burgess.iv added a subscriber: cfe-commits.
rsmith accepted this revision.Sep 2 2016, 2:56 PM
rsmith edited edge metadata.
rsmith added inline comments.
lib/Sema/SemaOverload.cpp
1813–1815 ↗(On Diff #69922)

Yuck, this violates our invariants: First should only ever be some kind of lvalue conversion (it should in this case be an ICK_Array_To_Pointer or ICK_Function_To_Pointer if the argument was an array or function, respectively, and otherwise ICK_Identity). Third should only ever be ICK_Identity or ICK_Qualification. It seems fine to always set Third to ICK_Identity and model the qualification change as part of the second "standard" conversion; this should still compare worse than any other sequence by the rank test.

Now I've noticed this, I'd like to see it fixed, but since this is pre-existing I don't mind whether we fix this as part of this change or separately.

5931 ↗(On Diff #69922)

You should check Second here rather than First.

This revision is now accepted and ready to land.Sep 2 2016, 2:56 PM
george.burgess.iv marked 2 inline comments as done.Sep 2 2016, 4:07 PM
george.burgess.iv added inline comments.
lib/Sema/SemaOverload.cpp
1813–1815 ↗(On Diff #69922)

I'll fix this in a follow-up commit; thanks for pointing it out!

This revision was automatically updated to reflect the committed changes.

This change breaks OpenCL(libclc). CLC expects signed/unsigned overloads to be non-ambiguous.
For example:
atomic_max(global int *p, int val);
atomic_max(global unsigned int *p, unsigned int val);

need to work and be unambiguous (they are implemented using different instructions different instructions).

george.burgess.iv marked an inline comment as done.Sep 6 2016, 2:57 PM

Thanks for the heads-up!

I'm assuming that below is an example of code that this patch broke?

void foo(int *a, int) __attribute__((overloadable));
void foo(unsigned int *a, unsigned int) __attribute__((overloadable));

void callFoo() {
  unsigned int i;
  foo(&i, 0u); // fine.
  foo(&i, 0); // now-ambiguous overload.
}

Thanks for the heads-up!

I'm assuming that below is an example of code that this patch broke?

void foo(int *a, int) __attribute__((overloadable));
void foo(unsigned int *a, unsigned int) __attribute__((overloadable));

void callFoo() {
  unsigned int i;
  foo(&i, 0u); // fine.
  foo(&i, 0); // now-ambiguous overload.
}

yes, you're correct. (I'm pretty sure I tried this before posting, I must have done something wrong...) I'll send libclc patch. The specs don't really talk about overloads so I think the behaviour is OK. Although I think that users will expect atomic_add(unsigned int *, 5) to work.without error.

thanks

Although I think that users will expect atomic_add(unsigned int *, 5) to work.without error

Totally agree; I was just making sure that I understood the problem. :)

I think, with a small tweak in semantics, we can make everything work without needing to patch existing code. Let me play around a little.

yaxunl added a subscriber: yaxunl.Sep 7 2016, 7:20 AM

This seems to cause regression for the following test:

__kernel void test_atomic_fn(volatile __global uint *destMemory, __global uint *oldValues)
{
    int  tid = get_global_id(0);

oldValues[tid] = atom_add( &destMemory[0], tid + 3 );
atom_add( &destMemory[0], tid + 3 );
   atom_add( &destMemory[0], tid + 3 );
   atom_add( &destMemory[0], tid + 3 );

}

/tmp/AMD_17652_49/t_17652_51.cl:6:19: error: call to 'atom_add' is ambiguous
        oldValues[tid] = atom_add( &destMemory[0], tid + 3 );
                         ^~~~~~~~
/srv/opencl_rocm/drivers/opencl/library/amdgcn/headers/build/lnx64a/B_dbg/<stdin>:14336:12: note: candidate function
int __ovld atom_add(volatile __global int *p, int val);
           ^
/srv/opencl_rocm/drivers/opencl/library/amdgcn/headers/build/lnx64a/B_dbg/<stdin>:14337:21: note: candidate function
unsigned int __ovld atom_add(volatile __global unsigned int *p, unsigned int val);
                    ^
/srv/opencl_rocm/drivers/opencl/library/amdgcn/headers/build/lnx64a/B_dbg/<stdin>:14345:13: note: candidate function
long __ovld atom_add(volatile __global long *p, long val);
            ^
/srv/opencl_rocm/drivers/opencl/library/amdgcn/headers/build/lnx64a/B_dbg/<stdin>:14346:22: note: candidate function
unsigned long __ovld atom_add(volatile __global unsigned long *p, unsigned long val);
                     ^
/srv/opencl_rocm/drivers/opencl/library/amdgcn/headers/build/lnx64a/B_dbg/<stdin>:14347:13: note: candidate function
long __ovld atom_add(volatile __local long *p, long val);
            ^
/srv/opencl_rocm/drivers/opencl/library/amdgcn/headers/build/lnx64a/B_dbg/<stdin>:14348:22: note: candidate function
unsigned long __ovld atom_add(volatile __local unsigned long *p, unsigned long val);
                     ^
/tmp/AMD_17652_49/t_17652_51.cl:7:2: error: call to 'atom_add' is ambiguous
        atom_add( &destMemory[0], tid + 3 );
        ^~~~~~~~

Is this intended?

Thanks.