Skip to content

Commit

Permalink
Merge branch 'branch-0.19' into benchmark-strings-replace-re
Browse files Browse the repository at this point in the history
  • Loading branch information
davidwendt committed Mar 18, 2021
2 parents 224d35b + 4723051 commit 6d9de5b
Show file tree
Hide file tree
Showing 97 changed files with 3,921 additions and 1,741 deletions.
2 changes: 1 addition & 1 deletion build.sh
Original file line number Diff line number Diff line change
Expand Up @@ -135,7 +135,7 @@ if hasArg clean; then
fi

if (( ${BUILD_ALL_GPU_ARCH} == 0 )); then
CUDF_CMAKE_CUDA_ARCHITECTURES="-DCMAKE_CUDA_ARCHITECTURES=ALL"
CUDF_CMAKE_CUDA_ARCHITECTURES="-DCMAKE_CUDA_ARCHITECTURES="
echo "Building for the architecture of the GPU in the system..."
else
CUDF_CMAKE_CUDA_ARCHITECTURES=""
Expand Down
1 change: 1 addition & 0 deletions conda/recipes/libcudf/meta.yaml
Original file line number Diff line number Diff line change
Expand Up @@ -134,6 +134,7 @@ test:
- test -f $PREFIX/include/cudf/lists/detail/copying.hpp
- test -f $PREFIX/include/cudf/lists/detail/sorting.hpp
- test -f $PREFIX/include/cudf/lists/count_elements.hpp
- test -f $PREFIX/include/cudf/lists/explode.hpp
- test -f $PREFIX/include/cudf/lists/drop_list_duplicates.hpp
- test -f $PREFIX/include/cudf/lists/extract.hpp
- test -f $PREFIX/include/cudf/lists/contains.hpp
Expand Down
76 changes: 46 additions & 30 deletions cpp/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -117,6 +117,8 @@ include(cmake/Modules/ConfigureCUDA.cmake)

# find zlib
find_package(ZLIB REQUIRED)
# find Threads (needed by cudftestutil)
find_package(Threads REQUIRED)
# add third party dependencies using CPM
include(cmake/thirdparty/CUDF_GetCPM.cmake)
# find boost
Expand All @@ -133,8 +135,12 @@ include(cmake/thirdparty/CUDF_GetArrow.cmake)
include(cmake/thirdparty/CUDF_GetDLPack.cmake)
# find libcu++
include(cmake/thirdparty/CUDF_GetLibcudacxx.cmake)
# find or install GoogleTest
include(cmake/thirdparty/CUDF_GetGTest.cmake)
# Stringify libcudf and libcudacxx headers used in JIT operations
include(cmake/Modules/StringifyJITHeaders.cmake)
# find cuFile
include(cmake/Modules/FindcuFile.cmake)

###################################################################################################
# - library targets -------------------------------------------------------------------------------
Expand Down Expand Up @@ -240,6 +246,7 @@ add_library(cudf
src/io/statistics/column_stats.cu
src/io/utilities/data_sink.cpp
src/io/utilities/datasource.cpp
src/io/utilities/file_io_utilities.cpp
src/io/utilities/parsing_utils.cu
src/io/utilities/type_conversion.cpp
src/jit/cache.cpp
Expand All @@ -256,6 +263,7 @@ add_library(cudf
src/lists/copying/gather.cu
src/lists/copying/segmented_gather.cu
src/lists/count_elements.cu
src/lists/explode.cu
src/lists/extract.cu
src/lists/drop_list_duplicates.cu
src/lists/lists_column_factories.cu
Expand Down Expand Up @@ -285,7 +293,6 @@ add_library(cudf
src/replace/nulls.cu
src/replace/replace.cu
src/reshape/byte_cast.cu
src/reshape/explode.cu
src/reshape/interleave_columns.cu
src/reshape/tile.cu
src/rolling/grouped_rolling.cu
Expand Down Expand Up @@ -417,7 +424,8 @@ target_include_directories(cudf
"$<BUILD_INTERFACE:${CUDF_SOURCE_DIR}/include>"
"$<BUILD_INTERFACE:${CUDF_GENERATED_INCLUDE_DIR}/include>"
PRIVATE "$<BUILD_INTERFACE:${CUDF_SOURCE_DIR}/src>"
INTERFACE "$<INSTALL_INTERFACE:include>")
INTERFACE "$<INSTALL_INTERFACE:include>"
"$<INSTALL_INTERFACE:include/libcudf/libcudaxx>")

# Add Conda library paths if specified
if(CONDA_LINK_DIRS)
Expand Down Expand Up @@ -464,6 +472,11 @@ else()
target_link_libraries(cudf PUBLIC CUDA::nvrtc CUDA::cudart CUDA::cuda_driver)
endif()

# Add cuFile interface if available
if(TARGET cuFile::cuFile_interface)
target_link_libraries(cudf PRIVATE cuFile::cuFile_interface)
endif()

file(WRITE "${CUDF_BINARY_DIR}/fatbin.ld"
[=[
SECTIONS
Expand All @@ -480,34 +493,37 @@ add_library(cudf::cudf ALIAS cudf)
# - tests and benchmarks --------------------------------------------------------------------------
###################################################################################################

if (CUDF_BUILD_TESTS OR CUDF_BUILD_BENCHMARKS)
# Find or install GoogleTest
CPMFindPackage(NAME GTest
VERSION 1.10.0
GIT_REPOSITORY https://github.com/google/googletest.git
GIT_TAG release-1.10.0
GIT_SHALLOW TRUE
OPTIONS "INSTALL_GTEST OFF"
# googletest >= 1.10.0 provides a cmake config file -- use it if it exists
FIND_PACKAGE_ARGUMENTS "CONFIG")
# Add GTest aliases if they don't already exist.
# Assumes if GTest::gtest doesn't exist, the others don't either.
# TODO: Is this always a valid assumption?
if(NOT TARGET GTest::gtest)
add_library(GTest::gtest ALIAS gtest)
add_library(GTest::gmock ALIAS gmock)
add_library(GTest::gtest_main ALIAS gtest_main)
add_library(GTest::gmock_main ALIAS gmock_main)
endif()
if(GTest_ADDED)
install(TARGETS gmock
gtest
gmock_main
gtest_main
DESTINATION lib
EXPORT cudf-targets)
endif()
endif()
###################################################################################################
# - build cudftestutil ----------------------------------------------------------------------------

add_library(cudftestutil STATIC
tests/utilities/base_fixture.cpp
tests/utilities/column_utilities.cu
tests/utilities/table_utilities.cu
tests/strings/utilities.cu)

target_compile_options(cudftestutil
PUBLIC "$<$<COMPILE_LANGUAGE:CXX>:${CUDF_CXX_FLAGS}>"
"$<$<COMPILE_LANGUAGE:CUDA>:${CUDF_CUDA_FLAGS}>"
)

target_compile_features(cudftestutil PUBLIC cxx_std_14 cuda_std_14)

target_link_libraries(cudftestutil
PUBLIC GTest::gmock
GTest::gtest
Threads::Threads
cudf)

target_include_directories(cudftestutil
PUBLIC "$<BUILD_INTERFACE:${CUDF_SOURCE_DIR}>"
"$<BUILD_INTERFACE:${CUDF_SOURCE_DIR}/src>")

install(TARGETS cudftestutil
DESTINATION lib
EXPORT cudf-targets)

add_library(cudf::cudftestutil ALIAS cudftestutil)

###################################################################################################
# - add tests -------------------------------------------------------------------------------------
Expand Down
2 changes: 2 additions & 0 deletions cpp/benchmarks/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -185,9 +185,11 @@ ConfigureBench(STRINGS_BENCH
string/convert_floats_benchmark.cpp
string/copy_benchmark.cpp
string/extract_benchmark.cpp
string/factory_benchmark.cu
string/filter_benchmark.cpp
string/find_benchmark.cpp
string/replace_benchmark.cpp
string/replace_re_benchmark.cpp
string/split_benchmark.cpp
string/substring_benchmark.cpp
string/url_decode_benchmark.cpp)
2 changes: 1 addition & 1 deletion cpp/benchmarks/fixture/benchmark_fixture.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -88,4 +88,4 @@ class benchmark : public ::benchmark::Fixture {
std::shared_ptr<rmm::mr::device_memory_resource> mr;
};

}; // namespace cudf
} // namespace cudf
93 changes: 93 additions & 0 deletions cpp/benchmarks/string/factory_benchmark.cu
Original file line number Diff line number Diff line change
@@ -0,0 +1,93 @@
/*
* Copyright (c) 2021, NVIDIA CORPORATION.
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/

#include "string_bench_args.hpp"

#include <benchmark/benchmark.h>
#include <benchmarks/common/generate_benchmark_input.hpp>
#include <benchmarks/fixture/benchmark_fixture.hpp>
#include <benchmarks/synchronization/synchronization.hpp>

#include <cudf/strings/string_view.cuh>
#include <cudf/strings/strings_column_view.hpp>
#include <cudf_test/column_wrapper.hpp>

#include <rmm/device_uvector.hpp>

#include <thrust/execution_policy.h>
#include <thrust/transform.h>

#include <limits>

namespace {
using string_pair = thrust::pair<char const*, cudf::size_type>;
struct string_view_to_pair {
__device__ string_pair operator()(thrust::pair<cudf::string_view, bool> const& p)
{
return (p.second) ? string_pair{p.first.data(), p.first.size_bytes()} : string_pair{nullptr, 0};
}
};
} // namespace

class StringsFactory : public cudf::benchmark {
};

static void BM_factory(benchmark::State& state)
{
cudf::size_type const n_rows{static_cast<cudf::size_type>(state.range(0))};
cudf::size_type const max_str_length{static_cast<cudf::size_type>(state.range(1))};
data_profile table_profile;
table_profile.set_distribution_params(
cudf::type_id::STRING, distribution_id::NORMAL, 0, max_str_length);
auto const table =
create_random_table({cudf::type_id::STRING}, 1, row_count{n_rows}, table_profile);
auto d_column = cudf::column_device_view::create(table->view().column(0));
rmm::device_vector<string_pair> pairs(d_column->size());
thrust::transform(thrust::device,
d_column->pair_begin<cudf::string_view, true>(),
d_column->pair_end<cudf::string_view, true>(),
pairs.data(),
string_view_to_pair{});

for (auto _ : state) {
cuda_event_timer raii(state, true, 0);
cudf::make_strings_column(pairs);
}

cudf::strings_column_view input(table->view().column(0));
state.SetBytesProcessed(state.iterations() * input.chars_size());
}

static void generate_bench_args(benchmark::internal::Benchmark* b)
{
int const min_rows = 1 << 12;
int const max_rows = 1 << 24;
int const row_mult = 8;
int const min_rowlen = 1 << 5;
int const max_rowlen = 1 << 13;
int const len_mult = 4;
generate_string_bench_args(b, min_rows, max_rows, row_mult, min_rowlen, max_rowlen, len_mult);
}

#define STRINGS_BENCHMARK_DEFINE(name) \
BENCHMARK_DEFINE_F(StringsFactory, name) \
(::benchmark::State & st) { BM_factory(st); } \
BENCHMARK_REGISTER_F(StringsFactory, name) \
->Apply(generate_bench_args) \
->UseManualTime() \
->Unit(benchmark::kMillisecond);

STRINGS_BENCHMARK_DEFINE(factory)
2 changes: 2 additions & 0 deletions cpp/benchmarks/string/string_bench_args.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -17,6 +17,8 @@

#include <benchmark/benchmark.h>

#include <cudf/types.hpp>

/**
* @brief Generate row count and row length argument ranges for a string benchmark.
*
Expand Down
93 changes: 93 additions & 0 deletions cpp/benchmarks/string/substring_benchmark.cpp
Original file line number Diff line number Diff line change
@@ -0,0 +1,93 @@
/*
* Copyright (c) 2021, NVIDIA CORPORATION.
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/

#include "string_bench_args.hpp"

#include <benchmark/benchmark.h>
#include <benchmarks/common/generate_benchmark_input.hpp>
#include <benchmarks/fixture/benchmark_fixture.hpp>
#include <benchmarks/synchronization/synchronization.hpp>

#include <cudf/scalar/scalar.hpp>
#include <cudf/strings/strings_column_view.hpp>
#include <cudf/strings/substring.hpp>
#include <cudf_test/base_fixture.hpp>
#include <cudf_test/column_wrapper.hpp>

#include <limits>

#include <thrust/iterator/constant_iterator.h>

class StringSubstring : public cudf::benchmark {
};

enum substring_type { position, multi_position, delimiter, multi_delimiter };

static void BM_substring(benchmark::State& state, substring_type rt)
{
cudf::size_type const n_rows{static_cast<cudf::size_type>(state.range(0))};
cudf::size_type const max_str_length{static_cast<cudf::size_type>(state.range(1))};
data_profile table_profile;
table_profile.set_distribution_params(
cudf::type_id::STRING, distribution_id::NORMAL, 0, max_str_length);
auto const table =
create_random_table({cudf::type_id::STRING}, 1, row_count{n_rows}, table_profile);
cudf::strings_column_view input(table->view().column(0));
auto starts_itr = thrust::constant_iterator<cudf::size_type>(1);
auto stops_itr = thrust::constant_iterator<cudf::size_type>(max_str_length / 2);
cudf::test::fixed_width_column_wrapper<int32_t> starts(starts_itr, starts_itr + n_rows);
cudf::test::fixed_width_column_wrapper<int32_t> stops(stops_itr, stops_itr + n_rows);
auto delim_itr = thrust::constant_iterator<std::string>(" ");
cudf::test::strings_column_wrapper delimiters(delim_itr, delim_itr + n_rows);

for (auto _ : state) {
cuda_event_timer raii(state, true, 0);
switch (rt) {
case position: cudf::strings::slice_strings(input, 1, max_str_length / 2); break;
case multi_position: cudf::strings::slice_strings(input, starts, stops); break;
case delimiter: cudf::strings::slice_strings(input, std::string{" "}, 1); break;
case multi_delimiter:
cudf::strings::slice_strings(input, cudf::strings_column_view(delimiters), 1);
break;
}
}

state.SetBytesProcessed(state.iterations() * input.chars_size());
}

static void generate_bench_args(benchmark::internal::Benchmark* b)
{
int const min_rows = 1 << 12;
int const max_rows = 1 << 24;
int const row_mult = 8;
int const min_rowlen = 1 << 5;
int const max_rowlen = 1 << 13;
int const len_mult = 4;
generate_string_bench_args(b, min_rows, max_rows, row_mult, min_rowlen, max_rowlen, len_mult);
}

#define STRINGS_BENCHMARK_DEFINE(name) \
BENCHMARK_DEFINE_F(StringSubstring, name) \
(::benchmark::State & st) { BM_substring(st, substring_type::name); } \
BENCHMARK_REGISTER_F(StringSubstring, name) \
->Apply(generate_bench_args) \
->UseManualTime() \
->Unit(benchmark::kMillisecond);

STRINGS_BENCHMARK_DEFINE(position)
STRINGS_BENCHMARK_DEFINE(multi_position)
STRINGS_BENCHMARK_DEFINE(delimiter)
STRINGS_BENCHMARK_DEFINE(multi_delimiter)
22 changes: 6 additions & 16 deletions cpp/cmake/Modules/ConfigureCUDA.cmake
Original file line number Diff line number Diff line change
Expand Up @@ -17,26 +17,16 @@
# Find the CUDAToolkit
find_package(CUDAToolkit REQUIRED)

# Must come after find_package(CUDAToolkit) because we symlink
# ccache as a compiler front-end for nvcc in gpuCI CPU builds.
enable_language(CUDA)

if(CMAKE_CUDA_COMPILER_VERSION)
# Compute the version. from CMAKE_CUDA_COMPILER_VERSION
string(REGEX REPLACE "([0-9]+)\\.([0-9]+).*" "\\1" CUDA_VERSION_MAJOR ${CMAKE_CUDA_COMPILER_VERSION})
string(REGEX REPLACE "([0-9]+)\\.([0-9]+).*" "\\2" CUDA_VERSION_MINOR ${CMAKE_CUDA_COMPILER_VERSION})
set(CUDA_VERSION "${CUDA_VERSION_MAJOR}.${CUDA_VERSION_MINOR}")
endif()

message(VERBOSE "CUDF: CUDA_VERSION_MAJOR: ${CUDA_VERSION_MAJOR}")
message(VERBOSE "CUDF: CUDA_VERSION_MINOR: ${CUDA_VERSION_MINOR}")
message(STATUS "CUDF: CUDA_VERSION: ${CUDA_VERSION}")

# Auto-detect available GPU compute architectures

include(${CUDF_SOURCE_DIR}/cmake/Modules/SetGPUArchs.cmake)
message(STATUS "CUDF: Building CUDF for GPU architectures: ${CMAKE_CUDA_ARCHITECTURES}")

# Must come after find_package(CUDAToolkit) because we symlink
# ccache as a compiler front-end for nvcc in gpuCI CPU builds.
# Must also come after we detect and potentially rewrite
# CMAKE_CUDA_ARCHITECTURES
enable_language(CUDA)

if(CMAKE_COMPILER_IS_GNUCXX)
list(APPEND CUDF_CXX_FLAGS -Wall -Werror -Wno-unknown-pragmas -Wno-error=deprecated-declarations)
if(CUDF_BUILD_TESTS OR CUDF_BUILD_BENCHMARKS)
Expand Down
Loading

0 comments on commit 6d9de5b

Please sign in to comment.