zig/test/nvptx.zig

104 lines
3.2 KiB
Zig
Raw Normal View History

2022-02-21 19:05:27 +00:00
const std = @import("std");
re-enable test-cases and get them all passing Instead of using `zig test` to build a special version of the compiler that runs all the test-cases, the zig build system is now used as much as possible - all with the basic steps found in the standard library. For incremental compilation tests (the ones that look like foo.0.zig, foo.1.zig, foo.2.zig, etc.), a special version of the compiler is compiled into a utility executable called "check-case" which checks exactly one sequence of incremental updates in an independent subprocess. Previously, all incremental and non-incremental test cases were done in the same test runner process. The compile error checking code is now simpler, but also a bit rudimentary, and so it additionally makes sure that the actual compile errors do not include *extra* messages, and it makes sure that the actual compile errors output in the same order as expected. It is also based on the "ends-with" property of each line rather than the previous logic, which frankly I didn't want to touch with a ten-meter pole. The compile error test cases have been updated to pass in light of these differences. Previously, 'error' mode with 0 compile errors was used to shoehorn in a different kind of test-case - one that only checks if a piece of code compiles without errors. Now there is a 'compile' mode of test-cases, and 'error' must be only used when there are greater than 0 errors. link test cases are updated to omit the target object format argument when calling checkObject since that is no longer needed. The test/stage2 directory is removed; the 2 files within are moved to be directly in the test/ directory.
2023-03-10 01:22:51 +00:00
const Cases = @import("src/Cases.zig");
2022-02-21 19:05:27 +00:00
zig build system: change target, compilation, and module APIs Introduce the concept of "target query" and "resolved target". A target query is what the user specifies, with some things left to default. A resolved target has the default things discovered and populated. In the future, std.zig.CrossTarget will be rename to std.Target.Query. Introduces `std.Build.resolveTargetQuery` to get from one to the other. The concept of `main_mod_path` is gone, no longer supported. You have to put the root source file at the module root now. * remove deprecated API * update build.zig for the breaking API changes in this branch * move std.Build.Step.Compile.BuildId to std.zig.BuildId * add more options to std.Build.ExecutableOptions, std.Build.ObjectOptions, std.Build.SharedLibraryOptions, std.Build.StaticLibraryOptions, and std.Build.TestOptions. * remove `std.Build.constructCMacro`. There is no use for this API. * deprecate `std.Build.Step.Compile.defineCMacro`. Instead, `std.Build.Module.addCMacro` is provided. - remove `std.Build.Step.Compile.defineCMacroRaw`. * deprecate `std.Build.Step.Compile.linkFrameworkNeeded` - use `std.Build.Module.linkFramework` * deprecate `std.Build.Step.Compile.linkFrameworkWeak` - use `std.Build.Module.linkFramework` * move more logic into `std.Build.Module` * allow `target` and `optimize` to be `null` when creating a Module. Along with other fields, those unspecified options will be inherited from parent `Module` when inserted into an import table. * the `target` field of `addExecutable` is now required. pass `b.host` to get the host target.
2023-12-03 04:51:34 +00:00
pub fn addCases(ctx: *Cases, b: *std.Build) !void {
const target = b.resolveTargetQuery(.{
.cpu_arch = .nvptx64,
.os_tag = .cuda,
});
2022-02-21 19:05:27 +00:00
{
zig build system: change target, compilation, and module APIs Introduce the concept of "target query" and "resolved target". A target query is what the user specifies, with some things left to default. A resolved target has the default things discovered and populated. In the future, std.zig.CrossTarget will be rename to std.Target.Query. Introduces `std.Build.resolveTargetQuery` to get from one to the other. The concept of `main_mod_path` is gone, no longer supported. You have to put the root source file at the module root now. * remove deprecated API * update build.zig for the breaking API changes in this branch * move std.Build.Step.Compile.BuildId to std.zig.BuildId * add more options to std.Build.ExecutableOptions, std.Build.ObjectOptions, std.Build.SharedLibraryOptions, std.Build.StaticLibraryOptions, and std.Build.TestOptions. * remove `std.Build.constructCMacro`. There is no use for this API. * deprecate `std.Build.Step.Compile.defineCMacro`. Instead, `std.Build.Module.addCMacro` is provided. - remove `std.Build.Step.Compile.defineCMacroRaw`. * deprecate `std.Build.Step.Compile.linkFrameworkNeeded` - use `std.Build.Module.linkFramework` * deprecate `std.Build.Step.Compile.linkFrameworkWeak` - use `std.Build.Module.linkFramework` * move more logic into `std.Build.Module` * allow `target` and `optimize` to be `null` when creating a Module. Along with other fields, those unspecified options will be inherited from parent `Module` when inserted into an import table. * the `target` field of `addExecutable` is now required. pass `b.host` to get the host target.
2023-12-03 04:51:34 +00:00
var case = addPtx(ctx, target, "simple addition and subtraction");
2022-02-21 19:05:27 +00:00
re-enable test-cases and get them all passing Instead of using `zig test` to build a special version of the compiler that runs all the test-cases, the zig build system is now used as much as possible - all with the basic steps found in the standard library. For incremental compilation tests (the ones that look like foo.0.zig, foo.1.zig, foo.2.zig, etc.), a special version of the compiler is compiled into a utility executable called "check-case" which checks exactly one sequence of incremental updates in an independent subprocess. Previously, all incremental and non-incremental test cases were done in the same test runner process. The compile error checking code is now simpler, but also a bit rudimentary, and so it additionally makes sure that the actual compile errors do not include *extra* messages, and it makes sure that the actual compile errors output in the same order as expected. It is also based on the "ends-with" property of each line rather than the previous logic, which frankly I didn't want to touch with a ten-meter pole. The compile error test cases have been updated to pass in light of these differences. Previously, 'error' mode with 0 compile errors was used to shoehorn in a different kind of test-case - one that only checks if a piece of code compiles without errors. Now there is a 'compile' mode of test-cases, and 'error' must be only used when there are greater than 0 errors. link test cases are updated to omit the target object format argument when calling checkObject since that is no longer needed. The test/stage2 directory is removed; the 2 files within are moved to be directly in the test/ directory.
2023-03-10 01:22:51 +00:00
case.addCompile(
2022-02-21 19:05:27 +00:00
\\fn add(a: i32, b: i32) i32 {
\\ return a + b;
\\}
\\
\\pub export fn add_and_substract(a: i32, out: *i32) callconv(.Kernel) void {
2022-02-21 19:05:27 +00:00
\\ const x = add(a, 7);
\\ var y = add(2, 0);
\\ y -= x;
\\ out.* = y;
\\}
);
}
{
zig build system: change target, compilation, and module APIs Introduce the concept of "target query" and "resolved target". A target query is what the user specifies, with some things left to default. A resolved target has the default things discovered and populated. In the future, std.zig.CrossTarget will be rename to std.Target.Query. Introduces `std.Build.resolveTargetQuery` to get from one to the other. The concept of `main_mod_path` is gone, no longer supported. You have to put the root source file at the module root now. * remove deprecated API * update build.zig for the breaking API changes in this branch * move std.Build.Step.Compile.BuildId to std.zig.BuildId * add more options to std.Build.ExecutableOptions, std.Build.ObjectOptions, std.Build.SharedLibraryOptions, std.Build.StaticLibraryOptions, and std.Build.TestOptions. * remove `std.Build.constructCMacro`. There is no use for this API. * deprecate `std.Build.Step.Compile.defineCMacro`. Instead, `std.Build.Module.addCMacro` is provided. - remove `std.Build.Step.Compile.defineCMacroRaw`. * deprecate `std.Build.Step.Compile.linkFrameworkNeeded` - use `std.Build.Module.linkFramework` * deprecate `std.Build.Step.Compile.linkFrameworkWeak` - use `std.Build.Module.linkFramework` * move more logic into `std.Build.Module` * allow `target` and `optimize` to be `null` when creating a Module. Along with other fields, those unspecified options will be inherited from parent `Module` when inserted into an import table. * the `target` field of `addExecutable` is now required. pass `b.host` to get the host target.
2023-12-03 04:51:34 +00:00
var case = addPtx(ctx, target, "read special registers");
2022-02-21 19:05:27 +00:00
re-enable test-cases and get them all passing Instead of using `zig test` to build a special version of the compiler that runs all the test-cases, the zig build system is now used as much as possible - all with the basic steps found in the standard library. For incremental compilation tests (the ones that look like foo.0.zig, foo.1.zig, foo.2.zig, etc.), a special version of the compiler is compiled into a utility executable called "check-case" which checks exactly one sequence of incremental updates in an independent subprocess. Previously, all incremental and non-incremental test cases were done in the same test runner process. The compile error checking code is now simpler, but also a bit rudimentary, and so it additionally makes sure that the actual compile errors do not include *extra* messages, and it makes sure that the actual compile errors output in the same order as expected. It is also based on the "ends-with" property of each line rather than the previous logic, which frankly I didn't want to touch with a ten-meter pole. The compile error test cases have been updated to pass in light of these differences. Previously, 'error' mode with 0 compile errors was used to shoehorn in a different kind of test-case - one that only checks if a piece of code compiles without errors. Now there is a 'compile' mode of test-cases, and 'error' must be only used when there are greater than 0 errors. link test cases are updated to omit the target object format argument when calling checkObject since that is no longer needed. The test/stage2 directory is removed; the 2 files within are moved to be directly in the test/ directory.
2023-03-10 01:22:51 +00:00
case.addCompile(
2022-10-04 06:31:36 +01:00
\\fn threadIdX() u32 {
\\ return asm ("mov.u32 \t%[r], %tid.x;"
2022-10-06 21:06:20 +01:00
\\ : [r] "=r" (-> u32),
2022-10-04 06:31:36 +01:00
\\ );
2022-02-21 19:05:27 +00:00
\\}
\\
\\pub export fn special_reg(a: []const i32, out: []i32) callconv(.Kernel) void {
\\ const i = threadIdX();
2022-02-21 19:05:27 +00:00
\\ out[i] = a[i] + 7;
\\}
);
}
{
zig build system: change target, compilation, and module APIs Introduce the concept of "target query" and "resolved target". A target query is what the user specifies, with some things left to default. A resolved target has the default things discovered and populated. In the future, std.zig.CrossTarget will be rename to std.Target.Query. Introduces `std.Build.resolveTargetQuery` to get from one to the other. The concept of `main_mod_path` is gone, no longer supported. You have to put the root source file at the module root now. * remove deprecated API * update build.zig for the breaking API changes in this branch * move std.Build.Step.Compile.BuildId to std.zig.BuildId * add more options to std.Build.ExecutableOptions, std.Build.ObjectOptions, std.Build.SharedLibraryOptions, std.Build.StaticLibraryOptions, and std.Build.TestOptions. * remove `std.Build.constructCMacro`. There is no use for this API. * deprecate `std.Build.Step.Compile.defineCMacro`. Instead, `std.Build.Module.addCMacro` is provided. - remove `std.Build.Step.Compile.defineCMacroRaw`. * deprecate `std.Build.Step.Compile.linkFrameworkNeeded` - use `std.Build.Module.linkFramework` * deprecate `std.Build.Step.Compile.linkFrameworkWeak` - use `std.Build.Module.linkFramework` * move more logic into `std.Build.Module` * allow `target` and `optimize` to be `null` when creating a Module. Along with other fields, those unspecified options will be inherited from parent `Module` when inserted into an import table. * the `target` field of `addExecutable` is now required. pass `b.host` to get the host target.
2023-12-03 04:51:34 +00:00
var case = addPtx(ctx, target, "address spaces");
2022-02-21 19:05:27 +00:00
re-enable test-cases and get them all passing Instead of using `zig test` to build a special version of the compiler that runs all the test-cases, the zig build system is now used as much as possible - all with the basic steps found in the standard library. For incremental compilation tests (the ones that look like foo.0.zig, foo.1.zig, foo.2.zig, etc.), a special version of the compiler is compiled into a utility executable called "check-case" which checks exactly one sequence of incremental updates in an independent subprocess. Previously, all incremental and non-incremental test cases were done in the same test runner process. The compile error checking code is now simpler, but also a bit rudimentary, and so it additionally makes sure that the actual compile errors do not include *extra* messages, and it makes sure that the actual compile errors output in the same order as expected. It is also based on the "ends-with" property of each line rather than the previous logic, which frankly I didn't want to touch with a ten-meter pole. The compile error test cases have been updated to pass in light of these differences. Previously, 'error' mode with 0 compile errors was used to shoehorn in a different kind of test-case - one that only checks if a piece of code compiles without errors. Now there is a 'compile' mode of test-cases, and 'error' must be only used when there are greater than 0 errors. link test cases are updated to omit the target object format argument when calling checkObject since that is no longer needed. The test/stage2 directory is removed; the 2 files within are moved to be directly in the test/ directory.
2023-03-10 01:22:51 +00:00
case.addCompile(
\\var x: i32 addrspace(.global) = 0;
2022-02-21 19:05:27 +00:00
\\
\\pub export fn increment(out: *i32) callconv(.Kernel) void {
2022-02-21 19:05:27 +00:00
\\ x += 1;
\\ out.* = x;
\\}
);
}
2022-10-04 06:31:36 +01:00
{
zig build system: change target, compilation, and module APIs Introduce the concept of "target query" and "resolved target". A target query is what the user specifies, with some things left to default. A resolved target has the default things discovered and populated. In the future, std.zig.CrossTarget will be rename to std.Target.Query. Introduces `std.Build.resolveTargetQuery` to get from one to the other. The concept of `main_mod_path` is gone, no longer supported. You have to put the root source file at the module root now. * remove deprecated API * update build.zig for the breaking API changes in this branch * move std.Build.Step.Compile.BuildId to std.zig.BuildId * add more options to std.Build.ExecutableOptions, std.Build.ObjectOptions, std.Build.SharedLibraryOptions, std.Build.StaticLibraryOptions, and std.Build.TestOptions. * remove `std.Build.constructCMacro`. There is no use for this API. * deprecate `std.Build.Step.Compile.defineCMacro`. Instead, `std.Build.Module.addCMacro` is provided. - remove `std.Build.Step.Compile.defineCMacroRaw`. * deprecate `std.Build.Step.Compile.linkFrameworkNeeded` - use `std.Build.Module.linkFramework` * deprecate `std.Build.Step.Compile.linkFrameworkWeak` - use `std.Build.Module.linkFramework` * move more logic into `std.Build.Module` * allow `target` and `optimize` to be `null` when creating a Module. Along with other fields, those unspecified options will be inherited from parent `Module` when inserted into an import table. * the `target` field of `addExecutable` is now required. pass `b.host` to get the host target.
2023-12-03 04:51:34 +00:00
var case = addPtx(ctx, target, "reduce in shared mem");
re-enable test-cases and get them all passing Instead of using `zig test` to build a special version of the compiler that runs all the test-cases, the zig build system is now used as much as possible - all with the basic steps found in the standard library. For incremental compilation tests (the ones that look like foo.0.zig, foo.1.zig, foo.2.zig, etc.), a special version of the compiler is compiled into a utility executable called "check-case" which checks exactly one sequence of incremental updates in an independent subprocess. Previously, all incremental and non-incremental test cases were done in the same test runner process. The compile error checking code is now simpler, but also a bit rudimentary, and so it additionally makes sure that the actual compile errors do not include *extra* messages, and it makes sure that the actual compile errors output in the same order as expected. It is also based on the "ends-with" property of each line rather than the previous logic, which frankly I didn't want to touch with a ten-meter pole. The compile error test cases have been updated to pass in light of these differences. Previously, 'error' mode with 0 compile errors was used to shoehorn in a different kind of test-case - one that only checks if a piece of code compiles without errors. Now there is a 'compile' mode of test-cases, and 'error' must be only used when there are greater than 0 errors. link test cases are updated to omit the target object format argument when calling checkObject since that is no longer needed. The test/stage2 directory is removed; the 2 files within are moved to be directly in the test/ directory.
2023-03-10 01:22:51 +00:00
case.addCompile(
2022-10-04 06:31:36 +01:00
\\fn threadIdX() u32 {
\\ return asm ("mov.u32 \t%[r], %tid.x;"
2022-10-06 21:06:20 +01:00
\\ : [r] "=r" (-> u32),
2022-10-04 06:31:36 +01:00
\\ );
\\}
\\
\\ var _sdata: [1024]f32 addrspace(.shared) = undefined;
\\ pub export fn reduceSum(d_x: []const f32, out: *f32) callconv(.Kernel) void {
\\ var sdata: *addrspace(.generic) [1024]f32 = @addrSpaceCast(&_sdata);
2022-10-04 06:31:36 +01:00
\\ const tid: u32 = threadIdX();
\\ var sum = d_x[tid];
\\ sdata[tid] = sum;
\\ asm volatile ("bar.sync \t0;");
\\ var s: u32 = 512;
\\ while (s > 0) : (s = s >> 1) {
\\ if (tid < s) {
\\ sum += sdata[tid + s];
\\ sdata[tid] = sum;
\\ }
\\ asm volatile ("bar.sync \t0;");
\\ }
\\
\\ if (tid == 0) {
\\ out.* = sum;
\\ }
\\ }
);
}
2022-02-21 19:05:27 +00:00
}
zig build system: change target, compilation, and module APIs Introduce the concept of "target query" and "resolved target". A target query is what the user specifies, with some things left to default. A resolved target has the default things discovered and populated. In the future, std.zig.CrossTarget will be rename to std.Target.Query. Introduces `std.Build.resolveTargetQuery` to get from one to the other. The concept of `main_mod_path` is gone, no longer supported. You have to put the root source file at the module root now. * remove deprecated API * update build.zig for the breaking API changes in this branch * move std.Build.Step.Compile.BuildId to std.zig.BuildId * add more options to std.Build.ExecutableOptions, std.Build.ObjectOptions, std.Build.SharedLibraryOptions, std.Build.StaticLibraryOptions, and std.Build.TestOptions. * remove `std.Build.constructCMacro`. There is no use for this API. * deprecate `std.Build.Step.Compile.defineCMacro`. Instead, `std.Build.Module.addCMacro` is provided. - remove `std.Build.Step.Compile.defineCMacroRaw`. * deprecate `std.Build.Step.Compile.linkFrameworkNeeded` - use `std.Build.Module.linkFramework` * deprecate `std.Build.Step.Compile.linkFrameworkWeak` - use `std.Build.Module.linkFramework` * move more logic into `std.Build.Module` * allow `target` and `optimize` to be `null` when creating a Module. Along with other fields, those unspecified options will be inherited from parent `Module` when inserted into an import table. * the `target` field of `addExecutable` is now required. pass `b.host` to get the host target.
2023-12-03 04:51:34 +00:00
fn addPtx(ctx: *Cases, target: std.Build.ResolvedTarget, name: []const u8) *Cases.Case {
re-enable test-cases and get them all passing Instead of using `zig test` to build a special version of the compiler that runs all the test-cases, the zig build system is now used as much as possible - all with the basic steps found in the standard library. For incremental compilation tests (the ones that look like foo.0.zig, foo.1.zig, foo.2.zig, etc.), a special version of the compiler is compiled into a utility executable called "check-case" which checks exactly one sequence of incremental updates in an independent subprocess. Previously, all incremental and non-incremental test cases were done in the same test runner process. The compile error checking code is now simpler, but also a bit rudimentary, and so it additionally makes sure that the actual compile errors do not include *extra* messages, and it makes sure that the actual compile errors output in the same order as expected. It is also based on the "ends-with" property of each line rather than the previous logic, which frankly I didn't want to touch with a ten-meter pole. The compile error test cases have been updated to pass in light of these differences. Previously, 'error' mode with 0 compile errors was used to shoehorn in a different kind of test-case - one that only checks if a piece of code compiles without errors. Now there is a 'compile' mode of test-cases, and 'error' must be only used when there are greater than 0 errors. link test cases are updated to omit the target object format argument when calling checkObject since that is no longer needed. The test/stage2 directory is removed; the 2 files within are moved to be directly in the test/ directory.
2023-03-10 01:22:51 +00:00
ctx.cases.append(.{
.name = name,
zig build system: change target, compilation, and module APIs Introduce the concept of "target query" and "resolved target". A target query is what the user specifies, with some things left to default. A resolved target has the default things discovered and populated. In the future, std.zig.CrossTarget will be rename to std.Target.Query. Introduces `std.Build.resolveTargetQuery` to get from one to the other. The concept of `main_mod_path` is gone, no longer supported. You have to put the root source file at the module root now. * remove deprecated API * update build.zig for the breaking API changes in this branch * move std.Build.Step.Compile.BuildId to std.zig.BuildId * add more options to std.Build.ExecutableOptions, std.Build.ObjectOptions, std.Build.SharedLibraryOptions, std.Build.StaticLibraryOptions, and std.Build.TestOptions. * remove `std.Build.constructCMacro`. There is no use for this API. * deprecate `std.Build.Step.Compile.defineCMacro`. Instead, `std.Build.Module.addCMacro` is provided. - remove `std.Build.Step.Compile.defineCMacroRaw`. * deprecate `std.Build.Step.Compile.linkFrameworkNeeded` - use `std.Build.Module.linkFramework` * deprecate `std.Build.Step.Compile.linkFrameworkWeak` - use `std.Build.Module.linkFramework` * move more logic into `std.Build.Module` * allow `target` and `optimize` to be `null` when creating a Module. Along with other fields, those unspecified options will be inherited from parent `Module` when inserted into an import table. * the `target` field of `addExecutable` is now required. pass `b.host` to get the host target.
2023-12-03 04:51:34 +00:00
.target = target,
re-enable test-cases and get them all passing Instead of using `zig test` to build a special version of the compiler that runs all the test-cases, the zig build system is now used as much as possible - all with the basic steps found in the standard library. For incremental compilation tests (the ones that look like foo.0.zig, foo.1.zig, foo.2.zig, etc.), a special version of the compiler is compiled into a utility executable called "check-case" which checks exactly one sequence of incremental updates in an independent subprocess. Previously, all incremental and non-incremental test cases were done in the same test runner process. The compile error checking code is now simpler, but also a bit rudimentary, and so it additionally makes sure that the actual compile errors do not include *extra* messages, and it makes sure that the actual compile errors output in the same order as expected. It is also based on the "ends-with" property of each line rather than the previous logic, which frankly I didn't want to touch with a ten-meter pole. The compile error test cases have been updated to pass in light of these differences. Previously, 'error' mode with 0 compile errors was used to shoehorn in a different kind of test-case - one that only checks if a piece of code compiles without errors. Now there is a 'compile' mode of test-cases, and 'error' must be only used when there are greater than 0 errors. link test cases are updated to omit the target object format argument when calling checkObject since that is no longer needed. The test/stage2 directory is removed; the 2 files within are moved to be directly in the test/ directory.
2023-03-10 01:22:51 +00:00
.updates = std.ArrayList(Cases.Update).init(ctx.cases.allocator),
.output_mode = .Obj,
re-enable test-cases and get them all passing Instead of using `zig test` to build a special version of the compiler that runs all the test-cases, the zig build system is now used as much as possible - all with the basic steps found in the standard library. For incremental compilation tests (the ones that look like foo.0.zig, foo.1.zig, foo.2.zig, etc.), a special version of the compiler is compiled into a utility executable called "check-case" which checks exactly one sequence of incremental updates in an independent subprocess. Previously, all incremental and non-incremental test cases were done in the same test runner process. The compile error checking code is now simpler, but also a bit rudimentary, and so it additionally makes sure that the actual compile errors do not include *extra* messages, and it makes sure that the actual compile errors output in the same order as expected. It is also based on the "ends-with" property of each line rather than the previous logic, which frankly I didn't want to touch with a ten-meter pole. The compile error test cases have been updated to pass in light of these differences. Previously, 'error' mode with 0 compile errors was used to shoehorn in a different kind of test-case - one that only checks if a piece of code compiles without errors. Now there is a 'compile' mode of test-cases, and 'error' must be only used when there are greater than 0 errors. link test cases are updated to omit the target object format argument when calling checkObject since that is no longer needed. The test/stage2 directory is removed; the 2 files within are moved to be directly in the test/ directory.
2023-03-10 01:22:51 +00:00
.deps = std.ArrayList(Cases.DepModule).init(ctx.cases.allocator),
.link_libc = false,
.backend = .llvm,
2022-10-06 21:06:20 +01:00
// Bug in Debug mode
.optimize_mode = .ReleaseSafe,
}) catch @panic("out of memory");
return &ctx.cases.items[ctx.cases.items.len - 1];
}