The use of nvcc to compile cuda kernels is distinct from the use of
cuda sdk libraries and linking against those libraries. We have
previously not bothered to distinguish these two cases because all
the filters that used cuda kernels also used the sdk. In the following
changes, I'm going to remove the sdk dependency from those filters,
but we need a way to ensure that nvcc is present and functioning, and
also a way to explicitly disable its use so that the filters are not
built.
Note that, unlike the cuda_sdk dependency, using nvcc to compile
a kernel does not cause a build to become non-free. Although nvcc
is distributed with the cuda sdk, and is EULA encumbered, the
compilation process we use does not introduce any EULA covered
code or libraries into the build. In this sense, using nvcc is just
like using any other proprietary compiler like msvc - compiling free
code doesn't suddently make it non-free.
There was previously some confusion on this topic, but the important
distinction is that we use nvcc to generate ptx files - these are
not compiled GPU binaries, but rather an intermediate assembly
representation that is JIT compiled (and I think linked with certain
nvidia library code) when you actually try and run the kernel. nvidia
use this technique to relax machine code compatibility between
hardware generations.
From here, we can make two observations:
* The ptx files that we include in libavfilter are aggregated rather
than linked, from the perspective of the (L)GPL
* No proprietary code is included with the ptx files. That code is
only linked in at the final compilation step at runtime.
Signed-off-by: Philip Langdale <phil...@overt.org>
---
configure | 28 ++++++++++++++++++++++++++++
1 file changed, 28 insertions(+)
diff --git a/configure b/configure
index bf40c1dcb9..2219eb1515 100755
--- a/configure
+++ b/configure
@@ -322,6 +322,7 @@ External library support:
--disable-amf disable AMF video encoding code [autodetect]
--disable-audiotoolbox disable Apple AudioToolbox code [autodetect]
--enable-cuda-sdk enable CUDA features that require the CUDA SDK [no]
+ --disable-cuda-nvcc disable Nvidia CUDA compiler [autodetect]
--disable-cuvid disable Nvidia CUVID support [autodetect]
--disable-d3d11va disable Microsoft Direct3D 11 video acceleration
code [autodetect]
--disable-dxva2 disable Microsoft DirectX 9 video acceleration code
[autodetect]
@@ -1001,6 +1002,10 @@ hostcc_o(){
eval printf '%s\\n' $HOSTCC_O
}
+nvcc_o(){
+ eval printf '%s\\n' $NVCC_O
+}
+
test_cc(){
log test_cc "$@"
cat > $TMPC
@@ -1022,6 +1027,13 @@ test_objcc(){
test_cmd $objcc -Werror=missing-prototypes $CPPFLAGS $CFLAGS $OBJCFLAGS
"$@" $OBJCC_C $(cc_o $TMPO) $TMPM
}
+test_nvcc(){
+ log test_nvcc "$@"
+ cat > $TMPCU
+ log_file $TMPCU
+ test_cmd $nvcc -ptx $NVCCFLAGS "$@" $NVCC_C $(nvcc_o $TMPO) $TMPCU
+}
+
test_cpp(){
log test_cpp "$@"
cat > $TMPC
@@ -1786,6 +1798,7 @@ HWACCEL_AUTODETECT_LIBRARY_LIST="
audiotoolbox
crystalhd
cuda
+ cuda_nvcc
cuvid
d3d11va
dxva2
@@ -4238,6 +4251,7 @@ tmpfile TMPCPP .cpp
tmpfile TMPE $EXESUF
tmpfile TMPH .h
tmpfile TMPM .m
+tmpfile TMPCU .cu
tmpfile TMPO .o
tmpfile TMPS .S
tmpfile TMPSH .sh
@@ -6641,6 +6655,20 @@ else
nvccflags="$nvccflags -m32"
fi
+check_nvcc() {
+ log check_nvcc "$@"
+ disable cuda_nvcc
+ test_nvcc <<EOF && enable cuda_nvcc
+extern "C" {
+ __global__ void hello(unsigned char *data) {}
+}
+EOF
+}
+
+if ! disabled cuda_nvcc; then
+ check_nvcc
+fi
+
check_optflags(){
check_cflags "$@"
enabled lto && check_ldflags "$@"
--
2.19.1
_______________________________________________
ffmpeg-devel mailing list
ffmpeg-devel@ffmpeg.org
https://ffmpeg.org/mailman/listinfo/ffmpeg-devel
