Skip to content

Commit

Permalink
Disentangle CHLO dialect from MHLO dialect
Browse files Browse the repository at this point in the history 
At the moment, CHLO dialect partially depends on MHLO dialect. In preparation for the migration from MLIR-HLO's CHLO to StableHLO's CHLO, which is part of bootstrapping of StableHLO, this CL severs the dependency from CHLO to MHLO.

Concretely:
  1) CHLO got its own attributes and chlo.constant. This stuff is not something that can be effectively shared across dialects. In the past, CHLO has been using MHLO attributes and mhlo.constant.
  2) Functionality shared between CHLO and MHLO dialects has been factored into base.{cc,h,td}. Some of the stuff from hlo_base* went there, some of the stuff went into hlo_ops_attrs.td and hlo_ops_enums.td.
  3) Functionality shared between MHLO and LMHLO dialects has been factored into hlo_ops_common.td, which is a newly created file that keeps company with hlo_ops_common.{cc,h}.
  4) Various fixups were applied to adapt the codebase to these changes.
PiperOrigin-RevId: 469400940
  • Loading branch information
Eugene Burmako authored and TensorFlow MLIR Team committed Aug 23, 2022
1 parent 4a15463 commit 257add0
Show file tree
Hide file tree
Showing 29 changed files with 757 additions and 377 deletions.
96 changes: 70 additions & 26 deletions BUILD
View file
Original file line number Diff line number Diff line change
Expand Up @@ -19,6 +19,24 @@ filegroup(
srcs = glob(["include/mlir-hlo/Dialect/mhlo/IR/*.td"]),
)

gentbl_cc_library(
name = "base_attr_interfaces_inc_gen",
strip_include_prefix = "include",
tbl_outs = [
(
["-gen-attr-interface-decls"],
"include/mlir-hlo/Dialect/mhlo/IR/base_attr_interfaces.h.inc",
),
(
["-gen-attr-interface-defs"],
"include/mlir-hlo/Dialect/mhlo/IR/base_attr_interfaces.cc.inc",
),
],
tblgen = "@llvm-project//mlir:mlir-tblgen",
td_file = "include/mlir-hlo/Dialect/mhlo/IR/base.td",
deps = [":hlo_ops_td_files"],
)

td_library(
name = "hlo_ops_td_files",
srcs = glob(["include/mlir-hlo/Dialect/mhlo/IR/*.td"]),
Expand Down Expand Up @@ -95,72 +113,88 @@ gentbl_cc_library(
)

gentbl_cc_library(
name = "hlo_ops_inc_gen",
strip_include_prefix = "include",
name = "chlo_ops_attrs_inc_gen",
tbl_outs = [
(
["-gen-op-decls"],
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.h.inc",
["-gen-attrdef-decls"],
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops_attrs.h.inc",
),
(
["-gen-op-defs"],
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.cc.inc",
["-gen-attrdef-defs"],
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops_attrs.cc.inc",
),
],
tblgen = "@llvm-project//mlir:mlir-tblgen",
td_file = "include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.td",
td_file = "include/mlir-hlo/Dialect/mhlo/IR/chlo_ops.td",
deps = [":hlo_ops_td_files"],
)

gentbl_cc_library(
name = "hlo_ops_base_inc_gen",
name = "chlo_ops_enums_inc_gen",
tbl_outs = [
(
["-gen-enum-decls"],
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops_enums.h.inc",
),
(
["-gen-enum-defs"],
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops_enums.cc.inc",
),
],
tblgen = "@llvm-project//mlir:mlir-tblgen",
td_file = "include/mlir-hlo/Dialect/mhlo/IR/chlo_ops.td",
deps = [":hlo_ops_td_files"],
)

gentbl_cc_library(
name = "hlo_ops_inc_gen",
strip_include_prefix = "include",
tbl_outs = [
(
["-gen-op-decls"],
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base.h.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.h.inc",
),
(
["-gen-op-defs"],
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.cc.inc",
),
],
tblgen = "@llvm-project//mlir:mlir-tblgen",
td_file = "include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base.td",
td_file = "include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.td",
deps = [":hlo_ops_td_files"],
)

gentbl_cc_library(
name = "hlo_ops_base_attrs_inc_gen",
name = "hlo_ops_attrs_inc_gen",
tbl_outs = [
(
["-gen-attrdef-decls"],
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base_attrs.h.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_attrs.h.inc",
),
(
["-gen-attrdef-defs"],
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base_attrs.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_attrs.cc.inc",
),
],
tblgen = "@llvm-project//mlir:mlir-tblgen",
td_file = "include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base.td",
td_file = "include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.td",
deps = [":hlo_ops_td_files"],
)

gentbl_cc_library(
name = "hlo_ops_base_enums_inc_gen",
name = "hlo_ops_enums_inc_gen",
tbl_outs = [
(
["-gen-enum-decls"],
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base_enums.h.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_enums.h.inc",
),
(
["-gen-enum-defs"],
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base_enums.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_enums.cc.inc",
),
],
tblgen = "@llvm-project//mlir:mlir-tblgen",
td_file = "include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base.td",
td_file = "include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.td",
deps = [":hlo_ops_td_files"],
)

Expand Down Expand Up @@ -425,32 +459,42 @@ cc_library(
cc_library(
name = "mlir_hlo",
srcs = [
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops.h.inc",
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops_attrs.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops_attrs.h.inc",
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops_enums.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops_enums.h.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.h.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base_attrs.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base_attrs.h.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base_enums.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base_enums.h.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_attrs.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_attrs.h.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_enums.cc.inc",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_enums.h.inc",
"lib/Dialect/mhlo/IR/base.cc",
"lib/Dialect/mhlo/IR/chlo_ops.cc",
"lib/Dialect/mhlo/IR/hlo_ops.cc",
"lib/utils/broadcast_utils.cc",
"lib/utils/hlo_utils.cc",
],
hdrs = [
"include/mlir-hlo/Dialect/mhlo/IR/base.h",
"include/mlir-hlo/Dialect/mhlo/IR/chlo_ops.h",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops.h",
"include/mlir-hlo/Dialect/mhlo/IR/hlo_ops_base.h",
"include/mlir-hlo/utils/broadcast_utils.h",
"include/mlir-hlo/utils/hlo_utils.h",
],
includes = ["include"],
deps = [
":base_attr_interfaces_inc_gen",
":canonicalize_inc_gen",
":chlo_ops_attrs_inc_gen",
":chlo_ops_enums_inc_gen",
":chlo_ops_inc_gen",
":convert_op_folder",
":hlo_ops_base_attrs_inc_gen",
":hlo_ops_base_inc_gen",
":hlo_ops_attrs_inc_gen",
":hlo_ops_common",
":hlo_ops_enums_inc_gen",
":hlo_ops_inc_gen",
":hlo_ops_pattern_gen",
"@llvm-project//llvm:Support",
Expand Down
2 changes: 1 addition & 1 deletion include/mlir-hlo/Dialect/lhlo/IR/lhlo_ops_base.td
View file
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@ limitations under the License.

include "mlir/Dialect/MemRef/IR/MemRefBase.td"
include "mlir/IR/OpBase.td"
include "mlir-hlo/Dialect/mhlo/IR/hlo_ops_base.td"
include "mlir-hlo/Dialect/mhlo/IR/hlo_ops_common.td"

//===----------------------------------------------------------------------===//
// LMHLO type definitions.
Expand Down
4 changes: 2 additions & 2 deletions include/mlir-hlo/Dialect/lhlo_gpu/IR/lhlo_gpu_ops.td
View file
Original file line number Diff line number Diff line change
Expand Up @@ -21,15 +21,15 @@ limitations under the License.

include "mlir/IR/OpBase.td"
include "mlir/Interfaces/SideEffectInterfaces.td"
include "mlir-hlo/Dialect/mhlo/IR/hlo_ops_base.td"
include "mlir-hlo/Dialect/mhlo/IR/base.td"
include "mlir-hlo/Dialect/lhlo/IR/lhlo_ops_base.td"
include "mlir-hlo/Dialect/lhlo_gpu/IR/lhlo_gpu_ops_base.td"
include "mlir-hlo/Dialect/lhlo_gpu/IR/lhlo_gpu_ops_enums.td"

class LHLOGPU_Op<string mnemonic, list<Trait> traits = []> :
Op<LmhloGpuDialect, mnemonic,
!listconcat([MemoryEffects<[MemRead, MemWrite]>], traits)>;

// Type for scratch buffers used by GPU library calls (memref<?xi8>)
def UntypedBuffer : MemRefRankOf<[I8], [1]>;

Expand Down
17 changes: 13 additions & 4 deletions include/mlir-hlo/Dialect/mhlo/IR/CMakeLists.txt
View file
Original file line number Diff line number Diff line change
Expand Up @@ -14,10 +14,19 @@
# limitations under the License.
#
# Need a separate function because of the .cc vs .cpp used in the one provided by MLIR
set(LLVM_TARGET_DEFINITIONS base.td)
mlir_tablegen(base_attr_interfaces.h.inc -gen-attr-interface-decls)
mlir_tablegen(base_attr_interfaces.cc.inc -gen-attr-interface-defs)
add_public_tablegen_target(MLIRHloBaseIncGen)

function(add_mlir_hlo_dialect dialect)
set(LLVM_TARGET_DEFINITIONS ${dialect}.td)
mlir_tablegen(${dialect}.h.inc -gen-op-decls)
mlir_tablegen(${dialect}.cc.inc -gen-op-defs)
mlir_tablegen(${dialect}_enums.h.inc -gen-enum-decls)
mlir_tablegen(${dialect}_enums.cc.inc -gen-enum-defs)
mlir_tablegen(${dialect}_attrs.h.inc -gen-attrdef-decls)
mlir_tablegen(${dialect}_attrs.cc.inc -gen-attrdef-defs)
add_public_tablegen_target(MLIR${dialect}IncGen)
add_dependencies(mlir-headers MLIR${dialect}IncGen)
endfunction()
Expand All @@ -27,8 +36,8 @@ add_mlir_hlo_dialect(chlo_ops)
set(LLVM_TARGET_DEFINITIONS hlo_ops.td)
mlir_tablegen(hlo_ops.h.inc -gen-op-decls)
mlir_tablegen(hlo_ops.cc.inc -gen-op-defs)
mlir_tablegen(hlo_ops_base_enums.h.inc -gen-enum-decls)
mlir_tablegen(hlo_ops_base_enums.cc.inc -gen-enum-defs)
mlir_tablegen(hlo_ops_base_attrs.h.inc -gen-attrdef-decls)
mlir_tablegen(hlo_ops_base_attrs.cc.inc -gen-attrdef-defs)
mlir_tablegen(hlo_ops_enums.h.inc -gen-enum-decls)
mlir_tablegen(hlo_ops_enums.cc.inc -gen-enum-defs)
mlir_tablegen(hlo_ops_attrs.h.inc -gen-attrdef-decls)
mlir_tablegen(hlo_ops_attrs.cc.inc -gen-attrdef-defs)
add_public_tablegen_target(MLIRhlo_opsIncGen)
Loading

0 comments on commit 257add0

Please sign in to comment.