[go: nahoru, domu]

Skip to content

Commit

Permalink
Replace instances of "blacklist" with "denylist" where possible. See …
Browse files Browse the repository at this point in the history
…Google Developer guidelines at https://developers.google.com/style/word-list#blacklist for more information.

PiperOrigin-RevId: 322242499
Change-Id: I66ea33f87811bbc734f538029d806240da91130f
  • Loading branch information
Karmel Allison authored and tensorflower-gardener committed Jul 20, 2020
1 parent 9f3c94a commit 18ebe82
Show file tree
Hide file tree
Showing 64 changed files with 572 additions and 559 deletions.
6 changes: 3 additions & 3 deletions RELEASE.md
Original file line number Diff line number Diff line change
Expand Up @@ -66,9 +66,9 @@
* Tracing and Debugging:
* <ADD RELEASE NOTES HERE>
* Other:
* We have replaced uses of "whitelist" with "allowlist" where possible.
Please see https://developers.google.com/style/word-list#blacklist for more
context.
* We have replaced uses of "whitelist" and "blacklist" with "allowlist"
and "denylist" where possible. Please see
https://developers.google.com/style/word-list#blacklist for more context.
* <ADD RELEASE NOTES HERE>

## Thanks to our Contributors
Expand Down
2 changes: 1 addition & 1 deletion tensorflow/compiler/jit/mark_for_compilation_pass_test.cc
Original file line number Diff line number Diff line change
Expand Up @@ -1829,7 +1829,7 @@ TEST(XlaCompilationTest, XLALiteAllowlist) {
}
EXPECT_TRUE(unknow_op.empty())
<< "Someone added support for a new TF opeations inside XLA. They must "
"be included in the XLALite allowlist or blacklist:\n"
"be included in the XLALite allowlist or denylist:\n"
<< absl::StrJoin(unknow_op, "\n");
}
} // namespace
Expand Down
2 changes: 1 addition & 1 deletion tensorflow/compiler/tests/eager_test.py
Original file line number Diff line number Diff line change
Expand Up @@ -311,7 +311,7 @@ def testConv(self):
if 'GPU' in self.device:
# TODO(b/32333178)
self.skipTest('Current implementation of RandomStandardNormal kernel '
'is very slow on GPU, and has been blacklisted.')
'is very slow on GPU, and has been denylisted.')
with self.test_scope():
data_format = 'channels_last'
conv = convolutional.Conv2D(
Expand Down
18 changes: 9 additions & 9 deletions tensorflow/compiler/tf2tensorrt/segment/segment.cc
Original file line number Diff line number Diff line change
Expand Up @@ -711,15 +711,15 @@ Status SegmentGraph(const Graph* tf_graph,
std::unordered_set<string> unsupported_ops;
int num_unsupported_ops = 0;

// Getting the operations blacklisted for conversion
string tftrt_op_blacklist_str;
// Getting the operations denylisted for conversion
string tftrt_op_denylist_str;
TF_CHECK_OK(
ReadStringFromEnvVar("TF_TRT_OP_BLACKLIST", "", &tftrt_op_blacklist_str));
ReadStringFromEnvVar("TF_TRT_OP_DENYLIST", "", &tftrt_op_denylist_str));

auto tftrt_op_blacklist = gtl::FlatSet<string>{}; // non-absl ok
auto tftrt_op_denylist = gtl::FlatSet<string>{}; // non-absl ok

for (const auto& x : str_util::Split(tftrt_op_blacklist_str, ",")) {
tftrt_op_blacklist.insert(x);
for (const auto& x : str_util::Split(tftrt_op_denylist_str, ",")) {
tftrt_op_denylist.insert(x);
}

// Parsing each node of the graph
Expand Down Expand Up @@ -761,13 +761,13 @@ Status SegmentGraph(const Graph* tf_graph,
const Status status = candidate_fn(node->tf_node());
if (!status.ok()) {
exclude_node(status.error_message());
} else if (tftrt_op_blacklist.count(node->tf_node()->type_string())) {
} else if (tftrt_op_denylist.count(node->tf_node()->type_string())) {
// WARNING verbosity since the user explicitly requests this behavior.
LOG_WARNING_WITH_PREFIX
<< "Blacklisted as TF-TRT candidate, "
<< "Denylisted as TF-TRT candidate, "
<< "(Op type: " << node->tf_node()->type_string() << "), "
<< "(Op name: " << node->name() << ")";
exclude_node("Blacklisted with the env var TF_TRT_OP_BLACKLIST");
exclude_node("Denylisted with the env var TF_TRT_OP_DENYLIST");
} else {
VLOG(2) << "Accepted as a TF-TRT candidate, "
<< "(Op type: " << node->tf_node()->type_string() << "), "
Expand Down
8 changes: 4 additions & 4 deletions tensorflow/compiler/xla/debug_options_flags.cc
Original file line number Diff line number Diff line change
Expand Up @@ -535,10 +535,10 @@ static void AllocateFlags() {
flag_values->xla_gpu_force_conv_nchw(),
"For cuDNN convolutions, always NCHW layouts."));
flag_objects->push_back(tensorflow::Flag(
"xla_gpu_algorithm_blacklist_path",
string_setter_for(&DebugOptions::set_xla_gpu_algorithm_blacklist_path),
flag_values->xla_gpu_algorithm_blacklist_path(),
"An AlgorithmBlacklist text proto file as a blacklist of convolutions to "
"xla_gpu_algorithm_denylist_path",
string_setter_for(&DebugOptions::set_xla_gpu_algorithm_denylist_path),
flag_values->xla_gpu_algorithm_denylist_path(),
"An AlgorithmDenylist text proto file as a denylist of convolutions to "
"avoid to use."));
flag_objects->push_back(tensorflow::Flag(
"xla_gpu_deterministic_reductions",
Expand Down
2 changes: 1 addition & 1 deletion tensorflow/compiler/xla/service/gpu/BUILD
Original file line number Diff line number Diff line change
Expand Up @@ -1676,7 +1676,7 @@ cc_library(
tf_cc_test(
name = "hlo_algorithm_blacklist_test",
srcs = ["hlo_algorithm_blacklist_test.cc"],
data = ["data/hlo_algorithm_blacklist.pbtxt"],
data = ["data/hlo_algorithm_denylist.pbtxt"],
tags = ["no_pip"],
deps = [
":hlo_algorithm_blacklist",
Expand Down
10 changes: 5 additions & 5 deletions tensorflow/compiler/xla/service/gpu/gpu_autotuning.proto
Original file line number Diff line number Diff line change
Expand Up @@ -15,19 +15,19 @@ message ConvInstructionLog {
repeated uint64 operand_addresses = 4;
}

message BlacklistedAlgorithm {
message DenylistedAlgorithm {
int64 id = 1;
bool tensor_ops = 2;
}

message AlgorithmBlacklistEntry {
message AlgorithmDenylistEntry {
string hlo = 1;
tensorflow.ComputeCapability cc = 2;
tensorflow.CudnnVersion cudnn_version = 3;
string blas_version = 5;
repeated BlacklistedAlgorithm algos = 4;
repeated DenylistedAlgorithm algos = 4;
}

message AlgorithmBlacklist {
repeated AlgorithmBlacklistEntry entries = 1;
message AlgorithmDenylist {
repeated AlgorithmDenylistEntry entries = 1;
}
24 changes: 11 additions & 13 deletions tensorflow/compiler/xla/service/gpu/gpu_conv_algorithm_picker.cc
Original file line number Diff line number Diff line change
Expand Up @@ -438,18 +438,17 @@ GpuConvAlgorithmPicker::PickBestAlgorithmNoCacheCuda(
(void)blas->GetVersion(&blas_version);
}

absl::Span<const AlgorithmDesc> blacklisted_algos =
GetBlacklistedConvAlgorithms(GetComputeCapability(stream_exec_),
GetCudnnVersion(stream_exec_), blas_version,
canonical_hlo);
absl::Span<const AlgorithmDesc> disabled_algos = GetDisabledConvAlgorithms(
GetComputeCapability(stream_exec_), GetCudnnVersion(stream_exec_),
blas_version, canonical_hlo);

for (const AlgorithmDesc& alg : GetAlgorithms(kind, stream_exec_)) {
XLA_SCOPED_LOGGING_TIMER_LEVEL(
absl::StrCat("CudnnConvAlgorithmPicker::PickBestAlgorithm algo ",
AlgorithmToString(alg)),
2);

if (absl::c_linear_search(blacklisted_algos, alg)) {
if (absl::c_linear_search(disabled_algos, alg)) {
LOG(INFO) << "Omitted potentially buggy algorithm "
<< AlgorithmToString(alg) << " for conv " << instr->ToString();
continue;
Expand Down Expand Up @@ -503,7 +502,7 @@ GpuConvAlgorithmPicker::PickBestAlgorithmNoCacheCuda(

if (!input_output_allocator_redzone_clear ||
!scratch_allocator_redzone_clear) {
AlgorithmBlacklist proto;
AlgorithmDenylist proto;
auto entry = proto.add_entries();
entry->set_hlo(canonical_hlo);
*entry->mutable_cc() = GetComputeCapability(stream_exec_);
Expand All @@ -513,13 +512,12 @@ GpuConvAlgorithmPicker::PickBestAlgorithmNoCacheCuda(
algo->set_id(alg.algo_id());
algo->set_tensor_ops(alg.tensor_ops_enabled());

LOG(ERROR)
<< "To blacklist this algorithm for this convolution, "
"copy-paste the following "
"proto to the blacklist file pointed by XLA_FLAGS "
"--xla_gpu_algorithm_blacklist_path="
<< GetDebugOptionsFromFlags().xla_gpu_algorithm_blacklist_path()
<< " : " << proto.ShortDebugString();
LOG(ERROR) << "To denylist this algorithm for this convolution, "
"copy-paste the following "
"proto to the denylist file pointed by XLA_FLAGS "
"--xla_gpu_algorithm_denylist_path="
<< GetDebugOptionsFromFlags().xla_gpu_algorithm_denylist_path()
<< " : " << proto.ShortDebugString();
continue;
}

Expand Down
22 changes: 10 additions & 12 deletions tensorflow/compiler/xla/service/gpu/hlo_algorithm_blacklist.cc
Original file line number Diff line number Diff line change
Expand Up @@ -24,7 +24,7 @@ limitations under the License.
namespace xla {
namespace gpu {

constexpr char kDefaultBlacklist[] = R"pb(
constexpr char kDefaultDenylist[] = R"pb(
entries {
hlo: "(f32[4,32,32,32]{2,1,3,0}, u8[0]{0}) custom-call(f32[4,32,32,32]{2,1,3,0}, f32[5,5,32,32]{1,0,2,3}), window={size=5x5 pad=2_2x2_2}, dim_labels=b01f_01io->b01f, custom_call_target=\"__cudnn$convForward\", backend_config=\"{conv_result_scale:1}\""
cc { major: 7 }
Expand All @@ -41,28 +41,26 @@ constexpr char kDefaultBlacklist[] = R"pb(
}
)pb";

absl::Span<const stream_executor::dnn::AlgorithmDesc>
GetBlacklistedConvAlgorithms(tensorflow::ComputeCapability cc,
tensorflow::CudnnVersion cudnn_version,
const std::string& blas_version,
const std::string& hlo) {
absl::Span<const stream_executor::dnn::AlgorithmDesc> GetDisabledConvAlgorithms(
tensorflow::ComputeCapability cc, tensorflow::CudnnVersion cudnn_version,
const std::string& blas_version, const std::string& hlo) {
// Key is the tuple of canonicalized hlo, compute capability major/minor,
// cudnn version major/minor/patch, blas version.
using MapType = absl::flat_hash_map<
std::tuple<std::string, int, int, int, int, int, std::string>,
std::vector<stream_executor::dnn::AlgorithmDesc>>;

static MapType* blacklist = [] {
static MapType* denylist = [] {
MapType* list = new MapType();
AlgorithmBlacklist proto;
AlgorithmDenylist proto;
std::string file_path =
GetDebugOptionsFromFlags().xla_gpu_algorithm_blacklist_path();
GetDebugOptionsFromFlags().xla_gpu_algorithm_denylist_path();
if (!file_path.empty()) {
TF_CHECK_OK(tensorflow::ReadTextProto(tensorflow::Env::Default(),
file_path, &proto));
} else {
CHECK(tensorflow::protobuf::TextFormat::ParseFromString(
std::string(kDefaultBlacklist), &proto));
std::string(kDefaultDenylist), &proto));
}
for (const auto& entry : proto.entries()) {
for (const auto& algo : entry.algos()) {
Expand All @@ -77,10 +75,10 @@ GetBlacklistedConvAlgorithms(tensorflow::ComputeCapability cc,
return list;
}();

auto iter = blacklist->find(std::make_tuple(
auto iter = denylist->find(std::make_tuple(
hlo, cc.major(), cc.minor(), cudnn_version.major(), cudnn_version.minor(),
cudnn_version.patch(), std::string(blas_version)));
if (iter != blacklist->end()) {
if (iter != denylist->end()) {
return iter->second;
}
return {};
Expand Down
14 changes: 6 additions & 8 deletions tensorflow/compiler/xla/service/gpu/hlo_algorithm_blacklist.h
Original file line number Diff line number Diff line change
Expand Up @@ -13,8 +13,8 @@ See the License for the specific language governing permissions and
limitations under the License.
==============================================================================*/

#ifndef TENSORFLOW_COMPILER_XLA_SERVICE_GPU_HLO_ALGORITHM_BLACKLIST_H_
#define TENSORFLOW_COMPILER_XLA_SERVICE_GPU_HLO_ALGORITHM_BLACKLIST_H_
#ifndef TENSORFLOW_COMPILER_XLA_SERVICE_GPU_HLO_ALGORITHM_DENYLIST_H_
#define TENSORFLOW_COMPILER_XLA_SERVICE_GPU_HLO_ALGORITHM_DENYLIST_H_

#include <vector>

Expand All @@ -24,13 +24,11 @@ limitations under the License.
namespace xla {
namespace gpu {

absl::Span<const stream_executor::dnn::AlgorithmDesc>
GetBlacklistedConvAlgorithms(tensorflow::ComputeCapability cc,
tensorflow::CudnnVersion cudnn_version,
const std::string& blas_version,
const std::string& hlo);
absl::Span<const stream_executor::dnn::AlgorithmDesc> GetDisabledConvAlgorithms(
tensorflow::ComputeCapability cc, tensorflow::CudnnVersion cudnn_version,
const std::string& blas_version, const std::string& hlo);

} // namespace gpu
} // namespace xla

#endif // TENSORFLOW_COMPILER_XLA_SERVICE_GPU_HLO_ALGORITHM_BLACKLIST_H_
#endif // TENSORFLOW_COMPILER_XLA_SERVICE_GPU_HLO_ALGORITHM_DENYLIST_H_
Original file line number Diff line number Diff line change
Expand Up @@ -26,30 +26,30 @@ namespace xla {
namespace gpu {
namespace {

class BlacklistTest : public testing::Test {
class DenylistTest : public testing::Test {
protected:
BlacklistTest() {
DenylistTest() {
tensorflow::setenv(
"XLA_FLAGS",
absl::StrCat(
"--xla_gpu_algorithm_blacklist_path=",
"--xla_gpu_algorithm_denylist_path=",
tensorflow::GetDataDependencyFilepath(tensorflow::io::JoinPath(
"tensorflow", "compiler", "xla", "service", "gpu", "data",
"hlo_algorithm_blacklist.pbtxt")))
"hlo_algorithm_denylist.pbtxt")))
.data(),
0);
}
};

TEST_F(BlacklistTest, DefaultTest) {
TEST_F(DenylistTest, DefaultTest) {
tensorflow::ComputeCapability cc;
cc.set_major(7);
cc.set_minor(0);
tensorflow::CudnnVersion cudnn_version;
cudnn_version.set_major(7);
cudnn_version.set_minor(6);
cudnn_version.set_patch(2);
auto list = GetBlacklistedConvAlgorithms(
auto list = GetDisabledConvAlgorithms(
cc, cudnn_version, /*blas_version=*/"9000",
R"((f16[256,112,112,64]{3,2,1,0}, u8[0]{0}) custom-call(f16[256,224,224,4]{3,2,1,0}, f16[7,7,4,64]{2,1,0,3}), window={size=7x7 stride=2x2 pad=3_3x3_3}, dim_labels=b01f_01io->b01f, custom_call_target="__cudnn$convForward", backend_config="{conv_result_scale:1}")");
ASSERT_EQ(4, list.size());
Expand All @@ -59,7 +59,7 @@ TEST_F(BlacklistTest, DefaultTest) {
EXPECT_EQ(stream_executor::dnn::AlgorithmDesc(1, true), list[3]);
}

TEST_F(BlacklistTest, NegativeTest) {
TEST_F(DenylistTest, NegativeTest) {
tensorflow::ComputeCapability cc;
cc.set_major(7);
cc.set_minor(0);
Expand All @@ -68,7 +68,7 @@ TEST_F(BlacklistTest, NegativeTest) {
cudnn_version.set_minor(6);
cudnn_version.set_minor(2);
auto list =
GetBlacklistedConvAlgorithms(cc, cudnn_version, "9000", R"(invalid hlo)");
GetDisabledConvAlgorithms(cc, cudnn_version, "9000", R"(invalid hlo)");
ASSERT_EQ(0, list.size());
}

Expand Down
Loading

0 comments on commit 18ebe82

Please sign in to comment.