diff --git a/.github/workflows/continuous-integration.yml b/.github/workflows/continuous-integration.yml new file mode 100644 index 00000000..f165597e --- /dev/null +++ b/.github/workflows/continuous-integration.yml @@ -0,0 +1,69 @@ +name: Continuous Integration + +on: + push: + branches: + - master + - develop + # Skip jobs when only documentation files are changed + paths-ignore: + - '**.md' + - '**.rst' + - 'docs/**' + pull_request: + paths-ignore: + - '**.md' + - '**.rst' + - 'docs/**' + +jobs: + build: + runs-on: ubuntu-latest + strategy: + matrix: + container: ["alicevision/popsift-deps:cuda10.2-ubuntu18.04", "alicevision/popsift-deps:cuda11.8.0-ubuntu20.04", "alicevision/popsift-deps:cuda12.1.0-ubuntu22.04"] + build_tpe: ["Release", "Debug"] + exclude: + # excludes debug on this one as it has a segmentation fault during the compilation (!) + - container: "alicevision/popsift-deps:cuda12.1.0-ubuntu22.04" + build_tpe: "Debug" + + container: + image: ${{ matrix.container }} + + env: + DEPS_INSTALL_DIR: /opt/ + BUILD_TYPE: ${{ matrix.build_tpe }} + CTEST_OUTPUT_ON_FAILURE: 1 + steps: + - uses: actions/checkout@v2 + + - name: Prepare File Tree + run: | + mkdir ./build + mkdir ./build_as_3rdparty + mkdir ../popsift_install + + - name: Configure CMake + working-directory: ./build + run: | + cmake .. \ + -DCMAKE_BUILD_TYPE=${BUILD_TYPE} \ + -DBUILD_SHARED_LIBS:BOOL=ON \ + -DCMAKE_PREFIX_PATH="${DEPS_INSTALL_DIR}" \ + -DPopSift_BUILD_DOCS:BOOL=OFF \ + -DCMAKE_INSTALL_PREFIX:PATH=$PWD/../../popsift_install + + - name: Build + working-directory: ./build + run: | + make -j$(nproc) install + + - name: Build As Third Party + working-directory: ./build_as_3rdparty + run: | + cmake ../src/application \ + -DBUILD_SHARED_LIBS:BOOL=ON \ + -DCMAKE_BUILD_TYPE=${BUILD_TYPE} \ + -DCMAKE_PREFIX_PATH:PATH="$PWD/../../popsift_install;${DEPS_INSTALL_DIR}" + make -j$(nproc) diff --git a/Dockerfile_deps b/Dockerfile_deps index bed8cc59..d5560d4e 100644 --- a/Dockerfile_deps +++ b/Dockerfile_deps @@ -7,7 +7,7 @@ LABEL maintainer="AliceVision Team alicevision@googlegroups.com" # see https://hub.docker.com/r/nvidia/cuda/ # # For example, to create a ubuntu 16.04 with cuda 8.0 for development, use -# docker build --build-arg CUDA_TAG=8.0 --tag alicevision/popsift-deps:cuda${CUDA_TAG}-ubuntu${OS_TAG} . +# docker build --build-arg CUDA_TAG=8.0 --tag alicevision/popsift-deps:cuda${CUDA_TAG}-ubuntu${OS_TAG} -f Dockerfile_deps . # # then execute with nvidia docker (https://github.com/nvidia/nvidia-docker/wiki/Installation-(version-2.0)) # docker run -it --runtime=nvidia popsift_deps @@ -32,12 +32,12 @@ RUN apt-get clean && apt-get update && apt-get install -y --no-install-recommend libboost-thread-dev \ && rm -rf /var/lib/apt/lists/* - # Manually install cmake +# Manually install cmake WORKDIR /tmp/cmake -ENV CMAKE_VERSION=3.17 +ENV CMAKE_VERSION=3.24 ENV CMAKE_VERSION_FULL=${CMAKE_VERSION}.2 -RUN wget https://cmake.org/files/v3.17/cmake-${CMAKE_VERSION_FULL}.tar.gz && \ - tar zxvf cmake-${CMAKE_VERSION_FULL}.tar.gz && \ +RUN wget https://cmake.org/files/v${CMAKE_VERSION}/cmake-${CMAKE_VERSION_FULL}.tar.gz && \ + tar zxf cmake-${CMAKE_VERSION_FULL}.tar.gz && \ cd cmake-${CMAKE_VERSION_FULL} && \ ./bootstrap --prefix=/usr/local -- -DCMAKE_BUILD_TYPE:STRING=Release -DCMAKE_USE_OPENSSL:BOOL=ON && \ make -j$(nproc) install && \ diff --git a/README.md b/README.md index 2f257845..418d1278 100644 --- a/README.md +++ b/README.md @@ -1,7 +1,8 @@ # PopSift -[![CII Best Practices](https://bestpractices.coreinfrastructure.org/projects/3728/badge)](https://bestpractices.coreinfrastructure.org/projects/3728) [![Codacy Badge](https://api.codacy.com/project/badge/Grade/8b0f7a68bc0d4df2ac89c6e732917caa)](https://app.codacy.com/manual/alicevision/popsift?utm_source=github.com&utm_medium=referral&utm_content=alicevision/popsift&utm_campaign=Badge_Grade_Settings) +[![CII Best Practices](https://bestpractices.coreinfrastructure.org/projects/3728/badge)](https://bestpractices.coreinfrastructure.org/projects/3728) +[![Codacy Badge](https://app.codacy.com/project/badge/Grade/64f9192b53df46b483e7cf5be7e2dddd)](https://app.codacy.com/gh/alicevision/popsift/dashboard?utm_source=gh&utm_medium=referral&utm_content=&utm_campaign=Badge_grade) PopSift is an open-source implementation of the SIFT algorithm in CUDA. PopSift tries to stick as closely as possible to David Lowe's famous paper [1], while extracting features from an image in real-time at least on an NVidia GTX 980 Ti GPU. @@ -100,10 +101,12 @@ In particular, users can choose to generate results very similar to VLFeat or re We acknowledge that there is at least one SIFT implementation that is vastly faster, but it makes considerable sacrifices in terms of accuracy and compatibility. ## Continuous integration: -- [![Build Status](https://travis-ci.org/alicevision/popsift.svg?branch=master)](https://travis-ci.org/alicevision/popsift) master branch. -- [![Build Status](https://travis-ci.org/alicevision/popsift.svg?branch=develop)](https://travis-ci.org/alicevision/popsift) develop branch. -- [![Build status](https://ci.appveyor.com/api/projects/status/rsm5269hs288c2ji/branch/develop?svg=true)](https://ci.appveyor.com/project/AliceVision/popsift/branch/develop) - develop branch. + +* ![Continuous Integration](https://github.com/alicevision/popsift/workflows/Continuous%20Integration/badge.svg?branch=master) master branch on Linux. + +* ![Continuous Integration](https://github.com/alicevision/popsift/workflows/Continuous%20Integration/badge.svg?branch=develop) develop branch on Linux. + +* [![Build status](https://ci.appveyor.com/api/projects/status/rsm5269hs288c2ji/branch/develop?svg=true)](https://ci.appveyor.com/project/AliceVision/popsift/branch/develop) develop branch on Windows. ## License diff --git a/cmake/ChooseCudaCC.cmake b/cmake/ChooseCudaCC.cmake index 425e8bd5..aba4eb91 100755 --- a/cmake/ChooseCudaCC.cmake +++ b/cmake/ChooseCudaCC.cmake @@ -65,7 +65,7 @@ function(chooseCudaCC SUPPORTED_CC SUPPORTED_GENCODE_FLAGS) set(CC_LIST_BY_SYSTEM_PROCESSOR "") if(CMAKE_SYSTEM_PROCESSOR IN_LIST OTHER_SUPPORTED_PROCESSORS) - list(APPEND CC_LIST_BY_SYSTEM_PROCESSOR "20;21;30;35;50;52;60;61;70;75;80;86") + list(APPEND CC_LIST_BY_SYSTEM_PROCESSOR "20;21;30;35;50;52;60;61;70;75;80;86;89;90") endif() if(CMAKE_SYSTEM_PROCESSOR IN_LIST TEGRA_SUPPORTED_PROCESSORS) list(APPEND CC_LIST_BY_SYSTEM_PROCESSOR "32;53;62;72") @@ -78,10 +78,17 @@ function(chooseCudaCC SUPPORTED_CC SUPPORTED_GENCODE_FLAGS) # Default setting of the CUDA CC versions to compile. # Shortening the lists saves a lot of compile time. # - set(CUDA_MIN_CC 20) - set(CUDA_MAX_CC 86) - if(CUDA_VERSION VERSION_GREATER_EQUAL 11.1) + + # The current version last time this list was updated was CUDA 12.1. + if(CUDA_VERSION VERSION_GREATER_EQUAL 12) + set(CUDA_MIN_CC 50) + set(CUDA_MAX_CC 90) + elseif(CUDA_VERSION VERSION_GREATER_EQUAL 11.8) + set(CUDA_MIN_CC 35) + set(CUDA_MAX_CC 90) + elseif(CUDA_VERSION VERSION_GREATER_EQUAL 11.1) set(CUDA_MIN_CC 35) + set(CUDA_MAX_CC 86) elseif(CUDA_VERSION_MAJOR GREATER_EQUAL 11) set(CUDA_MIN_CC 35) set(CUDA_MAX_CC 80) @@ -92,8 +99,10 @@ function(chooseCudaCC SUPPORTED_CC SUPPORTED_GENCODE_FLAGS) set(CUDA_MIN_CC 30) set(CUDA_MAX_CC 72) elseif(CUDA_VERSION_MAJOR GREATER_EQUAL 8) + set(CUDA_MIN_CC 20) set(CUDA_MAX_CC 62) elseif(CUDA_VERSION_MAJOR GREATER_EQUAL 7) + set(CUDA_MIN_CC 20) set(CUDA_MAX_CC 53) else() message(FATAL_ERROR "We do not support a CUDA SDK below version 7.0") diff --git a/src/popsift/common/device_prop.cu b/src/popsift/common/device_prop.cu index 8436a69b..a55821cd 100644 --- a/src/popsift/common/device_prop.cu +++ b/src/popsift/common/device_prop.cu @@ -274,7 +274,7 @@ bool device_prop_t::checkLimit_2DsurfLayered( int& width, int& height, int& laye std::cerr << __FILE__ << ":" << __LINE__ << ": CUDA device " << currentDevice << std::endl << " does not support layered 2D surfaces " << width - << " bytes wide." << endl; + << " pixels wide." << endl; } width = ptr->maxSurface2DLayered[0]; returnSuccess = false; diff --git a/src/popsift/common/device_prop.h b/src/popsift/common/device_prop.h index ed5db2b2..7a0b142d 100644 --- a/src/popsift/common/device_prop.h +++ b/src/popsift/common/device_prop.h @@ -91,8 +91,7 @@ class device_prop_t /** * @brief Check if a request exceeds the current CUDA device's limit in * surface2DLayered dimensions. surface2DLayered is the writable equivalent - * to texture2DLayered, but the width must be given in bytes, not elements. - * Since we use float, images cannot be as wide as expected. + * to texture2DLayered. * @param[in,out] width Desired width of the texture. * @param[in,out] height Desired height of the texture. * @param[in,out] layers Desired depth of the texture. diff --git a/src/popsift/common/plane_2d.h b/src/popsift/common/plane_2d.h index 2dad48cd..8202e7a5 100644 --- a/src/popsift/common/plane_2d.h +++ b/src/popsift/common/plane_2d.h @@ -170,10 +170,10 @@ template struct PitchPlane2D : public PlaneT PlaneBase::freeHost2D( this->data, mode ); } __host__ __device__ - inline short getPitchInBytes( ) const { return _pitchInBytes; } + inline size_t getPitchInBytes( ) const { return _pitchInBytes; } protected: - int _pitchInBytes; // pitch width in bytes + size_t _pitchInBytes; // pitch width in bytes }; /************************************************************* @@ -340,7 +340,7 @@ template class Plane2D : public PitchPlane2D __host__ __device__ inline short getHeight( ) const { return _rows; } __host__ __device__ - inline short getByteSize( ) const { return this->_pitchInBytes*_rows; } + inline size_t getByteSize( ) const { return this->_pitchInBytes * _rows; } __host__ inline void allocDev( int w, int h ) { _cols = w; diff --git a/src/popsift/popsift.cpp b/src/popsift/popsift.cpp index f7e983aa..09575772 100755 --- a/src/popsift/popsift.cpp +++ b/src/popsift/popsift.cpp @@ -187,12 +187,7 @@ PopSift::AllocTest PopSift::testTextureFit( int width, int height ) */ int depth = _config.levels + 3; - /* Surfaces have a limited width in bytes, not in elements. - * Our DOG pyramid stores 4/byte floats, so me must check for - * that width. - */ - int byteWidth = width * sizeof(float); - retval = _device_properties.checkLimit_2DsurfLayered( byteWidth, + retval = _device_properties.checkLimit_2DsurfLayered( width, height, depth, warn ); @@ -219,13 +214,13 @@ std::string PopSift::testTextureFitErrorString( AllocTest err, int width, int he { const float upscaleFactor = _config.getUpscaleFactor(); const float scaleFactor = 1.0f / powf( 2.0f, -upscaleFactor ); - int w = ceilf( width * scaleFactor ) * sizeof(float); + int w = ceilf( width * scaleFactor ); int h = ceilf( height * scaleFactor ); int d = _config.levels + 3; _device_properties.checkLimit_2DsurfLayered( w, h, d, false ); - w = w / scaleFactor / sizeof(float); + w = w / scaleFactor; h = h / scaleFactor; ostr << "E Cannot use" << (upscaleFactor==1 ? " default " : " ") diff --git a/src/popsift/s_filtergrid.cu b/src/popsift/s_filtergrid.cu index 301c6a96..a766c2de 100644 --- a/src/popsift/s_filtergrid.cu +++ b/src/popsift/s_filtergrid.cu @@ -19,9 +19,11 @@ #if ! POPSIFT_IS_DEFINED(POPSIFT_DISABLE_GRID_FILTER) #include +#include #include #include #include +#include #include #include #include