/aosp_15_r20/external/tensorflow/tensorflow/compiler/xla/service/cpu/ |
H A D | cpu_compiler.cc | 751 const llvm::Module& llvm_module) { in GetIRModuleHooks() argument 755 user_hook(llvm_module); in GetIRModuleHooks() 757 llvm_ir::DumpIrIfEnabled(*hlo_module_ptr, llvm_module, optimized); in GetIRModuleHooks() 759 return {[hook](const llvm::Module& llvm_module) { in GetIRModuleHooks() 760 return hook(/*optimized=*/false, llvm_module); in GetIRModuleHooks() 762 [hook](const llvm::Module& llvm_module) { in GetIRModuleHooks() 763 return hook(/*optimized=*/true, llvm_module); in GetIRModuleHooks() 767 Status VerifyLlvmModule(const llvm::Module& llvm_module) { in VerifyLlvmModule() argument 774 TF_RET_CHECK(!llvm::verifyModule(llvm_module, &err_stream)) in VerifyLlvmModule() 1128 auto llvm_module = in CompileLegacyCpuExecutable() local [all …]
|
H A D | mlir_emitter.cc | 98 llvm::Module *llvm_module = b->GetInsertBlock()->getParent()->getParent(); in EmitMlirFuncAndCall() local 125 mlir_llvm_module->setDataLayout(llvm_module->getDataLayout()); in EmitMlirFuncAndCall() 127 *llvm_module, std::move(mlir_llvm_module), llvm::Linker::None, in EmitMlirFuncAndCall() 135 llvm::Function *func = llvm_module->getFunction(func_name); in EmitMlirFuncAndCall()
|
H A D | ir_function.cc | 30 llvm::Module* llvm_module, const int64_t num_dynamic_loop_bounds) { in GetComputeFunctionParams() argument 31 llvm::Type* i8_ptr_type = llvm::Type::getInt8PtrTy(llvm_module->getContext()); in GetComputeFunctionParams() 34 llvm::Type::getInt64PtrTy(llvm_module->getContext()); in GetComputeFunctionParams() 48 llvm::Module* llvm_module, llvm::IRBuilder<>* b, in IrFunction() argument 51 llvm_module_(llvm_module), in IrFunction()
|
H A D | ir_function.h | 57 const HloModuleConfig& module_config, llvm::Module* llvm_module,
|
H A D | ir_emitter.h | 88 const BufferAssignment& assignment, llvm::Module* llvm_module,
|
H A D | ir_emitter.cc | 98 llvm::Module* llvm_module, in IrEmitter() argument 108 module_(llvm_module), in IrEmitter() 109 arch_type_(llvm::Triple(llvm_module->getTargetTriple()).getArch()), in IrEmitter() 110 b_(llvm_module->getContext()), in IrEmitter() 116 alias_analysis_(hlo_module, assignment, &llvm_module->getContext()), in IrEmitter()
|
/aosp_15_r20/external/tensorflow/tensorflow/compiler/xla/service/gpu/ |
H A D | gpu_compiler.cc | 934 llvm::Module* llvm_module, in RemoveUnusedAndUninitializedGlobals() argument 941 llvm_module->getGlobalVariable(info.symbol_name); in RemoveUnusedAndUninitializedGlobals() 952 std::unique_ptr<llvm::Module> llvm_module; member 975 results->llvm_module = std::make_unique<llvm::Module>("", *llvm_context); in CompileModuleToLlvmIrImpl() 976 results->llvm_module->setTargetTriple(target_triple); in CompileModuleToLlvmIrImpl() 977 results->llvm_module->setDataLayout(data_layout); in CompileModuleToLlvmIrImpl() 1038 &mlir_context, results->llvm_module.get()); in CompileModuleToLlvmIrImpl() 1058 RemoveUnusedAndUninitializedGlobals(ir_emitter_context.llvm_module(), in CompileModuleToLlvmIrImpl() 1099 std::unique_ptr<llvm::Module> llvm_module, in CompileToTargetBinary() argument 1107 llvm::Module* llvm_module, bool relocatable, in CompileToTargetBinary() [all …]
|
H A D | ir_emitter_context.h | 44 mlir::MLIRContext* mlir_context, llvm::Module* llvm_module) in IrEmitterContext() argument 52 llvm_module_(llvm_module) {} in IrEmitterContext() 71 llvm::Module* llvm_module() { return llvm_module_; } in llvm_module() function
|
H A D | nvptx_compiler.cc | 232 llvm::Module* llvm_module) { in MaybeLoadLLVMFromFile() argument 257 llvm::LLVMContext& context = llvm_module->getContext(); in MaybeLoadLLVMFromFile() 329 llvm::Module* llvm_module, in CompileTargetBinary() argument 348 MaybeLoadLLVMFromFile(debug_module, llvm_module); in CompileTargetBinary() 353 selected_module = llvm_module; in CompileTargetBinary()
|
H A D | gpu_compiler.h | 105 std::unique_ptr<llvm::Module> llvm_module, 144 llvm::Module* llvm_module, GpuVersion gpu_version, 193 std::unique_ptr<llvm::Module> llvm_module,
|
H A D | ir_emitter_nested.cc | 86 root_shape, ir_emitter_context_->llvm_module()->getDataLayout()); in CodegenNestedComputation() 98 ir_emitter_context_->llvm_module()); // The parent LLVM module. in CodegenNestedComputation() 232 ir_emitter_context_->llvm_module()->getGlobalList().push_back( in EmitConstants()
|
H A D | hlo_to_ir_bindings.h | 34 HloToIrBindings(llvm::IRBuilder<>* b, llvm::Module* llvm_module, in HloToIrBindings() argument 36 : is_nested_(is_nested), b_(b), module_(llvm_module) {} in HloToIrBindings()
|
H A D | amdgpu_compiler.cc | 135 llvm::Module* llvm_module, in CompileTargetBinary() argument 154 hsaco, amdgpu::CompileToHsaco(llvm_module, gpu_version, module_config, in CompileTargetBinary()
|
H A D | amdgpu_compiler.h | 47 const HloModuleConfig& module_config, llvm::Module* llvm_module,
|
H A D | nvptx_compiler.h | 53 const HloModuleConfig& module_config, llvm::Module* llvm_module,
|
H A D | ir_emitter_unnested.cc | 158 llvm::Module* llvm_module) { in AnnotateWithInt32Value() argument 160 llvm_module->getOrInsertNamedMetadata("nvvm.annotations"); in AnnotateWithInt32Value() 161 llvm::Function* ir_kernel = llvm_module->getFunction(kernel_name.c_str()); in AnnotateWithInt32Value() 162 llvm::LLVMContext& llvm_context = llvm_module->getContext(); in AnnotateWithInt32Value() 176 llvm::Module* llvm_module) { in AnnotateThunkLaunchDimensions() argument 183 kernel_name, llvm_module); in AnnotateThunkLaunchDimensions() 186 kernel_name, llvm_module); in AnnotateThunkLaunchDimensions() 190 kernel_name, llvm_module); in AnnotateThunkLaunchDimensions() 1601 std::unique_ptr<llvm::Module> llvm_module = mlir::translateModuleToLLVMIR( in EmitLaunchFunc() local 1603 if (!llvm_module) in EmitLaunchFunc() [all …]
|
/aosp_15_r20/external/tensorflow/tensorflow/compiler/xla/service/gpu/tests/ |
H A D | mlir_gpu_test_base.cc | 48 auto llvm_module = std::make_unique<llvm::Module>("", llvm_context); in CompileMlirModule() local 50 llvm_module->setTargetTriple(amdgpu::TargetTriple()); in CompileMlirModule() 51 llvm_module->setDataLayout(amdgpu::DataLayout()); in CompileMlirModule() 53 llvm_module->setTargetTriple(nvptx::TargetTriple()); in CompileMlirModule() 54 llvm_module->setDataLayout(nvptx::DataLayout()); in CompileMlirModule() 64 /*mlir_context=*/nullptr, llvm_module.get()); in CompileMlirModule() 71 std::move(llvm_module), &ir_emitter_context); in CompileMlirModule()
|
H A D | hlo_to_llvm_ir.cc | 76 TF_ASSIGN_OR_RETURN(std::unique_ptr<llvm::Module> llvm_module, in CompileAndPrintLlvmIr() 87 llvm_module->print(llvm::outs(), nullptr); in CompileAndPrintLlvmIr() 93 llvm_module.get(), cuda_compute_capability, in CompileAndPrintLlvmIr()
|
/aosp_15_r20/external/tensorflow/tensorflow/compiler/mlir/tools/kernel_gen/ |
H A D | tf_to_kernel.cc | 77 std::unique_ptr<llvm::Module> llvm_module = in EmitToBinary() local 80 auto target_machine = GetTargetMachine(llvm_module.get()); in EmitToBinary() 81 llvm_module->setDataLayout(target_machine->createDataLayout()); in EmitToBinary() 86 target_machine.get())(llvm_module.get())) { in EmitToBinary() 97 llvm::Triple(llvm_module->getTargetTriple()))); in EmitToBinary() 103 codegen_passes.run(*llvm_module); in EmitToBinary()
|
/aosp_15_r20/external/mesa3d/src/amd/vulkan/ |
H A D | radv_nir_to_llvm.c | 433 ac_compile_llvm_module(struct ac_llvm_compiler *ac_llvm, LLVMModuleRef llvm_module, struct radv_sha… in ac_compile_llvm_module() argument 442 ac_dump_module(llvm_module); in ac_compile_llvm_module() 447 char *llvm_ir = LLVMPrintModuleToString(llvm_module); in ac_compile_llvm_module() 452 int v = radv_llvm_compile(llvm_module, &elf_buffer, &elf_size, ac_llvm); in ac_compile_llvm_module() 457 LLVMContextRef ctx = LLVMGetModuleContext(llvm_module); in ac_compile_llvm_module() 458 LLVMDisposeModule(llvm_module); in ac_compile_llvm_module() 484 LLVMModuleRef llvm_module; in radv_compile_nir_shader() local 486 llvm_module = ac_translate_nir_to_llvm(ac_llvm, options, info, nir, nir_count, args); in radv_compile_nir_shader() 488 …ac_compile_llvm_module(ac_llvm, llvm_module, rbinary, radv_get_shader_name(info, nir[nir_count - 1… in radv_compile_nir_shader()
|
/aosp_15_r20/external/tensorflow/tensorflow/compiler/mlir/tfrt/transforms/lmhlo_to_gpu/ |
H A D | kernel_ops_pattern.cc | 220 const xla::HloModuleConfig& hlo_module_config, llvm::Module* llvm_module) { in Emit() argument 229 llvm_module->setTargetTriple(target_triple); in Emit() 230 llvm_module->setDataLayout(data_layout); in Emit() 236 func_op->getContext(), llvm_module); in Emit() 281 auto llvm_module = std::make_unique<llvm::Module>("", llvm_context); in Match() local 284 gpu_options, hlo_module_config, llvm_module.get()); in Match() 313 llvm_module.get(), gpu_version, hlo_module_config, libdevice_dir); in Match() 318 auto ptx = xla::gpu::nvptx::CompileToPtx(llvm_module.get(), in Match()
|
/aosp_15_r20/external/tensorflow/tensorflow/compiler/xla/mlir/transforms/runtime/ |
H A D | jit_compiler.cc | 255 auto llvm_module = translateModuleToLLVMIR(compiler->module(), *llvm_ctx); in Compile() local 256 if (!llvm_module) in Compile() 261 std::move(llvm_ctx), std::move(llvm_module), entrypoint, engine_options); in Compile()
|
/aosp_15_r20/external/tensorflow/tensorflow/compiler/xla/service/llvm_ir/ |
H A D | llvm_util.cc | 585 const llvm::Module& llvm_module, bool optimized, in DumpIrIfEnabled() argument 598 DumpModuleToString(llvm_module)); in DumpIrIfEnabled() 606 DumpModuleToString(*DropConstantInitializers(llvm_module))); in DumpIrIfEnabled()
|
H A D | llvm_util.h | 271 const llvm::Module& llvm_module, bool optimized,
|
/aosp_15_r20/external/tensorflow/tensorflow/compiler/xla/tests/ |
H A D | llvm_compiler_test.cc | 63 const HloModuleConfig& module_config, llvm::Module* llvm_module, in CompileTargetBinary() argument
|