[BUG] Cutlass' static library defines fewer symbols than the shared library

[joerowell]

Describe the bug I've recently been playing around with building cutlass as both a shared and static library. However, I've noticed that the static library often omits some symbols that are present in the shared library.

Steps/Code to reproduce bug I've encountered this issue on 3a8c01a18b. I've not tried any other cutlass versions, but I don't think there would be a difference.

First, make and navigate to a build directory from the cutlass root directory. Then, configure cutlass' build system using the following cmake command:

cmake .. -DCUTLASS_NVCC_ARCHS=90 -DCUTLASS_ENABLE_TESTS=OFF -DCUTLASS_UNITY_BUILD_ENABLED=ON -DCUTLASS_LIBRARY_OPERATIONS=gemm

The NVCC archs, tests and unity builds aren't that important for reproducibility: I've seen the issue occur in many different configurations of flags.

Then, compile both the static and the shared libraries via make.

make cutlass_library
make cutlass_library_static

This should succeed. Finally, run the following commands and notice the different outputs:

readelf -Ws tools/library/libcutlass.a --demangle | grep cutlass::library::from_string | wc -l ; I get 12
readelf -Ws tools/library/libcutlass.so --demangle | grep cutlass::library::from_string | wc -l ; I get 24

Moreover, if I actually look at the symbols themselves, I see that the static library doesn't actually define all of the symbols.

readelf -Ws tools/library/libcutlass.a --demangle | grep "cutlass::library::initialize_all(cutlass::library::Manifest&)

   9: 0000000000000000     0 NOTYPE  GLOBAL DEFAULT  UND cutlass::library::initialize_all(cutlass::library::Manifest&)
   7: 0000000000000000    27 FUNC    GLOBAL DEFAULT    1 cutlass::library::initialize_all(cutlass::library::Manifest&)

Whereas the shared library does define these:

readelf -Ws tools/library/libcutlass.so --demangle | grep "cutlass::library::intialize_all(cutlass::library::Manifest&)

 13441: 00000000005c5ee0    27 FUNC    GLOBAL DEFAULT   14 cutlass::library::initialize_all(cutlass::library::Manifest&)
 15449: 00000000005c5ee0    27 FUNC    GLOBAL DEFAULT   14 cutlass::library::initialize_all(cutlass::library::Manifest&)

This means that a third-party application that wants to statically link against cutlass is somewhat limited.

Expected behavior My expectation would be that the static library would export all non-explicitly hidden symbols. I appreciate there's some subtle linker behaviour that can influence how symbols are exported, but this seems unusual.

[joerowell]

We debugged this issue and found out that the static library doesn't appear to be link against the compiled versions of the autogenerated kernels i.e. the files in build/tools/library/ aren't explicitly linked against libcutlass.a.