Skip to content

Commit

Permalink
Merge pull request #3135 from trevorsm7/fea-move-nvtx-functions-to-na…
Browse files Browse the repository at this point in the history
…mespace

[REVIEW] Add nvtx functions to cudf namespace
  • Loading branch information
trevorsm7 authored Oct 22, 2019
2 parents 1d7e44a + 435a67f commit c57b315
Show file tree
Hide file tree
Showing 11 changed files with 143 additions and 20 deletions.
1 change: 1 addition & 0 deletions CHANGELOG.md
Original file line number Diff line number Diff line change
Expand Up @@ -41,6 +41,7 @@
- PR #3126 Round 2 of snappy decompression optimizations
- PR #3046 Define and implement new copying APIs `empty_like` and `allocate_like`
- PR #3128 Support MultiIndex in DataFrame.join
- PR #3135 Add nvtx utilities to cudf::nvtx namespace
- PR #3021 Java host side concat of serialized buffers
- PR #3138 Movey unary files to legacy

Expand Down
1 change: 1 addition & 0 deletions cpp/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -415,6 +415,7 @@ add_library(cudf
src/utilities/legacy/cuda_utils.cu
src/utilities/column_utils.cpp
src/utilities/legacy/error_utils.cpp
src/utilities/nvtx/nvtx_utils.cpp
src/utilities/nvtx/legacy/nvtx_utils.cpp
src/copying/copy.cpp
src/copying/legacy/copy.cpp
Expand Down
66 changes: 66 additions & 0 deletions cpp/include/cudf/utilities/nvtx_utils.hpp
Original file line number Diff line number Diff line change
@@ -0,0 +1,66 @@
#pragma once

#include <cstdint>

namespace cudf {
namespace nvtx {

enum class color : uint32_t {
GREEN = 0xff00ff00,
BLUE = 0xff0000ff,
YELLOW = 0xffffff00,
PURPLE = 0xffff00ff,
CYAN = 0xff00ffff,
RED = 0xffff0000,
WHITE = 0xffffffff,
DARK_GREEN = 0xff006600,
ORANGE = 0xffffa500,
};

constexpr color JOIN_COLOR = color::CYAN;
constexpr color GROUPBY_COLOR = color::GREEN;
constexpr color BINARY_OP_COLOR = color::YELLOW;
constexpr color PARTITION_COLOR = color::PURPLE;
constexpr color READ_CSV_COLOR = color::PURPLE;

/**---------------------------------------------------------------------------*
* @brief Start an NVTX range.
*
* This function is useful only for profiling with nvvp or Nsight Systems. It
* demarcates the begining of a user-defined range with a specified name and
* color that will show up in the timeline view of nvvp/Nsight Systems. Can be
* nested within other ranges.
*
* @throws cudf::logic_error if `name` is null
*
* @param[in] name The name of the NVTX range
* @param[in] color The color to use for the range
*---------------------------------------------------------------------------**/
void range_push(const char* name, color color);

/**---------------------------------------------------------------------------*
* @brief Start a NVTX range with a custom ARGB color code.
*
* This function is useful only for profiling with nvvp or Nsight Systems. It
* demarcates the begining of a user-defined range with a specified name and
* color that will show up in the timeline view of nvvp/Nsight Systems. Can be
* nested within other ranges.
*
* @throws cudf::logic_error if `name` is null
*
* @param[in] name The name of the NVTX range
* @param[in] color The ARGB hex color code to use to color this range (e.g., 0xFF00FF00)
*---------------------------------------------------------------------------**/
void range_push_hex(const char* name, uint32_t color);

/**---------------------------------------------------------------------------*
* @brief Ends the inner-most NVTX range.
*
* This function is useful only for profiling with nvvp or Nsight Systems. It
* will demarcate the end of the inner-most range, i.e., the most recent call to
* range_push.
*---------------------------------------------------------------------------**/
void range_pop();

} // namespace nvtx
} // namespace cudf
6 changes: 3 additions & 3 deletions cpp/src/binaryop/compiled/launcher.cuh
Original file line number Diff line number Diff line change
Expand Up @@ -17,7 +17,7 @@
#ifndef COMPILED_BINARY_OPS_LAUNCHER_H
#define COMPILED_BINARY_OPS_LAUNCHER_H

#include <utilities/nvtx/legacy/nvtx_utils.h>
#include <cudf/utilities/nvtx_utils.hpp>
#include <utilities/error_utils.hpp>
#include <cudf/cudf.h>

Expand Down Expand Up @@ -57,7 +57,7 @@ struct BinaryOp {
GDF_REQUIRE(lhs->size == output->size, GDF_COLUMN_SIZE_MISMATCH);
GDF_REQUIRE(lhs->dtype == rhs->dtype, GDF_UNSUPPORTED_DTYPE);

PUSH_RANGE("LIBGDF_BINARY_OP", BINARY_OP_COLOR);
nvtx::range_push("CUDF_BINARY_OP", nvtx::BINARY_OP_COLOR);
// find optimal blocksize
int mingridsize, blocksize;
CUDA_TRY(
Expand All @@ -82,7 +82,7 @@ struct BinaryOp {

cudaDeviceSynchronize();

POP_RANGE();
nvtx::range_pop();

CUDA_CHECK_LAST();
return GDF_SUCCESS;
Expand Down
1 change: 0 additions & 1 deletion cpp/src/groupby/legacy/groupby_without_aggregation.cu
Original file line number Diff line number Diff line change
Expand Up @@ -15,7 +15,6 @@
*/

#include <cudf/cudf.h>
#include <utilities/nvtx/legacy/nvtx_utils.h>
#include <cudf/legacy/copying.hpp>
#include <cudf/types.hpp>
#include <cudf/utilities/legacy/nvcategory_util.hpp>
Expand Down
6 changes: 3 additions & 3 deletions cpp/src/hash/legacy/hashing.cu
Original file line number Diff line number Diff line change
Expand Up @@ -23,7 +23,7 @@
#include <table/legacy/device_table.cuh>
#include "hash/hash_functions.cuh"
#include <utilities/int_fastdiv.h>
#include <utilities/nvtx/legacy/nvtx_utils.h>
#include <cudf/utilities/nvtx_utils.hpp>
#include <copying/legacy/scatter.hpp>
#include <cudf/types.hpp>

Expand Down Expand Up @@ -603,7 +603,7 @@ gdf_error gdf_hash_partition(int num_input_cols,
return GDF_COLUMN_SIZE_MISMATCH;
}

PUSH_RANGE("LIBGDF_HASH_PARTITION", PARTITION_COLOR);
cudf::nvtx::range_push("CUDF_HASH_PARTITION", cudf::nvtx::PARTITION_COLOR);

cudf::table input_table(input, num_input_cols);
cudf::table output_table(partitioned_output, num_input_cols);
Expand Down Expand Up @@ -638,7 +638,7 @@ gdf_error gdf_hash_partition(int num_input_cols,
gdf_status = GDF_INVALID_HASH_FUNCTION;
}

POP_RANGE();
cudf::nvtx::range_pop();

return gdf_status;
}
Expand Down
10 changes: 5 additions & 5 deletions cpp/src/join/legacy/joining.cu
Original file line number Diff line number Diff line change
Expand Up @@ -23,7 +23,7 @@
#include <utilities/column_utils.hpp>
#include <utilities/error_utils.hpp>
#include <cudf/utilities/legacy/type_dispatcher.hpp>
#include <utilities/nvtx/legacy/nvtx_utils.h>
#include <cudf/utilities/nvtx_utils.hpp>
#include <cudf/utilities/legacy/nvcategory_util.hpp>
#include <nvstrings/NVCategory.h>
#include <copying/legacy/gather.hpp>
Expand Down Expand Up @@ -214,7 +214,7 @@ void join_call(cudf::table const& left, cudf::table const& right,
gdf_method join_method = join_context->flag_method;
gdf_error gdf_error_code{GDF_SUCCESS};

PUSH_RANGE("LIBGDF_JOIN", JOIN_COLOR);
nvtx::range_push("CUDF_JOIN", nvtx::JOIN_COLOR);

switch(join_method)
{
Expand Down Expand Up @@ -245,7 +245,7 @@ void join_call(cudf::table const& left, cudf::table const& right,
CUDF_FAIL("Unsupported join Method");
}

POP_RANGE();
nvtx::range_pop();
}

/**---------------------------------------------------------------------------*
Expand Down Expand Up @@ -322,7 +322,7 @@ cudf::table construct_join_output_df(
gdf_column * left_indices,
gdf_column * right_indices) {

PUSH_RANGE("LIBGDF_JOIN_OUTPUT", JOIN_COLOR);
nvtx::range_push("CUDF_JOIN_OUTPUT", nvtx::JOIN_COLOR);
//create left and right input table with columns not joined on
std::vector<cudf::size_type> left_columns_in_common (columns_in_common.size());
std::vector<cudf::size_type> right_columns_in_common (columns_in_common.size());
Expand Down Expand Up @@ -406,7 +406,7 @@ cudf::table construct_join_output_df(
}

CHECK_STREAM(0);
POP_RANGE();
nvtx::range_pop();
return result;
}

Expand Down
6 changes: 3 additions & 3 deletions cpp/src/rolling/rolling.cu
Original file line number Diff line number Diff line change
Expand Up @@ -14,7 +14,7 @@
* limitations under the License.
*/

#include <utilities/nvtx/legacy/nvtx_utils.h>
#include <cudf/utilities/nvtx_utils.hpp>
#include <cudf/utilities/legacy/type_dispatcher.hpp>
#include <utilities/bit_util.cuh>
#include <bitmask/legacy/bit_mask.cuh>
Expand Down Expand Up @@ -154,7 +154,7 @@ struct rolling_window_launcher
typename std::enable_if_t<cudf::detail::is_supported<ColumnType, agg_op>(), std::nullptr_t> = nullptr>
void dispatch_aggregation_type(cudf::size_type nrows, cudaStream_t stream, TArgs... FArgs)
{
PUSH_RANGE("CUDF_ROLLING", GDF_ORANGE);
cudf::nvtx::range_push("CUDF_ROLLING", cudf::nvtx::color::ORANGE);

cudf::size_type block = 256;
cudf::size_type grid = (nrows + block-1) / block;
Expand All @@ -164,7 +164,7 @@ struct rolling_window_launcher
// check the stream for debugging
CHECK_STREAM(stream);

POP_RANGE();
cudf::nvtx::range_pop();
}

/**
Expand Down
6 changes: 3 additions & 3 deletions cpp/src/transpose/legacy/transpose.cu
Original file line number Diff line number Diff line change
@@ -1,6 +1,6 @@
/* Copyright 2018 NVIDIA Corporation. All rights reserved. */

#include <utilities/nvtx/legacy/nvtx_utils.h>
#include <cudf/utilities/nvtx_utils.hpp>
#include <cudf/utilities/legacy/type_dispatcher.hpp>
#include <rmm/thrust_rmm_allocator.h>
#include <bitmask/legacy/legacy_bitmask.hpp>
Expand Down Expand Up @@ -167,7 +167,7 @@ gdf_error gdf_transpose(cudf::size_type ncols, gdf_column** in_cols,
}
}

PUSH_RANGE("CUDF_TRANSPOSE", GDF_GREEN);
cudf::nvtx::range_push("CUDF_TRANSPOSE", cudf::nvtx::color::GREEN);

// Copy input columns `data` and `valid` pointers to device
std::vector<void*> in_columns_data(ncols);
Expand Down Expand Up @@ -213,6 +213,6 @@ gdf_error gdf_transpose(cudf::size_type ncols, gdf_column** in_cols,
out_cols[i]->null_count = out_columns_nullct[i];
}

POP_RANGE();
cudf::nvtx::range_pop();
return GDF_SUCCESS;
}
55 changes: 55 additions & 0 deletions cpp/src/utilities/nvtx/nvtx_utils.cpp
Original file line number Diff line number Diff line change
@@ -0,0 +1,55 @@
/*
* Copyright (c) 2019, 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 <cudf/utilities/nvtx_utils.hpp>
#include "utilities/error_utils.hpp"

#ifdef USE_NVTX
#include <nvToolsExt.h>
#endif

namespace cudf {
namespace nvtx {

void range_push(const char* name, color color)
{
range_push_hex(name, static_cast<uint32_t>(color));
}

void range_push_hex(const char* name, uint32_t color)
{
#ifdef USE_NVTX
CUDF_EXPECTS(name != nullptr, "Null name string.");

nvtxEventAttributes_t eventAttrib{};
eventAttrib.version = NVTX_VERSION;
eventAttrib.size = NVTX_EVENT_ATTRIB_STRUCT_SIZE;
eventAttrib.colorType = NVTX_COLOR_ARGB;
eventAttrib.color = color;
eventAttrib.messageType = NVTX_MESSAGE_TYPE_ASCII;
eventAttrib.message.ascii = name;
nvtxRangePushEx(&eventAttrib);
#endif
}

void range_pop()
{
#ifdef USE_NVTX
nvtxRangePop();
#endif
}

} // namespace nvtx
} // namespace cudf
5 changes: 3 additions & 2 deletions java/src/main/native/src/NvtxRangeJni.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -17,6 +17,7 @@
#include <cstddef>
#include <cstdint>
#include "cudf/cudf.h"
#include <cudf/utilities/nvtx_utils.hpp>

#include "jni_utils.hpp"

Expand All @@ -27,15 +28,15 @@ Java_ai_rapids_cudf_NvtxRange_push(JNIEnv *env, jclass clazz,
jstring name, jint color_bits) {
try {
cudf::jni::native_jstring range_name(env, name);
JNI_GDF_TRY(env, , gdf_nvtx_range_push_hex(range_name.get(), color_bits));
cudf::nvtx::range_push_hex(range_name.get(), color_bits);
}
CATCH_STD(env, );
}

JNIEXPORT void JNICALL
Java_ai_rapids_cudf_NvtxRange_pop(JNIEnv *env, jclass clazz) {
try {
JNI_GDF_TRY(env, , gdf_nvtx_range_pop());
cudf::nvtx::range_pop();
}
CATCH_STD(env, );
}
Expand Down

0 comments on commit c57b315

Please sign in to comment.