OpenE2K/rust
2
0
Fork 0
Code Issues Pull Requests Projects Releases Wiki Activity

Auto merge of #58791 - denzp:asm-compile-tests, r=alexcrichton

Browse Source 
Introduce assembly tests suite

The change introduces a new test suite - **Assembly** tests. The motivation behind this is an ability to perform end-to-end codegen testing with LLVM backend. Turned out, NVPTX backend sometimes missing common Rust features (`i128` and libcalls in the past, and still full atomics support) due to different reasons.

Prior to this change, basic NVPTX assembly tests were implemented within `run-make` suite. Now, it's easier to write additional and maintain existing tests for the target.

cc @gnzlbg @peterhj
cc @eddyb I adjusted mangling scheme expectation, so there is no need to change the tests for #57967
This commit is contained in:
bors 2019-03-20 17:54:27 +00:00
commit 82e2f3ec25
23 changed files with 328 additions and 141 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

1
src/bootstrap/builder.rs
View File

@ -374,6 +374,7 @@ impl<'a> Builder<'a> {
test::MirOpt,
test::Codegen,
test::CodegenUnits,
test::Assembly,
test::Incremental,
test::Debuginfo,
test::UiFullDeps,

6
src/bootstrap/test.rs
View File

@ -936,6 +936,12 @@ host_test!(RunMakeFullDeps {
suite: "run-make-fulldeps"
});
default_test!(Assembly {
path: "src/test/assembly",
mode: "assembly",
suite: "assembly"
});
#[derive(Debug, Copy, Clone, PartialEq, Eq, Hash)]
struct Compiletest {
compiler: Compiler,

3
src/ci/docker/test-various/Dockerfile
View File

@ -53,7 +53,8 @@ ENV WASM_SCRIPT python2.7 /checkout/x.py test --target $WASM_TARGETS \
ENV NVPTX_TARGETS=nvptx64-nvidia-cuda
ENV NVPTX_SCRIPT python2.7 /checkout/x.py test --target $NVPTX_TARGETS \
src/test/run-make
src/test/run-make \
src/test/assembly
ENV MUSL_TARGETS=x86_64-unknown-linux-musl \
CC_x86_64_unknown_linux_musl=x86_64-linux-musl-gcc \

8
src/test/assembly/auxiliary/breakpoint-panic-handler.rs Normal file
View File

@ -0,0 +1,8 @@
#![feature(core_intrinsics)]
#![no_std]
#[panic_handler]
unsafe fn breakpoint_panic_handler(_: &::core::panic::PanicInfo) -> ! {
core::intrinsics::breakpoint();
core::hint::unreachable_unchecked();
}

0
src/test/run-make/nvptx-dylib-crate/dep.rs → src/test/assembly/auxiliary/non-inline-dependency.rs
View File

12
src/test/assembly/nvptx-arch-default.rs Normal file
View File

@ -0,0 +1,12 @@
// assembly-output: ptx-linker
// compile-flags: --crate-type cdylib
// only-nvptx64
#![no_std]
// aux-build: breakpoint-panic-handler.rs
extern crate breakpoint_panic_handler;
// Verify default target arch with ptx-linker.
// CHECK: .target sm_30
// CHECK: .address_size 64

9
src/test/assembly/nvptx-arch-emit-asm.rs Normal file
View File

@ -0,0 +1,9 @@
// assembly-output: emit-asm
// compile-flags: --crate-type rlib
// only-nvptx64
#![no_std]
// Verify default arch without ptx-linker involved.
// CHECK: .target sm_30
// CHECK: .address_size 64

12
src/test/assembly/nvptx-arch-link-arg.rs Normal file
View File

@ -0,0 +1,12 @@
// assembly-output: ptx-linker
// compile-flags: --crate-type cdylib -C link-arg=--arch=sm_60
// only-nvptx64
#![no_std]
// aux-build: breakpoint-panic-handler.rs
extern crate breakpoint_panic_handler;
// Verify target arch override via `link-arg`.
// CHECK: .target sm_60
// CHECK: .address_size 64

12
src/test/assembly/nvptx-arch-target-cpu.rs Normal file
View File

@ -0,0 +1,12 @@
// assembly-output: ptx-linker
// compile-flags: --crate-type cdylib -C target-cpu=sm_50
// only-nvptx64
#![no_std]
// aux-build: breakpoint-panic-handler.rs
extern crate breakpoint_panic_handler;
// Verify target arch override via `target-cpu`.
// CHECK: .target sm_50
// CHECK: .address_size 64

85
src/test/assembly/nvptx-atomics.rs Normal file
View File

@ -0,0 +1,85 @@
// assembly-output: ptx-linker
// compile-flags: --crate-type cdylib
// only-nvptx64
#![feature(abi_ptx, core_intrinsics)]
#![no_std]
use core::intrinsics::*;
// aux-build: breakpoint-panic-handler.rs
extern crate breakpoint_panic_handler;
// Currently, LLVM NVPTX backend can only emit atomic instructions with
// `relaxed` (PTX default) ordering. But it's also useful to make sure
// the backend won't fail with other orders. Apparently, the backend
// doesn't support fences as well. As a workaround `llvm.nvvm.membar.*`
// could work, and perhaps on the long run, all the atomic operations
// should rather be provided by `core::arch::nvptx`.
// Also, PTX ISA doesn't have atomic `load`, `store` and `nand`.
// FIXME(denzp): add tests for `core::sync::atomic::*`.
#[no_mangle]
pub unsafe extern "ptx-kernel" fn atomics_kernel(a: *mut u32) {
// CHECK: atom.global.and.b32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
// CHECK: atom.global.and.b32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
atomic_and(a, 1);
atomic_and_relaxed(a, 1);
// CHECK: atom.global.cas.b32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1, 2;
// CHECK: atom.global.cas.b32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1, 2;
atomic_cxchg(a, 1, 2);
atomic_cxchg_relaxed(a, 1, 2);
// CHECK: atom.global.max.s32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
// CHECK: atom.global.max.s32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
atomic_max(a, 1);
atomic_max_relaxed(a, 1);
// CHECK: atom.global.min.s32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
// CHECK: atom.global.min.s32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
atomic_min(a, 1);
atomic_min_relaxed(a, 1);
// CHECK: atom.global.or.b32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
// CHECK: atom.global.or.b32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
atomic_or(a, 1);
atomic_or_relaxed(a, 1);
// CHECK: atom.global.max.u32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
// CHECK: atom.global.max.u32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
atomic_umax(a, 1);
atomic_umax_relaxed(a, 1);
// CHECK: atom.global.min.u32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
// CHECK: atom.global.min.u32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
atomic_umin(a, 1);
atomic_umin_relaxed(a, 1);
// CHECK: atom.global.add.u32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
// CHECK: atom.global.add.u32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
atomic_xadd(a, 1);
atomic_xadd_relaxed(a, 1);
// CHECK: atom.global.exch.b32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
// CHECK: atom.global.exch.b32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
atomic_xchg(a, 1);
atomic_xchg_relaxed(a, 1);
// CHECK: atom.global.xor.b32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
// CHECK: atom.global.xor.b32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], 1;
atomic_xor(a, 1);
atomic_xor_relaxed(a, 1);
// CHECK: mov.u32 %[[sub_0_arg:r[0-9]+]], 100;
// CHECK: neg.s32 temp, %[[sub_0_arg]];
// CHECK: atom.global.add.u32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], temp;
atomic_xsub(a, 100);
// CHECK: mov.u32 %[[sub_1_arg:r[0-9]+]], 200;
// CHECK: neg.s32 temp, %[[sub_1_arg]];
// CHECK: atom.global.add.u32 %{{r[0-9]+}}, [%{{rd[0-9]+}}], temp;
atomic_xsub_relaxed(a, 200);
}

27
src/test/assembly/nvptx-internalizing.rs Normal file
View File

@ -0,0 +1,27 @@
// assembly-output: ptx-linker
// compile-flags: --crate-type cdylib
// only-nvptx64
#![feature(abi_ptx)]
#![no_std]
// aux-build: breakpoint-panic-handler.rs
extern crate breakpoint_panic_handler;
// aux-build: non-inline-dependency.rs
extern crate non_inline_dependency as dep;
// Verify that no extra function declarations are present.
// CHECK-NOT: .func
// CHECK: .visible .entry top_kernel(
#[no_mangle]
pub unsafe extern "ptx-kernel" fn top_kernel(a: *const u32, b: *mut u32) {
// CHECK: add.s32 %{{r[0-9]+}}, %{{r[0-9]+}}, 5;
*b = *a + 5;
}
// Verify that no extra function definitions are here.
// CHECK-NOT: .func
// CHECK-NOT: .entry

39
src/test/assembly/nvptx-linking-binary.rs Normal file
View File

@ -0,0 +1,39 @@
// assembly-output: ptx-linker
// compile-flags: --crate-type bin
// only-nvptx64
#![feature(abi_ptx)]
#![no_main]
#![no_std]
// aux-build: breakpoint-panic-handler.rs
extern crate breakpoint_panic_handler;
// aux-build: non-inline-dependency.rs
extern crate non_inline_dependency as dep;
// Make sure declarations are there.
// CHECK: .func (.param .b32 func_retval0) wrapping_external_fn
// CHECK: .func (.param .b32 func_retval0) panicking_external_fn
// CHECK-LABEL: .visible .entry top_kernel(
#[no_mangle]
pub unsafe extern "ptx-kernel" fn top_kernel(a: *const u32, b: *mut u32) {
// CHECK: call.uni (retval0),
// CHECK-NEXT: wrapping_external_fn
// CHECK: ld.param.b32 %[[LHS:r[0-9]+]], [retval0+0];
let lhs = dep::wrapping_external_fn(*a);
// CHECK: call.uni (retval0),
// CHECK-NEXT: panicking_external_fn
// CHECK: ld.param.b32 %[[RHS:r[0-9]+]], [retval0+0];
let rhs = dep::panicking_external_fn(*a);
// CHECK: add.s32 %[[RES:r[0-9]+]], %[[RHS]], %[[LHS]];
// CHECK: st.global.u32 [%{{rd[0-9]+}}], %[[RES]];
*b = lhs + rhs;
}
// Verify that external function bodies are available.
// CHECK: .func (.param .b32 func_retval0) wrapping_external_fn
// CHECK: .func (.param .b32 func_retval0) panicking_external_fn

38
src/test/assembly/nvptx-linking-cdylib.rs Normal file
View File

@ -0,0 +1,38 @@
// assembly-output: ptx-linker
// compile-flags: --crate-type cdylib
// only-nvptx64
#![feature(abi_ptx)]
#![no_std]
// aux-build: breakpoint-panic-handler.rs
extern crate breakpoint_panic_handler;
// aux-build: non-inline-dependency.rs
extern crate non_inline_dependency as dep;
// Make sure declarations are there.
// CHECK: .func (.param .b32 func_retval0) wrapping_external_fn
// CHECK: .func (.param .b32 func_retval0) panicking_external_fn
// CHECK-LABEL: .visible .entry top_kernel(
#[no_mangle]
pub unsafe extern "ptx-kernel" fn top_kernel(a: *const u32, b: *mut u32) {
// CHECK: call.uni (retval0),
// CHECK-NEXT: wrapping_external_fn
// CHECK: ld.param.b32 %[[LHS:r[0-9]+]], [retval0+0];
let lhs = dep::wrapping_external_fn(*a);
// CHECK: call.uni (retval0),
// CHECK-NEXT: panicking_external_fn
// CHECK: ld.param.b32 %[[RHS:r[0-9]+]], [retval0+0];
let rhs = dep::panicking_external_fn(*a);
// CHECK: add.s32 %[[RES:r[0-9]+]], %[[RHS]], %[[LHS]];
// CHECK: st.global.u32 [%{{rd[0-9]+}}], %[[RES]];
*b = lhs + rhs;
}
// Verify that external function bodies are available.
// CHECK: .func (.param .b32 func_retval0) wrapping_external_fn
// CHECK: .func (.param .b32 func_retval0) panicking_external_fn

22
src/test/run-make/nvptx-emit-asm/kernel.rs → src/test/assembly/nvptx-safe-naming.rs
View File

@ -1,13 +1,15 @@
#![no_std]
#![deny(warnings)]
#![feature(abi_ptx)]
// assembly-output: ptx-linker
// compile-flags: --crate-type cdylib
// only-nvptx64
// Verify the default CUDA arch.
// CHECK: .target sm_30
// CHECK: .address_size 64
#![feature(abi_ptx)]
#![no_std]
// aux-build: breakpoint-panic-handler.rs
extern crate breakpoint_panic_handler;
// Verify function name doesn't contain unacceaptable characters.
// CHECK: .func (.param .b32 func_retval0) [[IMPL_FN:_ZN[a-zA-Z0-9$_]+square[a-zA-Z0-9$_]+]]
// CHECK: .func (.param .b32 func_retval0) [[IMPL_FN:[a-zA-Z0-9$_]+square[a-zA-Z0-9$_]+]](
// CHECK-LABEL: .visible .entry top_kernel(
#[no_mangle]
@ -33,9 +35,3 @@ pub mod deep {
}
}
}
// Verify that external function bodies are available.
// CHECK: .func (.param .b32 func_retval0) [[IMPL_FN]]
// CHECK: {
// CHECK: mul.lo.s32 %{{r[0-9]+}}, %{{r[0-9]+}}, %{{r[0-9]+}}
// CHECK: }

12
src/test/run-make/nvptx-binary-crate/Makefile
View File

@ -1,12 +0,0 @@
-include ../../run-make-fulldeps/tools.mk
ifeq ($(TARGET),nvptx64-nvidia-cuda)
all:
$(RUSTC) main.rs --crate-type="bin" --target $(TARGET) -O -C link-arg=--arch=sm_60 -o $(TMPDIR)/main.link_arg.ptx
$(RUSTC) main.rs --crate-type="bin" --target $(TARGET) -O -C target-cpu=sm_60 -o $(TMPDIR)/main.target_cpu.ptx
FileCheck main.rs --input-file $(TMPDIR)/main.link_arg.ptx
FileCheck main.rs --input-file $(TMPDIR)/main.target_cpu.ptx
else
all:
endif

28
src/test/run-make/nvptx-binary-crate/main.rs
View File

@ -1,28 +0,0 @@
#![no_std]
#![no_main]
#![deny(warnings)]
#![feature(abi_ptx, core_intrinsics)]
// Check the overriden CUDA arch.
// CHECK: .target sm_60
// CHECK: .address_size 64
// Verify that no extra function declarations are present.
// CHECK-NOT: .func
// CHECK-LABEL: .visible .entry top_kernel(
#[no_mangle]
pub unsafe extern "ptx-kernel" fn top_kernel(a: *const u32, b: *mut u32) {
// CHECK: add.s32 %{{r[0-9]+}}, %{{r[0-9]+}}, 5;
*b = *a + 5;
}
// Verify that no extra function definitions are there.
// CHECK-NOT: .func
// CHECK-NOT: .entry
#[panic_handler]
unsafe fn breakpoint_panic_handler(_: &::core::panic::PanicInfo) -> ! {
core::intrinsics::breakpoint();
core::hint::unreachable_unchecked();
}

10
src/test/run-make/nvptx-dylib-crate/Makefile
View File

@ -1,10 +0,0 @@
-include ../../run-make-fulldeps/tools.mk
ifeq ($(TARGET),nvptx64-nvidia-cuda)
all:
$(RUSTC) dep.rs --crate-type="rlib" --target $(TARGET)
$(RUSTC) kernel.rs --crate-type="cdylib" -O --target $(TARGET)
FileCheck kernel.rs --input-file $(TMPDIR)/kernel.ptx
else
all:
endif

59
src/test/run-make/nvptx-dylib-crate/kernel.rs
View File

@ -1,59 +0,0 @@
#![no_std]
#![deny(warnings)]
#![feature(abi_ptx, core_intrinsics)]
extern crate dep;
// Verify the default CUDA arch.
// CHECK: .target sm_30
// CHECK: .address_size 64
// Make sure declarations are there.
// CHECK: .func (.param .b32 func_retval0) wrapping_external_fn
// CHECK: .func (.param .b32 func_retval0) panicking_external_fn
// CHECK: .func [[PANIC_HANDLER:_ZN4core9panicking5panic[a-zA-Z0-9]+]]
// CHECK-LABEL: .visible .entry top_kernel(
#[no_mangle]
pub unsafe extern "ptx-kernel" fn top_kernel(a: *const u32, b: *mut u32) {
// CHECK: call.uni (retval0),
// CHECK-NEXT: wrapping_external_fn
// CHECK: ld.param.b32 %[[LHS:r[0-9]+]], [retval0+0];
let lhs = dep::wrapping_external_fn(*a);
// CHECK: call.uni (retval0),
// CHECK-NEXT: panicking_external_fn
// CHECK: ld.param.b32 %[[RHS:r[0-9]+]], [retval0+0];
let rhs = dep::panicking_external_fn(*a);
// CHECK: add.s32 %[[RES:r[0-9]+]], %[[RHS]], %[[LHS]];
// CHECK: st.global.u32 [%{{rd[0-9]+}}], %[[RES]];
*b = lhs + rhs;
}
// Verify that external function bodies are available.
// CHECK-LABEL: .func (.param .b32 func_retval0) wrapping_external_fn
// CHECK: {
// CHECK: st.param.b32 [func_retval0+0], %{{r[0-9]+}};
// CHECK: }
// Also verify panic behavior.
// CHECK-LABEL: .func (.param .b32 func_retval0) panicking_external_fn
// CHECK: {
// CHECK: %{{p[0-9]+}} bra [[PANIC_LABEL:[a-zA-Z0-9_]+]];
// CHECK: [[PANIC_LABEL]]:
// CHECK: call.uni
// CHECK: [[PANIC_HANDLER]]
// CHECK: }
// Verify whether out dummy panic formatter has a correct body.
// CHECK: .func [[PANIC_FMT:_ZN4core9panicking9panic_fmt[a-zA-Z0-9]+]]()
// CHECK: {
// CHECK: trap;
// CHECK: }
#[panic_handler]
unsafe fn breakpoint_panic_handler(_: &::core::panic::PanicInfo) -> ! {
core::intrinsics::breakpoint();
core::hint::unreachable_unchecked();
}

9
src/test/run-make/nvptx-emit-asm/Makefile
View File

@ -1,9 +0,0 @@
-include ../../run-make-fulldeps/tools.mk
ifeq ($(TARGET),nvptx64-nvidia-cuda)
all:
$(RUSTC) kernel.rs --crate-type="rlib" --emit asm,llvm-ir -O --target $(TARGET)
FileCheck kernel.rs --input-file $(TMPDIR)/kernel.s
else
all:
endif

3
src/tools/compiletest/src/common.rs
View File

@ -26,6 +26,7 @@ pub enum Mode {
Ui,
JsDocTest,
MirOpt,
Assembly,
}
impl Mode {
@ -62,6 +63,7 @@ impl FromStr for Mode {
"ui" => Ok(Ui),
"js-doc-test" => Ok(JsDocTest),
"mir-opt" => Ok(MirOpt),
"assembly" => Ok(Assembly),
_ => Err(()),
}
}
@ -86,6 +88,7 @@ impl fmt::Display for Mode {
Ui => "ui",
JsDocTest => "js-doc-test",
MirOpt => "mir-opt",
Assembly => "assembly",
};
fmt::Display::fmt(s, f)
}

12
src/tools/compiletest/src/header.rs
View File

@ -335,6 +335,7 @@ pub struct TestProps {
pub failure_status: i32,
pub run_rustfix: bool,
pub rustfix_only_machine_applicable: bool,
pub assembly_output: Option<String>,
}
impl TestProps {
@ -370,6 +371,7 @@ impl TestProps {
failure_status: -1,
run_rustfix: false,
rustfix_only_machine_applicable: false,
assembly_output: None,
}
}
@ -517,6 +519,10 @@ impl TestProps {
self.rustfix_only_machine_applicable =
config.parse_rustfix_only_machine_applicable(ln);
}
if self.assembly_output.is_none() {
self.assembly_output = config.parse_assembly_output(ln);
}
});
if self.failure_status == -1 {
@ -594,6 +600,7 @@ impl Config {
fn parse_aux_build(&self, line: &str) -> Option<String> {
self.parse_name_value_directive(line, "aux-build")
.map(|r| r.trim().to_string())
}
fn parse_compile_flags(&self, line: &str) -> Option<String> {
@ -676,6 +683,11 @@ impl Config {
self.parse_name_directive(line, "skip-codegen")
}
fn parse_assembly_output(&self, line: &str) -> Option<String> {
self.parse_name_value_directive(line, "assembly-output")
.map(|r| r.trim().to_string())
}
fn parse_env(&self, line: &str, name: &str) -> Option<(String, String)> {
self.parse_name_value_directive(line, name).map(|nv| {
// nv is either FOO or FOO=BAR

58
src/tools/compiletest/src/runtest.rs
View File

@ -4,7 +4,7 @@ use crate::common::{output_base_dir, output_base_name, output_testname_unique};
use crate::common::{Codegen, CodegenUnits, DebugInfoBoth, DebugInfoGdb, DebugInfoLldb, Rustdoc};
use crate::common::{CompileFail, Pretty, RunFail, RunPass, RunPassValgrind};
use crate::common::{Config, TestPaths};
use crate::common::{Incremental, MirOpt, RunMake, Ui, JsDocTest};
use crate::common::{Incremental, MirOpt, RunMake, Ui, JsDocTest, Assembly};
use diff;
use crate::errors::{self, Error, ErrorKind};
use filetime::FileTime;
@ -275,6 +275,7 @@ impl<'test> TestCx<'test> {
RunMake => self.run_rmake_test(),
RunPass | Ui => self.run_ui_test(),
MirOpt => self.run_mir_opt_test(),
Assembly => self.run_assembly_test(),
JsDocTest => self.run_js_doc_test(),
}
}
@ -1606,6 +1607,7 @@ impl<'test> TestCx<'test> {
|| self.config.target.contains("emscripten")
|| (self.config.target.contains("musl") && !aux_props.force_host)
|| self.config.target.contains("wasm32")
|| self.config.target.contains("nvptx")
{
// We primarily compile all auxiliary libraries as dynamic libraries
// to avoid code size bloat and large binaries as much as possible
@ -1805,7 +1807,7 @@ impl<'test> TestCx<'test> {
rustc.arg(dir_opt);
}
RunFail | RunPassValgrind | Pretty | DebugInfoBoth | DebugInfoGdb | DebugInfoLldb
| Codegen | Rustdoc | RunMake | CodegenUnits | JsDocTest => {
| Codegen | Rustdoc | RunMake | CodegenUnits | JsDocTest | Assembly => {
// do not use JSON output
}
}
@ -2100,12 +2102,37 @@ impl<'test> TestCx<'test> {
self.compose_and_run_compiler(rustc, None)
}
fn check_ir_with_filecheck(&self) -> ProcRes {
let irfile = self.output_base_name().with_extension("ll");
fn compile_test_and_save_assembly(&self) -> (ProcRes, PathBuf) {
// This works with both `--emit asm` (as default output name for the assembly)
// and `ptx-linker` because the latter can write output at requested location.
let output_path = self.output_base_name().with_extension("s");
let output_file = TargetLocation::ThisFile(output_path.clone());
let mut rustc = self.make_compile_args(&self.testpaths.file, output_file);
rustc.arg("-L").arg(self.aux_output_dir_name());
match self.props.assembly_output.as_ref().map(AsRef::as_ref) {
Some("emit-asm") => {
rustc.arg("--emit=asm");
}
Some("ptx-linker") => {
// No extra flags needed.
}
Some(_) => self.fatal("unknown 'assembly-output' header"),
None => self.fatal("missing 'assembly-output' header"),
}
(self.compose_and_run_compiler(rustc, None), output_path)
}
fn verify_with_filecheck(&self, output: &Path) -> ProcRes {
let mut filecheck = Command::new(self.config.llvm_filecheck.as_ref().unwrap());
filecheck
.arg("--input-file")
.arg(irfile)
.arg(output)
.arg(&self.testpaths.file);
// It would be more appropriate to make most of the arguments configurable through
// a comment-attribute similar to `compile-flags`. For example, --check-prefixes is a very
@ -2124,12 +2151,29 @@ impl<'test> TestCx<'test> {
self.fatal("missing --llvm-filecheck");
}
let mut proc_res = self.compile_test_and_save_ir();
let proc_res = self.compile_test_and_save_ir();
if !proc_res.status.success() {
self.fatal_proc_rec("compilation failed!", &proc_res);
}
proc_res = self.check_ir_with_filecheck();
let output_path = self.output_base_name().with_extension("ll");
let proc_res = self.verify_with_filecheck(&output_path);
if !proc_res.status.success() {
self.fatal_proc_rec("verification with 'FileCheck' failed", &proc_res);
}
}
fn run_assembly_test(&self) {
if self.config.llvm_filecheck.is_none() {
self.fatal("missing --llvm-filecheck");
}
let (proc_res, output_path) = self.compile_test_and_save_assembly();
if !proc_res.status.success() {
self.fatal_proc_rec("compilation failed!", &proc_res);
}
let proc_res = self.verify_with_filecheck(&output_path);
if !proc_res.status.success() {
self.fatal_proc_rec("verification with 'FileCheck' failed", &proc_res);
}

4
src/tools/compiletest/src/util.rs
View File

@ -41,7 +41,6 @@ const ARCH_TABLE: &'static [(&'static str, &'static str)] = &[
("armv7", "arm"),
("armv7s", "arm"),
("asmjs", "asmjs"),
("cuda", "cuda"),
("hexagon", "hexagon"),
("i386", "x86"),
("i586", "x86"),
@ -59,6 +58,7 @@ const ARCH_TABLE: &'static [(&'static str, &'static str)] = &[
("mipsisa64r6", "mips64"),
("mipsisa64r6el", "mips64"),
("msp430", "msp430"),
("nvptx64", "nvptx64"),
("powerpc", "powerpc"),
("powerpc64", "powerpc64"),
("powerpc64le", "powerpc64"),
@ -166,7 +166,7 @@ fn test_get_arch_failure() {
fn test_get_arch() {
assert_eq!("x86_64", get_arch("x86_64-unknown-linux-gnu"));
assert_eq!("x86_64", get_arch("amd64"));
assert_eq!("cuda", get_arch("nvptx64-nvidia-cuda"));
assert_eq!("nvptx64", get_arch("nvptx64-nvidia-cuda"));
}
#[test]