Commit 7cd8321b82bb142d6bc03b918e883ad8c4bcdffb | zig

author	Alex Rønne Petersen <alex@alexrp.com> 2026-08-03 14:11:28+02:00
committer	Alex Rønne Petersen <alex@alexrp.com> 2026-08-31 23:11:17+02:00
log	7cd8321b82bb142d6bc03b918e883ad8c4bcdffb
tree	e8fa748a606d7150fe4571c1b4777905b85cd74d
parent	107e01eb8e3260fe76b2a65f80c8fe707d6d67c0
signature	Signed by SSH key SHA256:7B/LJ7bpR1eX8aCXSr4mtd5M45VMPKcx9zY8e95b5QM

std.Target: update CPU models and features to LLVM 23

18 files changed, 2754 insertions(+), 442 deletions(-)

lib
- compiler
  - aro
    aro
    Compilation.zig
- std
  - Target.zig
  - Target
    aarch64.zig
    amdgcn.zig
    hexagon.zig
    mips.zig
    nvptx.zig
    powerpc.zig
    riscv.zig
    sparc.zig
    wasm.zig
    x86.zig
    xtensa.zig
src
- codegen
  - x86_64
    CodeGen.zig
- link
  - Wasm.zig
- target.zig
test
- tests.zig
tools
- update_cpu_features.zig

lib/compiler/aro/aro/Compilation.zig-1

...	...	@@ -560,7 +560,6 @@ fn generateSystemDefines(comp: Compilation, w: Io.Writer) !void {
560	560	.{ .amx_fp8, "__AMX_FP8__" },
561	561	.{ .amx_movrs, "__AMX_MOVRS__" },
562	562	.{ .amx_avx512, "__AMX_AVX512__" },
563		.{ .amx_tf32, "__AMX_TF32__" },
564	563	.{ .cmpccxadd, "__CMPCCXADD__" },
565	564	.{ .raoint, "__RAOINT__" },
566	565	.{ .avxifma, "__AVXIFMA__" },

lib/std/Target.zig+1-1

...	...	@@ -1238,7 +1238,7 @@ pub const Cpu = struct {
1238	1238	pub const Set = struct {
1239	1239	ints: [usize_count]usize,
1240	1240
1241		pub const needed_bit_count = 347;
	1241	pub const needed_bit_count = 378;
1242	1242	pub const byte_count = @divCeil(needed_bit_count, 8);
1243	1243	pub const usize_count = (byte_count + (@sizeOf(usize) - 1)) / @sizeOf(usize);
1244	1244	pub const Index = std.math.Log2Int(@Int(.unsigned, usize_count * @bitSizeOf(usize)));

lib/std/Target/aarch64.zig+191-33

...	...	@@ -10,6 +10,7 @@ pub const Feature = enum {
10	10	aes,
11	11	aggressive_fma,
12	12	aggressive_interleaving,
	13	align_cmp_csel_pairs,
13	14	alternate_sextload_cvt_f32_pattern,
14	15	altnzcv,
15	16	alu_lsl_fast,
...	...	@@ -38,7 +39,6 @@ pub const Feature = enum {
38	39	ccpp,
39	40	chk,
40	41	clrbhb,
41		cmh,
42	42	cmp_bcc_fusion,
43	43	cmpbr,
44	44	complxnum,
...	...	@@ -48,6 +48,7 @@ pub const Feature = enum {
48	48	crypto,
49	49	cssc,
50	50	d128,
	51	disable_distinct_predicate_dst_reg,
51	52	disable_fast_inc_vl,
52	53	disable_latency_sched_heuristic,
53	54	disable_ldp,
...	...	@@ -71,8 +72,10 @@ pub const Feature = enum {
71	72	f8f16mm,
72	73	f8f32mm,
73	74	faminmax,
	75	fast_ld1_single,
74	76	fgt,
75	77	fix_cortex_a53_835769,
	78	fixed_load_latency_4,
76	79	flagm,
77	80	fmv,
78	81	force_32bit_jump_tables,
...	...	@@ -95,14 +98,16 @@ pub const Feature = enum {
95	98	fuse_crypto_eor,
96	99	fuse_csel,
97	100	fuse_cset,
	101	fuse_fcsel,
98	102	fuse_literals,
99	103	gcie,
100	104	gcs,
101	105	harden_sls_blr,
102	106	harden_sls_nocomdat,
103	107	harden_sls_retbr,
	108	has_limited_64bit_vector_mul_bandwidth,
104	109	hbc,
105		hcx,
	110	hinte,
106	111	i8mm,
107	112	ite,
108	113	jsconv,
...	...	@@ -121,12 +126,13 @@ pub const Feature = enum {
121	126	mops,
122	127	mops_go,
123	128	mpam,
124		mpamv2,
125	129	mte,
126	130	mtetc,
127	131	neon,
128	132	nmi,
129	133	no_bti_at_return_twice,
	134	no_lfi_loads,
	135	no_lfi_stores,
130	136	no_neg_immediates,
131	137	no_sve_fp_ld1r,
132	138	no_zcz_fpr64,
...	...	@@ -138,7 +144,6 @@ pub const Feature = enum {
138	144	pan_rwv,
139	145	pauth,
140	146	pauth_lr,
141		pcdphint,
142	147	perfmon,
143	148	poe2,
144	149	pops,
...	...	@@ -328,6 +333,11 @@ pub const all_features = blk: {
328	333	.description = "Make use of aggressive interleaving during vectorization",
329	334	.dependencies = featureSet(&[_]Feature{}),
330	335	};
	336	result[@backingInt(Feature.align_cmp_csel_pairs)] = .{
	337	.llvm_name = "align-cmp-csel-pairs",
	338	.description = "Align certain CMP/FCMP and CSEL/FCSEL instruction pairs",
	339	.dependencies = featureSet(&[_]Feature{}),
	340	};
331	341	result[@backingInt(Feature.alternate_sextload_cvt_f32_pattern)] = .{
332	342	.llvm_name = "alternate-sextload-cvt-f32-pattern",
333	343	.description = "Use alternative pattern for sextload convert to f32",
...	...	@@ -476,11 +486,6 @@ pub const all_features = blk: {
476	486	.description = "Enable Clear BHB instruction",
477	487	.dependencies = featureSet(&[_]Feature{}),
478	488	};
479		result[@backingInt(Feature.cmh)] = .{
480		.llvm_name = "cmh",
481		.description = "Enable Armv9.7-A Contention Management Hints",
482		.dependencies = featureSet(&[_]Feature{}),
483		};
484	489	result[@backingInt(Feature.cmp_bcc_fusion)] = .{
485	490	.llvm_name = "cmp-bcc-fusion",
486	491	.description = "CPU fuses cmp+bcc operations",
...	...	@@ -533,6 +538,11 @@ pub const all_features = blk: {
533	538	.lse128,
534	539	}),
535	540	};
	541	result[@backingInt(Feature.disable_distinct_predicate_dst_reg)] = .{
	542	.llvm_name = "disable-distinct-predicate-dst-reg",
	543	.description = "Disabling selecting a distinct predicate register for the destination operand of instructions that take a governing predicate",
	544	.dependencies = featureSet(&[_]Feature{}),
	545	};
536	546	result[@backingInt(Feature.disable_fast_inc_vl)] = .{
537	547	.llvm_name = "disable-fast-inc-vl",
538	548	.description = "Do not prefer INC/DEC, ALL, { 1, 2, 4 } over ADDVL",
...	...	@@ -545,7 +555,7 @@ pub const all_features = blk: {
545	555	};
546	556	result[@backingInt(Feature.disable_ldp)] = .{
547	557	.llvm_name = "disable-ldp",
548		.description = "Do not emit ldp",
	558	.description = "Prefer not to emit ldp",
549	559	.dependencies = featureSet(&[_]Feature{}),
550	560	};
551	561	result[@backingInt(Feature.disable_maximize_scalable_bandwidth)] = .{
...	...	@@ -555,7 +565,7 @@ pub const all_features = blk: {
555	565	};
556	566	result[@backingInt(Feature.disable_stp)] = .{
557	567	.llvm_name = "disable-stp",
558		.description = "Do not emit stp",
	568	.description = "Prefer not to emit stp",
559	569	.dependencies = featureSet(&[_]Feature{}),
560	570	};
561	571	result[@backingInt(Feature.disable_unpredicated_ld_st_lower)] = .{
...	...	@@ -671,6 +681,11 @@ pub const all_features = blk: {
671	681	.neon,
672	682	}),
673	683	};
	684	result[@backingInt(Feature.fast_ld1_single)] = .{
	685	.llvm_name = "fast-ld1-single",
	686	.description = "Single-element LD1 to vector lane has same performance as regular load",
	687	.dependencies = featureSet(&[_]Feature{}),
	688	};
674	689	result[@backingInt(Feature.fgt)] = .{
675	690	.llvm_name = "fgt",
676	691	.description = "Enable fine grained virtualization traps extension",
...	...	@@ -681,6 +696,11 @@ pub const all_features = blk: {
681	696	.description = "Mitigate Cortex-A53 Erratum 835769",
682	697	.dependencies = featureSet(&[_]Feature{}),
683	698	};
	699	result[@backingInt(Feature.fixed_load_latency_4)] = .{
	700	.llvm_name = "fixed-load-latency-4",
	701	.description = "Use a fixed latency of 4 for loads instead of querying the scheduling model",
	702	.dependencies = featureSet(&[_]Feature{}),
	703	};
684	704	result[@backingInt(Feature.flagm)] = .{
685	705	.llvm_name = "flagm",
686	706	.description = "Enable Armv8.4-A Flag Manipulation instructions",
...	...	@@ -813,6 +833,11 @@ pub const all_features = blk: {
813	833	.description = "CPU can fuse CMP and CSET operations",
814	834	.dependencies = featureSet(&[_]Feature{}),
815	835	};
	836	result[@backingInt(Feature.fuse_fcsel)] = .{
	837	.llvm_name = "fuse-fcsel",
	838	.description = "CPU can fuse FCMP and FCSEL operations",
	839	.dependencies = featureSet(&[_]Feature{}),
	840	};
816	841	result[@backingInt(Feature.fuse_literals)] = .{
817	842	.llvm_name = "fuse-literals",
818	843	.description = "CPU fuses literal generation operations",
...	...	@@ -845,14 +870,19 @@ pub const all_features = blk: {
845	870	.description = "Harden against straight line speculation across RET and BR instructions",
846	871	.dependencies = featureSet(&[_]Feature{}),
847	872	};
	873	result[@backingInt(Feature.has_limited_64bit_vector_mul_bandwidth)] = .{
	874	.llvm_name = "has-limited-64bit-vector-mul-bandwidth",
	875	.description = "Has limited 64bit vector multiply bandwidth compared to scalar multiply",
	876	.dependencies = featureSet(&[_]Feature{}),
	877	};
848	878	result[@backingInt(Feature.hbc)] = .{
849	879	.llvm_name = "hbc",
850	880	.description = "Enable Armv8.8-A Hinted Conditional Branches Extension",
851	881	.dependencies = featureSet(&[_]Feature{}),
852	882	};
853		result[@backingInt(Feature.hcx)] = .{
854		.llvm_name = "hcx",
855		.description = "Enable Armv8.7-A HCRX_EL2 system register",
	883	result[@backingInt(Feature.hinte)] = .{
	884	.llvm_name = "hinte",
	885	.description = "Enable extended A64 hint instruction space",
856	886	.dependencies = featureSet(&[_]Feature{}),
857	887	};
858	888	result[@backingInt(Feature.i8mm)] = .{
...	...	@@ -962,11 +992,6 @@ pub const all_features = blk: {
962	992	.description = "Enable Armv8.4-A Memory system Partitioning and Monitoring extension",
963	993	.dependencies = featureSet(&[_]Feature{}),
964	994	};
965		result[@backingInt(Feature.mpamv2)] = .{
966		.llvm_name = "mpamv2",
967		.description = "Enable Armv9.7-A MPAMv2 Lookaside Buffer Invalidate instructions",
968		.dependencies = featureSet(&[_]Feature{}),
969		};
970	995	result[@backingInt(Feature.mte)] = .{
971	996	.llvm_name = "mte",
972	997	.description = "Enable Memory Tagging Extension",
...	...	@@ -996,6 +1021,16 @@ pub const all_features = blk: {
996	1021	.description = "Don't place a BTI instruction after a return-twice",
997	1022	.dependencies = featureSet(&[_]Feature{}),
998	1023	};
	1024	result[@backingInt(Feature.no_lfi_loads)] = .{
	1025	.llvm_name = "no-lfi-loads",
	1026	.description = "Disable LFI sandboxing for load instructions",
	1027	.dependencies = featureSet(&[_]Feature{}),
	1028	};
	1029	result[@backingInt(Feature.no_lfi_stores)] = .{
	1030	.llvm_name = "no-lfi-stores",
	1031	.description = "Disable LFI sandboxing for store instructions",
	1032	.dependencies = featureSet(&[_]Feature{}),
	1033	};
999	1034	result[@backingInt(Feature.no_neg_immediates)] = .{
1000	1035	.llvm_name = "no-neg-immediates",
1001	1036	.description = "Convert immediates and instructions to their negated or complemented equivalent when the immediate does not fit in the encoding.",
...	...	@@ -1063,11 +1098,6 @@ pub const all_features = blk: {
1063	1098	.description = "Enable Armv9.5-A PAC enhancements",
1064	1099	.dependencies = featureSet(&[_]Feature{}),
1065	1100	};
1066		result[@backingInt(Feature.pcdphint)] = .{
1067		.llvm_name = "pcdphint",
1068		.description = "Enable Armv9.6-A Producer Consumer Data Placement hints",
1069		.dependencies = featureSet(&[_]Feature{}),
1070		};
1071	1101	result[@backingInt(Feature.perfmon)] = .{
1072	1102	.llvm_name = "perfmon",
1073	1103	.description = "Enable Armv8.0-A PMUv3 Performance Monitors extension",
...	...	@@ -1813,7 +1843,6 @@ pub const all_features = blk: {
1813	1843	.llvm_name = "v8.7a",
1814	1844	.description = "Support ARM v8.7a architecture",
1815	1845	.dependencies = featureSet(&[_]Feature{
1816		.hcx,
1817	1846	.spe_eef,
1818	1847	.v8_6a,
1819	1848	.wfxt,
...	...	@@ -1931,6 +1960,7 @@ pub const all_features = blk: {
1931	1960	.description = "Support ARM v9.6a architecture",
1932	1961	.dependencies = featureSet(&[_]Feature{
1933	1962	.cmpbr,
	1963	.hinte,
1934	1964	.lsui,
1935	1965	.occmo,
1936	1966	.v9_5a,
...	...	@@ -2265,12 +2295,14 @@ pub const cpu = struct {
2265	2295	.features = featureSet(&[_]Feature{
2266	2296	.aes,
2267	2297	.aggressive_fma,
	2298	.align_cmp_csel_pairs,
2268	2299	.alternate_sextload_cvt_f32_pattern,
2269	2300	.altnzcv,
2270	2301	.arith_bcc_fusion,
2271	2302	.arith_cbz_fusion,
2272	2303	.ccdp,
2273	2304	.disable_latency_sched_heuristic,
	2305	.fast_ld1_single,
2274	2306	.fp16fml,
2275	2307	.fptoint,
2276	2308	.fuse_address,
...	...	@@ -2301,10 +2333,12 @@ pub const cpu = struct {
2301	2333	.llvm_name = "apple-a15",
2302	2334	.features = featureSet(&[_]Feature{
2303	2335	.aes,
	2336	.align_cmp_csel_pairs,
2304	2337	.alternate_sextload_cvt_f32_pattern,
2305	2338	.arith_bcc_fusion,
2306	2339	.arith_cbz_fusion,
2307	2340	.disable_latency_sched_heuristic,
	2341	.fast_ld1_single,
2308	2342	.fp16fml,
2309	2343	.fpac,
2310	2344	.fuse_address,
...	...	@@ -2332,10 +2366,12 @@ pub const cpu = struct {
2332	2366	.llvm_name = "apple-a16",
2333	2367	.features = featureSet(&[_]Feature{
2334	2368	.aes,
	2369	.align_cmp_csel_pairs,
2335	2370	.alternate_sextload_cvt_f32_pattern,
2336	2371	.arith_bcc_fusion,
2337	2372	.arith_cbz_fusion,
2338	2373	.disable_latency_sched_heuristic,
	2374	.fast_ld1_single,
2339	2375	.fp16fml,
2340	2376	.fpac,
2341	2377	.fuse_address,
...	...	@@ -2344,8 +2380,8 @@ pub const cpu = struct {
2344	2380	.fuse_arith_logic,
2345	2381	.fuse_crypto_eor,
2346	2382	.fuse_csel,
	2383	.fuse_fcsel,
2347	2384	.fuse_literals,
2348		.hcx,
2349	2385	.max_interleave_factor_4,
2350	2386	.no_zcz_fpr64,
2351	2387	.perfmon,
...	...	@@ -2364,10 +2400,12 @@ pub const cpu = struct {
2364	2400	.llvm_name = "apple-a17",
2365	2401	.features = featureSet(&[_]Feature{
2366	2402	.aes,
	2403	.align_cmp_csel_pairs,
2367	2404	.alternate_sextload_cvt_f32_pattern,
2368	2405	.arith_bcc_fusion,
2369	2406	.arith_cbz_fusion,
2370	2407	.disable_latency_sched_heuristic,
	2408	.fast_ld1_single,
2371	2409	.fp16fml,
2372	2410	.fpac,
2373	2411	.fuse_address,
...	...	@@ -2376,8 +2414,8 @@ pub const cpu = struct {
2376	2414	.fuse_arith_logic,
2377	2415	.fuse_crypto_eor,
2378	2416	.fuse_csel,
	2417	.fuse_fcsel,
2379	2418	.fuse_literals,
2380		.hcx,
2381	2419	.max_interleave_factor_4,
2382	2420	.no_zcz_fpr64,
2383	2421	.perfmon,
...	...	@@ -2396,10 +2434,12 @@ pub const cpu = struct {
2396	2434	.llvm_name = "apple-a18",
2397	2435	.features = featureSet(&[_]Feature{
2398	2436	.aes,
	2437	.align_cmp_csel_pairs,
2399	2438	.alternate_sextload_cvt_f32_pattern,
2400	2439	.arith_bcc_fusion,
2401	2440	.arith_cbz_fusion,
2402	2441	.disable_latency_sched_heuristic,
	2442	.fast_ld1_single,
2403	2443	.fp16fml,
2404	2444	.fpac,
2405	2445	.fuse_address,
...	...	@@ -2408,6 +2448,7 @@ pub const cpu = struct {
2408	2448	.fuse_arith_logic,
2409	2449	.fuse_crypto_eor,
2410	2450	.fuse_csel,
	2451	.fuse_fcsel,
2411	2452	.fuse_literals,
2412	2453	.max_interleave_factor_4,
2413	2454	.no_zcz_fpr64,
...	...	@@ -2429,11 +2470,13 @@ pub const cpu = struct {
2429	2470	.llvm_name = "apple-a19",
2430	2471	.features = featureSet(&[_]Feature{
2431	2472	.aes,
	2473	.align_cmp_csel_pairs,
2432	2474	.alternate_sextload_cvt_f32_pattern,
2433	2475	.arith_bcc_fusion,
2434	2476	.arith_cbz_fusion,
2435	2477	.cssc,
2436	2478	.disable_latency_sched_heuristic,
	2479	.fast_ld1_single,
2437	2480	.fp16fml,
2438	2481	.fpac,
2439	2482	.fuse_address,
...	...	@@ -2442,6 +2485,7 @@ pub const cpu = struct {
2442	2485	.fuse_arith_logic,
2443	2486	.fuse_crypto_eor,
2444	2487	.fuse_csel,
	2488	.fuse_fcsel,
2445	2489	.fuse_literals,
2446	2490	.hbc,
2447	2491	.max_interleave_factor_4,
...	...	@@ -2541,12 +2585,14 @@ pub const cpu = struct {
2541	2585	.features = featureSet(&[_]Feature{
2542	2586	.aes,
2543	2587	.aggressive_fma,
	2588	.align_cmp_csel_pairs,
2544	2589	.alternate_sextload_cvt_f32_pattern,
2545	2590	.altnzcv,
2546	2591	.arith_bcc_fusion,
2547	2592	.arith_cbz_fusion,
2548	2593	.ccdp,
2549	2594	.disable_latency_sched_heuristic,
	2595	.fast_ld1_single,
2550	2596	.fp16fml,
2551	2597	.fptoint,
2552	2598	.fuse_address,
...	...	@@ -2577,10 +2623,12 @@ pub const cpu = struct {
2577	2623	.llvm_name = "apple-m2",
2578	2624	.features = featureSet(&[_]Feature{
2579	2625	.aes,
	2626	.align_cmp_csel_pairs,
2580	2627	.alternate_sextload_cvt_f32_pattern,
2581	2628	.arith_bcc_fusion,
2582	2629	.arith_cbz_fusion,
2583	2630	.disable_latency_sched_heuristic,
	2631	.fast_ld1_single,
2584	2632	.fp16fml,
2585	2633	.fpac,
2586	2634	.fuse_address,
...	...	@@ -2608,10 +2656,12 @@ pub const cpu = struct {
2608	2656	.llvm_name = "apple-m3",
2609	2657	.features = featureSet(&[_]Feature{
2610	2658	.aes,
	2659	.align_cmp_csel_pairs,
2611	2660	.alternate_sextload_cvt_f32_pattern,
2612	2661	.arith_bcc_fusion,
2613	2662	.arith_cbz_fusion,
2614	2663	.disable_latency_sched_heuristic,
	2664	.fast_ld1_single,
2615	2665	.fp16fml,
2616	2666	.fpac,
2617	2667	.fuse_address,
...	...	@@ -2620,8 +2670,8 @@ pub const cpu = struct {
2620	2670	.fuse_arith_logic,
2621	2671	.fuse_crypto_eor,
2622	2672	.fuse_csel,
	2673	.fuse_fcsel,
2623	2674	.fuse_literals,
2624		.hcx,
2625	2675	.max_interleave_factor_4,
2626	2676	.no_zcz_fpr64,
2627	2677	.perfmon,
...	...	@@ -2640,10 +2690,12 @@ pub const cpu = struct {
2640	2690	.llvm_name = "apple-m4",
2641	2691	.features = featureSet(&[_]Feature{
2642	2692	.aes,
	2693	.align_cmp_csel_pairs,
2643	2694	.alternate_sextload_cvt_f32_pattern,
2644	2695	.arith_bcc_fusion,
2645	2696	.arith_cbz_fusion,
2646	2697	.disable_latency_sched_heuristic,
	2698	.fast_ld1_single,
2647	2699	.fp16fml,
2648	2700	.fpac,
2649	2701	.fuse_address,
...	...	@@ -2652,6 +2704,7 @@ pub const cpu = struct {
2652	2704	.fuse_arith_logic,
2653	2705	.fuse_crypto_eor,
2654	2706	.fuse_csel,
	2707	.fuse_fcsel,
2655	2708	.fuse_literals,
2656	2709	.max_interleave_factor_4,
2657	2710	.no_zcz_fpr64,
...	...	@@ -2673,11 +2726,13 @@ pub const cpu = struct {
2673	2726	.llvm_name = "apple-m5",
2674	2727	.features = featureSet(&[_]Feature{
2675	2728	.aes,
	2729	.align_cmp_csel_pairs,
2676	2730	.alternate_sextload_cvt_f32_pattern,
2677	2731	.arith_bcc_fusion,
2678	2732	.arith_cbz_fusion,
2679	2733	.cssc,
2680	2734	.disable_latency_sched_heuristic,
	2735	.fast_ld1_single,
2681	2736	.fp16fml,
2682	2737	.fpac,
2683	2738	.fuse_address,
...	...	@@ -2686,6 +2741,7 @@ pub const cpu = struct {
2686	2741	.fuse_arith_logic,
2687	2742	.fuse_crypto_eor,
2688	2743	.fuse_csel,
	2744	.fuse_fcsel,
2689	2745	.fuse_literals,
2690	2746	.hbc,
2691	2747	.max_interleave_factor_4,
...	...	@@ -2712,10 +2768,12 @@ pub const cpu = struct {
2712	2768	.llvm_name = "apple-s10",
2713	2769	.features = featureSet(&[_]Feature{
2714	2770	.aes,
	2771	.align_cmp_csel_pairs,
2715	2772	.alternate_sextload_cvt_f32_pattern,
2716	2773	.arith_bcc_fusion,
2717	2774	.arith_cbz_fusion,
2718	2775	.disable_latency_sched_heuristic,
	2776	.fast_ld1_single,
2719	2777	.fp16fml,
2720	2778	.fpac,
2721	2779	.fuse_address,
...	...	@@ -2724,8 +2782,8 @@ pub const cpu = struct {
2724	2782	.fuse_arith_logic,
2725	2783	.fuse_crypto_eor,
2726	2784	.fuse_csel,
	2785	.fuse_fcsel,
2727	2786	.fuse_literals,
2728		.hcx,
2729	2787	.max_interleave_factor_4,
2730	2788	.no_zcz_fpr64,
2731	2789	.perfmon,
...	...	@@ -2864,10 +2922,12 @@ pub const cpu = struct {
2864	2922	.llvm_name = "apple-s9",
2865	2923	.features = featureSet(&[_]Feature{
2866	2924	.aes,
	2925	.align_cmp_csel_pairs,
2867	2926	.alternate_sextload_cvt_f32_pattern,
2868	2927	.arith_bcc_fusion,
2869	2928	.arith_cbz_fusion,
2870	2929	.disable_latency_sched_heuristic,
	2930	.fast_ld1_single,
2871	2931	.fp16fml,
2872	2932	.fpac,
2873	2933	.fuse_address,
...	...	@@ -2876,8 +2936,8 @@ pub const cpu = struct {
2876	2936	.fuse_arith_logic,
2877	2937	.fuse_crypto_eor,
2878	2938	.fuse_csel,
	2939	.fuse_fcsel,
2879	2940	.fuse_literals,
2880		.hcx,
2881	2941	.max_interleave_factor_4,
2882	2942	.no_zcz_fpr64,
2883	2943	.perfmon,
...	...	@@ -2891,6 +2951,34 @@ pub const cpu = struct {
2891	2951	.zcz_gpr64,
2892	2952	}),
2893	2953	};
	2954	pub const armagicpu: CpuModel = .{
	2955	.name = "armagicpu",
	2956	.llvm_name = "armagicpu",
	2957	.features = featureSet(&[_]Feature{
	2958	.alu_lsl_fast,
	2959	.avoid_ldapur,
	2960	.brbe,
	2961	.enable_select_opt,
	2962	.ete,
	2963	.fp16fml,
	2964	.fpac,
	2965	.fuse_adrp_add,
	2966	.fuse_aes,
	2967	.fuse_csel,
	2968	.fuse_cset,
	2969	.has_limited_64bit_vector_mul_bandwidth,
	2970	.ls64,
	2971	.mte,
	2972	.perfmon,
	2973	.predictable_select_expensive,
	2974	.rand,
	2975	.spe,
	2976	.sve_bitperm,
	2977	.use_fixed_over_scalable_if_equal_cost,
	2978	.use_postra_scheduler,
	2979	.v9_2a,
	2980	}),
	2981	};
2894	2982	pub const c1_nano: CpuModel = .{
2895	2983	.name = "c1_nano",
2896	2984	.llvm_name = "c1-nano",
...	...	@@ -2987,6 +3075,7 @@ pub const cpu = struct {
2987	3075	.fuse_aes,
2988	3076	.fuse_csel,
2989	3077	.fuse_cset,
	3078	.max_interleave_factor_4,
2990	3079	.mte,
2991	3080	.perfmon,
2992	3081	.predictable_select_expensive,
...	...	@@ -3016,7 +3105,6 @@ pub const cpu = struct {
3016	3105	.features = featureSet(&[_]Feature{
3017	3106	.alu_lsl_fast,
3018	3107	.bf16,
3019		.disable_maximize_scalable_bandwidth,
3020	3108	.enable_select_opt,
3021	3109	.ete,
3022	3110	.fp16fml,
...	...	@@ -3025,11 +3113,13 @@ pub const cpu = struct {
3025	3113	.fuse_aes,
3026	3114	.fuse_csel,
3027	3115	.fuse_cset,
	3116	.has_limited_64bit_vector_mul_bandwidth,
3028	3117	.i8mm,
3029	3118	.mte,
3030	3119	.perfmon,
3031	3120	.predictable_select_expensive,
3032	3121	.sve_bitperm,
	3122	.use_fixed_over_scalable_if_equal_cost,
3033	3123	.use_postra_scheduler,
3034	3124	.v9a,
3035	3125	}),
...	...	@@ -3562,6 +3652,7 @@ pub const cpu = struct {
3562	3652	.fullfp16,
3563	3653	.fuse_adrp_add,
3564	3654	.fuse_aes,
	3655	.max_interleave_factor_4,
3565	3656	.perfmon,
3566	3657	.predictable_select_expensive,
3567	3658	.rcpc,
...	...	@@ -3587,6 +3678,7 @@ pub const cpu = struct {
3587	3678	.fuse_adrp_add,
3588	3679	.fuse_aes,
3589	3680	.lse2,
	3681	.max_interleave_factor_4,
3590	3682	.pauth,
3591	3683	.perfmon,
3592	3684	.predictable_select_expensive,
...	...	@@ -3612,6 +3704,7 @@ pub const cpu = struct {
3612	3704	.fuse_adrp_add,
3613	3705	.fuse_aes,
3614	3706	.i8mm,
	3707	.max_interleave_factor_4,
3615	3708	.mte,
3616	3709	.perfmon,
3617	3710	.predictable_select_expensive,
...	...	@@ -3635,6 +3728,7 @@ pub const cpu = struct {
3635	3728	.fuse_adrp_add,
3636	3729	.fuse_aes,
3637	3730	.i8mm,
	3731	.max_interleave_factor_4,
3638	3732	.mte,
3639	3733	.perfmon,
3640	3734	.predictable_select_expensive,
...	...	@@ -3659,6 +3753,7 @@ pub const cpu = struct {
3659	3753	.fuse_aes,
3660	3754	.fuse_csel,
3661	3755	.fuse_cset,
	3756	.max_interleave_factor_4,
3662	3757	.mte,
3663	3758	.perfmon,
3664	3759	.predictable_select_expensive,
...	...	@@ -3683,6 +3778,7 @@ pub const cpu = struct {
3683	3778	.fuse_aes,
3684	3779	.fuse_csel,
3685	3780	.fuse_cset,
	3781	.max_interleave_factor_4,
3686	3782	.mte,
3687	3783	.perfmon,
3688	3784	.predictable_select_expensive,
...	...	@@ -3898,6 +3994,7 @@ pub const cpu = struct {
3898	3994	.fuse_aes,
3899	3995	.fuse_csel,
3900	3996	.fuse_cset,
	3997	.max_interleave_factor_4,
3901	3998	.mte,
3902	3999	.perfmon,
3903	4000	.predictable_select_expensive,
...	...	@@ -3917,6 +4014,7 @@ pub const cpu = struct {
3917	4014	.features = featureSet(&[_]Feature{
3918	4015	.enable_select_opt,
3919	4016	.ete,
	4017	.fixed_load_latency_4,
3920	4018	.fuse_adrp_add,
3921	4019	.fuse_aes,
3922	4020	.neon,
...	...	@@ -3955,6 +4053,36 @@ pub const cpu = struct {
3955	4053	.v9a,
3956	4054	}),
3957	4055	};
	4056	pub const hip12: CpuModel = .{
	4057	.name = "hip12",
	4058	.llvm_name = "hip12",
	4059	.features = featureSet(&[_]Feature{
	4060	.addr_lsl_slow_14,
	4061	.alu_lsl_fast,
	4062	.brbe,
	4063	.cmp_bcc_fusion,
	4064	.ete,
	4065	.fp16fml,
	4066	.fpac,
	4067	.fuse_aes,
	4068	.hbc,
	4069	.ls64,
	4070	.nmi,
	4071	.perfmon,
	4072	.rand,
	4073	.rcpc3,
	4074	.rme,
	4075	.spe,
	4076	.store_pair_suppress,
	4077	.sve2,
	4078	.sve_aes,
	4079	.sve_bitperm,
	4080	.sve_sha3,
	4081	.sve_sm4,
	4082	.use_postra_scheduler,
	4083	.v8_7a,
	4084	}),
	4085	};
3958	4086	pub const kryo: CpuModel = .{
3959	4087	.name = "kryo",
3960	4088	.llvm_name = "kryo",
...	...	@@ -4045,7 +4173,6 @@ pub const cpu = struct {
4045	4173	.features = featureSet(&[_]Feature{
4046	4174	.alu_lsl_fast,
4047	4175	.bf16,
4048		.disable_maximize_scalable_bandwidth,
4049	4176	.enable_select_opt,
4050	4177	.ete,
4051	4178	.fp16fml,
...	...	@@ -4054,11 +4181,13 @@ pub const cpu = struct {
4054	4181	.fuse_aes,
4055	4182	.fuse_csel,
4056	4183	.fuse_cset,
	4184	.has_limited_64bit_vector_mul_bandwidth,
4057	4185	.i8mm,
4058	4186	.mte,
4059	4187	.perfmon,
4060	4188	.predictable_select_expensive,
4061	4189	.sve_bitperm,
	4190	.use_fixed_over_scalable_if_equal_cost,
4062	4191	.use_postra_scheduler,
4063	4192	.v9a,
4064	4193	}),
...	...	@@ -4169,6 +4298,7 @@ pub const cpu = struct {
4169	4298	.rand,
4170	4299	.spe,
4171	4300	.sve_bitperm,
	4301	.use_fixed_over_scalable_if_equal_cost,
4172	4302	.use_postra_scheduler,
4173	4303	.v9_2a,
4174	4304	}),
...	...	@@ -4188,14 +4318,15 @@ pub const cpu = struct {
4188	4318	.fuse_aes,
4189	4319	.fuse_csel,
4190	4320	.fuse_cset,
	4321	.has_limited_64bit_vector_mul_bandwidth,
4191	4322	.ls64,
4192		.max_interleave_factor_4,
4193	4323	.mte,
4194	4324	.perfmon,
4195	4325	.predictable_select_expensive,
4196	4326	.rand,
4197	4327	.spe,
4198	4328	.sve_bitperm,
	4329	.use_fixed_over_scalable_if_equal_cost,
4199	4330	.use_postra_scheduler,
4200	4331	.v9_2a,
4201	4332	}),
...	...	@@ -4248,6 +4379,33 @@ pub const cpu = struct {
4248	4379	.v8_6a,
4249	4380	}),
4250	4381	};
	4382	pub const rigel: CpuModel = .{
	4383	.name = "rigel",
	4384	.llvm_name = "rigel",
	4385	.features = featureSet(&[_]Feature{
	4386	.brbe,
	4387	.chk,
	4388	.ete,
	4389	.faminmax,
	4390	.fp16fml,
	4391	.fp8dot2,
	4392	.fp8dot4,
	4393	.fp8fma,
	4394	.fpac,
	4395	.ls64,
	4396	.lut,
	4397	.mte,
	4398	.olympus,
	4399	.perfmon,
	4400	.rand,
	4401	.spe,
	4402	.sve_aes,
	4403	.sve_bitperm,
	4404	.sve_sha3,
	4405	.sve_sm4,
	4406	.v9_2a,
	4407	}),
	4408	};
4251	4409	pub const saphira: CpuModel = .{
4252	4410	.name = "saphira",
4253	4411	.llvm_name = "saphira",

lib/std/Target/amdgcn.zig+998-133

...	...	@@ -6,6 +6,7 @@ const CpuModel = std.Target.Cpu.Model;
6	6
7	7	pub const Feature = enum {
8	8	@"1024_addressable_vgprs",
	9	@"1536_physical_vgprs",
9	10	@"16_bit_insts",
10	11	@"45_bit_num_records_buffer_resource",
11	12	@"64_bit_literals",
...	...	@@ -18,12 +19,12 @@ pub const Feature = enum {
18	19	addressablelocalmemorysize327680,
19	20	addressablelocalmemorysize65536,
20	21	agent_scope_fine_grained_remote_memory_atomics,
21		allocate1_5xvgprs,
22	22	aperture_regs,
23	23	architected_flat_scratch,
24	24	architected_sgprs,
25	25	ashr_pk_insts,
26	26	assembler_permissive_wavesize,
	27	asynccnt,
27	28	atomic_buffer_global_pk_add_f16_insts,
28	29	atomic_buffer_global_pk_add_f16_no_rtn_insts,
29	30	atomic_buffer_pk_add_bf16_inst,
...	...	@@ -75,9 +76,12 @@ pub const Feature = enum {
75	76	dpp,
76	77	dpp8,
77	78	dpp_64bit,
	79	dpp_row_bcast,
78	80	dpp_src1_sgpr,
	81	dpp_wavefront_shifts,
79	82	ds128,
80	83	ds_src2_insts,
	84	dx10_clamp_and_ieee_mode,
81	85	emulated_system_scope_atomics,
82	86	extended_image_insts,
83	87	f16bf16_to_fp6bf6_cvt_scale_insts,
...	...	@@ -91,13 +95,17 @@ pub const Feature = enum {
91	95	flat_global_insts,
92	96	flat_gvs_mode,
93	97	flat_inst_offsets,
	98	flat_offset_bits_12,
	99	flat_offset_bits_24,
94	100	flat_scratch,
95	101	flat_scratch_insts,
96	102	flat_segment_offset_bug,
	103	flat_signed_offset,
97	104	fma_mix_bf16_insts,
98	105	fma_mix_insts,
99	106	fmacf64_inst,
100	107	fmaf,
	108	formatted_mubuf_insts,
101	109	fp4_cvt_scale_insts,
102	110	fp64,
103	111	fp6bf6_cvt_scale_insts,
...	...	@@ -116,10 +124,15 @@ pub const Feature = enum {
116	124	gfx10_b_encoding,
117	125	gfx10_insts,
118	126	gfx11,
	127	gfx11_7_insts,
119	128	gfx11_insts,
120	129	gfx12,
121	130	gfx1250_insts,
	131	gfx1251_gemm_insts,
	132	gfx125x_lowest_rate_wmma,
122	133	gfx12_insts,
	134	gfx13,
	135	gfx13_insts,
123	136	gfx7_gfx8_gfx9_insts,
124	137	gfx8_insts,
125	138	gfx9,
...	...	@@ -135,6 +148,8 @@ pub const Feature = enum {
135	148	image_insts,
136	149	image_store_d16_bug,
137	150	inst_fwd_prefetch_bug,
	151	instcachelinesize128,
	152	instcachelinesize64,
138	153	int_clamp_insts,
139	154	inv_2pi_inline_imm,
140	155	kernarg_preload,
...	...	@@ -149,6 +164,7 @@ pub const Feature = enum {
149	164	mad_intra_fwd_bug,
150	165	mad_mac_f32_insts,
151	166	mad_mix_insts,
	167	mad_nc_64_32_insts,
152	168	mad_u32_inst,
153	169	mai_insts,
154	170	max_hard_clause_length_32,
...	...	@@ -161,22 +177,31 @@ pub const Feature = enum {
161	177	mfma_inline_literal_bug,
162	178	mimg_r128,
163	179	min3_max3_pkf16,
	180	min_max_i64_insts,
164	181	minimum3_maximum3_f16,
165	182	minimum3_maximum3_f32,
166	183	minimum3_maximum3_pkf16,
167	184	movrel,
	185	mqsad_insts,
	186	mqsad_pk_insts,
168	187	msaa_load_dst_sel_bug,
	188	msad_insts,
	189	mtbuf_insts,
169	190	negative_scratch_offset_bug,
170	191	negative_unaligned_scratch_offset_bug,
171	192	no_data_dep_hazard,
	193	no_f16_pseudo_scalar_trans_inline_constants,
172	194	no_sdst_cmpx,
173	195	nsa_clause_bug,
174	196	nsa_encoding,
175	197	nsa_to_vmem_bug,
176	198	offset_3f_bug,
177	199	packed_fp32_ops,
	200	packed_fp64_ops,
178	201	packed_tid,
	202	packed_u64_ops,
179	203	partial_nsa_encoding,
	204	permlane16_insts,
180	205	permlane16_swap,
181	206	permlane32_swap,
182	207	pk_add_min_max_insts,
...	...	@@ -185,23 +210,25 @@ pub const Feature = enum {
185	210	precise_memory,
186	211	priv_enabled_trap2_nop_bug,
187	212	prng_inst,
188		promote_alloca,
189	213	prt_strict_null,
190	214	pseudo_scalar_trans,
191	215	qsad_insts,
192	216	r128_a16,
193	217	real_true16,
194		relaxed_buffer_oob_mode,
195	218	required_export_priority,
196	219	requires_cov6,
	220	requires_initial_unclause_vmem,
197	221	restricted_soffset,
	222	s_barrier_leave_imm,
198	223	s_memrealtime,
199	224	s_memtime_inst,
200	225	s_wakeup_barrier_inst,
	226	s_wakeup_imm,
201	227	sad_insts,
202	228	safe_cu_prefetch,
203	229	safe_smem_prefetch,
204	230	salu_float,
	231	salu_minimum_maximum_insts,
205	232	scalar_atomics,
206	233	scalar_dwordx3_loads,
207	234	scalar_flat_scratch_insts,
...	...	@@ -219,13 +246,17 @@ pub const Feature = enum {
219	246	shader_cycles_hi_lo_registers,
220	247	shader_cycles_register,
221	248	si_scheduler,
	249	smem_prefetch_insts,
222	250	smem_to_vector_write_hazard,
223	251	southern_islands,
224	252	sramecc,
225	253	sramecc_support,
	254	swmmac_gfx1200_insts,
	255	swmmac_gfx1250_insts,
226	256	tanh_insts,
227	257	tensor_cvt_lut_insts,
228		tgsplit,
	258	tgsplit_support,
	259	trans_coexecution_hazard,
229	260	transpose_load_f4f6_insts,
230	261	trap_handler,
231	262	trig_reduced_range,
...	...	@@ -236,7 +267,10 @@ pub const Feature = enum {
236	267	unaligned_scratch_access,
237	268	unpacked_d16_vmem,
238	269	unsafe_ds_offset_folding,
	270	use_add_pc64_inst,
239	271	user_sgpr_init16_bug,
	272	v_mov_b64_inst,
	273	v_mul_u64_inst,
240	274	valu_trans_use_hazard,
241	275	vcmpx_exec_war_hazard,
242	276	vcmpx_permlane_hazard,
...	...	@@ -249,6 +283,7 @@ pub const Feature = enum {
249	283	volcanic_islands,
250	284	vop3_literal,
251	285	vop3p,
	286	vop3px2_increments_va_vdst_twice,
252	287	vopd,
253	288	vscnt,
254	289	wait_xcnt,
...	...	@@ -256,8 +291,12 @@ pub const Feature = enum {
256	291	wavefrontsize16,
257	292	wavefrontsize32,
258	293	wavefrontsize64,
	294	wmma_128b_insts,
	295	wmma_256b_insts,
	296	wmma_coexecution_hazards,
259	297	xf32_insts,
260	298	xnack,
	299	xnack_on_off_modes,
261	300	xnack_support,
262	301	};
263	302
...	...	@@ -276,6 +315,11 @@ pub const all_features = blk: {
276	315	.description = "Has 1024 addressable VGPRs",
277	316	.dependencies = featureSet(&[_]Feature{}),
278	317	};
	318	result[@backingInt(Feature.@"1536_physical_vgprs")] = .{
	319	.llvm_name = "1536-physical-vgprs",
	320	.description = "Has 1536 physical VGPRs per SIMD",
	321	.dependencies = featureSet(&[_]Feature{}),
	322	};
279	323	result[@backingInt(Feature.@"16_bit_insts")] = .{
280	324	.llvm_name = "16-bit-insts",
281	325	.description = "Has i16/f16 instructions",
...	...	@@ -336,11 +380,6 @@ pub const all_features = blk: {
336	380	.description = "Agent (device) scoped atomic operations, excluding those directly supported by PCIe (i.e. integer atomic add, exchange, and compare-and-swap), are functional for allocations in host or peer device memory.",
337	381	.dependencies = featureSet(&[_]Feature{}),
338	382	};
339		result[@backingInt(Feature.allocate1_5xvgprs)] = .{
340		.llvm_name = "allocate1_5xvgprs",
341		.description = "Has 50% more physical VGPRs and 50% larger allocation granule",
342		.dependencies = featureSet(&[_]Feature{}),
343		};
344	383	result[@backingInt(Feature.aperture_regs)] = .{
345	384	.llvm_name = "aperture-regs",
346	385	.description = "Has Memory Aperture Base and Size Registers",
...	...	@@ -363,7 +402,12 @@ pub const all_features = blk: {
363	402	};
364	403	result[@backingInt(Feature.assembler_permissive_wavesize)] = .{
365	404	.llvm_name = "assembler-permissive-wavesize",
366		.description = "allow parsing wave32 and wave64 variants of instructions",
	405	.description = "Allow parsing wave32 and wave64 variants of instructions",
	406	.dependencies = featureSet(&[_]Feature{}),
	407	};
	408	result[@backingInt(Feature.asynccnt)] = .{
	409	.llvm_name = "asynccnt",
	410	.description = "Has separate asynccnt counter",
367	411	.dependencies = featureSet(&[_]Feature{}),
368	412	};
369	413	result[@backingInt(Feature.atomic_buffer_global_pk_add_f16_insts)] = .{
...	...	@@ -637,11 +681,21 @@ pub const all_features = blk: {
637	681	.description = "Support DPP (Data Parallel Primitives) extension in DP ALU",
638	682	.dependencies = featureSet(&[_]Feature{}),
639	683	};
	684	result[@backingInt(Feature.dpp_row_bcast)] = .{
	685	.llvm_name = "dpp-row-bcast",
	686	.description = "Support DPP row_bcast15/row_bcast31",
	687	.dependencies = featureSet(&[_]Feature{}),
	688	};
640	689	result[@backingInt(Feature.dpp_src1_sgpr)] = .{
641	690	.llvm_name = "dpp-src1-sgpr",
642	691	.description = "Support SGPR for Src1 of DPP instructions",
643	692	.dependencies = featureSet(&[_]Feature{}),
644	693	};
	694	result[@backingInt(Feature.dpp_wavefront_shifts)] = .{
	695	.llvm_name = "dpp-wavefront-shifts",
	696	.description = "Support DPP wave_shl/wave_rol/wave_shr/wave_ror",
	697	.dependencies = featureSet(&[_]Feature{}),
	698	};
645	699	result[@backingInt(Feature.ds128)] = .{
646	700	.llvm_name = "enable-ds128",
647	701	.description = "Use ds_{read\|write}_b128",
...	...	@@ -652,6 +706,11 @@ pub const all_features = blk: {
652	706	.description = "Has ds_*_src2 instructions",
653	707	.dependencies = featureSet(&[_]Feature{}),
654	708	};
	709	result[@backingInt(Feature.dx10_clamp_and_ieee_mode)] = .{
	710	.llvm_name = "dx10-clamp-and-ieee-mode",
	711	.description = "Target has DX10_CLAMP and IEEE_MODE kernel descriptor bits",
	712	.dependencies = featureSet(&[_]Feature{}),
	713	};
655	714	result[@backingInt(Feature.emulated_system_scope_atomics)] = .{
656	715	.llvm_name = "emulated-system-scope-atomics",
657	716	.description = "System scope atomics unsupported by the PCI-e are emulated in HW via CAS loop and functional.",
...	...	@@ -723,6 +782,16 @@ pub const all_features = blk: {
723	782	.description = "Flat instructions have immediate offset addressing mode",
724	783	.dependencies = featureSet(&[_]Feature{}),
725	784	};
	785	result[@backingInt(Feature.flat_offset_bits_12)] = .{
	786	.llvm_name = "flat-offset-bits-12",
	787	.description = "Number of bits for flat offset encoding",
	788	.dependencies = featureSet(&[_]Feature{}),
	789	};
	790	result[@backingInt(Feature.flat_offset_bits_24)] = .{
	791	.llvm_name = "flat-offset-bits-24",
	792	.description = "Number of bits for flat offset encoding",
	793	.dependencies = featureSet(&[_]Feature{}),
	794	};
726	795	result[@backingInt(Feature.flat_scratch)] = .{
727	796	.llvm_name = "enable-flat-scratch",
728	797	.description = "Use scratch_* flat memory instructions to access scratch",
...	...	@@ -740,6 +809,11 @@ pub const all_features = blk: {
740	809	.description = "GFX10 bug where inst_offset is ignored when flat instructions access global memory",
741	810	.dependencies = featureSet(&[_]Feature{}),
742	811	};
	812	result[@backingInt(Feature.flat_signed_offset)] = .{
	813	.llvm_name = "flat-signed-offset",
	814	.description = "Immediate offset of FLAT instructions are always signed",
	815	.dependencies = featureSet(&[_]Feature{}),
	816	};
743	817	result[@backingInt(Feature.fma_mix_bf16_insts)] = .{
744	818	.llvm_name = "fma-mix-bf16-insts",
745	819	.description = "Has v_fma_mix_f32_bf16, v_fma_mixlo_bf16, v_fma_mixhi_bf16 instructions",
...	...	@@ -760,6 +834,11 @@ pub const all_features = blk: {
760	834	.description = "Enable single precision FMA (not as fast as mul+add, but fused)",
761	835	.dependencies = featureSet(&[_]Feature{}),
762	836	};
	837	result[@backingInt(Feature.formatted_mubuf_insts)] = .{
	838	.llvm_name = "formatted-mubuf-insts",
	839	.description = "Has formatted memory untyped buffer instructions.",
	840	.dependencies = featureSet(&[_]Feature{}),
	841	};
763	842	result[@backingInt(Feature.fp4_cvt_scale_insts)] = .{
764	843	.llvm_name = "fp4-cvt-scale-insts",
765	844	.description = "Has fp4 conversion scale instructions",
...	...	@@ -841,13 +920,16 @@ pub const all_features = blk: {
841	920	.default_component_zero,
842	921	.dpp,
843	922	.dpp8,
	923	.dx10_clamp_and_ieee_mode,
844	924	.extended_image_insts,
845	925	.fast_denormal_f32,
846	926	.fast_fmaf,
847	927	.flat_global_insts,
848	928	.flat_inst_offsets,
	929	.flat_offset_bits_12,
849	930	.flat_scratch_insts,
850	931	.fma_mix_insts,
	932	.formatted_mubuf_insts,
851	933	.fp64,
852	934	.g16,
853	935	.gds,
...	...	@@ -856,14 +938,20 @@ pub const all_features = blk: {
856	938	.gfx9_insts,
857	939	.gws,
858	940	.image_insts,
	941	.instcachelinesize64,
859	942	.int_clamp_insts,
860	943	.inv_2pi_inline_imm,
861	944	.lerp_inst,
862	945	.max_hard_clause_length_63,
863	946	.mimg_r128,
864	947	.movrel,
	948	.mqsad_insts,
	949	.mqsad_pk_insts,
	950	.msad_insts,
	951	.mtbuf_insts,
865	952	.no_data_dep_hazard,
866	953	.no_sdst_cmpx,
	954	.permlane16_insts,
867	955	.pk_fmac_f16_inst,
868	956	.qsad_insts,
869	957	.s_memrealtime,
...	...	@@ -929,6 +1017,7 @@ pub const all_features = blk: {
929	1017	.flat_inst_offsets,
930	1018	.flat_scratch_insts,
931	1019	.fma_mix_insts,
	1020	.formatted_mubuf_insts,
932	1021	.fp64,
933	1022	.g16,
934	1023	.gds,
...	...	@@ -940,14 +1029,20 @@ pub const all_features = blk: {
940	1029	.gfx8_insts,
941	1030	.gfx9_insts,
942	1031	.gws,
	1032	.instcachelinesize128,
943	1033	.int_clamp_insts,
944	1034	.inv_2pi_inline_imm,
945	1035	.lerp_inst,
946	1036	.max_hard_clause_length_32,
947	1037	.mimg_r128,
948	1038	.movrel,
	1039	.mqsad_insts,
	1040	.mqsad_pk_insts,
	1041	.msad_insts,
	1042	.mtbuf_insts,
949	1043	.no_data_dep_hazard,
950	1044	.no_sdst_cmpx,
	1045	.permlane16_insts,
951	1046	.pk_fmac_f16_inst,
952	1047	.qsad_insts,
953	1048	.sad_insts,
...	...	@@ -962,6 +1057,11 @@ pub const all_features = blk: {
962	1057	.vscnt,
963	1058	}),
964	1059	};
	1060	result[@backingInt(Feature.gfx11_7_insts)] = .{
	1061	.llvm_name = "gfx11-7-insts",
	1062	.description = "Additional instructions for GFX11.7",
	1063	.dependencies = featureSet(&[_]Feature{}),
	1064	};
965	1065	result[@backingInt(Feature.gfx11_insts)] = .{
966	1066	.llvm_name = "gfx11-insts",
967	1067	.description = "Additional instructions for GFX11+",
...	...	@@ -986,7 +1086,9 @@ pub const all_features = blk: {
986	1086	.fast_fmaf,
987	1087	.flat_global_insts,
988	1088	.flat_inst_offsets,
	1089	.flat_offset_bits_24,
989	1090	.flat_scratch_insts,
	1091	.flat_signed_offset,
990	1092	.fma_mix_insts,
991	1093	.fp64,
992	1094	.g16,
...	...	@@ -999,6 +1101,7 @@ pub const all_features = blk: {
999	1101	.gfx8_insts,
1000	1102	.gfx9_insts,
1001	1103	.ieee_minimum_maximum_insts,
	1104	.instcachelinesize128,
1002	1105	.int_clamp_insts,
1003	1106	.inv_2pi_inline_imm,
1004	1107	.max_hard_clause_length_32,
...	...	@@ -1008,7 +1111,9 @@ pub const all_features = blk: {
1008	1111	.movrel,
1009	1112	.no_data_dep_hazard,
1010	1113	.no_sdst_cmpx,
	1114	.permlane16_insts,
1011	1115	.pk_fmac_f16_inst,
	1116	.salu_minimum_maximum_insts,
1012	1117	.true16,
1013	1118	.unaligned_buffer_access,
1014	1119	.unaligned_ds_access,
...	...	@@ -1024,11 +1129,91 @@ pub const all_features = blk: {
1024	1129	.description = "Additional instructions for GFX1250+",
1025	1130	.dependencies = featureSet(&[_]Feature{}),
1026	1131	};
	1132	result[@backingInt(Feature.gfx1251_gemm_insts)] = .{
	1133	.llvm_name = "gfx1251-gemm-insts",
	1134	.description = "Has additional gfx1251 DGEMM instructions",
	1135	.dependencies = featureSet(&[_]Feature{}),
	1136	};
	1137	result[@backingInt(Feature.gfx125x_lowest_rate_wmma)] = .{
	1138	.llvm_name = "gfx125x-lowest-rate-wmma",
	1139	.description = "Has the lowest rate wmma in the gfx125x family, mainly for the gfx12-5-generic",
	1140	.dependencies = featureSet(&[_]Feature{}),
	1141	};
1027	1142	result[@backingInt(Feature.gfx12_insts)] = .{
1028	1143	.llvm_name = "gfx12-insts",
1029	1144	.description = "Additional instructions for GFX12+",
1030	1145	.dependencies = featureSet(&[_]Feature{}),
1031	1146	};
	1147	result[@backingInt(Feature.gfx13)] = .{
	1148	.llvm_name = "gfx13",
	1149	.description = "GFX13 GPU generation",
	1150	.dependencies = featureSet(&[_]Feature{
	1151	.@"16_bit_insts",
	1152	.a16,
	1153	.add_no_carry_insts,
	1154	.agent_scope_fine_grained_remote_memory_atomics,
	1155	.aperture_regs,
	1156	.atomic_fmin_fmax_flat_f32,
	1157	.atomic_fmin_fmax_global_f32,
	1158	.ci_insts,
	1159	.default_component_broadcast,
	1160	.dpp,
	1161	.dpp8,
	1162	.extended_image_insts,
	1163	.fast_denormal_f32,
	1164	.fast_fmaf,
	1165	.flat_global_insts,
	1166	.flat_inst_offsets,
	1167	.flat_offset_bits_24,
	1168	.flat_scratch_insts,
	1169	.flat_signed_offset,
	1170	.fma_mix_insts,
	1171	.formatted_mubuf_insts,
	1172	.fp64,
	1173	.g16,
	1174	.gfx10_3_insts,
	1175	.gfx10_a_encoding,
	1176	.gfx10_b_encoding,
	1177	.gfx10_insts,
	1178	.gfx11_insts,
	1179	.gfx12_insts,
	1180	.gfx13_insts,
	1181	.gfx8_insts,
	1182	.gfx9_insts,
	1183	.ieee_minimum_maximum_insts,
	1184	.image_insts,
	1185	.instcachelinesize128,
	1186	.int_clamp_insts,
	1187	.inv_2pi_inline_imm,
	1188	.max_hard_clause_length_32,
	1189	.mimg_r128,
	1190	.minimum3_maximum3_f16,
	1191	.minimum3_maximum3_f32,
	1192	.movrel,
	1193	.mtbuf_insts,
	1194	.no_data_dep_hazard,
	1195	.no_sdst_cmpx,
	1196	.permlane16_insts,
	1197	.pk_fmac_f16_inst,
	1198	.salu_minimum_maximum_insts,
	1199	.true16,
	1200	.unaligned_buffer_access,
	1201	.unaligned_ds_access,
	1202	.unaligned_scratch_access,
	1203	.vop3_literal,
	1204	.vop3p,
	1205	.vopd,
	1206	.vscnt,
	1207	}),
	1208	};
	1209	result[@backingInt(Feature.gfx13_insts)] = .{
	1210	.llvm_name = "gfx13-insts",
	1211	.description = "Additional instructions for GFX13+",
	1212	.dependencies = featureSet(&[_]Feature{
	1213	.s_barrier_leave_imm,
	1214	.s_wakeup_imm,
	1215	}),
	1216	};
1032	1217	result[@backingInt(Feature.gfx7_gfx8_gfx9_insts)] = .{
1033	1218	.llvm_name = "gfx7-gfx8-gfx9-insts",
1034	1219	.description = "Instructions shared in GFX7, GFX8, GFX9",
...	...	@@ -1054,20 +1239,29 @@ pub const all_features = blk: {
1054	1239	.cvt_pknorm_vop3_insts,
1055	1240	.default_component_zero,
1056	1241	.dpp,
	1242	.dpp_row_bcast,
	1243	.dpp_wavefront_shifts,
	1244	.dx10_clamp_and_ieee_mode,
1057	1245	.fast_denormal_f32,
1058	1246	.fast_fmaf,
1059	1247	.flat_global_insts,
1060	1248	.flat_inst_offsets,
1061	1249	.flat_scratch_insts,
	1250	.formatted_mubuf_insts,
1062	1251	.fp64,
1063	1252	.gcn3_encoding,
1064	1253	.gfx7_gfx8_gfx9_insts,
1065	1254	.gfx8_insts,
1066	1255	.gfx9_insts,
1067	1256	.gws,
	1257	.instcachelinesize64,
1068	1258	.int_clamp_insts,
1069	1259	.inv_2pi_inline_imm,
1070	1260	.lerp_inst,
	1261	.mqsad_insts,
	1262	.mqsad_pk_insts,
	1263	.msad_insts,
	1264	.mtbuf_insts,
1071	1265	.negative_scratch_offset_bug,
1072	1266	.qsad_insts,
1073	1267	.r128_a16,
...	...	@@ -1089,7 +1283,7 @@ pub const all_features = blk: {
1089	1283	.vmem_write_vgpr_in_order,
1090	1284	.vop3p,
1091	1285	.wavefrontsize64,
1092		.xnack_support,
	1286	.xnack_on_off_modes,
1093	1287	}),
1094	1288	};
1095	1289	result[@backingInt(Feature.gfx90a_insts)] = .{
...	...	@@ -1142,7 +1336,7 @@ pub const all_features = blk: {
1142	1336	};
1143	1337	result[@backingInt(Feature.ieee_minimum_maximum_insts)] = .{
1144	1338	.llvm_name = "ieee-minimum-maximum-insts",
1145		.description = "Has v_minimum/maximum_f16/f32/f64, v_minimummaximum/maximumminimum_f16/f32 and v_pk_minimum/maximum_f16 instructions",
	1339	.description = "Has v_minimum/maximum_f16/f32/f64, v_minimummaximum/maximumminimum_f16/f32 andv_pk_minimum/maximum_f16 instructions",
1146	1340	.dependencies = featureSet(&[_]Feature{}),
1147	1341	};
1148	1342	result[@backingInt(Feature.image_gather4_d16_bug)] = .{
...	...	@@ -1165,6 +1359,16 @@ pub const all_features = blk: {
1165	1359	.description = "S_INST_PREFETCH instruction causes shader to hang",
1166	1360	.dependencies = featureSet(&[_]Feature{}),
1167	1361	};
	1362	result[@backingInt(Feature.instcachelinesize128)] = .{
	1363	.llvm_name = "instcachelinesize128",
	1364	.description = "Instruction cache line size in bytes.",
	1365	.dependencies = featureSet(&[_]Feature{}),
	1366	};
	1367	result[@backingInt(Feature.instcachelinesize64)] = .{
	1368	.llvm_name = "instcachelinesize64",
	1369	.description = "Instruction cache line size in bytes.",
	1370	.dependencies = featureSet(&[_]Feature{}),
	1371	};
1168	1372	result[@backingInt(Feature.int_clamp_insts)] = .{
1169	1373	.llvm_name = "int-clamp-insts",
1170	1374	.description = "Support clamp for integer destination",
...	...	@@ -1235,6 +1439,11 @@ pub const all_features = blk: {
1235	1439	.description = "Has v_mad_mix_f32, v_mad_mixlo_f16, v_mad_mixhi_f16 instructions",
1236	1440	.dependencies = featureSet(&[_]Feature{}),
1237	1441	};
	1442	result[@backingInt(Feature.mad_nc_64_32_insts)] = .{
	1443	.llvm_name = "mad-nc-64-32-insts",
	1444	.description = "Has v_mad_nc_{u64_u32\|i64_i32} instructions",
	1445	.dependencies = featureSet(&[_]Feature{}),
	1446	};
1238	1447	result[@backingInt(Feature.mad_u32_inst)] = .{
1239	1448	.llvm_name = "mad-u32-inst",
1240	1449	.description = "Has v_mad_u32 instruction",
...	...	@@ -1295,6 +1504,11 @@ pub const all_features = blk: {
1295	1504	.description = "Has v_pk_min3_num_f16 and v_pk_max3_num_f16 instructions",
1296	1505	.dependencies = featureSet(&[_]Feature{}),
1297	1506	};
	1507	result[@backingInt(Feature.min_max_i64_insts)] = .{
	1508	.llvm_name = "min-max-i64-insts",
	1509	.description = "Has v_{min\|max}_{i\|u}64 instructions",
	1510	.dependencies = featureSet(&[_]Feature{}),
	1511	};
1298	1512	result[@backingInt(Feature.minimum3_maximum3_f16)] = .{
1299	1513	.llvm_name = "minimum3-maximum3-f16",
1300	1514	.description = "Has v_minimum3_f16 and v_maximum3_f16 instructions",
...	...	@@ -1315,11 +1529,31 @@ pub const all_features = blk: {
1315	1529	.description = "Has v_movrel*_b32 instructions",
1316	1530	.dependencies = featureSet(&[_]Feature{}),
1317	1531	};
	1532	result[@backingInt(Feature.mqsad_insts)] = .{
	1533	.llvm_name = "mqsad-insts",
	1534	.description = "Has v_mqsad_u32_u8 instruction",
	1535	.dependencies = featureSet(&[_]Feature{}),
	1536	};
	1537	result[@backingInt(Feature.mqsad_pk_insts)] = .{
	1538	.llvm_name = "mqsad-pk-insts",
	1539	.description = "Has v_mqsad_pk_u16_u8 instruction",
	1540	.dependencies = featureSet(&[_]Feature{}),
	1541	};
1318	1542	result[@backingInt(Feature.msaa_load_dst_sel_bug)] = .{
1319	1543	.llvm_name = "msaa-load-dst-sel-bug",
1320	1544	.description = "MSAA loads not honoring dst_sel bug",
1321	1545	.dependencies = featureSet(&[_]Feature{}),
1322	1546	};
	1547	result[@backingInt(Feature.msad_insts)] = .{
	1548	.llvm_name = "msad-insts",
	1549	.description = "Has v_msad_u8 instruction",
	1550	.dependencies = featureSet(&[_]Feature{}),
	1551	};
	1552	result[@backingInt(Feature.mtbuf_insts)] = .{
	1553	.llvm_name = "mtbuf-insts",
	1554	.description = "Has memory typed buffer instructions.",
	1555	.dependencies = featureSet(&[_]Feature{}),
	1556	};
1323	1557	result[@backingInt(Feature.negative_scratch_offset_bug)] = .{
1324	1558	.llvm_name = "negative-scratch-offset-bug",
1325	1559	.description = "Negative immediate offsets in scratch instructions with an SGPR offset page fault on GFX9",
...	...	@@ -1327,7 +1561,7 @@ pub const all_features = blk: {
1327	1561	};
1328	1562	result[@backingInt(Feature.negative_unaligned_scratch_offset_bug)] = .{
1329	1563	.llvm_name = "negative-unaligned-scratch-offset-bug",
1330		.description = "Scratch instructions with a VGPR offset and a negative immediate offset that is not a multiple of 4 read wrong memory on GFX10",
	1564	.description = "Scratch instructions with a VGPR offset and a negative immediate offset thatis not a multiple of 4 read wrong memory on GFX10",
1331	1565	.dependencies = featureSet(&[_]Feature{}),
1332	1566	};
1333	1567	result[@backingInt(Feature.no_data_dep_hazard)] = .{
...	...	@@ -1335,6 +1569,11 @@ pub const all_features = blk: {
1335	1569	.description = "Does not need SW waitstates",
1336	1570	.dependencies = featureSet(&[_]Feature{}),
1337	1571	};
	1572	result[@backingInt(Feature.no_f16_pseudo_scalar_trans_inline_constants)] = .{
	1573	.llvm_name = "no-f16-pseudo-scalar-trans-inline-constants",
	1574	.description = "Inline constants are not supported for F16 pseudo scalar transcendentals",
	1575	.dependencies = featureSet(&[_]Feature{}),
	1576	};
1338	1577	result[@backingInt(Feature.no_sdst_cmpx)] = .{
1339	1578	.llvm_name = "no-sdst-cmpx",
1340	1579	.description = "V_CMPX does not write VCC/SGPR in addition to EXEC",
...	...	@@ -1365,16 +1604,31 @@ pub const all_features = blk: {
1365	1604	.description = "Support packed fp32 instructions",
1366	1605	.dependencies = featureSet(&[_]Feature{}),
1367	1606	};
	1607	result[@backingInt(Feature.packed_fp64_ops)] = .{
	1608	.llvm_name = "packed-fp64-ops",
	1609	.description = "Support packed fp64 instructions",
	1610	.dependencies = featureSet(&[_]Feature{}),
	1611	};
1368	1612	result[@backingInt(Feature.packed_tid)] = .{
1369	1613	.llvm_name = "packed-tid",
1370	1614	.description = "Workitem IDs are packed into v0 at kernel launch",
1371	1615	.dependencies = featureSet(&[_]Feature{}),
1372	1616	};
	1617	result[@backingInt(Feature.packed_u64_ops)] = .{
	1618	.llvm_name = "packed-u64-ops",
	1619	.description = "Support packed uint64 instructions",
	1620	.dependencies = featureSet(&[_]Feature{}),
	1621	};
1373	1622	result[@backingInt(Feature.partial_nsa_encoding)] = .{
1374	1623	.llvm_name = "partial-nsa-encoding",
1375	1624	.description = "Support partial NSA encoding for image instructions",
1376	1625	.dependencies = featureSet(&[_]Feature{}),
1377	1626	};
	1627	result[@backingInt(Feature.permlane16_insts)] = .{
	1628	.llvm_name = "permlane16-insts",
	1629	.description = "Has v_permlane16_b32/v_permlanex16_b32 instructions",
	1630	.dependencies = featureSet(&[_]Feature{}),
	1631	};
1378	1632	result[@backingInt(Feature.permlane16_swap)] = .{
1379	1633	.llvm_name = "permlane16-swap",
1380	1634	.description = "Has v_permlane16_swap_b32 instructions",
...	...	@@ -1415,11 +1669,6 @@ pub const all_features = blk: {
1415	1669	.description = "Has v_prng_b32 instruction",
1416	1670	.dependencies = featureSet(&[_]Feature{}),
1417	1671	};
1418		result[@backingInt(Feature.promote_alloca)] = .{
1419		.llvm_name = "promote-alloca",
1420		.description = "Enable promote alloca pass",
1421		.dependencies = featureSet(&[_]Feature{}),
1422		};
1423	1672	result[@backingInt(Feature.prt_strict_null)] = .{
1424	1673	.llvm_name = "enable-prt-strict-null",
1425	1674	.description = "Enable zeroing of result registers for sparse texture fetches",
...	...	@@ -1445,11 +1694,6 @@ pub const all_features = blk: {
1445	1694	.description = "Use true 16-bit registers",
1446	1695	.dependencies = featureSet(&[_]Feature{}),
1447	1696	};
1448		result[@backingInt(Feature.relaxed_buffer_oob_mode)] = .{
1449		.llvm_name = "relaxed-buffer-oob-mode",
1450		.description = "Disable strict out-of-bounds buffer guarantees. An OOB access may potentially cause an adjacent access to be treated as if it were also OOB",
1451		.dependencies = featureSet(&[_]Feature{}),
1452		};
1453	1697	result[@backingInt(Feature.required_export_priority)] = .{
1454	1698	.llvm_name = "required-export-priority",
1455	1699	.description = "Export priority must be explicitly manipulated on GFX11.5",
...	...	@@ -1460,11 +1704,21 @@ pub const all_features = blk: {
1460	1704	.description = "Target Requires Code Object V6",
1461	1705	.dependencies = featureSet(&[_]Feature{}),
1462	1706	};
	1707	result[@backingInt(Feature.requires_initial_unclause_vmem)] = .{
	1708	.llvm_name = "requires-initial-unclause-vmem",
	1709	.description = "Hardware entrypoints require first VMEM instruction to be unclaused",
	1710	.dependencies = featureSet(&[_]Feature{}),
	1711	};
1463	1712	result[@backingInt(Feature.restricted_soffset)] = .{
1464	1713	.llvm_name = "restricted-soffset",
1465	1714	.description = "Has restricted SOffset (immediate not supported).",
1466	1715	.dependencies = featureSet(&[_]Feature{}),
1467	1716	};
	1717	result[@backingInt(Feature.s_barrier_leave_imm)] = .{
	1718	.llvm_name = "s-barrier-leave-imm",
	1719	.description = "s_barrier_leave takes an immediate operand",
	1720	.dependencies = featureSet(&[_]Feature{}),
	1721	};
1468	1722	result[@backingInt(Feature.s_memrealtime)] = .{
1469	1723	.llvm_name = "s-memrealtime",
1470	1724	.description = "Has s_memrealtime instruction",
...	...	@@ -1480,6 +1734,11 @@ pub const all_features = blk: {
1480	1734	.description = "Has s_wakeup_barrier instruction.",
1481	1735	.dependencies = featureSet(&[_]Feature{}),
1482	1736	};
	1737	result[@backingInt(Feature.s_wakeup_imm)] = .{
	1738	.llvm_name = "s-wakeup-imm",
	1739	.description = "s_wakeup takes an immediate operand",
	1740	.dependencies = featureSet(&[_]Feature{}),
	1741	};
1483	1742	result[@backingInt(Feature.sad_insts)] = .{
1484	1743	.llvm_name = "sad-insts",
1485	1744	.description = "Has v_sad* instructions",
...	...	@@ -1500,6 +1759,11 @@ pub const all_features = blk: {
1500	1759	.description = "Has SALU floating point instructions",
1501	1760	.dependencies = featureSet(&[_]Feature{}),
1502	1761	};
	1762	result[@backingInt(Feature.salu_minimum_maximum_insts)] = .{
	1763	.llvm_name = "salu-minimum-maximum-insts",
	1764	.description = "Has s_minimum/maximum_f16/f32 instructions",
	1765	.dependencies = featureSet(&[_]Feature{}),
	1766	};
1503	1767	result[@backingInt(Feature.scalar_atomics)] = .{
1504	1768	.llvm_name = "scalar-atomics",
1505	1769	.description = "Has atomic scalar memory instructions",
...	...	@@ -1564,16 +1828,23 @@ pub const all_features = blk: {
1564	1828	.cvt_pknorm_vop2_insts,
1565	1829	.default_component_zero,
1566	1830	.ds_src2_insts,
	1831	.dx10_clamp_and_ieee_mode,
1567	1832	.extended_image_insts,
	1833	.formatted_mubuf_insts,
1568	1834	.fp64,
1569	1835	.gds,
1570	1836	.gfx7_gfx8_gfx9_insts,
1571	1837	.gws,
1572	1838	.image_insts,
	1839	.instcachelinesize64,
1573	1840	.lerp_inst,
1574	1841	.mad_mac_f32_insts,
1575	1842	.mimg_r128,
1576	1843	.movrel,
	1844	.mqsad_insts,
	1845	.mqsad_pk_insts,
	1846	.msad_insts,
	1847	.mtbuf_insts,
1577	1848	.qsad_insts,
1578	1849	.s_memtime_inst,
1579	1850	.sad_insts,
...	...	@@ -1613,6 +1884,11 @@ pub const all_features = blk: {
1613	1884	.description = "Enable SI Machine Scheduler",
1614	1885	.dependencies = featureSet(&[_]Feature{}),
1615	1886	};
	1887	result[@backingInt(Feature.smem_prefetch_insts)] = .{
	1888	.llvm_name = "smem-prefetch-insts",
	1889	.description = "Has s_prefetch_inst, s_prefetch_inst_pc_rel, s_prefetch_data and s_prefetch_data_pc_rel instructions",
	1890	.dependencies = featureSet(&[_]Feature{}),
	1891	};
1616	1892	result[@backingInt(Feature.smem_to_vector_write_hazard)] = .{
1617	1893	.llvm_name = "smem-to-vector-write-hazard",
1618	1894	.description = "s_load_dword followed by v_cmp page faults",
...	...	@@ -1629,16 +1905,22 @@ pub const all_features = blk: {
1629	1905	.cvt_pknorm_vop2_insts,
1630	1906	.default_component_zero,
1631	1907	.ds_src2_insts,
	1908	.dx10_clamp_and_ieee_mode,
1632	1909	.extended_image_insts,
	1910	.formatted_mubuf_insts,
1633	1911	.fp64,
1634	1912	.gds,
1635	1913	.gws,
1636	1914	.image_insts,
	1915	.instcachelinesize64,
1637	1916	.ldsbankcount32,
1638	1917	.lerp_inst,
1639	1918	.mad_mac_f32_insts,
1640	1919	.mimg_r128,
1641	1920	.movrel,
	1921	.mqsad_pk_insts,
	1922	.msad_insts,
	1923	.mtbuf_insts,
1642	1924	.s_memtime_inst,
1643	1925	.sad_insts,
1644	1926	.trig_reduced_range,
...	...	@@ -1656,6 +1938,16 @@ pub const all_features = blk: {
1656	1938	.description = "Hardware supports SRAMECC",
1657	1939	.dependencies = featureSet(&[_]Feature{}),
1658	1940	};
	1941	result[@backingInt(Feature.swmmac_gfx1200_insts)] = .{
	1942	.llvm_name = "swmmac-gfx1200-insts",
	1943	.description = "Has GFX1200 SWMMAC instructions",
	1944	.dependencies = featureSet(&[_]Feature{}),
	1945	};
	1946	result[@backingInt(Feature.swmmac_gfx1250_insts)] = .{
	1947	.llvm_name = "swmmac-gfx1250-insts",
	1948	.description = "Has GFX1250 SWMMAC instructions",
	1949	.dependencies = featureSet(&[_]Feature{}),
	1950	};
1659	1951	result[@backingInt(Feature.tanh_insts)] = .{
1660	1952	.llvm_name = "tanh-insts",
1661	1953	.description = "Has v_tanh_f32/f16 instructions",
...	...	@@ -1666,9 +1958,14 @@ pub const all_features = blk: {
1666	1958	.description = "Has v_perm_pk16* instructions",
1667	1959	.dependencies = featureSet(&[_]Feature{}),
1668	1960	};
1669		result[@backingInt(Feature.tgsplit)] = .{
1670		.llvm_name = "tgsplit",
1671		.description = "Enable threadgroup split execution",
	1961	result[@backingInt(Feature.tgsplit_support)] = .{
	1962	.llvm_name = "tgsplit-support",
	1963	.description = "Hardware support for threadgroup split execution mode",
	1964	.dependencies = featureSet(&[_]Feature{}),
	1965	};
	1966	result[@backingInt(Feature.trans_coexecution_hazard)] = .{
	1967	.llvm_name = "trans-coexecution-hazard",
	1968	.description = "Hazard when TRANS co-executes with other VALU instruction",
1672	1969	.dependencies = featureSet(&[_]Feature{}),
1673	1970	};
1674	1971	result[@backingInt(Feature.transpose_load_f4f6_insts)] = .{
...	...	@@ -1721,11 +2018,26 @@ pub const all_features = blk: {
1721	2018	.description = "Force using DS instruction immediate offsets on SI",
1722	2019	.dependencies = featureSet(&[_]Feature{}),
1723	2020	};
	2021	result[@backingInt(Feature.use_add_pc64_inst)] = .{
	2022	.llvm_name = "use-add-pc64-inst",
	2023	.description = "Use s_add_pc_i64 instruction.",
	2024	.dependencies = featureSet(&[_]Feature{}),
	2025	};
1724	2026	result[@backingInt(Feature.user_sgpr_init16_bug)] = .{
1725	2027	.llvm_name = "user-sgpr-init16-bug",
1726	2028	.description = "Bug requiring at least 16 user+system SGPRs to be enabled",
1727	2029	.dependencies = featureSet(&[_]Feature{}),
1728	2030	};
	2031	result[@backingInt(Feature.v_mov_b64_inst)] = .{
	2032	.llvm_name = "v-mov-b64-inst",
	2033	.description = "Has v_mov_b64 instruction",
	2034	.dependencies = featureSet(&[_]Feature{}),
	2035	};
	2036	result[@backingInt(Feature.v_mul_u64_inst)] = .{
	2037	.llvm_name = "v-mul-u64-inst",
	2038	.description = "Has v_mul_u64 instruction.",
	2039	.dependencies = featureSet(&[_]Feature{}),
	2040	};
1729	2041	result[@backingInt(Feature.valu_trans_use_hazard)] = .{
1730	2042	.llvm_name = "valu-trans-use-hazard",
1731	2043	.description = "Hazard when TRANS instructions are closely followed by a use of the result",
...	...	@@ -1758,7 +2070,7 @@ pub const all_features = blk: {
1758	2070	};
1759	2071	result[@backingInt(Feature.vmem_to_lds_load_insts)] = .{
1760	2072	.llvm_name = "vmem-to-lds-load-insts",
1761		.description = "The platform has memory to lds instructions (global_load w/lds bit set, buffer_load w/lds bit set or global_load_lds. This does not include scratch_load_lds.",
	2073	.description = "The platform has memory to lds instructions (global_load w/lds bit set, buffer_loadw/lds bit set or global_load_lds. This does not include scratch_load_lds.",
1762	2074	.dependencies = featureSet(&[_]Feature{}),
1763	2075	};
1764	2076	result[@backingInt(Feature.vmem_to_scalar_write_hazard)] = .{
...	...	@@ -1782,10 +2094,14 @@ pub const all_features = blk: {
1782	2094	.cvt_pknorm_vop2_insts,
1783	2095	.default_component_zero,
1784	2096	.dpp,
	2097	.dpp_row_bcast,
	2098	.dpp_wavefront_shifts,
1785	2099	.ds_src2_insts,
	2100	.dx10_clamp_and_ieee_mode,
1786	2101	.extended_image_insts,
1787	2102	.fast_denormal_f32,
1788	2103	.flat_address_space,
	2104	.formatted_mubuf_insts,
1789	2105	.fp64,
1790	2106	.gcn3_encoding,
1791	2107	.gds,
...	...	@@ -1793,12 +2109,17 @@ pub const all_features = blk: {
1793	2109	.gfx8_insts,
1794	2110	.gws,
1795	2111	.image_insts,
	2112	.instcachelinesize64,
1796	2113	.int_clamp_insts,
1797	2114	.inv_2pi_inline_imm,
1798	2115	.lerp_inst,
1799	2116	.mad_mac_f32_insts,
1800	2117	.mimg_r128,
1801	2118	.movrel,
	2119	.mqsad_insts,
	2120	.mqsad_pk_insts,
	2121	.msad_insts,
	2122	.mtbuf_insts,
1802	2123	.qsad_insts,
1803	2124	.s_memrealtime,
1804	2125	.s_memtime_inst,
...	...	@@ -1824,6 +2145,11 @@ pub const all_features = blk: {
1824	2145	.description = "Has VOP3P packed instructions",
1825	2146	.dependencies = featureSet(&[_]Feature{}),
1826	2147	};
	2148	result[@backingInt(Feature.vop3px2_increments_va_vdst_twice)] = .{
	2149	.llvm_name = "vop3px2-increments-va-vdst-twice",
	2150	.description = "VOP3PX2 encoded instructions increment the VaVdst counter twice",
	2151	.dependencies = featureSet(&[_]Feature{}),
	2152	};
1827	2153	result[@backingInt(Feature.vopd)] = .{
1828	2154	.llvm_name = "vopd",
1829	2155	.description = "Has VOPD dual issue wave32 instructions",
...	...	@@ -1859,6 +2185,21 @@ pub const all_features = blk: {
1859	2185	.description = "The number of threads per wavefront",
1860	2186	.dependencies = featureSet(&[_]Feature{}),
1861	2187	};
	2188	result[@backingInt(Feature.wmma_128b_insts)] = .{
	2189	.llvm_name = "wmma-128b-insts",
	2190	.description = "Has WMMA instructions where A and B matrices do not have duplicated data",
	2191	.dependencies = featureSet(&[_]Feature{}),
	2192	};
	2193	result[@backingInt(Feature.wmma_256b_insts)] = .{
	2194	.llvm_name = "wmma-256b-insts",
	2195	.description = "Has WMMA instructions where A and B matrices have duplicated data",
	2196	.dependencies = featureSet(&[_]Feature{}),
	2197	};
	2198	result[@backingInt(Feature.wmma_coexecution_hazards)] = .{
	2199	.llvm_name = "wmma-coexecution-hazards",
	2200	.description = "Hazards when WMMA co-executes with other WMMA/VALU instruction",
	2201	.dependencies = featureSet(&[_]Feature{}),
	2202	};
1862	2203	result[@backingInt(Feature.xf32_insts)] = .{
1863	2204	.llvm_name = "xf32-insts",
1864	2205	.description = "Has instructions that support xf32 format, such as v_mfma_f32_16x16x8_xf32 and v_mfma_f32_32x32x4_xf32",
...	...	@@ -1867,7 +2208,16 @@ pub const all_features = blk: {
1867	2208	result[@backingInt(Feature.xnack)] = .{
1868	2209	.llvm_name = "xnack",
1869	2210	.description = "Enable XNACK support",
1870		.dependencies = featureSet(&[_]Feature{}),
	2211	.dependencies = featureSet(&[_]Feature{
	2212	.xnack_support,
	2213	}),
	2214	};
	2215	result[@backingInt(Feature.xnack_on_off_modes)] = .{
	2216	.llvm_name = "xnack-on-off-modes",
	2217	.description = "Target supports XNACK on/off modes",
	2218	.dependencies = featureSet(&[_]Feature{
	2219	.xnack_support,
	2220	}),
1871	2221	};
1872	2222	result[@backingInt(Feature.xnack_support)] = .{
1873	2223	.llvm_name = "xnack-support",
...	...	@@ -1900,7 +2250,7 @@ pub const cpu = struct {
1900	2250	.ldsbankcount32,
1901	2251	.unpacked_d16_vmem,
1902	2252	.volcanic_islands,
1903		.xnack_support,
	2253	.xnack_on_off_modes,
1904	2254	}),
1905	2255	};
1906	2256	pub const fiji: CpuModel = .{
...	...	@@ -1951,7 +2301,7 @@ pub const cpu = struct {
1951	2301	.vcmpx_exec_war_hazard,
1952	2302	.vcmpx_permlane_hazard,
1953	2303	.vmem_to_scalar_write_hazard,
1954		.xnack_support,
	2304	.xnack_on_off_modes,
1955	2305	}),
1956	2306	};
1957	2307	pub const gfx1011: CpuModel = .{
...	...	@@ -1987,7 +2337,7 @@ pub const cpu = struct {
1987	2337	.vcmpx_exec_war_hazard,
1988	2338	.vcmpx_permlane_hazard,
1989	2339	.vmem_to_scalar_write_hazard,
1990		.xnack_support,
	2340	.xnack_on_off_modes,
1991	2341	}),
1992	2342	};
1993	2343	pub const gfx1012: CpuModel = .{
...	...	@@ -2023,7 +2373,7 @@ pub const cpu = struct {
2023	2373	.vcmpx_exec_war_hazard,
2024	2374	.vcmpx_permlane_hazard,
2025	2375	.vmem_to_scalar_write_hazard,
2026		.xnack_support,
	2376	.xnack_on_off_modes,
2027	2377	}),
2028	2378	};
2029	2379	pub const gfx1013: CpuModel = .{
...	...	@@ -2054,7 +2404,7 @@ pub const cpu = struct {
2054	2404	.vcmpx_exec_war_hazard,
2055	2405	.vcmpx_permlane_hazard,
2056	2406	.vmem_to_scalar_write_hazard,
2057		.xnack_support,
	2407	.xnack_on_off_modes,
2058	2408	}),
2059	2409	};
2060	2410	pub const gfx1030: CpuModel = .{
...	...	@@ -2232,7 +2582,7 @@ pub const cpu = struct {
2232	2582	.vcmpx_exec_war_hazard,
2233	2583	.vcmpx_permlane_hazard,
2234	2584	.vmem_to_scalar_write_hazard,
2235		.xnack_support,
	2585	.xnack_on_off_modes,
2236	2586	}),
2237	2587	};
2238	2588	pub const gfx10_3_generic: CpuModel = .{
...	...	@@ -2261,7 +2611,7 @@ pub const cpu = struct {
2261	2611	.name = "gfx1100",
2262	2612	.llvm_name = "gfx1100",
2263	2613	.features = featureSet(&[_]Feature{
2264		.allocate1_5xvgprs,
	2614	.@"1536_physical_vgprs",
2265	2615	.architected_flat_scratch,
2266	2616	.atomic_fadd_no_rtn_insts,
2267	2617	.atomic_fadd_rtn_insts,
...	...	@@ -2274,6 +2624,7 @@ pub const cpu = struct {
2274	2624	.dot7_insts,
2275	2625	.dot8_insts,
2276	2626	.dot9_insts,
	2627	.dx10_clamp_and_ieee_mode,
2277	2628	.flat_atomic_fadd_f32_inst,
2278	2629	.gfx11,
2279	2630	.image_insts,
...	...	@@ -2290,13 +2641,14 @@ pub const cpu = struct {
2290	2641	.user_sgpr_init16_bug,
2291	2642	.valu_trans_use_hazard,
2292	2643	.vcmpx_permlane_hazard,
	2644	.wmma_256b_insts,
2293	2645	}),
2294	2646	};
2295	2647	pub const gfx1101: CpuModel = .{
2296	2648	.name = "gfx1101",
2297	2649	.llvm_name = "gfx1101",
2298	2650	.features = featureSet(&[_]Feature{
2299		.allocate1_5xvgprs,
	2651	.@"1536_physical_vgprs",
2300	2652	.architected_flat_scratch,
2301	2653	.atomic_fadd_no_rtn_insts,
2302	2654	.atomic_fadd_rtn_insts,
...	...	@@ -2309,6 +2661,7 @@ pub const cpu = struct {
2309	2661	.dot7_insts,
2310	2662	.dot8_insts,
2311	2663	.dot9_insts,
	2664	.dx10_clamp_and_ieee_mode,
2312	2665	.flat_atomic_fadd_f32_inst,
2313	2666	.gfx11,
2314	2667	.image_insts,
...	...	@@ -2324,6 +2677,7 @@ pub const cpu = struct {
2324	2677	.shader_cycles_register,
2325	2678	.valu_trans_use_hazard,
2326	2679	.vcmpx_permlane_hazard,
	2680	.wmma_256b_insts,
2327	2681	}),
2328	2682	};
2329	2683	pub const gfx1102: CpuModel = .{
...	...	@@ -2342,6 +2696,7 @@ pub const cpu = struct {
2342	2696	.dot7_insts,
2343	2697	.dot8_insts,
2344	2698	.dot9_insts,
	2699	.dx10_clamp_and_ieee_mode,
2345	2700	.flat_atomic_fadd_f32_inst,
2346	2701	.gfx11,
2347	2702	.image_insts,
...	...	@@ -2358,6 +2713,7 @@ pub const cpu = struct {
2358	2713	.user_sgpr_init16_bug,
2359	2714	.valu_trans_use_hazard,
2360	2715	.vcmpx_permlane_hazard,
	2716	.wmma_256b_insts,
2361	2717	}),
2362	2718	};
2363	2719	pub const gfx1103: CpuModel = .{
...	...	@@ -2376,6 +2732,7 @@ pub const cpu = struct {
2376	2732	.dot7_insts,
2377	2733	.dot8_insts,
2378	2734	.dot9_insts,
	2735	.dx10_clamp_and_ieee_mode,
2379	2736	.flat_atomic_fadd_f32_inst,
2380	2737	.gfx11,
2381	2738	.image_insts,
...	...	@@ -2391,6 +2748,7 @@ pub const cpu = struct {
2391	2748	.shader_cycles_register,
2392	2749	.valu_trans_use_hazard,
2393	2750	.vcmpx_permlane_hazard,
	2751	.wmma_256b_insts,
2394	2752	}),
2395	2753	};
2396	2754	pub const gfx1150: CpuModel = .{
...	...	@@ -2410,6 +2768,7 @@ pub const cpu = struct {
2410	2768	.dot8_insts,
2411	2769	.dot9_insts,
2412	2770	.dpp_src1_sgpr,
	2771	.dx10_clamp_and_ieee_mode,
2413	2772	.flat_atomic_fadd_f32_inst,
2414	2773	.gfx11,
2415	2774	.image_insts,
...	...	@@ -2424,13 +2783,14 @@ pub const cpu = struct {
2424	2783	.salu_float,
2425	2784	.shader_cycles_register,
2426	2785	.vcmpx_permlane_hazard,
	2786	.wmma_256b_insts,
2427	2787	}),
2428	2788	};
2429	2789	pub const gfx1151: CpuModel = .{
2430	2790	.name = "gfx1151",
2431	2791	.llvm_name = "gfx1151",
2432	2792	.features = featureSet(&[_]Feature{
2433		.allocate1_5xvgprs,
	2793	.@"1536_physical_vgprs",
2434	2794	.architected_flat_scratch,
2435	2795	.atomic_fadd_no_rtn_insts,
2436	2796	.atomic_fadd_rtn_insts,
...	...	@@ -2444,6 +2804,7 @@ pub const cpu = struct {
2444	2804	.dot8_insts,
2445	2805	.dot9_insts,
2446	2806	.dpp_src1_sgpr,
	2807	.dx10_clamp_and_ieee_mode,
2447	2808	.flat_atomic_fadd_f32_inst,
2448	2809	.gfx11,
2449	2810	.image_insts,
...	...	@@ -2458,6 +2819,7 @@ pub const cpu = struct {
2458	2819	.salu_float,
2459	2820	.shader_cycles_register,
2460	2821	.vcmpx_permlane_hazard,
	2822	.wmma_256b_insts,
2461	2823	}),
2462	2824	};
2463	2825	pub const gfx1152: CpuModel = .{
...	...	@@ -2477,6 +2839,7 @@ pub const cpu = struct {
2477	2839	.dot8_insts,
2478	2840	.dot9_insts,
2479	2841	.dpp_src1_sgpr,
	2842	.dx10_clamp_and_ieee_mode,
2480	2843	.flat_atomic_fadd_f32_inst,
2481	2844	.gfx11,
2482	2845	.image_insts,
...	...	@@ -2491,6 +2854,7 @@ pub const cpu = struct {
2491	2854	.salu_float,
2492	2855	.shader_cycles_register,
2493	2856	.vcmpx_permlane_hazard,
	2857	.wmma_256b_insts,
2494	2858	}),
2495	2859	};
2496	2860	pub const gfx1153: CpuModel = .{
...	...	@@ -2510,6 +2874,7 @@ pub const cpu = struct {
2510	2874	.dot8_insts,
2511	2875	.dot9_insts,
2512	2876	.dpp_src1_sgpr,
	2877	.dx10_clamp_and_ieee_mode,
2513	2878	.flat_atomic_fadd_f32_inst,
2514	2879	.gfx11,
2515	2880	.image_insts,
...	...	@@ -2523,11 +2888,12 @@ pub const cpu = struct {
2523	2888	.salu_float,
2524	2889	.shader_cycles_register,
2525	2890	.vcmpx_permlane_hazard,
	2891	.wmma_256b_insts,
2526	2892	}),
2527	2893	};
2528		pub const gfx11_generic: CpuModel = .{
2529		.name = "gfx11_generic",
2530		.llvm_name = "gfx11-generic",
	2894	pub const gfx1154: CpuModel = .{
	2895	.name = "gfx1154",
	2896	.llvm_name = "gfx1154",
2531	2897	.features = featureSet(&[_]Feature{
2532	2898	.architected_flat_scratch,
2533	2899	.atomic_fadd_no_rtn_insts,
...	...	@@ -2541,47 +2907,32 @@ pub const cpu = struct {
2541	2907	.dot7_insts,
2542	2908	.dot8_insts,
2543	2909	.dot9_insts,
	2910	.dpp_src1_sgpr,
	2911	.dx10_clamp_and_ieee_mode,
2544	2912	.flat_atomic_fadd_f32_inst,
2545	2913	.gfx11,
2546	2914	.image_insts,
2547	2915	.ldsbankcount32,
2548		.mad_intra_fwd_bug,
2549	2916	.memory_atomic_fadd_f32_denormal_support,
2550		.msaa_load_dst_sel_bug,
2551	2917	.nsa_encoding,
2552	2918	.packed_tid,
2553	2919	.partial_nsa_encoding,
2554		.priv_enabled_trap2_nop_bug,
2555	2920	.real_true16,
2556	2921	.required_export_priority,
2557		.requires_cov6,
	2922	.salu_float,
2558	2923	.shader_cycles_register,
2559		.user_sgpr_init16_bug,
2560		.valu_trans_use_hazard,
2561	2924	.vcmpx_permlane_hazard,
	2925	.wmma_256b_insts,
2562	2926	}),
2563	2927	};
2564		pub const gfx1200: CpuModel = .{
2565		.name = "gfx1200",
2566		.llvm_name = "gfx1200",
	2928	pub const gfx1170: CpuModel = .{
	2929	.name = "gfx1170",
	2930	.llvm_name = "gfx1170",
2567	2931	.features = featureSet(&[_]Feature{
2568		.addressablelocalmemorysize65536,
2569		.allocate1_5xvgprs,
2570	2932	.architected_flat_scratch,
2571		.architected_sgprs,
2572		.atomic_buffer_global_pk_add_f16_insts,
2573		.atomic_buffer_pk_add_bf16_inst,
2574		.atomic_ds_pk_add_16_insts,
2575	2933	.atomic_fadd_no_rtn_insts,
2576	2934	.atomic_fadd_rtn_insts,
2577		.atomic_flat_pk_add_16_insts,
2578		.atomic_global_pk_add_bf16_inst,
2579	2935	.back_off_barrier,
2580		.bvh_dual_bvh_8_insts,
2581		.cube_insts,
2582		.cvt_norm_insts,
2583		.cvt_pknorm_vop2_insts,
2584		.cvt_pknorm_vop3_insts,
2585	2936	.d16_write_vgpr32,
2586	2937	.dl_insts,
2587	2938	.dot10_insts,
...	...	@@ -2591,49 +2942,36 @@ pub const cpu = struct {
2591	2942	.dot8_insts,
2592	2943	.dot9_insts,
2593	2944	.dpp_src1_sgpr,
2594		.extended_image_insts,
2595	2945	.flat_atomic_fadd_f32_inst,
2596	2946	.fp8_conversion_insts,
2597		.gfx12,
	2947	.gfx11,
	2948	.gfx11_7_insts,
	2949	.ieee_minimum_maximum_insts,
2598	2950	.image_insts,
2599	2951	.ldsbankcount32,
2600		.lerp_inst,
2601	2952	.memory_atomic_fadd_f32_denormal_support,
	2953	.minimum3_maximum3_f16,
	2954	.minimum3_maximum3_f32,
	2955	.no_f16_pseudo_scalar_trans_inline_constants,
2602	2956	.nsa_encoding,
2603	2957	.packed_tid,
2604	2958	.partial_nsa_encoding,
2605		.pseudo_scalar_trans,
2606		.qsad_insts,
2607		.restricted_soffset,
2608		.sad_insts,
	2959	.real_true16,
2609	2960	.salu_float,
2610		.scalar_dwordx3_loads,
2611		.shader_cycles_hi_lo_registers,
	2961	.shader_cycles_register,
	2962	.swmmac_gfx1200_insts,
2612	2963	.vcmpx_permlane_hazard,
2613		.waits_before_system_scope_stores,
	2964	.wmma_128b_insts,
2614	2965	}),
2615	2966	};
2616		pub const gfx1201: CpuModel = .{
2617		.name = "gfx1201",
2618		.llvm_name = "gfx1201",
	2967	pub const gfx1171: CpuModel = .{
	2968	.name = "gfx1171",
	2969	.llvm_name = "gfx1171",
2619	2970	.features = featureSet(&[_]Feature{
2620		.addressablelocalmemorysize65536,
2621		.allocate1_5xvgprs,
2622	2971	.architected_flat_scratch,
2623		.architected_sgprs,
2624		.atomic_buffer_global_pk_add_f16_insts,
2625		.atomic_buffer_pk_add_bf16_inst,
2626		.atomic_ds_pk_add_16_insts,
2627	2972	.atomic_fadd_no_rtn_insts,
2628	2973	.atomic_fadd_rtn_insts,
2629		.atomic_flat_pk_add_16_insts,
2630		.atomic_global_pk_add_bf16_inst,
2631	2974	.back_off_barrier,
2632		.bvh_dual_bvh_8_insts,
2633		.cube_insts,
2634		.cvt_norm_insts,
2635		.cvt_pknorm_vop2_insts,
2636		.cvt_pknorm_vop3_insts,
2637	2975	.d16_write_vgpr32,
2638	2976	.dl_insts,
2639	2977	.dot10_insts,
...	...	@@ -2643,61 +2981,304 @@ pub const cpu = struct {
2643	2981	.dot8_insts,
2644	2982	.dot9_insts,
2645	2983	.dpp_src1_sgpr,
2646		.extended_image_insts,
2647	2984	.flat_atomic_fadd_f32_inst,
2648	2985	.fp8_conversion_insts,
2649		.gfx12,
	2986	.gfx11,
	2987	.gfx11_7_insts,
	2988	.ieee_minimum_maximum_insts,
2650	2989	.image_insts,
2651	2990	.ldsbankcount32,
2652		.lerp_inst,
2653	2991	.memory_atomic_fadd_f32_denormal_support,
	2992	.minimum3_maximum3_f16,
	2993	.minimum3_maximum3_f32,
	2994	.no_f16_pseudo_scalar_trans_inline_constants,
2654	2995	.nsa_encoding,
2655	2996	.packed_tid,
2656	2997	.partial_nsa_encoding,
2657		.pseudo_scalar_trans,
2658		.qsad_insts,
2659		.restricted_soffset,
2660		.sad_insts,
	2998	.real_true16,
2661	2999	.salu_float,
2662		.scalar_dwordx3_loads,
2663		.shader_cycles_hi_lo_registers,
	3000	.shader_cycles_register,
	3001	.swmmac_gfx1200_insts,
2664	3002	.vcmpx_permlane_hazard,
2665		.waits_before_system_scope_stores,
	3003	.wmma_128b_insts,
2666	3004	}),
2667	3005	};
2668		pub const gfx1250: CpuModel = .{
2669		.name = "gfx1250",
2670		.llvm_name = "gfx1250",
	3006	pub const gfx1172: CpuModel = .{
	3007	.name = "gfx1172",
	3008	.llvm_name = "gfx1172",
2671	3009	.features = featureSet(&[_]Feature{
2672		.@"1024_addressable_vgprs",
2673		.@"45_bit_num_records_buffer_resource",
2674		.@"64_bit_literals",
2675		.add_min_max_insts,
2676		.add_sub_u64_insts,
2677		.addressablelocalmemorysize327680,
2678	3010	.architected_flat_scratch,
2679		.architected_sgprs,
2680		.ashr_pk_insts,
2681		.atomic_buffer_global_pk_add_f16_insts,
2682		.atomic_buffer_pk_add_bf16_inst,
2683		.atomic_ds_pk_add_16_insts,
2684	3011	.atomic_fadd_no_rtn_insts,
2685	3012	.atomic_fadd_rtn_insts,
2686		.atomic_flat_pk_add_16_insts,
2687		.atomic_fmin_fmax_flat_f64,
2688		.atomic_fmin_fmax_global_f64,
2689		.atomic_global_pk_add_bf16_inst,
2690		.bf16_cvt_insts,
2691		.bf16_pk_insts,
2692		.bf16_trans_insts,
2693		.bitop3_insts,
2694		.clusters,
2695		.cube_insts,
2696		.cumode,
2697		.cvt_norm_insts,
2698		.cvt_pk_f16_f32_inst,
2699		.cvt_pknorm_vop2_insts,
2700		.cvt_pknorm_vop3_insts,
	3013	.back_off_barrier,
	3014	.d16_write_vgpr32,
	3015	.dl_insts,
	3016	.dot10_insts,
	3017	.dot11_insts,
	3018	.dot12_insts,
	3019	.dot7_insts,
	3020	.dot8_insts,
	3021	.dot9_insts,
	3022	.dpp_src1_sgpr,
	3023	.flat_atomic_fadd_f32_inst,
	3024	.fp8_conversion_insts,
	3025	.gfx11,
	3026	.gfx11_7_insts,
	3027	.ieee_minimum_maximum_insts,
	3028	.image_insts,
	3029	.ldsbankcount32,
	3030	.memory_atomic_fadd_f32_denormal_support,
	3031	.minimum3_maximum3_f16,
	3032	.minimum3_maximum3_f32,
	3033	.no_f16_pseudo_scalar_trans_inline_constants,
	3034	.nsa_encoding,
	3035	.packed_tid,
	3036	.partial_nsa_encoding,
	3037	.real_true16,
	3038	.salu_float,
	3039	.shader_cycles_register,
	3040	.swmmac_gfx1200_insts,
	3041	.vcmpx_permlane_hazard,
	3042	.wmma_128b_insts,
	3043	}),
	3044	};
	3045	pub const gfx11_7_generic: CpuModel = .{
	3046	.name = "gfx11_7_generic",
	3047	.llvm_name = "gfx11-7-generic",
	3048	.features = featureSet(&[_]Feature{
	3049	.architected_flat_scratch,
	3050	.atomic_fadd_no_rtn_insts,
	3051	.atomic_fadd_rtn_insts,
	3052	.back_off_barrier,
	3053	.d16_write_vgpr32,
	3054	.dl_insts,
	3055	.dot10_insts,
	3056	.dot11_insts,
	3057	.dot12_insts,
	3058	.dot7_insts,
	3059	.dot8_insts,
	3060	.dot9_insts,
	3061	.dpp_src1_sgpr,
	3062	.flat_atomic_fadd_f32_inst,
	3063	.fp8_conversion_insts,
	3064	.gfx11,
	3065	.gfx11_7_insts,
	3066	.ieee_minimum_maximum_insts,
	3067	.image_insts,
	3068	.ldsbankcount32,
	3069	.memory_atomic_fadd_f32_denormal_support,
	3070	.minimum3_maximum3_f16,
	3071	.minimum3_maximum3_f32,
	3072	.no_f16_pseudo_scalar_trans_inline_constants,
	3073	.nsa_encoding,
	3074	.packed_tid,
	3075	.partial_nsa_encoding,
	3076	.real_true16,
	3077	.requires_cov6,
	3078	.salu_float,
	3079	.shader_cycles_register,
	3080	.swmmac_gfx1200_insts,
	3081	.vcmpx_permlane_hazard,
	3082	.wmma_128b_insts,
	3083	}),
	3084	};
	3085	pub const gfx11_generic: CpuModel = .{
	3086	.name = "gfx11_generic",
	3087	.llvm_name = "gfx11-generic",
	3088	.features = featureSet(&[_]Feature{
	3089	.architected_flat_scratch,
	3090	.atomic_fadd_no_rtn_insts,
	3091	.atomic_fadd_rtn_insts,
	3092	.back_off_barrier,
	3093	.d16_write_vgpr32,
	3094	.dl_insts,
	3095	.dot10_insts,
	3096	.dot12_insts,
	3097	.dot5_insts,
	3098	.dot7_insts,
	3099	.dot8_insts,
	3100	.dot9_insts,
	3101	.dx10_clamp_and_ieee_mode,
	3102	.flat_atomic_fadd_f32_inst,
	3103	.gfx11,
	3104	.image_insts,
	3105	.ldsbankcount32,
	3106	.mad_intra_fwd_bug,
	3107	.memory_atomic_fadd_f32_denormal_support,
	3108	.msaa_load_dst_sel_bug,
	3109	.nsa_encoding,
	3110	.packed_tid,
	3111	.partial_nsa_encoding,
	3112	.priv_enabled_trap2_nop_bug,
	3113	.real_true16,
	3114	.required_export_priority,
	3115	.requires_cov6,
	3116	.shader_cycles_register,
	3117	.user_sgpr_init16_bug,
	3118	.valu_trans_use_hazard,
	3119	.vcmpx_permlane_hazard,
	3120	.wmma_256b_insts,
	3121	}),
	3122	};
	3123	pub const gfx1200: CpuModel = .{
	3124	.name = "gfx1200",
	3125	.llvm_name = "gfx1200",
	3126	.features = featureSet(&[_]Feature{
	3127	.@"1536_physical_vgprs",
	3128	.addressablelocalmemorysize65536,
	3129	.architected_flat_scratch,
	3130	.architected_sgprs,
	3131	.atomic_buffer_global_pk_add_f16_insts,
	3132	.atomic_buffer_pk_add_bf16_inst,
	3133	.atomic_ds_pk_add_16_insts,
	3134	.atomic_fadd_no_rtn_insts,
	3135	.atomic_fadd_rtn_insts,
	3136	.atomic_flat_pk_add_16_insts,
	3137	.atomic_global_pk_add_bf16_inst,
	3138	.back_off_barrier,
	3139	.bvh_dual_bvh_8_insts,
	3140	.cube_insts,
	3141	.cvt_norm_insts,
	3142	.cvt_pknorm_vop2_insts,
	3143	.cvt_pknorm_vop3_insts,
	3144	.d16_write_vgpr32,
	3145	.dl_insts,
	3146	.dot10_insts,
	3147	.dot11_insts,
	3148	.dot12_insts,
	3149	.dot7_insts,
	3150	.dot8_insts,
	3151	.dot9_insts,
	3152	.dpp_src1_sgpr,
	3153	.extended_image_insts,
	3154	.flat_atomic_fadd_f32_inst,
	3155	.formatted_mubuf_insts,
	3156	.fp8_conversion_insts,
	3157	.gfx12,
	3158	.image_insts,
	3159	.ldsbankcount32,
	3160	.lerp_inst,
	3161	.memory_atomic_fadd_f32_denormal_support,
	3162	.mqsad_insts,
	3163	.mqsad_pk_insts,
	3164	.msad_insts,
	3165	.mtbuf_insts,
	3166	.no_f16_pseudo_scalar_trans_inline_constants,
	3167	.nsa_encoding,
	3168	.packed_tid,
	3169	.partial_nsa_encoding,
	3170	.pseudo_scalar_trans,
	3171	.qsad_insts,
	3172	.real_true16,
	3173	.restricted_soffset,
	3174	.sad_insts,
	3175	.salu_float,
	3176	.scalar_dwordx3_loads,
	3177	.shader_cycles_hi_lo_registers,
	3178	.smem_prefetch_insts,
	3179	.swmmac_gfx1200_insts,
	3180	.vcmpx_permlane_hazard,
	3181	.waits_before_system_scope_stores,
	3182	.wmma_128b_insts,
	3183	}),
	3184	};
	3185	pub const gfx1201: CpuModel = .{
	3186	.name = "gfx1201",
	3187	.llvm_name = "gfx1201",
	3188	.features = featureSet(&[_]Feature{
	3189	.@"1536_physical_vgprs",
	3190	.addressablelocalmemorysize65536,
	3191	.architected_flat_scratch,
	3192	.architected_sgprs,
	3193	.atomic_buffer_global_pk_add_f16_insts,
	3194	.atomic_buffer_pk_add_bf16_inst,
	3195	.atomic_ds_pk_add_16_insts,
	3196	.atomic_fadd_no_rtn_insts,
	3197	.atomic_fadd_rtn_insts,
	3198	.atomic_flat_pk_add_16_insts,
	3199	.atomic_global_pk_add_bf16_inst,
	3200	.back_off_barrier,
	3201	.bvh_dual_bvh_8_insts,
	3202	.cube_insts,
	3203	.cvt_norm_insts,
	3204	.cvt_pknorm_vop2_insts,
	3205	.cvt_pknorm_vop3_insts,
	3206	.d16_write_vgpr32,
	3207	.dl_insts,
	3208	.dot10_insts,
	3209	.dot11_insts,
	3210	.dot12_insts,
	3211	.dot7_insts,
	3212	.dot8_insts,
	3213	.dot9_insts,
	3214	.dpp_src1_sgpr,
	3215	.extended_image_insts,
	3216	.flat_atomic_fadd_f32_inst,
	3217	.formatted_mubuf_insts,
	3218	.fp8_conversion_insts,
	3219	.gfx12,
	3220	.image_insts,
	3221	.ldsbankcount32,
	3222	.lerp_inst,
	3223	.memory_atomic_fadd_f32_denormal_support,
	3224	.mqsad_insts,
	3225	.mqsad_pk_insts,
	3226	.msad_insts,
	3227	.mtbuf_insts,
	3228	.no_f16_pseudo_scalar_trans_inline_constants,
	3229	.nsa_encoding,
	3230	.packed_tid,
	3231	.partial_nsa_encoding,
	3232	.pseudo_scalar_trans,
	3233	.qsad_insts,
	3234	.real_true16,
	3235	.restricted_soffset,
	3236	.sad_insts,
	3237	.salu_float,
	3238	.scalar_dwordx3_loads,
	3239	.shader_cycles_hi_lo_registers,
	3240	.smem_prefetch_insts,
	3241	.swmmac_gfx1200_insts,
	3242	.vcmpx_permlane_hazard,
	3243	.waits_before_system_scope_stores,
	3244	.wmma_128b_insts,
	3245	}),
	3246	};
	3247	pub const gfx1250: CpuModel = .{
	3248	.name = "gfx1250",
	3249	.llvm_name = "gfx1250",
	3250	.features = featureSet(&[_]Feature{
	3251	.@"1024_addressable_vgprs",
	3252	.@"45_bit_num_records_buffer_resource",
	3253	.@"64_bit_literals",
	3254	.add_min_max_insts,
	3255	.add_sub_u64_insts,
	3256	.addressablelocalmemorysize327680,
	3257	.architected_flat_scratch,
	3258	.architected_sgprs,
	3259	.ashr_pk_insts,
	3260	.asynccnt,
	3261	.atomic_buffer_global_pk_add_f16_insts,
	3262	.atomic_buffer_pk_add_bf16_inst,
	3263	.atomic_ds_pk_add_16_insts,
	3264	.atomic_fadd_no_rtn_insts,
	3265	.atomic_fadd_rtn_insts,
	3266	.atomic_flat_pk_add_16_insts,
	3267	.atomic_fmin_fmax_flat_f64,
	3268	.atomic_fmin_fmax_global_f64,
	3269	.atomic_global_pk_add_bf16_inst,
	3270	.back_off_barrier,
	3271	.bf16_cvt_insts,
	3272	.bf16_pk_insts,
	3273	.bf16_trans_insts,
	3274	.bitop3_insts,
	3275	.clusters,
	3276	.cube_insts,
	3277	.cumode,
	3278	.cvt_norm_insts,
	3279	.cvt_pk_f16_f32_inst,
	3280	.cvt_pknorm_vop2_insts,
	3281	.cvt_pknorm_vop3_insts,
2701	3282	.d16_write_vgpr32,
2702	3283	.dl_insts,
2703	3284	.dot7_insts,
...	...	@@ -2719,12 +3300,18 @@ pub const cpu = struct {
2719	3300	.ldsbankcount32,
2720	3301	.lerp_inst,
2721	3302	.lshl_add_u64_inst,
	3303	.mad_nc_64_32_insts,
2722	3304	.mad_u32_inst,
2723	3305	.max_hard_clause_length_63,
2724	3306	.mcast_load_insts,
2725	3307	.memory_atomic_fadd_f32_denormal_support,
2726	3308	.min3_max3_pkf16,
	3309	.min_max_i64_insts,
2727	3310	.minimum3_maximum3_pkf16,
	3311	.mqsad_insts,
	3312	.mqsad_pk_insts,
	3313	.msad_insts,
	3314	.no_f16_pseudo_scalar_trans_inline_constants,
2728	3315	.packed_fp32_ops,
2729	3316	.packed_tid,
2730	3317	.permlane16_swap,
...	...	@@ -2732,6 +3319,8 @@ pub const cpu = struct {
2732	3319	.prng_inst,
2733	3320	.pseudo_scalar_trans,
2734	3321	.qsad_insts,
	3322	.real_true16,
	3323	.requires_initial_unclause_vmem,
2735	3324	.restricted_soffset,
2736	3325	.s_wakeup_barrier_inst,
2737	3326	.sad_insts,
...	...	@@ -2740,17 +3329,24 @@ pub const cpu = struct {
2740	3329	.setprio_inc_wg_inst,
2741	3330	.setreg_vgpr_msb_fixup,
2742	3331	.shader_cycles_hi_lo_registers,
	3332	.smem_prefetch_insts,
2743	3333	.sramecc_support,
	3334	.swmmac_gfx1200_insts,
	3335	.swmmac_gfx1250_insts,
2744	3336	.tanh_insts,
2745	3337	.tensor_cvt_lut_insts,
	3338	.trans_coexecution_hazard,
2746	3339	.transpose_load_f4f6_insts,
	3340	.v_mov_b64_inst,
	3341	.v_mul_u64_inst,
2747	3342	.vcmpx_permlane_hazard,
2748	3343	.vgpr_align2,
2749	3344	.vmem_pref_insts,
	3345	.vop3px2_increments_va_vdst_twice,
2750	3346	.wait_xcnt,
2751	3347	.wavefrontsize32,
	3348	.wmma_coexecution_hazards,
2752	3349	.xnack,
2753		.xnack_support,
2754	3350	}),
2755	3351	};
2756	3352	pub const gfx1251: CpuModel = .{
...	...	@@ -2766,6 +3362,7 @@ pub const cpu = struct {
2766	3362	.architected_flat_scratch,
2767	3363	.architected_sgprs,
2768	3364	.ashr_pk_insts,
	3365	.asynccnt,
2769	3366	.atomic_buffer_global_pk_add_f16_insts,
2770	3367	.atomic_buffer_pk_add_bf16_inst,
2771	3368	.atomic_ds_pk_add_16_insts,
...	...	@@ -2775,6 +3372,7 @@ pub const cpu = struct {
2775	3372	.atomic_fmin_fmax_flat_f64,
2776	3373	.atomic_fmin_fmax_global_f64,
2777	3374	.atomic_global_pk_add_bf16_inst,
	3375	.back_off_barrier,
2778	3376	.bf16_cvt_insts,
2779	3377	.bf16_pk_insts,
2780	3378	.bf16_trans_insts,
...	...	@@ -2800,27 +3398,40 @@ pub const cpu = struct {
2800	3398	.fmacf64_inst,
2801	3399	.fp8_conversion_insts,
2802	3400	.fp8e5m3_insts,
	3401	.full_rate_64_ops,
2803	3402	.gfx12,
2804	3403	.gfx1250_insts,
	3404	.gfx1251_gemm_insts,
	3405	.gfx125x_lowest_rate_wmma,
2805	3406	.globally_addressable_scratch,
2806	3407	.kernarg_preload,
2807	3408	.lds_barrier_arrive_atomic,
2808	3409	.ldsbankcount32,
2809	3410	.lerp_inst,
2810	3411	.lshl_add_u64_inst,
	3412	.mad_nc_64_32_insts,
2811	3413	.mad_u32_inst,
2812	3414	.max_hard_clause_length_63,
2813	3415	.mcast_load_insts,
2814	3416	.memory_atomic_fadd_f32_denormal_support,
2815	3417	.min3_max3_pkf16,
	3418	.min_max_i64_insts,
2816	3419	.minimum3_maximum3_pkf16,
	3420	.mqsad_insts,
	3421	.mqsad_pk_insts,
	3422	.msad_insts,
	3423	.no_f16_pseudo_scalar_trans_inline_constants,
2817	3424	.packed_fp32_ops,
	3425	.packed_fp64_ops,
2818	3426	.packed_tid,
	3427	.packed_u64_ops,
2819	3428	.permlane16_swap,
2820	3429	.pk_add_min_max_insts,
2821	3430	.prng_inst,
2822	3431	.pseudo_scalar_trans,
2823	3432	.qsad_insts,
	3433	.real_true16,
	3434	.requires_initial_unclause_vmem,
2824	3435	.restricted_soffset,
2825	3436	.s_wakeup_barrier_inst,
2826	3437	.sad_insts,
...	...	@@ -2828,25 +3439,127 @@ pub const cpu = struct {
2828	3439	.scalar_dwordx3_loads,
2829	3440	.setprio_inc_wg_inst,
2830	3441	.shader_cycles_hi_lo_registers,
	3442	.smem_prefetch_insts,
2831	3443	.sramecc_support,
	3444	.swmmac_gfx1200_insts,
	3445	.swmmac_gfx1250_insts,
2832	3446	.tanh_insts,
2833	3447	.tensor_cvt_lut_insts,
	3448	.trans_coexecution_hazard,
2834	3449	.transpose_load_f4f6_insts,
	3450	.use_add_pc64_inst,
	3451	.v_mov_b64_inst,
	3452	.v_mul_u64_inst,
2835	3453	.vcmpx_permlane_hazard,
2836	3454	.vgpr_align2,
2837	3455	.vmem_pref_insts,
	3456	.vop3px2_increments_va_vdst_twice,
2838	3457	.wait_xcnt,
2839	3458	.wavefrontsize32,
	3459	.wmma_coexecution_hazards,
	3460	.xnack,
	3461	}),
	3462	};
	3463	pub const gfx12_5_generic: CpuModel = .{
	3464	.name = "gfx12_5_generic",
	3465	.llvm_name = "gfx12-5-generic",
	3466	.features = featureSet(&[_]Feature{
	3467	.@"1024_addressable_vgprs",
	3468	.@"45_bit_num_records_buffer_resource",
	3469	.@"64_bit_literals",
	3470	.add_min_max_insts,
	3471	.add_sub_u64_insts,
	3472	.addressablelocalmemorysize327680,
	3473	.architected_flat_scratch,
	3474	.architected_sgprs,
	3475	.ashr_pk_insts,
	3476	.asynccnt,
	3477	.atomic_buffer_global_pk_add_f16_insts,
	3478	.atomic_buffer_pk_add_bf16_inst,
	3479	.atomic_ds_pk_add_16_insts,
	3480	.atomic_fadd_no_rtn_insts,
	3481	.atomic_fadd_rtn_insts,
	3482	.atomic_flat_pk_add_16_insts,
	3483	.atomic_fmin_fmax_flat_f64,
	3484	.atomic_fmin_fmax_global_f64,
	3485	.atomic_global_pk_add_bf16_inst,
	3486	.back_off_barrier,
	3487	.bf16_cvt_insts,
	3488	.bf16_pk_insts,
	3489	.bf16_trans_insts,
	3490	.bitop3_insts,
	3491	.clusters,
	3492	.cumode,
	3493	.cvt_pk_f16_f32_inst,
	3494	.d16_write_vgpr32,
	3495	.dl_insts,
	3496	.dot7_insts,
	3497	.dot8_insts,
	3498	.dpp_src1_sgpr,
	3499	.emulated_system_scope_atomics,
	3500	.flat_atomic_fadd_f32_inst,
	3501	.flat_buffer_global_fadd_f64_inst,
	3502	.flat_gvs_mode,
	3503	.fma_mix_bf16_insts,
	3504	.fmacf64_inst,
	3505	.fp8_conversion_insts,
	3506	.fp8e5m3_insts,
	3507	.gfx12,
	3508	.gfx1250_insts,
	3509	.gfx125x_lowest_rate_wmma,
	3510	.globally_addressable_scratch,
	3511	.kernarg_preload,
	3512	.lds_barrier_arrive_atomic,
	3513	.ldsbankcount32,
	3514	.lshl_add_u64_inst,
	3515	.mad_nc_64_32_insts,
	3516	.mad_u32_inst,
	3517	.max_hard_clause_length_63,
	3518	.mcast_load_insts,
	3519	.memory_atomic_fadd_f32_denormal_support,
	3520	.min3_max3_pkf16,
	3521	.min_max_i64_insts,
	3522	.minimum3_maximum3_pkf16,
	3523	.no_f16_pseudo_scalar_trans_inline_constants,
	3524	.packed_fp32_ops,
	3525	.packed_tid,
	3526	.permlane16_swap,
	3527	.pk_add_min_max_insts,
	3528	.prng_inst,
	3529	.pseudo_scalar_trans,
	3530	.real_true16,
	3531	.requires_cov6,
	3532	.requires_initial_unclause_vmem,
	3533	.restricted_soffset,
	3534	.s_wakeup_barrier_inst,
	3535	.salu_float,
	3536	.scalar_dwordx3_loads,
	3537	.setprio_inc_wg_inst,
	3538	.setreg_vgpr_msb_fixup,
	3539	.shader_cycles_hi_lo_registers,
	3540	.smem_prefetch_insts,
	3541	.sramecc_support,
	3542	.tanh_insts,
	3543	.tensor_cvt_lut_insts,
	3544	.trans_coexecution_hazard,
	3545	.transpose_load_f4f6_insts,
	3546	.v_mov_b64_inst,
	3547	.v_mul_u64_inst,
	3548	.vcmpx_permlane_hazard,
	3549	.vgpr_align2,
	3550	.vmem_pref_insts,
	3551	.wait_xcnt,
	3552	.wavefrontsize32,
	3553	.wmma_coexecution_hazards,
2840	3554	.xnack,
2841		.xnack_support,
2842	3555	}),
2843	3556	};
2844	3557	pub const gfx12_generic: CpuModel = .{
2845	3558	.name = "gfx12_generic",
2846	3559	.llvm_name = "gfx12-generic",
2847	3560	.features = featureSet(&[_]Feature{
	3561	.@"1536_physical_vgprs",
2848	3562	.addressablelocalmemorysize65536,
2849		.allocate1_5xvgprs,
2850	3563	.architected_flat_scratch,
2851	3564	.architected_sgprs,
2852	3565	.atomic_buffer_global_pk_add_f16_insts,
...	...	@@ -2873,25 +3586,170 @@ pub const cpu = struct {
2873	3586	.dpp_src1_sgpr,
2874	3587	.extended_image_insts,
2875	3588	.flat_atomic_fadd_f32_inst,
	3589	.formatted_mubuf_insts,
2876	3590	.fp8_conversion_insts,
2877	3591	.gfx12,
2878	3592	.image_insts,
2879	3593	.ldsbankcount32,
2880	3594	.lerp_inst,
2881	3595	.memory_atomic_fadd_f32_denormal_support,
	3596	.mqsad_insts,
	3597	.mqsad_pk_insts,
	3598	.msad_insts,
	3599	.mtbuf_insts,
	3600	.no_f16_pseudo_scalar_trans_inline_constants,
2882	3601	.nsa_encoding,
2883	3602	.packed_tid,
2884	3603	.partial_nsa_encoding,
2885	3604	.pseudo_scalar_trans,
2886	3605	.qsad_insts,
	3606	.real_true16,
2887	3607	.requires_cov6,
2888	3608	.restricted_soffset,
2889	3609	.sad_insts,
2890	3610	.salu_float,
2891	3611	.scalar_dwordx3_loads,
2892	3612	.shader_cycles_hi_lo_registers,
	3613	.smem_prefetch_insts,
	3614	.swmmac_gfx1200_insts,
2893	3615	.vcmpx_permlane_hazard,
2894	3616	.waits_before_system_scope_stores,
	3617	.wmma_128b_insts,
	3618	}),
	3619	};
	3620	pub const gfx1310: CpuModel = .{
	3621	.name = "gfx1310",
	3622	.llvm_name = "gfx1310",
	3623	.features = featureSet(&[_]Feature{
	3624	.@"1536_physical_vgprs",
	3625	.@"64_bit_literals",
	3626	.addressablelocalmemorysize65536,
	3627	.architected_flat_scratch,
	3628	.architected_sgprs,
	3629	.ashr_pk_insts,
	3630	.atomic_buffer_global_pk_add_f16_insts,
	3631	.atomic_buffer_pk_add_bf16_inst,
	3632	.atomic_ds_pk_add_16_insts,
	3633	.atomic_fadd_no_rtn_insts,
	3634	.atomic_fadd_rtn_insts,
	3635	.atomic_flat_pk_add_16_insts,
	3636	.atomic_fmin_fmax_flat_f64,
	3637	.atomic_fmin_fmax_global_f64,
	3638	.atomic_global_pk_add_bf16_inst,
	3639	.bf16_cvt_insts,
	3640	.bf16_pk_insts,
	3641	.bf16_trans_insts,
	3642	.bitop3_insts,
	3643	.clusters,
	3644	.cube_insts,
	3645	.cvt_norm_insts,
	3646	.cvt_pk_f16_f32_inst,
	3647	.cvt_pknorm_vop2_insts,
	3648	.cvt_pknorm_vop3_insts,
	3649	.dl_insts,
	3650	.dot7_insts,
	3651	.dot8_insts,
	3652	.dpp_src1_sgpr,
	3653	.f16bf16_to_fp6bf6_cvt_scale_insts,
	3654	.flat_atomic_fadd_f32_inst,
	3655	.flat_gvs_mode,
	3656	.fma_mix_bf16_insts,
	3657	.fmacf64_inst,
	3658	.fp8_conversion_insts,
	3659	.gfx1250_insts,
	3660	.gfx13,
	3661	.globally_addressable_scratch,
	3662	.ldsbankcount32,
	3663	.lerp_inst,
	3664	.mqsad_insts,
	3665	.mqsad_pk_insts,
	3666	.msad_insts,
	3667	.nsa_encoding,
	3668	.packed_tid,
	3669	.partial_nsa_encoding,
	3670	.permlane16_swap,
	3671	.prng_inst,
	3672	.pseudo_scalar_trans,
	3673	.qsad_insts,
	3674	.restricted_soffset,
	3675	.s_wakeup_barrier_inst,
	3676	.sad_insts,
	3677	.salu_float,
	3678	.scalar_dwordx3_loads,
	3679	.shader_cycles_register,
	3680	.smem_prefetch_insts,
	3681	.tanh_insts,
	3682	.tensor_cvt_lut_insts,
	3683	.transpose_load_f4f6_insts,
	3684	.vcmpx_permlane_hazard,
	3685	}),
	3686	};
	3687	pub const gfx13_generic: CpuModel = .{
	3688	.name = "gfx13_generic",
	3689	.llvm_name = "gfx13-generic",
	3690	.features = featureSet(&[_]Feature{
	3691	.@"1536_physical_vgprs",
	3692	.@"64_bit_literals",
	3693	.addressablelocalmemorysize65536,
	3694	.architected_flat_scratch,
	3695	.architected_sgprs,
	3696	.ashr_pk_insts,
	3697	.atomic_buffer_global_pk_add_f16_insts,
	3698	.atomic_buffer_pk_add_bf16_inst,
	3699	.atomic_ds_pk_add_16_insts,
	3700	.atomic_fadd_no_rtn_insts,
	3701	.atomic_fadd_rtn_insts,
	3702	.atomic_flat_pk_add_16_insts,
	3703	.atomic_fmin_fmax_flat_f64,
	3704	.atomic_fmin_fmax_global_f64,
	3705	.atomic_global_pk_add_bf16_inst,
	3706	.bf16_cvt_insts,
	3707	.bf16_pk_insts,
	3708	.bf16_trans_insts,
	3709	.bitop3_insts,
	3710	.clusters,
	3711	.cube_insts,
	3712	.cvt_norm_insts,
	3713	.cvt_pk_f16_f32_inst,
	3714	.cvt_pknorm_vop2_insts,
	3715	.cvt_pknorm_vop3_insts,
	3716	.dl_insts,
	3717	.dot7_insts,
	3718	.dot8_insts,
	3719	.dpp_src1_sgpr,
	3720	.f16bf16_to_fp6bf6_cvt_scale_insts,
	3721	.flat_atomic_fadd_f32_inst,
	3722	.flat_gvs_mode,
	3723	.fma_mix_bf16_insts,
	3724	.fmacf64_inst,
	3725	.fp8_conversion_insts,
	3726	.gfx1250_insts,
	3727	.gfx13,
	3728	.globally_addressable_scratch,
	3729	.ldsbankcount32,
	3730	.lerp_inst,
	3731	.mqsad_insts,
	3732	.mqsad_pk_insts,
	3733	.msad_insts,
	3734	.nsa_encoding,
	3735	.packed_tid,
	3736	.partial_nsa_encoding,
	3737	.permlane16_swap,
	3738	.prng_inst,
	3739	.pseudo_scalar_trans,
	3740	.qsad_insts,
	3741	.requires_cov6,
	3742	.restricted_soffset,
	3743	.s_wakeup_barrier_inst,
	3744	.sad_insts,
	3745	.salu_float,
	3746	.scalar_dwordx3_loads,
	3747	.shader_cycles_register,
	3748	.smem_prefetch_insts,
	3749	.tanh_insts,
	3750	.tensor_cvt_lut_insts,
	3751	.transpose_load_f4f6_insts,
	3752	.vcmpx_permlane_hazard,
2895	3753	}),
2896	3754	};
2897	3755	pub const gfx600: CpuModel = .{
...	...	@@ -2977,7 +3835,7 @@ pub const cpu = struct {
2977	3835	.ldsbankcount32,
2978	3836	.unpacked_d16_vmem,
2979	3837	.volcanic_islands,
2980		.xnack_support,
	3838	.xnack_on_off_modes,
2981	3839	}),
2982	3840	};
2983	3841	pub const gfx802: CpuModel = .{
...	...	@@ -3017,7 +3875,7 @@ pub const cpu = struct {
3017	3875	.image_store_d16_bug,
3018	3876	.ldsbankcount16,
3019	3877	.volcanic_islands,
3020		.xnack_support,
	3878	.xnack_on_off_modes,
3021	3879	}),
3022	3880	};
3023	3881	pub const gfx900: CpuModel = .{
...	...	@@ -3175,6 +4033,7 @@ pub const cpu = struct {
3175	4033	.packed_tid,
3176	4034	.pk_fmac_f16_inst,
3177	4035	.sramecc_support,
	4036	.tgsplit_support,
3178	4037	.vgpr_align2,
3179	4038	}),
3180	4039	};
...	...	@@ -3239,6 +4098,8 @@ pub const cpu = struct {
3239	4098	.packed_tid,
3240	4099	.pk_fmac_f16_inst,
3241	4100	.sramecc_support,
	4101	.tgsplit_support,
	4102	.v_mov_b64_inst,
3242	4103	.vgpr_align2,
3243	4104	.xf32_insts,
3244	4105	}),
...	...	@@ -3295,6 +4156,8 @@ pub const cpu = struct {
3295	4156	.pk_fmac_f16_inst,
3296	4157	.prng_inst,
3297	4158	.sramecc_support,
	4159	.tgsplit_support,
	4160	.v_mov_b64_inst,
3298	4161	.vgpr_align2,
3299	4162	}),
3300	4163	};
...	...	@@ -3342,6 +4205,8 @@ pub const cpu = struct {
3342	4205	.pk_fmac_f16_inst,
3343	4206	.requires_cov6,
3344	4207	.sramecc_support,
	4208	.tgsplit_support,
	4209	.v_mov_b64_inst,
3345	4210	.vgpr_align2,
3346	4211	}),
3347	4212	};
...	...	@@ -3452,7 +4317,7 @@ pub const cpu = struct {
3452	4317	.image_store_d16_bug,
3453	4318	.ldsbankcount16,
3454	4319	.volcanic_islands,
3455		.xnack_support,
	4320	.xnack_on_off_modes,
3456	4321	}),
3457	4322	};
3458	4323	pub const tahiti: CpuModel = .{

lib/std/Target/hexagon.zig+204

...	...	@@ -34,7 +34,41 @@ pub const Feature = enum {
34	34	nvs,
35	35	packets,
36	36	prev65,
	37	reserved_r10,
	38	reserved_r11,
	39	reserved_r12,
	40	reserved_r13,
	41	reserved_r14,
	42	reserved_r15,
	43	reserved_r16,
	44	reserved_r17,
	45	reserved_r18,
37	46	reserved_r19,
	47	reserved_r20,
	48	reserved_r21,
	49	reserved_r22,
	50	reserved_r23,
	51	reserved_r24,
	52	reserved_r25,
	53	reserved_r26,
	54	reserved_r27,
	55	reserved_r28,
	56	reserved_r6,
	57	reserved_r7,
	58	reserved_r8,
	59	reserved_r9,
	60	scs_reg_r16,
	61	scs_reg_r17,
	62	scs_reg_r18,
	63	scs_reg_r19,
	64	scs_reg_r20,
	65	scs_reg_r21,
	66	scs_reg_r22,
	67	scs_reg_r23,
	68	scs_reg_r24,
	69	scs_reg_r25,
	70	scs_reg_r26,
	71	scs_reg_r27,
38	72	small_data,
39	73	tinycore,
40	74	v5,
...	...	@@ -241,11 +275,181 @@ pub const all_features = blk: {
241	275	.description = "Support features deprecated in v65",
242	276	.dependencies = featureSet(&[_]Feature{}),
243	277	};
	278	result[@backingInt(Feature.reserved_r10)] = .{
	279	.llvm_name = "reserved-r10",
	280	.description = "Reserve register R10",
	281	.dependencies = featureSet(&[_]Feature{}),
	282	};
	283	result[@backingInt(Feature.reserved_r11)] = .{
	284	.llvm_name = "reserved-r11",
	285	.description = "Reserve register R11",
	286	.dependencies = featureSet(&[_]Feature{}),
	287	};
	288	result[@backingInt(Feature.reserved_r12)] = .{
	289	.llvm_name = "reserved-r12",
	290	.description = "Reserve register R12",
	291	.dependencies = featureSet(&[_]Feature{}),
	292	};
	293	result[@backingInt(Feature.reserved_r13)] = .{
	294	.llvm_name = "reserved-r13",
	295	.description = "Reserve register R13",
	296	.dependencies = featureSet(&[_]Feature{}),
	297	};
	298	result[@backingInt(Feature.reserved_r14)] = .{
	299	.llvm_name = "reserved-r14",
	300	.description = "Reserve register R14",
	301	.dependencies = featureSet(&[_]Feature{}),
	302	};
	303	result[@backingInt(Feature.reserved_r15)] = .{
	304	.llvm_name = "reserved-r15",
	305	.description = "Reserve register R15",
	306	.dependencies = featureSet(&[_]Feature{}),
	307	};
	308	result[@backingInt(Feature.reserved_r16)] = .{
	309	.llvm_name = "reserved-r16",
	310	.description = "Reserve register R16",
	311	.dependencies = featureSet(&[_]Feature{}),
	312	};
	313	result[@backingInt(Feature.reserved_r17)] = .{
	314	.llvm_name = "reserved-r17",
	315	.description = "Reserve register R17",
	316	.dependencies = featureSet(&[_]Feature{}),
	317	};
	318	result[@backingInt(Feature.reserved_r18)] = .{
	319	.llvm_name = "reserved-r18",
	320	.description = "Reserve register R18",
	321	.dependencies = featureSet(&[_]Feature{}),
	322	};
244	323	result[@backingInt(Feature.reserved_r19)] = .{
245	324	.llvm_name = "reserved-r19",
246	325	.description = "Reserve register R19",
247	326	.dependencies = featureSet(&[_]Feature{}),
248	327	};
	328	result[@backingInt(Feature.reserved_r20)] = .{
	329	.llvm_name = "reserved-r20",
	330	.description = "Reserve register R20",
	331	.dependencies = featureSet(&[_]Feature{}),
	332	};
	333	result[@backingInt(Feature.reserved_r21)] = .{
	334	.llvm_name = "reserved-r21",
	335	.description = "Reserve register R21",
	336	.dependencies = featureSet(&[_]Feature{}),
	337	};
	338	result[@backingInt(Feature.reserved_r22)] = .{
	339	.llvm_name = "reserved-r22",
	340	.description = "Reserve register R22",
	341	.dependencies = featureSet(&[_]Feature{}),
	342	};
	343	result[@backingInt(Feature.reserved_r23)] = .{
	344	.llvm_name = "reserved-r23",
	345	.description = "Reserve register R23",
	346	.dependencies = featureSet(&[_]Feature{}),
	347	};
	348	result[@backingInt(Feature.reserved_r24)] = .{
	349	.llvm_name = "reserved-r24",
	350	.description = "Reserve register R24",
	351	.dependencies = featureSet(&[_]Feature{}),
	352	};
	353	result[@backingInt(Feature.reserved_r25)] = .{
	354	.llvm_name = "reserved-r25",
	355	.description = "Reserve register R25",
	356	.dependencies = featureSet(&[_]Feature{}),
	357	};
	358	result[@backingInt(Feature.reserved_r26)] = .{
	359	.llvm_name = "reserved-r26",
	360	.description = "Reserve register R26",
	361	.dependencies = featureSet(&[_]Feature{}),
	362	};
	363	result[@backingInt(Feature.reserved_r27)] = .{
	364	.llvm_name = "reserved-r27",
	365	.description = "Reserve register R27",
	366	.dependencies = featureSet(&[_]Feature{}),
	367	};
	368	result[@backingInt(Feature.reserved_r28)] = .{
	369	.llvm_name = "reserved-r28",
	370	.description = "Reserve register R28",
	371	.dependencies = featureSet(&[_]Feature{}),
	372	};
	373	result[@backingInt(Feature.reserved_r6)] = .{
	374	.llvm_name = "reserved-r6",
	375	.description = "Reserve register R6",
	376	.dependencies = featureSet(&[_]Feature{}),
	377	};
	378	result[@backingInt(Feature.reserved_r7)] = .{
	379	.llvm_name = "reserved-r7",
	380	.description = "Reserve register R7",
	381	.dependencies = featureSet(&[_]Feature{}),
	382	};
	383	result[@backingInt(Feature.reserved_r8)] = .{
	384	.llvm_name = "reserved-r8",
	385	.description = "Reserve register R8",
	386	.dependencies = featureSet(&[_]Feature{}),
	387	};
	388	result[@backingInt(Feature.reserved_r9)] = .{
	389	.llvm_name = "reserved-r9",
	390	.description = "Reserve register R9",
	391	.dependencies = featureSet(&[_]Feature{}),
	392	};
	393	result[@backingInt(Feature.scs_reg_r16)] = .{
	394	.llvm_name = "scs-reg-r16",
	395	.description = "Use register R16 as the shadow call stack pointer",
	396	.dependencies = featureSet(&[_]Feature{}),
	397	};
	398	result[@backingInt(Feature.scs_reg_r17)] = .{
	399	.llvm_name = "scs-reg-r17",
	400	.description = "Use register R17 as the shadow call stack pointer",
	401	.dependencies = featureSet(&[_]Feature{}),
	402	};
	403	result[@backingInt(Feature.scs_reg_r18)] = .{
	404	.llvm_name = "scs-reg-r18",
	405	.description = "Use register R18 as the shadow call stack pointer",
	406	.dependencies = featureSet(&[_]Feature{}),
	407	};
	408	result[@backingInt(Feature.scs_reg_r19)] = .{
	409	.llvm_name = "scs-reg-r19",
	410	.description = "Use register R19 as the shadow call stack pointer",
	411	.dependencies = featureSet(&[_]Feature{}),
	412	};
	413	result[@backingInt(Feature.scs_reg_r20)] = .{
	414	.llvm_name = "scs-reg-r20",
	415	.description = "Use register R20 as the shadow call stack pointer",
	416	.dependencies = featureSet(&[_]Feature{}),
	417	};
	418	result[@backingInt(Feature.scs_reg_r21)] = .{
	419	.llvm_name = "scs-reg-r21",
	420	.description = "Use register R21 as the shadow call stack pointer",
	421	.dependencies = featureSet(&[_]Feature{}),
	422	};
	423	result[@backingInt(Feature.scs_reg_r22)] = .{
	424	.llvm_name = "scs-reg-r22",
	425	.description = "Use register R22 as the shadow call stack pointer",
	426	.dependencies = featureSet(&[_]Feature{}),
	427	};
	428	result[@backingInt(Feature.scs_reg_r23)] = .{
	429	.llvm_name = "scs-reg-r23",
	430	.description = "Use register R23 as the shadow call stack pointer",
	431	.dependencies = featureSet(&[_]Feature{}),
	432	};
	433	result[@backingInt(Feature.scs_reg_r24)] = .{
	434	.llvm_name = "scs-reg-r24",
	435	.description = "Use register R24 as the shadow call stack pointer",
	436	.dependencies = featureSet(&[_]Feature{}),
	437	};
	438	result[@backingInt(Feature.scs_reg_r25)] = .{
	439	.llvm_name = "scs-reg-r25",
	440	.description = "Use register R25 as the shadow call stack pointer",
	441	.dependencies = featureSet(&[_]Feature{}),
	442	};
	443	result[@backingInt(Feature.scs_reg_r26)] = .{
	444	.llvm_name = "scs-reg-r26",
	445	.description = "Use register R26 as the shadow call stack pointer",
	446	.dependencies = featureSet(&[_]Feature{}),
	447	};
	448	result[@backingInt(Feature.scs_reg_r27)] = .{
	449	.llvm_name = "scs-reg-r27",
	450	.description = "Use register R27 as the shadow call stack pointer",
	451	.dependencies = featureSet(&[_]Feature{}),
	452	};
249	453	result[@backingInt(Feature.small_data)] = .{
250	454	.llvm_name = "small-data",
251	455	.description = "Allow GP-relative addressing of global variables",

lib/std/Target/mips.zig+23

...	...	@@ -13,6 +13,7 @@ pub const Feature = enum {
13	13	dspr2,
14	14	dspr3,
15	15	eva,
	16	fix_r5900,
16	17	fp64,
17	18	fpxx,
18	19	ginv,
...	...	@@ -52,6 +53,7 @@ pub const Feature = enum {
52	53	notraps,
53	54	p5600,
54	55	ptr64,
	56	r5900,
55	57	single_float,
56	58	soft_float,
57	59	strict_align,
...	...	@@ -121,6 +123,11 @@ pub const all_features = blk: {
121	123	.description = "Mips EVA ASE",
122	124	.dependencies = featureSet(&[_]Feature{}),
123	125	};
	126	result[@backingInt(Feature.fix_r5900)] = .{
	127	.llvm_name = "fix-r5900",
	128	.description = "Enable R5900 short loop erratum fix",
	129	.dependencies = featureSet(&[_]Feature{}),
	130	};
124	131	result[@backingInt(Feature.fp64)] = .{
125	132	.llvm_name = "fp64",
126	133	.description = "Support 64-bit FP registers",
...	...	@@ -372,6 +379,15 @@ pub const all_features = blk: {
372	379	.description = "Pointers are 64-bit wide",
373	380	.dependencies = featureSet(&[_]Feature{}),
374	381	};
	382	result[@backingInt(Feature.r5900)] = .{
	383	.llvm_name = "r5900",
	384	.description = "R5900 (PS2 Emotion Engine) Support",
	385	.dependencies = featureSet(&[_]Feature{
	386	.fix_r5900,
	387	.mips3,
	388	.single_float,
	389	}),
	390	};
375	391	result[@backingInt(Feature.single_float)] = .{
376	392	.llvm_name = "single-float",
377	393	.description = "Only supports single precision float",
...	...	@@ -596,4 +612,11 @@ pub const cpu = struct {
596	612	.soft_float,
597	613	}),
598	614	};
	615	pub const r5900: CpuModel = .{
	616	.name = "r5900",
	617	.llvm_name = "r5900",
	618	.features = featureSet(&[_]Feature{
	619	.r5900,
	620	}),
	621	};
599	622	};

lib/std/Target/nvptx.zig+18

...	...	@@ -36,6 +36,9 @@ pub const Feature = enum {
36	36	ptx87,
37	37	ptx88,
38	38	ptx90,
	39	ptx91,
	40	ptx92,
	41	ptx93,
39	42	sm_100,
40	43	sm_100a,
41	44	sm_100f,
...	...	@@ -242,6 +245,21 @@ pub const all_features = blk: {
242	245	.description = "Use PTX version 90",
243	246	.dependencies = featureSet(&[_]Feature{}),
244	247	};
	248	result[@backingInt(Feature.ptx91)] = .{
	249	.llvm_name = "ptx91",
	250	.description = "Use PTX version 91",
	251	.dependencies = featureSet(&[_]Feature{}),
	252	};
	253	result[@backingInt(Feature.ptx92)] = .{
	254	.llvm_name = "ptx92",
	255	.description = "Use PTX version 92",
	256	.dependencies = featureSet(&[_]Feature{}),
	257	};
	258	result[@backingInt(Feature.ptx93)] = .{
	259	.llvm_name = "ptx93",
	260	.description = "Use PTX version 93",
	261	.dependencies = featureSet(&[_]Feature{}),
	262	};
245	263	result[@backingInt(Feature.sm_100)] = .{
246	264	.llvm_name = "sm_100",
247	265	.description = "Target SM 100",

lib/std/Target/powerpc.zig+17-2

...	...	@@ -43,6 +43,7 @@ pub const Feature = enum {
43	43	fuse_wideimm,
44	44	fuse_zeromove,
45	45	fusion,
	46	future_vector,
46	47	hard_float,
47	48	htm,
48	49	icbt,
...	...	@@ -83,6 +84,7 @@ pub const Feature = enum {
83	84	spe,
84	85	stfiwx,
85	86	two_const_nr,
	87	use_ptrgl_helper,
86	88	vectors_use_two_units,
87	89	vsx,
88	90	};
...	...	@@ -342,6 +344,14 @@ pub const all_features = blk: {
342	344	.description = "Target supports instruction fusion",
343	345	.dependencies = featureSet(&[_]Feature{}),
344	346	};
	347	result[@backingInt(Feature.future_vector)] = .{
	348	.llvm_name = "future-vector",
	349	.description = "Enable Future vector instructions",
	350	.dependencies = featureSet(&[_]Feature{
	351	.isa_future_instructions,
	352	.power10_vector,
	353	}),
	354	};
345	355	result[@backingInt(Feature.hard_float)] = .{
346	356	.llvm_name = "hard-float",
347	357	.description = "Enable floating-point instructions",
...	...	@@ -441,6 +451,7 @@ pub const all_features = blk: {
441	451	.description = "32Byte load and store instructions",
442	452	.dependencies = featureSet(&[_]Feature{
443	453	.isa_v30_instructions,
	454	.vsx,
444	455	}),
445	456	};
446	457	result[@backingInt(Feature.partword_atomics)] = .{
...	...	@@ -580,6 +591,11 @@ pub const all_features = blk: {
580	591	.description = "Requires two constant Newton-Raphson computation",
581	592	.dependencies = featureSet(&[_]Feature{}),
582	593	};
	594	result[@backingInt(Feature.use_ptrgl_helper)] = .{
	595	.llvm_name = "use-ptrgl-helper",
	596	.description = "Use ._ptrgl for indirect calls",
	597	.dependencies = featureSet(&[_]Feature{}),
	598	};
583	599	result[@backingInt(Feature.vectors_use_two_units)] = .{
584	600	.llvm_name = "vectors-use-two-units",
585	601	.description = "Vectors use two units",
...	...	@@ -803,8 +819,8 @@ pub const cpu = struct {
803	819	.fuse_logical_add,
804	820	.fuse_sha3,
805	821	.fuse_store,
	822	.future_vector,
806	823	.icbt,
807		.isa_future_instructions,
808	824	.isa_v206_instructions,
809	825	.isel,
810	826	.ldbrx,
...	...	@@ -814,7 +830,6 @@ pub const cpu = struct {
814	830	.partword_atomics,
815	831	.pcrelative_memops,
816	832	.popcntd,
817		.power10_vector,
818	833	.ppc_postra_sched,
819	834	.ppc_prera_sched,
820	835	.predictable_select_expensive,

lib/std/Target/riscv.zig+673-134

...	...	@@ -8,13 +8,8 @@ pub const Feature = enum {
8	8	@"32bit",
9	9	@"64bit",
10	10	a,
11		add_load_fusion,
12		addi_load_fusion,
13	11	andes45,
14		auipc_addi_fusion,
15		auipc_load_fusion,
16	12	b,
17		bfext_fusion,
18	13	c,
19	14	conditional_cmv_fusion,
20	15	d,
...	...	@@ -25,6 +20,7 @@ pub const Feature = enum {
25	20	disable_postmisched_store_clustering,
26	21	dlen_factor_2,
27	22	e,
	23	enable_select_opt,
28	24	enable_vsetvli_sched_heuristic,
29	25	exact_asm,
30	26	experimental,
...	...	@@ -32,26 +28,57 @@ pub const Feature = enum {
32	28	experimental_rvm23u32,
33	29	experimental_smpmpmt,
34	30	experimental_svukte,
35		experimental_xrivosvisni,
36		experimental_xrivosvizip,
	31	experimental_xqccmt,
37	32	experimental_xsfmclic,
38	33	experimental_xsfsclic,
	34	experimental_y,
39	35	experimental_zibi,
40	36	experimental_zicfilp,
41	37	experimental_zicfiss,
	38	experimental_zvabd,
42	39	experimental_zvbc32e,
	40	experimental_zvdot4a8i,
	41	experimental_zvfbdota32f,
43	42	experimental_zvfbfa,
44	43	experimental_zvfofp8min,
	44	experimental_zvfqwbdota8f,
	45	experimental_zvfqwdota8f,
	46	experimental_zvfwbdota16bf,
	47	experimental_zvfwdota16bf,
45	48	experimental_zvkgs,
46		experimental_zvqdotq,
	49	experimental_zvqwbdota16i,
	50	experimental_zvqwbdota8i,
	51	experimental_zvqwdota16i,
	52	experimental_zvqwdota8i,
	53	experimental_zvvfmm,
	54	experimental_zvvmm,
	55	experimental_zvvmtls,
	56	experimental_zvvmttls,
	57	experimental_zvzip,
47	58	f,
48	59	forced_atomics,
	60	fusion_add_load,
	61	fusion_add_mem,
	62	fusion_addi_load,
	63	fusion_auipc_addi,
	64	fusion_auipc_load,
	65	fusion_bfext,
	66	fusion_ld_add,
	67	fusion_logic_imm_reg,
	68	fusion_logic_reg_imm,
	69	fusion_logic_reg_reg,
	70	fusion_lui_addi,
	71	fusion_lui_load,
	72	fusion_mul_add,
	73	fusion_qc_e_li_load_store,
	74	fusion_shift_bit_extract,
	75	fusion_shifted_zextw,
	76	fusion_shxadd_load,
	77	fusion_zexth,
	78	fusion_zextw,
49	79	h,
50	80	i,
51		ld_add_fusion,
52	81	log_vrgather,
53		lui_addi_fusion,
54		lui_load_fusion,
55	82	m,
56	83	mips_p8700,
57	84	no_default_unroll,
...	...	@@ -67,6 +94,7 @@ pub const Feature = enum {
67	94	optimized_zero_stride_load,
68	95	permissive_zalrsc,
69	96	predictable_select_expensive,
	97	prefer_ascending_load_store,
70	98	prefer_vsetvli_over_read_vlenb,
71	99	prefer_w_inst,
72	100	q,
...	...	@@ -118,17 +146,16 @@ pub const Feature = enum {
118	146	sha,
119	147	shcounterenw,
120	148	shgatpa,
121		shifted_zextw_fusion,
122	149	shlcofideleg,
123	150	short_forward_branch_ialu,
124	151	short_forward_branch_iload,
125	152	short_forward_branch_iminmax,
	153	short_forward_branch_imm,
126	154	short_forward_branch_imul,
127	155	shtvala,
128	156	shvsatpa,
129	157	shvstvala,
130	158	shvstvecd,
131		shxadd_load_fusion,
132	159	single_element_vec_fp64,
133	160	smaia,
134	161	smcdeleg,
...	...	@@ -165,6 +192,7 @@ pub const Feature = enum {
165	192	svinval,
166	193	svnapot,
167	194	svpbmt,
	195	svrsw60t59b,
168	196	svvptc,
169	197	tagged_globals,
170	198	unaligned_scalar_mem,
...	...	@@ -174,6 +202,7 @@ pub const Feature = enum {
174	202	ventana_veyron,
175	203	vl_dependent_latency,
176	204	vxrm_pipeline_flush,
	205	xaifet,
177	206	xandesbfhcvt,
178	207	xandesperf,
179	208	xandesvbfhcvt,
...	...	@@ -181,6 +210,7 @@ pub const Feature = enum {
181	210	xandesvpackfph,
182	211	xandesvsinth,
183	212	xandesvsintload,
	213	xcheriot,
184	214	xcvalu,
185	215	xcvbi,
186	216	xcvbitmanip,
...	...	@@ -215,6 +245,7 @@ pub const Feature = enum {
215	245	xsfcease,
216	246	xsfmm128t,
217	247	xsfmm16t,
	248	xsfmm32a,
218	249	xsfmm32a16f,
219	250	xsfmm32a32f,
220	251	xsfmm32a8f,
...	...	@@ -236,6 +267,7 @@ pub const Feature = enum {
236	267	xsifivecdiscarddlone,
237	268	xsifivecflushdlone,
238	269	xsmtvdot,
	270	xsmtvdotii,
239	271	xtheadba,
240	272	xtheadbb,
241	273	xtheadbs,
...	...	@@ -275,8 +307,6 @@ pub const Feature = enum {
275	307	zcmp,
276	308	zcmt,
277	309	zdinx,
278		zexth_fusion,
279		zextw_fusion,
280	310	zfa,
281	311	zfbfmin,
282	312	zfh,
...	...	@@ -290,6 +320,7 @@ pub const Feature = enum {
290	320	zicboz,
291	321	ziccamoa,
292	322	ziccamoc,
	323	ziccid,
293	324	ziccif,
294	325	zicclsm,
295	326	ziccrse,
...	...	@@ -301,7 +332,7 @@ pub const Feature = enum {
301	332	zihintpause,
302	333	zihpm,
303	334	zilsd,
304		zilsd_4byte_align,
	335	zilsd_word_align,
305	336	zimop,
306	337	zk,
307	338	zkn,
...	...	@@ -382,31 +413,11 @@ pub const all_features = blk: {
382	413	.zalrsc,
383	414	}),
384	415	};
385		result[@backingInt(Feature.add_load_fusion)] = .{
386		.llvm_name = "add-load-fusion",
387		.description = "Enable ADD(.UW) + load macrofusion",
388		.dependencies = featureSet(&[_]Feature{}),
389		};
390		result[@backingInt(Feature.addi_load_fusion)] = .{
391		.llvm_name = "addi-load-fusion",
392		.description = "Enable ADDI + load macrofusion",
393		.dependencies = featureSet(&[_]Feature{}),
394		};
395	416	result[@backingInt(Feature.andes45)] = .{
396	417	.llvm_name = "andes45",
397	418	.description = "Andes 45-Series processors",
398	419	.dependencies = featureSet(&[_]Feature{}),
399	420	};
400		result[@backingInt(Feature.auipc_addi_fusion)] = .{
401		.llvm_name = "auipc-addi-fusion",
402		.description = "Enable AUIPC+ADDI macrofusion",
403		.dependencies = featureSet(&[_]Feature{}),
404		};
405		result[@backingInt(Feature.auipc_load_fusion)] = .{
406		.llvm_name = "auipc-load-fusion",
407		.description = "Enable AUIPC + load macrofusion",
408		.dependencies = featureSet(&[_]Feature{}),
409		};
410	421	result[@backingInt(Feature.b)] = .{
411	422	.llvm_name = "b",
412	423	.description = "'B' (the collection of the Zba, Zbb, Zbs extensions)",
...	...	@@ -416,11 +427,6 @@ pub const all_features = blk: {
416	427	.zbs,
417	428	}),
418	429	};
419		result[@backingInt(Feature.bfext_fusion)] = .{
420		.llvm_name = "bfext-fusion",
421		.description = "Enable SLLI+SRLI (bitfield extract) macrofusion",
422		.dependencies = featureSet(&[_]Feature{}),
423		};
424	430	result[@backingInt(Feature.c)] = .{
425	431	.llvm_name = "c",
426	432	.description = "'C' (Compressed Instructions)",
...	...	@@ -475,6 +481,11 @@ pub const all_features = blk: {
475	481	.description = "'E' (Embedded Instruction Set with 16 GPRs)",
476	482	.dependencies = featureSet(&[_]Feature{}),
477	483	};
	484	result[@backingInt(Feature.enable_select_opt)] = .{
	485	.llvm_name = "enable-select-opt",
	486	.description = "Enable the select optimize pass for select loop heuristics",
	487	.dependencies = featureSet(&[_]Feature{}),
	488	};
478	489	result[@backingInt(Feature.enable_vsetvli_sched_heuristic)] = .{
479	490	.llvm_name = "enable-vsetvli-sched-heuristic",
480	491	.description = "Enable vsetvli-based scheduling heuristic",
...	...	@@ -493,7 +504,12 @@ pub const all_features = blk: {
493	504	result[@backingInt(Feature.experimental_p)] = .{
494	505	.llvm_name = "experimental-p",
495	506	.description = "'P' ('Base P' (Packed SIMD))",
496		.dependencies = featureSet(&[_]Feature{}),
	507	.dependencies = featureSet(&[_]Feature{
	508	.zba,
	509	.zbb,
	510	.zicsr,
	511	.zmmul,
	512	}),
497	513	};
498	514	result[@backingInt(Feature.experimental_rvm23u32)] = .{
499	515	.llvm_name = "experimental-rvm23u32",
...	...	@@ -522,15 +538,13 @@ pub const all_features = blk: {
522	538	.description = "'Svukte' (Address-Independent Latency of User-Mode Faults to Supervisor Addresses)",
523	539	.dependencies = featureSet(&[_]Feature{}),
524	540	};
525		result[@backingInt(Feature.experimental_xrivosvisni)] = .{
526		.llvm_name = "experimental-xrivosvisni",
527		.description = "'XRivosVisni' (Rivos Vector Integer Small New)",
528		.dependencies = featureSet(&[_]Feature{}),
529		};
530		result[@backingInt(Feature.experimental_xrivosvizip)] = .{
531		.llvm_name = "experimental-xrivosvizip",
532		.description = "'XRivosVizip' (Rivos Vector Register Zips)",
533		.dependencies = featureSet(&[_]Feature{}),
	541	result[@backingInt(Feature.experimental_xqccmt)] = .{
	542	.llvm_name = "experimental-xqccmt",
	543	.description = "'Xqccmt' (Qualcomm 16-bit Table Jump)",
	544	.dependencies = featureSet(&[_]Feature{
	545	.zca,
	546	.zicsr,
	547	}),
534	548	};
535	549	result[@backingInt(Feature.experimental_xsfmclic)] = .{
536	550	.llvm_name = "experimental-xsfmclic",
...	...	@@ -542,6 +556,11 @@ pub const all_features = blk: {
542	556	.description = "'XSfsclic' (SiFive CLIC Supervisor-mode CSRs)",
543	557	.dependencies = featureSet(&[_]Feature{}),
544	558	};
	559	result[@backingInt(Feature.experimental_y)] = .{
	560	.llvm_name = "experimental-y",
	561	.description = "'Y' ('Base Y' (CHERI))",
	562	.dependencies = featureSet(&[_]Feature{}),
	563	};
545	564	result[@backingInt(Feature.experimental_zibi)] = .{
546	565	.llvm_name = "experimental-zibi",
547	566	.description = "'Zibi' (Branch with Immediate)",
...	...	@@ -562,6 +581,11 @@ pub const all_features = blk: {
562	581	.zimop,
563	582	}),
564	583	};
	584	result[@backingInt(Feature.experimental_zvabd)] = .{
	585	.llvm_name = "experimental-zvabd",
	586	.description = "'Zvabd' (Vector Absolute Difference)",
	587	.dependencies = featureSet(&[_]Feature{}),
	588	};
565	589	result[@backingInt(Feature.experimental_zvbc32e)] = .{
566	590	.llvm_name = "experimental-zvbc32e",
567	591	.description = "'Zvbc32e' (Vector Carryless Multiplication with 32-bits elements)",
...	...	@@ -569,6 +593,20 @@ pub const all_features = blk: {
569	593	.zve32x,
570	594	}),
571	595	};
	596	result[@backingInt(Feature.experimental_zvdot4a8i)] = .{
	597	.llvm_name = "experimental-zvdot4a8i",
	598	.description = "'Zvdot4a8i' (Vector 4-element Dot Product of packed 8-bit Integers)",
	599	.dependencies = featureSet(&[_]Feature{
	600	.zve32x,
	601	}),
	602	};
	603	result[@backingInt(Feature.experimental_zvfbdota32f)] = .{
	604	.llvm_name = "experimental-zvfbdota32f",
	605	.description = "'Zvfbdota32f' (FP32 batched dot-product extension)",
	606	.dependencies = featureSet(&[_]Feature{
	607	.zve32f,
	608	}),
	609	};
572	610	result[@backingInt(Feature.experimental_zvfbfa)] = .{
573	611	.llvm_name = "experimental-zvfbfa",
574	612	.description = "'Zvfbfa' (Additional BF16 vector compute support)",
...	...	@@ -584,6 +622,34 @@ pub const all_features = blk: {
584	622	.zve32f,
585	623	}),
586	624	};
	625	result[@backingInt(Feature.experimental_zvfqwbdota8f)] = .{
	626	.llvm_name = "experimental-zvfqwbdota8f",
	627	.description = "'Zvfqwbdota8f' (OCP FP8 batched dot-product extension)",
	628	.dependencies = featureSet(&[_]Feature{
	629	.zve32f,
	630	}),
	631	};
	632	result[@backingInt(Feature.experimental_zvfqwdota8f)] = .{
	633	.llvm_name = "experimental-zvfqwdota8f",
	634	.description = "'Zvfqwdota8f' (OCP FP8 Dot-Product)",
	635	.dependencies = featureSet(&[_]Feature{
	636	.zve32f,
	637	}),
	638	};
	639	result[@backingInt(Feature.experimental_zvfwbdota16bf)] = .{
	640	.llvm_name = "experimental-zvfwbdota16bf",
	641	.description = "'Zvfwbdota16bf' (BF16 batched dot-product extension)",
	642	.dependencies = featureSet(&[_]Feature{
	643	.zve32f,
	644	}),
	645	};
	646	result[@backingInt(Feature.experimental_zvfwdota16bf)] = .{
	647	.llvm_name = "experimental-zvfwdota16bf",
	648	.description = "'Zvfwdota16bf' (BF16 Dot-Product)",
	649	.dependencies = featureSet(&[_]Feature{
	650	.zve32f,
	651	}),
	652	};
587	653	result[@backingInt(Feature.experimental_zvkgs)] = .{
588	654	.llvm_name = "experimental-zvkgs",
589	655	.description = "'Zvkgs' (Vector-Scalar GCM instructions for Cryptography)",
...	...	@@ -591,9 +657,65 @@ pub const all_features = blk: {
591	657	.zvkg,
592	658	}),
593	659	};
594		result[@backingInt(Feature.experimental_zvqdotq)] = .{
595		.llvm_name = "experimental-zvqdotq",
596		.description = "'Zvqdotq' (Vector quad widening 4D Dot Product)",
	660	result[@backingInt(Feature.experimental_zvqwbdota16i)] = .{
	661	.llvm_name = "experimental-zvqwbdota16i",
	662	.description = "'Zvqwbdota16i' (16-bit integer batched dot-product extension)",
	663	.dependencies = featureSet(&[_]Feature{
	664	.zve64x,
	665	}),
	666	};
	667	result[@backingInt(Feature.experimental_zvqwbdota8i)] = .{
	668	.llvm_name = "experimental-zvqwbdota8i",
	669	.description = "'Zvqwbdota8i' (8-bit integer batched dot-product extension)",
	670	.dependencies = featureSet(&[_]Feature{
	671	.zve32x,
	672	}),
	673	};
	674	result[@backingInt(Feature.experimental_zvqwdota16i)] = .{
	675	.llvm_name = "experimental-zvqwdota16i",
	676	.description = "'Zvqwdota16i' (16-bit Integer Dot-Product)",
	677	.dependencies = featureSet(&[_]Feature{
	678	.zve64x,
	679	}),
	680	};
	681	result[@backingInt(Feature.experimental_zvqwdota8i)] = .{
	682	.llvm_name = "experimental-zvqwdota8i",
	683	.description = "'Zvqwdota8i' (8-bit Integer Dot-Product)",
	684	.dependencies = featureSet(&[_]Feature{
	685	.zve32x,
	686	}),
	687	};
	688	result[@backingInt(Feature.experimental_zvvfmm)] = .{
	689	.llvm_name = "experimental-zvvfmm",
	690	.description = "'Zvvfmm' (Floating-Point Matrix Multiply-Accumulate)",
	691	.dependencies = featureSet(&[_]Feature{
	692	.zve32f,
	693	}),
	694	};
	695	result[@backingInt(Feature.experimental_zvvmm)] = .{
	696	.llvm_name = "experimental-zvvmm",
	697	.description = "'Zvvmm' (Integer Matrix Multiply-Accumulate)",
	698	.dependencies = featureSet(&[_]Feature{
	699	.zve32x,
	700	}),
	701	};
	702	result[@backingInt(Feature.experimental_zvvmtls)] = .{
	703	.llvm_name = "experimental-zvvmtls",
	704	.description = "'Zvvmtls' (Matrix Tile Load/Store)",
	705	.dependencies = featureSet(&[_]Feature{
	706	.zve32x,
	707	}),
	708	};
	709	result[@backingInt(Feature.experimental_zvvmttls)] = .{
	710	.llvm_name = "experimental-zvvmttls",
	711	.description = "'Zvvmttls' (Transposing Matrix Tile Load/Store)",
	712	.dependencies = featureSet(&[_]Feature{
	713	.zve32x,
	714	}),
	715	};
	716	result[@backingInt(Feature.experimental_zvzip)] = .{
	717	.llvm_name = "experimental-zvzip",
	718	.description = "'Zvzip' (Vector Reordering Structured Data)",
597	719	.dependencies = featureSet(&[_]Feature{
598	720	.zve32x,
599	721	}),
...	...	@@ -610,36 +732,116 @@ pub const all_features = blk: {
610	732	.description = "Assume that lock-free native-width atomics are available",
611	733	.dependencies = featureSet(&[_]Feature{}),
612	734	};
613		result[@backingInt(Feature.h)] = .{
614		.llvm_name = "h",
615		.description = "'H' (Hypervisor)",
	735	result[@backingInt(Feature.fusion_add_load)] = .{
	736	.llvm_name = "fusion-add-load",
	737	.description = "Enable ADD(.UW) + load macrofusion",
616	738	.dependencies = featureSet(&[_]Feature{}),
617	739	};
618		result[@backingInt(Feature.i)] = .{
619		.llvm_name = "i",
620		.description = "'I' (Base Integer Instruction Set)",
	740	result[@backingInt(Feature.fusion_add_mem)] = .{
	741	.llvm_name = "fusion-add-mem",
	742	.description = "Enable ADD+LOAD/STORE macrofusion",
	743	.dependencies = featureSet(&[_]Feature{}),
	744	};
	745	result[@backingInt(Feature.fusion_addi_load)] = .{
	746	.llvm_name = "fusion-addi-load",
	747	.description = "Enable ADDI + load macrofusion",
	748	.dependencies = featureSet(&[_]Feature{}),
	749	};
	750	result[@backingInt(Feature.fusion_auipc_addi)] = .{
	751	.llvm_name = "fusion-auipc-addi",
	752	.description = "Enable AUIPC+ADDI macrofusion",
	753	.dependencies = featureSet(&[_]Feature{}),
	754	};
	755	result[@backingInt(Feature.fusion_auipc_load)] = .{
	756	.llvm_name = "fusion-auipc-load",
	757	.description = "Enable AUIPC + load macrofusion",
	758	.dependencies = featureSet(&[_]Feature{}),
	759	};
	760	result[@backingInt(Feature.fusion_bfext)] = .{
	761	.llvm_name = "fusion-bfext",
	762	.description = "Enable SLLI+SRLI (bitfield extract) macrofusion",
621	763	.dependencies = featureSet(&[_]Feature{}),
622	764	};
623		result[@backingInt(Feature.ld_add_fusion)] = .{
624		.llvm_name = "ld-add-fusion",
	765	result[@backingInt(Feature.fusion_ld_add)] = .{
	766	.llvm_name = "fusion-ld-add",
625	767	.description = "Enable LD+ADD macrofusion",
626	768	.dependencies = featureSet(&[_]Feature{}),
627	769	};
628		result[@backingInt(Feature.log_vrgather)] = .{
629		.llvm_name = "log-vrgather",
630		.description = "Has vrgather.vv with LMUL*log2(LMUL) latency",
	770	result[@backingInt(Feature.fusion_logic_imm_reg)] = .{
	771	.llvm_name = "fusion-logic-imm-reg",
	772	.description = "Enable ANDI/ORI/XORI+AND/OR/XOR macrofusion",
	773	.dependencies = featureSet(&[_]Feature{}),
	774	};
	775	result[@backingInt(Feature.fusion_logic_reg_imm)] = .{
	776	.llvm_name = "fusion-logic-reg-imm",
	777	.description = "Enable AND/OR/XOR+ANDI/ORI/XORI macrofusion",
	778	.dependencies = featureSet(&[_]Feature{}),
	779	};
	780	result[@backingInt(Feature.fusion_logic_reg_reg)] = .{
	781	.llvm_name = "fusion-logic-reg-reg",
	782	.description = "Enable AND/OR/XOR+AND/OR/XOR macrofusion",
631	783	.dependencies = featureSet(&[_]Feature{}),
632	784	};
633		result[@backingInt(Feature.lui_addi_fusion)] = .{
634		.llvm_name = "lui-addi-fusion",
	785	result[@backingInt(Feature.fusion_lui_addi)] = .{
	786	.llvm_name = "fusion-lui-addi",
635	787	.description = "Enable LUI+ADDI macro fusion",
636	788	.dependencies = featureSet(&[_]Feature{}),
637	789	};
638		result[@backingInt(Feature.lui_load_fusion)] = .{
639		.llvm_name = "lui-load-fusion",
	790	result[@backingInt(Feature.fusion_lui_load)] = .{
	791	.llvm_name = "fusion-lui-load",
640	792	.description = "Enable LUI + load macrofusion",
641	793	.dependencies = featureSet(&[_]Feature{}),
642	794	};
	795	result[@backingInt(Feature.fusion_mul_add)] = .{
	796	.llvm_name = "fusion-mul-add",
	797	.description = "Enable MUL+ADD macrofusion",
	798	.dependencies = featureSet(&[_]Feature{}),
	799	};
	800	result[@backingInt(Feature.fusion_qc_e_li_load_store)] = .{
	801	.llvm_name = "fusion-qc-e-li-load-store",
	802	.description = "Enable QC.E.LI + Load/Store macrofusion",
	803	.dependencies = featureSet(&[_]Feature{}),
	804	};
	805	result[@backingInt(Feature.fusion_shift_bit_extract)] = .{
	806	.llvm_name = "fusion-shift-bit-extract",
	807	.description = "Enable SLLI+SRLI/SRAI macrofusion",
	808	.dependencies = featureSet(&[_]Feature{}),
	809	};
	810	result[@backingInt(Feature.fusion_shifted_zextw)] = .{
	811	.llvm_name = "fusion-shifted-zextw",
	812	.description = "Enable SLLI+SRLI to be fused when computing (shifted) word zero extension",
	813	.dependencies = featureSet(&[_]Feature{}),
	814	};
	815	result[@backingInt(Feature.fusion_shxadd_load)] = .{
	816	.llvm_name = "fusion-shxadd-load",
	817	.description = "Enable SH(1\|2\|3)ADD(.UW) + load macrofusion",
	818	.dependencies = featureSet(&[_]Feature{}),
	819	};
	820	result[@backingInt(Feature.fusion_zexth)] = .{
	821	.llvm_name = "fusion-zexth",
	822	.description = "Enable SLLI+SRLI to be fused to zero extension of halfword",
	823	.dependencies = featureSet(&[_]Feature{}),
	824	};
	825	result[@backingInt(Feature.fusion_zextw)] = .{
	826	.llvm_name = "fusion-zextw",
	827	.description = "Enable SLLI+SRLI to be fused to zero extension of word",
	828	.dependencies = featureSet(&[_]Feature{}),
	829	};
	830	result[@backingInt(Feature.h)] = .{
	831	.llvm_name = "h",
	832	.description = "'H' (Hypervisor)",
	833	.dependencies = featureSet(&[_]Feature{}),
	834	};
	835	result[@backingInt(Feature.i)] = .{
	836	.llvm_name = "i",
	837	.description = "'I' (Base Integer Instruction Set)",
	838	.dependencies = featureSet(&[_]Feature{}),
	839	};
	840	result[@backingInt(Feature.log_vrgather)] = .{
	841	.llvm_name = "log-vrgather",
	842	.description = "Has vrgather.vv with LMUL*log2(LMUL) latency",
	843	.dependencies = featureSet(&[_]Feature{}),
	844	};
643	845	result[@backingInt(Feature.m)] = .{
644	846	.llvm_name = "m",
645	847	.description = "'M' (Integer Multiplication and Division)",
...	...	@@ -654,7 +856,7 @@ pub const all_features = blk: {
654	856	};
655	857	result[@backingInt(Feature.no_default_unroll)] = .{
656	858	.llvm_name = "no-default-unroll",
657		.description = "Disable default unroll preference.",
	859	.description = "Disable default unroll preference",
658	860	.dependencies = featureSet(&[_]Feature{}),
659	861	};
660	862	result[@backingInt(Feature.no_sink_splat_operands)] = .{
...	...	@@ -664,7 +866,7 @@ pub const all_features = blk: {
664	866	};
665	867	result[@backingInt(Feature.no_trailing_seq_cst_fence)] = .{
666	868	.llvm_name = "no-trailing-seq-cst-fence",
667		.description = "Disable trailing fence for seq-cst store.",
	869	.description = "Disable trailing fence for seq-cst store",
668	870	.dependencies = featureSet(&[_]Feature{}),
669	871	};
670	872	result[@backingInt(Feature.optimized_nf2_segment_load_store)] = .{
...	...	@@ -717,6 +919,11 @@ pub const all_features = blk: {
717	919	.description = "Prefer likely predicted branches over selects",
718	920	.dependencies = featureSet(&[_]Feature{}),
719	921	};
	922	result[@backingInt(Feature.prefer_ascending_load_store)] = .{
	923	.llvm_name = "prefer-ascending-load-store",
	924	.description = "Prefer ascending load/store order for better clustering and performance",
	925	.dependencies = featureSet(&[_]Feature{}),
	926	};
720	927	result[@backingInt(Feature.prefer_vsetvli_over_read_vlenb)] = .{
721	928	.llvm_name = "prefer-vsetvli-over-read-vlenb",
722	929	.description = "Prefer vsetvli over read vlenb CSR to calculate VLEN",
...	...	@@ -736,7 +943,7 @@ pub const all_features = blk: {
736	943	};
737	944	result[@backingInt(Feature.relax)] = .{
738	945	.llvm_name = "relax",
739		.description = "Enable Linker relaxation.",
	946	.description = "Enable Linker relaxation",
740	947	.dependencies = featureSet(&[_]Feature{}),
741	948	};
742	949	result[@backingInt(Feature.reserve_x1)] = .{
...	...	@@ -1187,7 +1394,7 @@ pub const all_features = blk: {
1187	1394	};
1188	1395	result[@backingInt(Feature.save_restore)] = .{
1189	1396	.llvm_name = "save-restore",
1190		.description = "Enable save/restore.",
	1397	.description = "Enable save/restore",
1191	1398	.dependencies = featureSet(&[_]Feature{}),
1192	1399	};
1193	1400	result[@backingInt(Feature.sdext)] = .{
...	...	@@ -1224,11 +1431,6 @@ pub const all_features = blk: {
1224	1431	.description = "'Shgatpa' (SvNNx4 mode supported for all modes supported by satp, as well as Bare)",
1225	1432	.dependencies = featureSet(&[_]Feature{}),
1226	1433	};
1227		result[@backingInt(Feature.shifted_zextw_fusion)] = .{
1228		.llvm_name = "shifted-zextw-fusion",
1229		.description = "Enable SLLI+SRLI to be fused when computing (shifted) word zero extension",
1230		.dependencies = featureSet(&[_]Feature{}),
1231		};
1232	1434	result[@backingInt(Feature.shlcofideleg)] = .{
1233	1435	.llvm_name = "shlcofideleg",
1234	1436	.description = "'Shlcofideleg' (Delegating LCOFI Interrupts to VS-mode)",
...	...	@@ -1253,6 +1455,13 @@ pub const all_features = blk: {
1253	1455	.short_forward_branch_ialu,
1254	1456	}),
1255	1457	};
	1458	result[@backingInt(Feature.short_forward_branch_imm)] = .{
	1459	.llvm_name = "short-forward-branch-imm",
	1460	.description = "Enable short forward branch optimization for branches with immediates",
	1461	.dependencies = featureSet(&[_]Feature{
	1462	.short_forward_branch_ialu,
	1463	}),
	1464	};
1256	1465	result[@backingInt(Feature.short_forward_branch_imul)] = .{
1257	1466	.llvm_name = "short-forward-branch-imul",
1258	1467	.description = "Enable short forward branch optimization for MUL instruction",
...	...	@@ -1280,11 +1489,6 @@ pub const all_features = blk: {
1280	1489	.description = "'Shvstvecd' (vstvec supports Direct mode)",
1281	1490	.dependencies = featureSet(&[_]Feature{}),
1282	1491	};
1283		result[@backingInt(Feature.shxadd_load_fusion)] = .{
1284		.llvm_name = "shxadd-load-fusion",
1285		.description = "Enable SH(1\|2\|3)ADD(.UW) + load macrofusion",
1286		.dependencies = featureSet(&[_]Feature{}),
1287		};
1288	1492	result[@backingInt(Feature.single_element_vec_fp64)] = .{
1289	1493	.llvm_name = "single-element-vec-fp64",
1290	1494	.description = "Certain vector FP64 operations produce a single result element per cycle",
...	...	@@ -1473,6 +1677,11 @@ pub const all_features = blk: {
1473	1677	.description = "'Svpbmt' (Page-Based Memory Types)",
1474	1678	.dependencies = featureSet(&[_]Feature{}),
1475	1679	};
	1680	result[@backingInt(Feature.svrsw60t59b)] = .{
	1681	.llvm_name = "svrsw60t59b",
	1682	.description = "'Svrsw60t59b' (PTE Reserved-for-Software Bits 60-59)",
	1683	.dependencies = featureSet(&[_]Feature{}),
	1684	};
1476	1685	result[@backingInt(Feature.svvptc)] = .{
1477	1686	.llvm_name = "svvptc",
1478	1687	.description = "'Svvptc' (Obviating Memory-Management Instructions after Marking PTEs Valid)",
...	...	@@ -1521,6 +1730,13 @@ pub const all_features = blk: {
1521	1730	.description = "VXRM writes causes pipeline flush",
1522	1731	.dependencies = featureSet(&[_]Feature{}),
1523	1732	};
	1733	result[@backingInt(Feature.xaifet)] = .{
	1734	.llvm_name = "xaifet",
	1735	.description = "'XAIFET' (AI Foundry ET Extension)",
	1736	.dependencies = featureSet(&[_]Feature{
	1737	.f,
	1738	}),
	1739	};
1524	1740	result[@backingInt(Feature.xandesbfhcvt)] = .{
1525	1741	.llvm_name = "xandesbfhcvt",
1526	1742	.description = "'XAndesBFHCvt' (Andes Scalar BFLOAT16 Conversion Extension)",
...	...	@@ -1568,6 +1784,15 @@ pub const all_features = blk: {
1568	1784	.zve32x,
1569	1785	}),
1570	1786	};
	1787	result[@backingInt(Feature.xcheriot)] = .{
	1788	.llvm_name = "xcheriot",
	1789	.description = "'XCheriot' (CHERIoT extension)",
	1790	.dependencies = featureSet(&[_]Feature{
	1791	.e,
	1792	.m,
	1793	.zca,
	1794	}),
	1795	};
1571	1796	result[@backingInt(Feature.xcvalu)] = .{
1572	1797	.llvm_name = "xcvalu",
1573	1798	.description = "'XCValu' (CORE-V ALU Operations)",
...	...	@@ -1787,6 +2012,15 @@ pub const all_features = blk: {
1787	2012	.zvl64b,
1788	2013	}),
1789	2014	};
	2015	result[@backingInt(Feature.xsfmm32a)] = .{
	2016	.llvm_name = "xsfmm32a",
	2017	.description = "'XSfmm32a' (TEW=32-bit accumulation operands - int: 8b; float: fp16, bf16, fp32)",
	2018	.dependencies = featureSet(&[_]Feature{
	2019	.xsfmm32a16f,
	2020	.xsfmm32a32f,
	2021	.xsfmm32a8i,
	2022	}),
	2023	};
1790	2024	result[@backingInt(Feature.xsfmm32a16f)] = .{
1791	2025	.llvm_name = "xsfmm32a16f",
1792	2026	.description = "'XSfmm32a16f' (TEW=32-bit accumulation, operands - float: 16b, widen=2 (IEEE, BF))",
...	...	@@ -1938,6 +2172,13 @@ pub const all_features = blk: {
1938	2172	.zve32f,
1939	2173	}),
1940	2174	};
	2175	result[@backingInt(Feature.xsmtvdotii)] = .{
	2176	.llvm_name = "xsmtvdotii",
	2177	.description = "'XSMTVDotII' (SpacemiT Vector Extension for Matrix 2.0)",
	2178	.dependencies = featureSet(&[_]Feature{
	2179	.zve32f,
	2180	}),
	2181	};
1941	2182	result[@backingInt(Feature.xtheadba)] = .{
1942	2183	.llvm_name = "xtheadba",
1943	2184	.description = "'XTHeadBa' (T-Head address calculation instructions)",
...	...	@@ -2069,7 +2310,9 @@ pub const all_features = blk: {
2069	2310	result[@backingInt(Feature.zbc)] = .{
2070	2311	.llvm_name = "zbc",
2071	2312	.description = "'Zbc' (Carry-Less Multiplication)",
2072		.dependencies = featureSet(&[_]Feature{}),
	2313	.dependencies = featureSet(&[_]Feature{
	2314	.zbkc,
	2315	}),
2073	2316	};
2074	2317	result[@backingInt(Feature.zbkb)] = .{
2075	2318	.llvm_name = "zbkb",
...	...	@@ -2165,16 +2408,6 @@ pub const all_features = blk: {
2165	2408	.zfinx,
2166	2409	}),
2167	2410	};
2168		result[@backingInt(Feature.zexth_fusion)] = .{
2169		.llvm_name = "zexth-fusion",
2170		.description = "Enable SLLI+SRLI to be fused to zero extension of halfword",
2171		.dependencies = featureSet(&[_]Feature{}),
2172		};
2173		result[@backingInt(Feature.zextw_fusion)] = .{
2174		.llvm_name = "zextw-fusion",
2175		.description = "Enable SLLI+SRLI to be fused to zero extension of word",
2176		.dependencies = featureSet(&[_]Feature{}),
2177		};
2178	2411	result[@backingInt(Feature.zfa)] = .{
2179	2412	.llvm_name = "zfa",
2180	2413	.description = "'Zfa' (Additional Floating-Point)",
...	...	@@ -2254,6 +2487,13 @@ pub const all_features = blk: {
2254	2487	.description = "'Ziccamoc' (Main Memory Supports Atomics in Zacas)",
2255	2488	.dependencies = featureSet(&[_]Feature{}),
2256	2489	};
	2490	result[@backingInt(Feature.ziccid)] = .{
	2491	.llvm_name = "ziccid",
	2492	.description = "'Ziccid' (Instruction/Data Coherence and Consistency)",
	2493	.dependencies = featureSet(&[_]Feature{
	2494	.ziccif,
	2495	}),
	2496	};
2257	2497	result[@backingInt(Feature.ziccif)] = .{
2258	2498	.llvm_name = "ziccif",
2259	2499	.description = "'Ziccif' (Main Memory Supports Instruction Fetch with Atomicity Requirement)",
...	...	@@ -2313,8 +2553,8 @@ pub const all_features = blk: {
2313	2553	.description = "'Zilsd' (Load/Store Pair Instructions)",
2314	2554	.dependencies = featureSet(&[_]Feature{}),
2315	2555	};
2316		result[@backingInt(Feature.zilsd_4byte_align)] = .{
2317		.llvm_name = "zilsd-4byte-align",
	2556	result[@backingInt(Feature.zilsd_word_align)] = .{
	2557	.llvm_name = "zilsd-word-align",
2318	2558	.description = "Allow 4-byte alignment for Zilsd LD/SD instructions",
2319	2559	.dependencies = featureSet(&[_]Feature{}),
2320	2560	};
...	...	@@ -2543,6 +2783,7 @@ pub const all_features = blk: {
2543	2783	.description = "'Zvknhb' (Vector SHA-2 (SHA-256 and SHA-512))",
2544	2784	.dependencies = featureSet(&[_]Feature{
2545	2785	.zve64x,
	2786	.zvknha,
2546	2787	}),
2547	2788	};
2548	2789	result[@backingInt(Feature.zvks)] = .{
...	...	@@ -2681,6 +2922,16 @@ pub const all_features = blk: {
2681	2922	};
2682	2923
2683	2924	pub const cpu = struct {
	2925	pub const an_erbium: CpuModel = .{
	2926	.name = "an_erbium",
	2927	.llvm_name = "an-erbium",
	2928	.features = featureSet(&[_]Feature{
	2929	.@"64bit",
	2930	.c,
	2931	.m,
	2932	.xaifet,
	2933	}),
	2934	};
2684	2935	pub const andes_45_series: CpuModel = .{
2685	2936	.name = "andes_45_series",
2686	2937	.llvm_name = "andes-45-series",
...	...	@@ -2833,6 +3084,16 @@ pub const cpu = struct {
2833	3084	.m,
2834	3085	}),
2835	3086	};
	3087	pub const et_soc1: CpuModel = .{
	3088	.name = "et_soc1",
	3089	.llvm_name = "et-soc1",
	3090	.features = featureSet(&[_]Feature{
	3091	.@"64bit",
	3092	.c,
	3093	.m,
	3094	.xaifet,
	3095	}),
	3096	};
2836	3097	pub const generic: CpuModel = .{
2837	3098	.name = "generic",
2838	3099	.llvm_name = "generic",
...	...	@@ -2848,6 +3109,7 @@ pub const cpu = struct {
2848	3109	.llvm_name = "generic-rv32",
2849	3110	.features = featureSet(&[_]Feature{
2850	3111	.@"32bit",
	3112	.enable_select_opt,
2851	3113	.i,
2852	3114	.optimized_nf2_segment_load_store,
2853	3115	}),
...	...	@@ -2857,6 +3119,7 @@ pub const cpu = struct {
2857	3119	.llvm_name = "generic-rv64",
2858	3120	.features = featureSet(&[_]Feature{
2859	3121	.@"64bit",
	3122	.enable_select_opt,
2860	3123	.i,
2861	3124	.optimized_nf2_segment_load_store,
2862	3125	}),
...	...	@@ -3020,13 +3283,13 @@ pub const cpu = struct {
3020	3283	.features = featureSet(&[_]Feature{
3021	3284	.@"64bit",
3022	3285	.a,
3023		.auipc_addi_fusion,
3024	3286	.b,
3025	3287	.c,
3026	3288	.conditional_cmv_fusion,
3027	3289	.d,
	3290	.fusion_auipc_addi,
	3291	.fusion_lui_addi,
3028	3292	.i,
3029		.lui_addi_fusion,
3030	3293	.m,
3031	3294	.no_default_unroll,
3032	3295	.unaligned_scalar_mem,
...	...	@@ -3055,12 +3318,12 @@ pub const cpu = struct {
3055	3318	.features = featureSet(&[_]Feature{
3056	3319	.@"64bit",
3057	3320	.a,
3058		.auipc_addi_fusion,
3059	3321	.b,
3060	3322	.c,
3061	3323	.conditional_cmv_fusion,
	3324	.fusion_auipc_addi,
	3325	.fusion_lui_addi,
3062	3326	.i,
3063		.lui_addi_fusion,
3064	3327	.m,
3065	3328	.no_default_unroll,
3066	3329	.no_sink_splat_operands,
...	...	@@ -3099,12 +3362,13 @@ pub const cpu = struct {
3099	3362	.features = featureSet(&[_]Feature{
3100	3363	.@"64bit",
3101	3364	.a,
3102		.auipc_addi_fusion,
3103	3365	.c,
3104	3366	.conditional_cmv_fusion,
3105	3367	.d,
	3368	.enable_select_opt,
	3369	.fusion_auipc_addi,
	3370	.fusion_lui_addi,
3106	3371	.i,
3107		.lui_addi_fusion,
3108	3372	.m,
3109	3373	.no_default_unroll,
3110	3374	.use_postra_scheduler,
...	...	@@ -3119,12 +3383,12 @@ pub const cpu = struct {
3119	3383	.features = featureSet(&[_]Feature{
3120	3384	.@"64bit",
3121	3385	.a,
3122		.auipc_addi_fusion,
3123	3386	.b,
3124	3387	.c,
3125	3388	.conditional_cmv_fusion,
	3389	.fusion_auipc_addi,
	3390	.fusion_lui_addi,
3126	3391	.i,
3127		.lui_addi_fusion,
3128	3392	.m,
3129	3393	.no_default_unroll,
3130	3394	.no_sink_splat_operands,
...	...	@@ -3155,18 +3419,18 @@ pub const cpu = struct {
3155	3419	.zvksg,
3156	3420	}),
3157	3421	};
3158		pub const sifive_p870: CpuModel = .{
3159		.name = "sifive_p870",
3160		.llvm_name = "sifive-p870",
	3422	pub const sifive_p870_d: CpuModel = .{
	3423	.name = "sifive_p870_d",
	3424	.llvm_name = "sifive-p870-d",
3161	3425	.features = featureSet(&[_]Feature{
3162	3426	.@"64bit",
3163	3427	.a,
3164		.auipc_addi_fusion,
3165	3428	.b,
3166	3429	.c,
3167	3430	.conditional_cmv_fusion,
	3431	.fusion_auipc_addi,
	3432	.fusion_lui_addi,
3168	3433	.i,
3169		.lui_addi_fusion,
3170	3434	.m,
3171	3435	.no_default_unroll,
3172	3436	.no_sink_splat_operands,
...	...	@@ -3294,6 +3558,103 @@ pub const cpu = struct {
3294	3558	.zifencei,
3295	3559	}),
3296	3560	};
	3561	pub const sifive_x160: CpuModel = .{
	3562	.name = "sifive_x160",
	3563	.llvm_name = "sifive-x160",
	3564	.features = featureSet(&[_]Feature{
	3565	.@"32bit",
	3566	.a,
	3567	.b,
	3568	.c,
	3569	.dlen_factor_2,
	3570	.experimental_zicfilp,
	3571	.experimental_zvdot4a8i,
	3572	.experimental_zvfbfa,
	3573	.i,
	3574	.m,
	3575	.no_default_unroll,
	3576	.optimized_nf2_segment_load_store,
	3577	.optimized_zero_stride_load,
	3578	.short_forward_branch_ialu,
	3579	.use_postra_scheduler,
	3580	.vl_dependent_latency,
	3581	.xsfcease,
	3582	.zawrs,
	3583	.zce,
	3584	.zcf,
	3585	.zcmop,
	3586	.zfa,
	3587	.zfh,
	3588	.zicbom,
	3589	.zicbop,
	3590	.zicboz,
	3591	.zicond,
	3592	.zifencei,
	3593	.zihintntl,
	3594	.zihintpause,
	3595	.zihpm,
	3596	.zimop,
	3597	.zkt,
	3598	.zvbb,
	3599	.zvfbfwma,
	3600	.zvfh,
	3601	.zvkt,
	3602	.zvl128b,
	3603	}),
	3604	};
	3605	pub const sifive_x180: CpuModel = .{
	3606	.name = "sifive_x180",
	3607	.llvm_name = "sifive-x180",
	3608	.features = featureSet(&[_]Feature{
	3609	.@"64bit",
	3610	.a,
	3611	.b,
	3612	.c,
	3613	.dlen_factor_2,
	3614	.experimental_zicfilp,
	3615	.experimental_zvdot4a8i,
	3616	.experimental_zvfbfa,
	3617	.i,
	3618	.m,
	3619	.no_default_unroll,
	3620	.optimized_nf2_segment_load_store,
	3621	.optimized_zero_stride_load,
	3622	.short_forward_branch_ialu,
	3623	.use_postra_scheduler,
	3624	.v,
	3625	.vl_dependent_latency,
	3626	.xsfcease,
	3627	.xsfvfbfexp16e,
	3628	.xsfvfexp16e,
	3629	.xsfvfexp32e,
	3630	.xsfvfexpa,
	3631	.xsifivecflushdlone,
	3632	.za64rs,
	3633	.zawrs,
	3634	.zcb,
	3635	.zcd,
	3636	.zcmop,
	3637	.zfa,
	3638	.zfh,
	3639	.zic64b,
	3640	.zicbom,
	3641	.zicbop,
	3642	.zicboz,
	3643	.ziccamoa,
	3644	.ziccif,
	3645	.ziccrse,
	3646	.zicond,
	3647	.zifencei,
	3648	.zihintntl,
	3649	.zihintpause,
	3650	.zihpm,
	3651	.zimop,
	3652	.zkt,
	3653	.zvbb,
	3654	.zvfbfwma,
	3655	.zvkt,
	3656	}),
	3657	};
3297	3658	pub const sifive_x280: CpuModel = .{
3298	3659	.name = "sifive_x280",
3299	3660	.llvm_name = "sifive-x280",
...	...	@@ -3372,7 +3733,7 @@ pub const cpu = struct {
3372	3733	};
3373	3734	pub const spacemit_a100: CpuModel = .{
3374	3735	.name = "spacemit_a100",
3375		.llvm_name = null,
	3736	.llvm_name = "spacemit-a100",
3376	3737	.features = featureSet(&[_]Feature{
3377	3738	.@"64bit",
3378	3739	.a,
...	...	@@ -3405,10 +3766,10 @@ pub const cpu = struct {
3405	3766	.unaligned_scalar_mem,
3406	3767	.v,
3407	3768	.vxrm_pipeline_flush,
	3769	.xsmtvdotii,
3408	3770	.za64rs,
3409	3771	.zawrs,
3410	3772	.zbc,
3411		.zbkc,
3412	3773	.zcb,
3413	3774	.zcmop,
3414	3775	.zfa,
...	...	@@ -3433,7 +3794,6 @@ pub const cpu = struct {
3433	3794	.zvfbfwma,
3434	3795	.zvfh,
3435	3796	.zvkng,
3436		.zvknha,
3437	3797	.zvksc,
3438	3798	.zvksg,
3439	3799	.zvl1024b,
...	...	@@ -3441,27 +3801,75 @@ pub const cpu = struct {
3441	3801	};
3442	3802	pub const spacemit_x100: CpuModel = .{
3443	3803	.name = "spacemit_x100",
3444		.llvm_name = null,
	3804	.llvm_name = "spacemit-x100",
3445	3805	.features = featureSet(&[_]Feature{
	3806	.@"64bit",
	3807	.a,
	3808	.b,
	3809	.c,
3446	3810	.dlen_factor_2,
	3811	.fusion_add_mem,
	3812	.fusion_logic_imm_reg,
	3813	.fusion_logic_reg_imm,
	3814	.fusion_logic_reg_reg,
	3815	.fusion_mul_add,
	3816	.fusion_shift_bit_extract,
	3817	.i,
	3818	.m,
	3819	.no_default_unroll,
3447	3820	.optimized_nf2_segment_load_store,
3448	3821	.optimized_nf3_segment_load_store,
3449	3822	.optimized_nf4_segment_load_store,
3450		.rva23s64,
	3823	.sha,
3451	3824	.smepmp,
3452	3825	.smnpm,
3453	3826	.smstateen,
	3827	.ssccptr,
	3828	.sscofpmf,
	3829	.sscounterenw,
	3830	.ssnpm,
3454	3831	.sspm,
	3832	.sstc,
	3833	.sstvala,
	3834	.sstvecd,
	3835	.ssu64xl,
	3836	.supm,
	3837	.svade,
	3838	.svbare,
	3839	.svinval,
	3840	.svnapot,
	3841	.svpbmt,
3455	3842	.unaligned_scalar_mem,
	3843	.v,
3456	3844	.vxrm_pipeline_flush,
3457	3845	.xsmtvdot,
	3846	.za64rs,
	3847	.zawrs,
3458	3848	.zbc,
3459		.zbkc,
	3849	.zcb,
	3850	.zcmop,
	3851	.zfa,
3460	3852	.zfh,
	3853	.zic64b,
	3854	.zicbom,
	3855	.zicbop,
	3856	.zicboz,
	3857	.ziccamoa,
	3858	.ziccif,
	3859	.zicclsm,
	3860	.ziccrse,
	3861	.zicntr,
	3862	.zicond,
	3863	.zifencei,
	3864	.zihintntl,
	3865	.zihintpause,
	3866	.zihpm,
	3867	.zimop,
	3868	.zkt,
	3869	.zvbb,
3461	3870	.zvfbfwma,
3462	3871	.zvfh,
3463	3872	.zvkng,
3464		.zvknha,
3465	3873	.zvksc,
3466	3874	.zvksg,
3467	3875	.zvl256b,
...	...	@@ -3476,11 +3884,13 @@ pub const cpu = struct {
3476	3884	.b,
3477	3885	.c,
3478	3886	.dlen_factor_2,
	3887	.enable_select_opt,
3479	3888	.i,
3480	3889	.m,
3481	3890	.optimized_nf2_segment_load_store,
3482	3891	.optimized_nf3_segment_load_store,
3483	3892	.optimized_nf4_segment_load_store,
	3893	.prefer_ascending_load_store,
3484	3894	.ssccptr,
3485	3895	.sscofpmf,
3486	3896	.sscounterenw,
...	...	@@ -3498,7 +3908,6 @@ pub const cpu = struct {
3498	3908	.xsmtvdot,
3499	3909	.za64rs,
3500	3910	.zbc,
3501		.zbkc,
3502	3911	.zfh,
3503	3912	.zic64b,
3504	3913	.zicbom,
...	...	@@ -3652,9 +4061,9 @@ pub const cpu = struct {
3652	4061	.zkn,
3653	4062	}),
3654	4063	};
3655		pub const tt_ascalon_d8: CpuModel = .{
3656		.name = "tt_ascalon_d8",
3657		.llvm_name = "tt-ascalon-d8",
	4064	pub const tt_ascalon_x: CpuModel = .{
	4065	.name = "tt_ascalon_x",
	4066	.llvm_name = "tt-ascalon-x",
3658	4067	.features = featureSet(&[_]Feature{
3659	4068	.@"64bit",
3660	4069	.a,
...	...	@@ -3674,6 +4083,10 @@ pub const cpu = struct {
3674	4083	.optimized_zero_stride_load,
3675	4084	.sha,
3676	4085	.smaia,
	4086	.smmpm,
	4087	.smnpm,
	4088	.smrnmi,
	4089	.smstateen,
3677	4090	.ssaia,
3678	4091	.ssccptr,
3679	4092	.sscofpmf,
...	...	@@ -3715,6 +4128,7 @@ pub const cpu = struct {
3715	4128	.zihintpause,
3716	4129	.zihpm,
3717	4130	.zimop,
	4131	.zkr,
3718	4132	.zkt,
3719	4133	.zvbb,
3720	4134	.zvbc,
...	...	@@ -3730,27 +4144,27 @@ pub const cpu = struct {
3730	4144	.features = featureSet(&[_]Feature{
3731	4145	.@"64bit",
3732	4146	.a,
3733		.add_load_fusion,
3734		.auipc_addi_fusion,
3735		.auipc_load_fusion,
3736	4147	.c,
3737	4148	.d,
3738	4149	.disable_misched_load_clustering,
3739	4150	.disable_postmisched_load_clustering,
3740	4151	.disable_postmisched_store_clustering,
	4152	.fusion_add_load,
	4153	.fusion_auipc_addi,
	4154	.fusion_auipc_load,
	4155	.fusion_lui_addi,
	4156	.fusion_lui_load,
	4157	.fusion_shifted_zextw,
	4158	.fusion_zexth,
	4159	.fusion_zextw,
3741	4160	.i,
3742		.lui_addi_fusion,
3743		.lui_load_fusion,
3744	4161	.m,
3745		.shifted_zextw_fusion,
3746	4162	.ventana_veyron,
3747	4163	.xventanacondops,
3748	4164	.zba,
3749	4165	.zbb,
3750	4166	.zbc,
3751	4167	.zbs,
3752		.zexth_fusion,
3753		.zextw_fusion,
3754	4168	.zicbom,
3755	4169	.zicbop,
3756	4170	.zicboz,
...	...	@@ -3768,11 +4182,13 @@ pub const cpu = struct {
3768	4182	.a,
3769	4183	.b,
3770	4184	.c,
	4185	.fusion_shifted_zextw,
	4186	.fusion_zexth,
	4187	.fusion_zextw,
3771	4188	.i,
3772	4189	.m,
3773	4190	.no_default_unroll,
3774	4191	.sha,
3775		.shifted_zextw_fusion,
3776	4192	.smaia,
3777	4193	.smcsrind,
3778	4194	.smdbltrp,
...	...	@@ -3806,8 +4222,6 @@ pub const cpu = struct {
3806	4222	.zbc,
3807	4223	.zcb,
3808	4224	.zcmop,
3809		.zexth_fusion,
3810		.zextw_fusion,
3811	4225	.zfa,
3812	4226	.zfh,
3813	4227	.zic64b,
...	...	@@ -3841,17 +4255,17 @@ pub const cpu = struct {
3841	4255	.a,
3842	4256	.c,
3843	4257	.d,
	4258	.fusion_shifted_zextw,
	4259	.fusion_zexth,
	4260	.fusion_zextw,
3844	4261	.i,
3845	4262	.m,
3846	4263	.no_default_unroll,
3847		.shifted_zextw_fusion,
3848	4264	.svinval,
3849	4265	.zba,
3850	4266	.zbb,
3851	4267	.zbc,
3852	4268	.zbs,
3853		.zexth_fusion,
3854		.zextw_fusion,
3855	4269	.zicbom,
3856	4270	.zicboz,
3857	4271	.zifencei,
...	...	@@ -3860,4 +4274,129 @@ pub const cpu = struct {
3860	4274	.zksh,
3861	4275	}),
3862	4276	};
	4277	pub const xt_c910v2: CpuModel = .{
	4278	.name = "xt_c910v2",
	4279	.llvm_name = "xt-c910v2",
	4280	.features = featureSet(&[_]Feature{
	4281	.@"64bit",
	4282	.a,
	4283	.b,
	4284	.c,
	4285	.i,
	4286	.m,
	4287	.ssccptr,
	4288	.sscofpmf,
	4289	.sscounterenw,
	4290	.sstc,
	4291	.sstvala,
	4292	.sstvecd,
	4293	.ssu64xl,
	4294	.svade,
	4295	.svbare,
	4296	.svinval,
	4297	.svnapot,
	4298	.svpbmt,
	4299	.unaligned_scalar_mem,
	4300	.use_postra_scheduler,
	4301	.vxrm_pipeline_flush,
	4302	.xtheadba,
	4303	.xtheadbb,
	4304	.xtheadbs,
	4305	.xtheadcmo,
	4306	.xtheadcondmov,
	4307	.xtheadfmemidx,
	4308	.xtheadmac,
	4309	.xtheadmemidx,
	4310	.xtheadmempair,
	4311	.xtheadsync,
	4312	.za64rs,
	4313	.zawrs,
	4314	.zbc,
	4315	.zcb,
	4316	.zcd,
	4317	.zcmop,
	4318	.zfa,
	4319	.zfbfmin,
	4320	.zfh,
	4321	.zic64b,
	4322	.zicbom,
	4323	.zicbop,
	4324	.zicboz,
	4325	.ziccamoa,
	4326	.ziccif,
	4327	.zicclsm,
	4328	.ziccrse,
	4329	.zicntr,
	4330	.zicond,
	4331	.zifencei,
	4332	.zihintntl,
	4333	.zihintpause,
	4334	.zihpm,
	4335	.zimop,
	4336	.zkt,
	4337	}),
	4338	};
	4339	pub const xt_c920v2: CpuModel = .{
	4340	.name = "xt_c920v2",
	4341	.llvm_name = "xt-c920v2",
	4342	.features = featureSet(&[_]Feature{
	4343	.@"64bit",
	4344	.a,
	4345	.b,
	4346	.c,
	4347	.i,
	4348	.m,
	4349	.ssccptr,
	4350	.sscofpmf,
	4351	.sscounterenw,
	4352	.sstc,
	4353	.sstvala,
	4354	.sstvecd,
	4355	.ssu64xl,
	4356	.svade,
	4357	.svbare,
	4358	.svinval,
	4359	.svnapot,
	4360	.svpbmt,
	4361	.unaligned_scalar_mem,
	4362	.use_postra_scheduler,
	4363	.vxrm_pipeline_flush,
	4364	.xtheadba,
	4365	.xtheadbb,
	4366	.xtheadbs,
	4367	.xtheadcmo,
	4368	.xtheadcondmov,
	4369	.xtheadfmemidx,
	4370	.xtheadmac,
	4371	.xtheadmemidx,
	4372	.xtheadmempair,
	4373	.xtheadsync,
	4374	.xtheadvdot,
	4375	.za64rs,
	4376	.zawrs,
	4377	.zbc,
	4378	.zcb,
	4379	.zcd,
	4380	.zcmop,
	4381	.zfa,
	4382	.zfh,
	4383	.zic64b,
	4384	.zicbom,
	4385	.zicbop,
	4386	.zicboz,
	4387	.ziccamoa,
	4388	.ziccif,
	4389	.ziccrse,
	4390	.zicntr,
	4391	.zicond,
	4392	.zifencei,
	4393	.zihintntl,
	4394	.zihintpause,
	4395	.zihpm,
	4396	.zimop,
	4397	.zkt,
	4398	.zvfbfwma,
	4399	.zvfh,
	4400	}),
	4401	};
3863	4402	};

lib/std/Target/sparc.zig+1

...	...	@@ -462,6 +462,7 @@ pub const cpu = struct {
462	462	.name = "leon3",
463	463	.llvm_name = "leon3",
464	464	.features = featureSet(&[_]Feature{
	465	.hasleoncasa,
465	466	.hasumacsmac,
466	467	.leon,
467	468	}),

lib/std/Target/wasm.zig+13-4

...	...	@@ -19,6 +19,7 @@ pub const Feature = enum {
19	19	nontrapping_bulk_memory_len0,
20	20	nontrapping_fptoint,
21	21	reference_types,
	22	relaxed_atomics,
22	23	relaxed_simd,
23	24	sign_ext,
24	25	simd128,
...	...	@@ -60,7 +61,10 @@ pub const all_features = blk: {
60	61	result[@backingInt(Feature.exception_handling)] = .{
61	62	.llvm_name = "exception-handling",
62	63	.description = "Enable Wasm exception handling",
63		.dependencies = featureSet(&[_]Feature{}),
	64	.dependencies = featureSet(&[_]Feature{
	65	.multivalue,
	66	.reference_types,
	67	}),
64	68	};
65	69	result[@backingInt(Feature.extended_const)] = .{
66	70	.llvm_name = "extended-const",
...	...	@@ -75,7 +79,9 @@ pub const all_features = blk: {
75	79	result[@backingInt(Feature.gc)] = .{
76	80	.llvm_name = "gc",
77	81	.description = "Enable wasm gc",
78		.dependencies = featureSet(&[_]Feature{}),
	82	.dependencies = featureSet(&[_]Feature{
	83	.reference_types,
	84	}),
79	85	};
80	86	result[@backingInt(Feature.multimemory)] = .{
81	87	.llvm_name = "multimemory",
...	...	@@ -111,6 +117,11 @@ pub const all_features = blk: {
111	117	.call_indirect_overlong,
112	118	}),
113	119	};
	120	result[@backingInt(Feature.relaxed_atomics)] = .{
	121	.llvm_name = "relaxed-atomics",
	122	.description = "Enable relaxed-atomics proposal",
	123	.dependencies = featureSet(&[_]Feature{}),
	124	};
114	125	result[@backingInt(Feature.relaxed_simd)] = .{
115	126	.llvm_name = "relaxed-simd",
116	127	.description = "Enable relaxed-simd instructions",
...	...	@@ -156,10 +167,8 @@ pub const cpu = struct {
156	167	.fp16,
157	168	.gc,
158	169	.multimemory,
159		.multivalue,
160	170	.mutable_globals,
161	171	.nontrapping_fptoint,
162		.reference_types,
163	172	.relaxed_simd,
164	173	.sign_ext,
165	174	.simd128,

lib/std/Target/x86.zig+533-16

...	...	@@ -20,7 +20,6 @@ pub const Feature = enum {
20	20	amx_fp8,
21	21	amx_int8,
22	22	amx_movrs,
23		amx_tf32,
24	23	amx_tile,
25	24	avx,
26	25	avx10_1,
...	...	@@ -28,6 +27,7 @@ pub const Feature = enum {
28	27	avx2,
29	28	avx512bf16,
30	29	avx512bitalg,
	30	avx512bmm,
31	31	avx512bw,
32	32	avx512cd,
33	33	avx512dq,
...	...	@@ -67,13 +67,17 @@ pub const Feature = enum {
67	67	enqcmd,
68	68	ermsb,
69	69	f16c,
	70	false_deps_bls,
	71	false_deps_compress,
	72	false_deps_expand,
70	73	false_deps_getmant,
71		false_deps_lzcnt_tzcnt,
	74	false_deps_lzcnt,
72	75	false_deps_mulc,
73	76	false_deps_mullq,
74	77	false_deps_perm,
75	78	false_deps_popcnt,
76	79	false_deps_range,
	80	false_deps_tzcnt,
77	81	fast_11bytenop,
78	82	fast_15bytenop,
79	83	fast_7bytenop,
...	...	@@ -105,6 +109,7 @@ pub const Feature = enum {
105	109	idivq_to_divl,
106	110	inline_asm_use_gpr32,
107	111	invpcid,
	112	jmpabs,
108	113	kl,
109	114	lea_sp,
110	115	lea_uses_ag,
...	...	@@ -137,6 +142,7 @@ pub const Feature = enum {
137	142	prefer_legacy_setcc,
138	143	prefer_mask_registers,
139	144	prefer_movmsk_over_vtest,
	145	prefer_ndd_mem,
140	146	prefer_no_gather,
141	147	prefer_no_scatter,
142	148	prefetchi,
...	...	@@ -149,6 +155,31 @@ pub const Feature = enum {
149	155	rdpru,
150	156	rdrnd,
151	157	rdseed,
	158	reserve_edi,
	159	reserve_r10,
	160	reserve_r11,
	161	reserve_r12,
	162	reserve_r13,
	163	reserve_r14,
	164	reserve_r15,
	165	reserve_r16,
	166	reserve_r17,
	167	reserve_r18,
	168	reserve_r19,
	169	reserve_r20,
	170	reserve_r21,
	171	reserve_r22,
	172	reserve_r23,
	173	reserve_r24,
	174	reserve_r25,
	175	reserve_r26,
	176	reserve_r27,
	177	reserve_r28,
	178	reserve_r29,
	179	reserve_r30,
	180	reserve_r31,
	181	reserve_r8,
	182	reserve_r9,
152	183	retpoline,
153	184	retpoline_external_thunk,
154	185	retpoline_indirect_branches,
...	...	@@ -164,6 +195,7 @@ pub const Feature = enum {
164	195	shstk,
165	196	slow_3ops_lea,
166	197	slow_incdec,
	198	slow_indirect_call,
167	199	slow_lea,
168	200	slow_pmaddwd,
169	201	slow_pmulld,
...	...	@@ -313,13 +345,6 @@ pub const all_features = blk: {
313	345	.amx_tile,
314	346	}),
315	347	};
316		result[@backingInt(Feature.amx_tf32)] = .{
317		.llvm_name = "amx-tf32",
318		.description = "Support AMX-TF32 instructions",
319		.dependencies = featureSet(&[_]Feature{
320		.amx_tile,
321		}),
322		};
323	348	result[@backingInt(Feature.amx_tile)] = .{
324	349	.llvm_name = "amx-tile",
325	350	.description = "Support AMX-TILE instructions",
...	...	@@ -377,6 +402,13 @@ pub const all_features = blk: {
377	402	.avx512bw,
378	403	}),
379	404	};
	405	result[@backingInt(Feature.avx512bmm)] = .{
	406	.llvm_name = "avx512bmm",
	407	.description = "Enable AVX512 Bit Matrix Multiply",
	408	.dependencies = featureSet(&[_]Feature{
	409	.avx512bw,
	410	}),
	411	};
380	412	result[@backingInt(Feature.avx512bw)] = .{
381	413	.llvm_name = "avx512bw",
382	414	.description = "Enable AVX-512 Byte and Word Instructions",
...	...	@@ -616,14 +648,29 @@ pub const all_features = blk: {
616	648	.avx,
617	649	}),
618	650	};
	651	result[@backingInt(Feature.false_deps_bls)] = .{
	652	.llvm_name = "false-deps-bls",
	653	.description = "BLSR/BLSI/BLSMSK have a false dependency on dest register",
	654	.dependencies = featureSet(&[_]Feature{}),
	655	};
	656	result[@backingInt(Feature.false_deps_compress)] = .{
	657	.llvm_name = "false-deps-compress",
	658	.description = "V[P]COMPRESSPD/PS/Q/D/W/B with zero-masking has a false dependency on dest register",
	659	.dependencies = featureSet(&[_]Feature{}),
	660	};
	661	result[@backingInt(Feature.false_deps_expand)] = .{
	662	.llvm_name = "false-deps-expand",
	663	.description = "V[P]EXPANDPD/PS/Q/D/W/B with zero-masking has a false dependency on dest register",
	664	.dependencies = featureSet(&[_]Feature{}),
	665	};
619	666	result[@backingInt(Feature.false_deps_getmant)] = .{
620	667	.llvm_name = "false-deps-getmant",
621	668	.description = "VGETMANTSS/SD/SH and VGETMANDPS/PD(memory version) has a false dependency on dest register",
622	669	.dependencies = featureSet(&[_]Feature{}),
623	670	};
624		result[@backingInt(Feature.false_deps_lzcnt_tzcnt)] = .{
625		.llvm_name = "false-deps-lzcnt-tzcnt",
626		.description = "LZCNT/TZCNT have a false dependency on dest register",
	671	result[@backingInt(Feature.false_deps_lzcnt)] = .{
	672	.llvm_name = "false-deps-lzcnt",
	673	.description = "LZCNT has a false dependency on dest register",
627	674	.dependencies = featureSet(&[_]Feature{}),
628	675	};
629	676	result[@backingInt(Feature.false_deps_mulc)] = .{
...	...	@@ -651,6 +698,11 @@ pub const all_features = blk: {
651	698	.description = "VRANGEPD/PS/SD/SS has a false dependency on dest register",
652	699	.dependencies = featureSet(&[_]Feature{}),
653	700	};
	701	result[@backingInt(Feature.false_deps_tzcnt)] = .{
	702	.llvm_name = "false-deps-tzcnt",
	703	.description = "TZCNT has a false dependency on dest register",
	704	.dependencies = featureSet(&[_]Feature{}),
	705	};
654	706	result[@backingInt(Feature.fast_11bytenop)] = .{
655	707	.llvm_name = "fast-11bytenop",
656	708	.description = "Target can quickly decode up to 11 byte NOPs",
...	...	@@ -813,6 +865,11 @@ pub const all_features = blk: {
813	865	.description = "Invalidate Process-Context Identifier",
814	866	.dependencies = featureSet(&[_]Feature{}),
815	867	};
	868	result[@backingInt(Feature.jmpabs)] = .{
	869	.llvm_name = "jmpabs",
	870	.description = "Support 64-bit absolute JMP",
	871	.dependencies = featureSet(&[_]Feature{}),
	872	};
816	873	result[@backingInt(Feature.kl)] = .{
817	874	.llvm_name = "kl",
818	875	.description = "Support Key Locker kl Instructions",
...	...	@@ -977,6 +1034,11 @@ pub const all_features = blk: {
977	1034	.description = "Prefer movmsk over vtest instruction",
978	1035	.dependencies = featureSet(&[_]Feature{}),
979	1036	};
	1037	result[@backingInt(Feature.prefer_ndd_mem)] = .{
	1038	.llvm_name = "prefer-ndd-mem",
	1039	.description = "Prefer NDD memory addressing",
	1040	.dependencies = featureSet(&[_]Feature{}),
	1041	};
980	1042	result[@backingInt(Feature.prefer_no_gather)] = .{
981	1043	.llvm_name = "prefer-no-gather",
982	1044	.description = "Prefer no gather instructions",
...	...	@@ -1037,6 +1099,131 @@ pub const all_features = blk: {
1037	1099	.description = "Support RDSEED instruction",
1038	1100	.dependencies = featureSet(&[_]Feature{}),
1039	1101	};
	1102	result[@backingInt(Feature.reserve_edi)] = .{
	1103	.llvm_name = "reserve-edi",
	1104	.description = "Reserve EDI, making it unavailable as a GPR",
	1105	.dependencies = featureSet(&[_]Feature{}),
	1106	};
	1107	result[@backingInt(Feature.reserve_r10)] = .{
	1108	.llvm_name = "reserve-r10",
	1109	.description = "Reserve R10, making it unavailable as a GPR",
	1110	.dependencies = featureSet(&[_]Feature{}),
	1111	};
	1112	result[@backingInt(Feature.reserve_r11)] = .{
	1113	.llvm_name = "reserve-r11",
	1114	.description = "Reserve R11, making it unavailable as a GPR",
	1115	.dependencies = featureSet(&[_]Feature{}),
	1116	};
	1117	result[@backingInt(Feature.reserve_r12)] = .{
	1118	.llvm_name = "reserve-r12",
	1119	.description = "Reserve R12, making it unavailable as a GPR",
	1120	.dependencies = featureSet(&[_]Feature{}),
	1121	};
	1122	result[@backingInt(Feature.reserve_r13)] = .{
	1123	.llvm_name = "reserve-r13",
	1124	.description = "Reserve R13, making it unavailable as a GPR",
	1125	.dependencies = featureSet(&[_]Feature{}),
	1126	};
	1127	result[@backingInt(Feature.reserve_r14)] = .{
	1128	.llvm_name = "reserve-r14",
	1129	.description = "Reserve R14, making it unavailable as a GPR",
	1130	.dependencies = featureSet(&[_]Feature{}),
	1131	};
	1132	result[@backingInt(Feature.reserve_r15)] = .{
	1133	.llvm_name = "reserve-r15",
	1134	.description = "Reserve R15, making it unavailable as a GPR",
	1135	.dependencies = featureSet(&[_]Feature{}),
	1136	};
	1137	result[@backingInt(Feature.reserve_r16)] = .{
	1138	.llvm_name = "reserve-r16",
	1139	.description = "Reserve R16, making it unavailable as a GPR",
	1140	.dependencies = featureSet(&[_]Feature{}),
	1141	};
	1142	result[@backingInt(Feature.reserve_r17)] = .{
	1143	.llvm_name = "reserve-r17",
	1144	.description = "Reserve R17, making it unavailable as a GPR",
	1145	.dependencies = featureSet(&[_]Feature{}),
	1146	};
	1147	result[@backingInt(Feature.reserve_r18)] = .{
	1148	.llvm_name = "reserve-r18",
	1149	.description = "Reserve R18, making it unavailable as a GPR",
	1150	.dependencies = featureSet(&[_]Feature{}),
	1151	};
	1152	result[@backingInt(Feature.reserve_r19)] = .{
	1153	.llvm_name = "reserve-r19",
	1154	.description = "Reserve R19, making it unavailable as a GPR",
	1155	.dependencies = featureSet(&[_]Feature{}),
	1156	};
	1157	result[@backingInt(Feature.reserve_r20)] = .{
	1158	.llvm_name = "reserve-r20",
	1159	.description = "Reserve R20, making it unavailable as a GPR",
	1160	.dependencies = featureSet(&[_]Feature{}),
	1161	};
	1162	result[@backingInt(Feature.reserve_r21)] = .{
	1163	.llvm_name = "reserve-r21",
	1164	.description = "Reserve R21, making it unavailable as a GPR",
	1165	.dependencies = featureSet(&[_]Feature{}),
	1166	};
	1167	result[@backingInt(Feature.reserve_r22)] = .{
	1168	.llvm_name = "reserve-r22",
	1169	.description = "Reserve R22, making it unavailable as a GPR",
	1170	.dependencies = featureSet(&[_]Feature{}),
	1171	};
	1172	result[@backingInt(Feature.reserve_r23)] = .{
	1173	.llvm_name = "reserve-r23",
	1174	.description = "Reserve R23, making it unavailable as a GPR",
	1175	.dependencies = featureSet(&[_]Feature{}),
	1176	};
	1177	result[@backingInt(Feature.reserve_r24)] = .{
	1178	.llvm_name = "reserve-r24",
	1179	.description = "Reserve R24, making it unavailable as a GPR",
	1180	.dependencies = featureSet(&[_]Feature{}),
	1181	};
	1182	result[@backingInt(Feature.reserve_r25)] = .{
	1183	.llvm_name = "reserve-r25",
	1184	.description = "Reserve R25, making it unavailable as a GPR",
	1185	.dependencies = featureSet(&[_]Feature{}),
	1186	};
	1187	result[@backingInt(Feature.reserve_r26)] = .{
	1188	.llvm_name = "reserve-r26",
	1189	.description = "Reserve R26, making it unavailable as a GPR",
	1190	.dependencies = featureSet(&[_]Feature{}),
	1191	};
	1192	result[@backingInt(Feature.reserve_r27)] = .{
	1193	.llvm_name = "reserve-r27",
	1194	.description = "Reserve R27, making it unavailable as a GPR",
	1195	.dependencies = featureSet(&[_]Feature{}),
	1196	};
	1197	result[@backingInt(Feature.reserve_r28)] = .{
	1198	.llvm_name = "reserve-r28",
	1199	.description = "Reserve R28, making it unavailable as a GPR",
	1200	.dependencies = featureSet(&[_]Feature{}),
	1201	};
	1202	result[@backingInt(Feature.reserve_r29)] = .{
	1203	.llvm_name = "reserve-r29",
	1204	.description = "Reserve R29, making it unavailable as a GPR",
	1205	.dependencies = featureSet(&[_]Feature{}),
	1206	};
	1207	result[@backingInt(Feature.reserve_r30)] = .{
	1208	.llvm_name = "reserve-r30",
	1209	.description = "Reserve R30, making it unavailable as a GPR",
	1210	.dependencies = featureSet(&[_]Feature{}),
	1211	};
	1212	result[@backingInt(Feature.reserve_r31)] = .{
	1213	.llvm_name = "reserve-r31",
	1214	.description = "Reserve R31, making it unavailable as a GPR",
	1215	.dependencies = featureSet(&[_]Feature{}),
	1216	};
	1217	result[@backingInt(Feature.reserve_r8)] = .{
	1218	.llvm_name = "reserve-r8",
	1219	.description = "Reserve R8, making it unavailable as a GPR",
	1220	.dependencies = featureSet(&[_]Feature{}),
	1221	};
	1222	result[@backingInt(Feature.reserve_r9)] = .{
	1223	.llvm_name = "reserve-r9",
	1224	.description = "Reserve R9, making it unavailable as a GPR",
	1225	.dependencies = featureSet(&[_]Feature{}),
	1226	};
1040	1227	result[@backingInt(Feature.retpoline)] = .{
1041	1228	.llvm_name = "retpoline",
1042	1229	.description = "Remove speculation of indirect branches from the generated code, either by avoiding them entirely or lowering them with a speculation blocking construct",
...	...	@@ -1123,6 +1310,11 @@ pub const all_features = blk: {
1123	1310	.description = "INC and DEC instructions are slower than ADD and SUB",
1124	1311	.dependencies = featureSet(&[_]Feature{}),
1125	1312	};
	1313	result[@backingInt(Feature.slow_indirect_call)] = .{
	1314	.llvm_name = "slow-indirect-call",
	1315	.description = "Indirect calls from memory are slow",
	1316	.dependencies = featureSet(&[_]Feature{}),
	1317	};
1126	1318	result[@backingInt(Feature.slow_lea)] = .{
1127	1319	.llvm_name = "slow-lea",
1128	1320	.description = "LEA instruction with certain arguments is slow",
...	...	@@ -2039,8 +2231,9 @@ pub const cpu = struct {
2039	2231	.cx16,
2040	2232	.ermsb,
2041	2233	.f16c,
2042		.false_deps_lzcnt_tzcnt,
	2234	.false_deps_lzcnt,
2043	2235	.false_deps_popcnt,
	2236	.false_deps_tzcnt,
2044	2237	.fast_15bytenop,
2045	2238	.fast_scalar_fsqrt,
2046	2239	.fast_shld_rotate,
...	...	@@ -2159,6 +2352,239 @@ pub const cpu = struct {
2159	2352	.x87,
2160	2353	}),
2161	2354	};
	2355	pub const c86_4g_m4: CpuModel = .{
	2356	.name = "c86_4g_m4",
	2357	.llvm_name = "c86-4g-m4",
	2358	.features = featureSet(&[_]Feature{
	2359	.@"64bit",
	2360	.adx,
	2361	.aes,
	2362	.allow_light_256_bit,
	2363	.avx2,
	2364	.bmi,
	2365	.bmi2,
	2366	.clflushopt,
	2367	.clzero,
	2368	.cmov,
	2369	.cx16,
	2370	.f16c,
	2371	.fast_15bytenop,
	2372	.fast_bextr,
	2373	.fast_imm16,
	2374	.fast_lzcnt,
	2375	.fast_movbe,
	2376	.fast_scalar_fsqrt,
	2377	.fast_scalar_shift_masks,
	2378	.fast_variable_perlane_shuffle,
	2379	.fast_vector_fsqrt,
	2380	.fma,
	2381	.fsgsbase,
	2382	.fxsr,
	2383	.idivq_to_divl,
	2384	.lzcnt,
	2385	.mmx,
	2386	.movbe,
	2387	.mwaitx,
	2388	.nopl,
	2389	.pclmul,
	2390	.popcnt,
	2391	.prfchw,
	2392	.rdrnd,
	2393	.rdseed,
	2394	.sahf,
	2395	.sbb_dep_breaking,
	2396	.sha,
	2397	.slow_shld,
	2398	.sse4a,
	2399	.vzeroupper,
	2400	.x87,
	2401	.xsavec,
	2402	.xsaveopt,
	2403	.xsaves,
	2404	}),
	2405	};
	2406	pub const c86_4g_m6: CpuModel = .{
	2407	.name = "c86_4g_m6",
	2408	.llvm_name = "c86-4g-m6",
	2409	.features = featureSet(&[_]Feature{
	2410	.@"64bit",
	2411	.adx,
	2412	.aes,
	2413	.allow_light_256_bit,
	2414	.avx2,
	2415	.bmi,
	2416	.bmi2,
	2417	.clflushopt,
	2418	.clzero,
	2419	.cmov,
	2420	.cx16,
	2421	.f16c,
	2422	.fast_15bytenop,
	2423	.fast_bextr,
	2424	.fast_imm16,
	2425	.fast_lzcnt,
	2426	.fast_movbe,
	2427	.fast_scalar_fsqrt,
	2428	.fast_scalar_shift_masks,
	2429	.fast_variable_perlane_shuffle,
	2430	.fast_vector_fsqrt,
	2431	.fma,
	2432	.fsgsbase,
	2433	.fxsr,
	2434	.idivq_to_divl,
	2435	.lzcnt,
	2436	.mmx,
	2437	.movbe,
	2438	.mwaitx,
	2439	.nopl,
	2440	.pclmul,
	2441	.popcnt,
	2442	.prfchw,
	2443	.rdrnd,
	2444	.rdseed,
	2445	.sahf,
	2446	.sbb_dep_breaking,
	2447	.sha,
	2448	.slow_shld,
	2449	.sse4a,
	2450	.vzeroupper,
	2451	.x87,
	2452	.xsavec,
	2453	.xsaveopt,
	2454	.xsaves,
	2455	}),
	2456	};
	2457	pub const c86_4g_m7: CpuModel = .{
	2458	.name = "c86_4g_m7",
	2459	.llvm_name = "c86-4g-m7",
	2460	.features = featureSet(&[_]Feature{
	2461	.@"64bit",
	2462	.adx,
	2463	.allow_light_256_bit,
	2464	.avx512bf16,
	2465	.avx512bitalg,
	2466	.avx512cd,
	2467	.avx512dq,
	2468	.avx512ifma,
	2469	.avx512vbmi,
	2470	.avx512vbmi2,
	2471	.avx512vl,
	2472	.avx512vnni,
	2473	.avx512vpopcntdq,
	2474	.bmi,
	2475	.bmi2,
	2476	.branchfusion,
	2477	.clflushopt,
	2478	.clwb,
	2479	.clzero,
	2480	.cmov,
	2481	.cx16,
	2482	.fast_15bytenop,
	2483	.fast_bextr,
	2484	.fast_imm16,
	2485	.fast_lzcnt,
	2486	.fast_movbe,
	2487	.fast_scalar_fsqrt,
	2488	.fast_scalar_shift_masks,
	2489	.fast_variable_perlane_shuffle,
	2490	.fast_vector_fsqrt,
	2491	.fsgsbase,
	2492	.fxsr,
	2493	.gfni,
	2494	.idivq_to_divl,
	2495	.lzcnt,
	2496	.mmx,
	2497	.movbe,
	2498	.mwaitx,
	2499	.nopl,
	2500	.popcnt,
	2501	.prefer_256_bit,
	2502	.prefer_no_gather,
	2503	.prefer_no_scatter,
	2504	.prfchw,
	2505	.rdrnd,
	2506	.rdseed,
	2507	.sahf,
	2508	.sbb_dep_breaking,
	2509	.sha,
	2510	.slow_shld,
	2511	.sse4a,
	2512	.vaes,
	2513	.vpclmulqdq,
	2514	.vzeroupper,
	2515	.wbnoinvd,
	2516	.x87,
	2517	.xsavec,
	2518	.xsaveopt,
	2519	.xsaves,
	2520	}),
	2521	};
	2522	pub const c86_4g_m8: CpuModel = .{
	2523	.name = "c86_4g_m8",
	2524	.llvm_name = "c86-4g-m8",
	2525	.features = featureSet(&[_]Feature{
	2526	.@"64bit",
	2527	.adx,
	2528	.allow_light_256_bit,
	2529	.avx512bf16,
	2530	.avx512bitalg,
	2531	.avx512cd,
	2532	.avx512dq,
	2533	.avx512ifma,
	2534	.avx512vbmi,
	2535	.avx512vbmi2,
	2536	.avx512vl,
	2537	.avx512vnni,
	2538	.avx512vpopcntdq,
	2539	.bmi,
	2540	.bmi2,
	2541	.branchfusion,
	2542	.clflushopt,
	2543	.clwb,
	2544	.clzero,
	2545	.cmov,
	2546	.cx16,
	2547	.fast_15bytenop,
	2548	.fast_bextr,
	2549	.fast_imm16,
	2550	.fast_lzcnt,
	2551	.fast_movbe,
	2552	.fast_scalar_fsqrt,
	2553	.fast_scalar_shift_masks,
	2554	.fast_variable_perlane_shuffle,
	2555	.fast_vector_fsqrt,
	2556	.fsgsbase,
	2557	.fxsr,
	2558	.gfni,
	2559	.idivq_to_divl,
	2560	.lzcnt,
	2561	.mmx,
	2562	.movbe,
	2563	.mwaitx,
	2564	.nopl,
	2565	.popcnt,
	2566	.prefer_256_bit,
	2567	.prefer_no_gather,
	2568	.prefer_no_scatter,
	2569	.prfchw,
	2570	.rdrnd,
	2571	.rdseed,
	2572	.sahf,
	2573	.sbb_dep_breaking,
	2574	.sha,
	2575	.shstk,
	2576	.slow_shld,
	2577	.sse4a,
	2578	.vaes,
	2579	.vpclmulqdq,
	2580	.vzeroupper,
	2581	.wbnoinvd,
	2582	.x87,
	2583	.xsavec,
	2584	.xsaveopt,
	2585	.xsaves,
	2586	}),
	2587	};
2162	2588	pub const cannonlake: CpuModel = .{
2163	2589	.name = "cannonlake",
2164	2590	.llvm_name = "cannonlake",
...	...	@@ -2467,7 +2893,6 @@ pub const cpu = struct {
2467	2893	.amx_fp8,
2468	2894	.amx_int8,
2469	2895	.amx_movrs,
2470		.amx_tf32,
2471	2896	.avx10_2,
2472	2897	.avxifma,
2473	2898	.avxneconvert,
...	...	@@ -2505,6 +2930,7 @@ pub const cpu = struct {
2505	2930	.gfni,
2506	2931	.idivq_to_divl,
2507	2932	.invpcid,
	2933	.jmpabs,
2508	2934	.lzcnt,
2509	2935	.macrofusion,
2510	2936	.mmx,
...	...	@@ -2523,6 +2949,7 @@ pub const cpu = struct {
2523	2949	.popcnt,
2524	2950	.ppx,
2525	2951	.prefer_256_bit,
	2952	.prefer_legacy_setcc,
2526	2953	.prefetchi,
2527	2954	.prfchw,
2528	2955	.ptwrite,
...	...	@@ -3056,8 +3483,9 @@ pub const cpu = struct {
3056	3483	.cx16,
3057	3484	.ermsb,
3058	3485	.f16c,
3059		.false_deps_lzcnt_tzcnt,
	3486	.false_deps_lzcnt,
3060	3487	.false_deps_popcnt,
	3488	.false_deps_tzcnt,
3061	3489	.fast_15bytenop,
3062	3490	.fast_scalar_fsqrt,
3063	3491	.fast_shld_rotate,
...	...	@@ -3679,6 +4107,7 @@ pub const cpu = struct {
3679	4107	.hreset,
3680	4108	.idivq_to_divl,
3681	4109	.invpcid,
	4110	.jmpabs,
3682	4111	.lzcnt,
3683	4112	.macrofusion,
3684	4113	.mmx,
...	...	@@ -3696,6 +4125,7 @@ pub const cpu = struct {
3696	4125	.pku,
3697	4126	.popcnt,
3698	4127	.ppx,
	4128	.prefer_legacy_setcc,
3699	4129	.prefer_movmsk_over_vtest,
3700	4130	.prefetchi,
3701	4131	.prfchw,
...	...	@@ -4813,8 +5243,9 @@ pub const cpu = struct {
4813	5243	.cmov,
4814	5244	.cx16,
4815	5245	.f16c,
4816		.false_deps_lzcnt_tzcnt,
	5246	.false_deps_lzcnt,
4817	5247	.false_deps_popcnt,
	5248	.false_deps_tzcnt,
4818	5249	.fast_15bytenop,
4819	5250	.fast_scalar_fsqrt,
4820	5251	.fast_shld_rotate,
...	...	@@ -5086,6 +5517,8 @@ pub const cpu = struct {
5086	5517	.clzero,
5087	5518	.cmov,
5088	5519	.cx16,
	5520	.false_deps_compress,
	5521	.false_deps_expand,
5089	5522	.fast_15bytenop,
5090	5523	.fast_bextr,
5091	5524	.fast_dpwssd,
...	...	@@ -5159,6 +5592,10 @@ pub const cpu = struct {
5159	5592	.clzero,
5160	5593	.cmov,
5161	5594	.cx16,
	5595	.false_deps_bls,
	5596	.false_deps_compress,
	5597	.false_deps_expand,
	5598	.false_deps_tzcnt,
5162	5599	.fast_15bytenop,
5163	5600	.fast_bextr,
5164	5601	.fast_dpwssd,
...	...	@@ -5195,6 +5632,7 @@ pub const cpu = struct {
5195	5632	.sbb_dep_breaking,
5196	5633	.sha,
5197	5634	.shstk,
	5635	.slow_indirect_call,
5198	5636	.smap,
5199	5637	.smep,
5200	5638	.sse4a,
...	...	@@ -5208,4 +5646,83 @@ pub const cpu = struct {
5208	5646	.xsaves,
5209	5647	}),
5210	5648	};
	5649	pub const znver6: CpuModel = .{
	5650	.name = "znver6",
	5651	.llvm_name = "znver6",
	5652	.features = featureSet(&[_]Feature{
	5653	.@"64bit",
	5654	.adx,
	5655	.allow_light_256_bit,
	5656	.avx512bf16,
	5657	.avx512bitalg,
	5658	.avx512bmm,
	5659	.avx512cd,
	5660	.avx512dq,
	5661	.avx512fp16,
	5662	.avx512ifma,
	5663	.avx512vbmi,
	5664	.avx512vbmi2,
	5665	.avx512vl,
	5666	.avx512vnni,
	5667	.avx512vp2intersect,
	5668	.avx512vpopcntdq,
	5669	.avxifma,
	5670	.avxneconvert,
	5671	.avxvnni,
	5672	.avxvnniint8,
	5673	.bmi,
	5674	.bmi2,
	5675	.branchfusion,
	5676	.clflushopt,
	5677	.clwb,
	5678	.clzero,
	5679	.cmov,
	5680	.cx16,
	5681	.fast_15bytenop,
	5682	.fast_bextr,
	5683	.fast_dpwssd,
	5684	.fast_imm16,
	5685	.fast_lzcnt,
	5686	.fast_movbe,
	5687	.fast_scalar_fsqrt,
	5688	.fast_scalar_shift_masks,
	5689	.fast_variable_perlane_shuffle,
	5690	.fast_vector_fsqrt,
	5691	.fsgsbase,
	5692	.fsrm,
	5693	.fxsr,
	5694	.gfni,
	5695	.idivq_to_divl,
	5696	.invpcid,
	5697	.lzcnt,
	5698	.macrofusion,
	5699	.mmx,
	5700	.movbe,
	5701	.movdir64b,
	5702	.movdiri,
	5703	.mwaitx,
	5704	.nopl,
	5705	.pku,
	5706	.popcnt,
	5707	.prefetchi,
	5708	.prfchw,
	5709	.rdpid,
	5710	.rdpru,
	5711	.rdrnd,
	5712	.rdseed,
	5713	.sahf,
	5714	.sbb_dep_breaking,
	5715	.sha,
	5716	.shstk,
	5717	.sse4a,
	5718	.vaes,
	5719	.vpclmulqdq,
	5720	.vzeroupper,
	5721	.wbnoinvd,
	5722	.x87,
	5723	.xsavec,
	5724	.xsaveopt,
	5725	.xsaves,
	5726	}),
	5727	};
5211	5728	};

lib/std/Target/xtensa.zig+35

...	...	@@ -13,6 +13,7 @@ pub const Feature = enum {
13	13	density,
14	14	dfpaccel,
15	15	div32,
	16	esp32s2ops,
16	17	exception,
17	18	extendedl32r,
18	19	forced_atomics,
...	...	@@ -93,6 +94,11 @@ pub const all_features = blk: {
93	94	.description = "Enable Xtensa Div32 option",
94	95	.dependencies = featureSet(&[_]Feature{}),
95	96	};
	97	result[@backingInt(Feature.esp32s2ops)] = .{
	98	.llvm_name = "esp32s2ops",
	99	.description = "Support Xtensa esp32-s2 ISA extension",
	100	.dependencies = featureSet(&[_]Feature{}),
	101	};
96	102	result[@backingInt(Feature.exception)] = .{
97	103	.llvm_name = "exception",
98	104	.description = "Enable Xtensa Exception option",
...	...	@@ -291,6 +297,35 @@ pub const cpu = struct {
291	297	.windowed,
292	298	}),
293	299	};
	300	pub const esp32s2: CpuModel = .{
	301	.name = "esp32s2",
	302	.llvm_name = "esp32s2",
	303	.features = featureSet(&[_]Feature{
	304	.clamps,
	305	.coprocessor,
	306	.dcache,
	307	.debug,
	308	.density,
	309	.div32,
	310	.esp32s2ops,
	311	.exception,
	312	.highpriinterrupts_level7,
	313	.interrupt,
	314	.minmax,
	315	.miscsr,
	316	.mul16,
	317	.mul32,
	318	.mul32high,
	319	.nsa,
	320	.prid,
	321	.regprotect,
	322	.rvector,
	323	.sext,
	324	.threadptr,
	325	.timers3,
	326	.windowed,
	327	}),
	328	};
294	329	pub const esp8266: CpuModel = .{
295	330	.name = "esp8266",
296	331	.llvm_name = "esp8266",

src/codegen/x86_64/CodeGen.zig+19-18

...	...	@@ -67564,7 +67564,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
67564	67564	.{ ._, ._, .sub, .dst0b, .sia(32, .src0, .sub_bit_size), ._, ._ },
67565	67565	} },
67566	67566	}, .{
67567		.required_features = .{ .false_deps_lzcnt_tzcnt, .lzcnt, null, null },
	67567	.required_features = .{ .false_deps_lzcnt, .lzcnt, null, null },
67568	67568	.src_constraints = .{ .{ .exact_int = 16 }, .any, .any },
67569	67569	.patterns = &.{
67570	67570	.{ .src = .{ .to_mut_gpr, .none, .none } },
...	...	@@ -67600,7 +67600,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
67600	67600	.{ ._, ._, .sub, .dst0b, .sia(16, .src0, .sub_bit_size), ._, ._ },
67601	67601	} },
67602	67602	}, .{
67603		.required_features = .{ .false_deps_lzcnt_tzcnt, .lzcnt, null, null },
	67603	.required_features = .{ .false_deps_lzcnt, .lzcnt, null, null },
67604	67604	.src_constraints = .{ .{ .unsigned_int = .word }, .any, .any },
67605	67605	.patterns = &.{
67606	67606	.{ .src = .{ .to_mut_gpr, .none, .none } },
...	...	@@ -67625,7 +67625,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
67625	67625	.{ ._, ._, .sub, .dst0b, .sia(16, .src0, .sub_bit_size), ._, ._ },
67626	67626	} },
67627	67627	}, .{
67628		.required_features = .{ .false_deps_lzcnt_tzcnt, .lzcnt, null, null },
	67628	.required_features = .{ .false_deps_lzcnt, .lzcnt, null, null },
67629	67629	.src_constraints = .{ .{ .exact_int = 32 }, .any, .any },
67630	67630	.patterns = &.{
67631	67631	.{ .src = .{ .to_mut_gpr, .none, .none } },
...	...	@@ -67661,7 +67661,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
67661	67661	.{ ._, ._, .sub, .dst0b, .sia(32, .src0, .sub_bit_size), ._, ._ },
67662	67662	} },
67663	67663	}, .{
67664		.required_features = .{ .false_deps_lzcnt_tzcnt, .lzcnt, null, null },
	67664	.required_features = .{ .false_deps_lzcnt, .lzcnt, null, null },
67665	67665	.src_constraints = .{ .{ .unsigned_int = .dword }, .any, .any },
67666	67666	.patterns = &.{
67667	67667	.{ .src = .{ .to_mut_gpr, .none, .none } },
...	...	@@ -67686,7 +67686,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
67686	67686	.{ ._, ._, .sub, .dst0b, .sia(32, .src0, .sub_bit_size), ._, ._ },
67687	67687	} },
67688	67688	}, .{
67689		.required_features = .{ .@"64bit", .false_deps_lzcnt_tzcnt, .lzcnt, null },
	67689	.required_features = .{ .@"64bit", .false_deps_lzcnt, .lzcnt, null },
67690	67690	.src_constraints = .{ .{ .exact_int = 64 }, .any, .any },
67691	67691	.patterns = &.{
67692	67692	.{ .src = .{ .to_mut_gpr, .none, .none } },
...	...	@@ -67724,7 +67724,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
67724	67724	.{ ._, ._, .sub, .dst0b, .sia(64, .src0, .sub_bit_size), ._, ._ },
67725	67725	} },
67726	67726	}, .{
67727		.required_features = .{ .@"64bit", .false_deps_lzcnt_tzcnt, .lzcnt, null },
	67727	.required_features = .{ .@"64bit", .false_deps_lzcnt, .lzcnt, null },
67728	67728	.src_constraints = .{ .{ .unsigned_int = .qword }, .any, .any },
67729	67729	.patterns = &.{
67730	67730	.{ .src = .{ .to_mut_gpr, .none, .none } },
...	...	@@ -68581,7 +68581,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
68581	68581	.{ ._, ._, .sub, .dst0b, .tmp0b, ._, ._ },
68582	68582	} },
68583	68583	}, .{
68584		.required_features = .{ .@"64bit", .false_deps_lzcnt_tzcnt, .lzcnt, null },
	68584	.required_features = .{ .@"64bit", .false_deps_lzcnt, .lzcnt, null },
68585	68585	.src_constraints = .{ .{ .unsigned_or_exact_remainder_int = .{ .of = .xword, .is = .qword } }, .any, .any },
68586	68586	.patterns = &.{
68587	68587	.{ .src = .{ .to_mem, .none, .none } },
...	...	@@ -68709,7 +68709,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
68709	68709	.{ ._, ._, .neg, .dst0d, ._, ._, ._ },
68710	68710	} },
68711	68711	}, .{
68712		.required_features = .{ .@"64bit", .false_deps_lzcnt_tzcnt, .lzcnt, null },
	68712	.required_features = .{ .@"64bit", .false_deps_lzcnt, .lzcnt, null },
68713	68713	.src_constraints = .{ .{ .unsigned_or_exact_remainder_int = .{ .of = .xword, .is = .xword } }, .any, .any },
68714	68714	.patterns = &.{
68715	68715	.{ .src = .{ .to_mem, .none, .none } },
...	...	@@ -68837,7 +68837,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
68837	68837	.{ ._, ._, .neg, .dst0d, ._, ._, ._ },
68838	68838	} },
68839	68839	}, .{
68840		.required_features = .{ .@"64bit", .false_deps_lzcnt_tzcnt, .lzcnt, null },
	68840	.required_features = .{ .@"64bit", .false_deps_lzcnt, .lzcnt, null },
68841	68841	.src_constraints = .{ .{ .remainder_int = .{ .of = .xword, .is = .qword } }, .any, .any },
68842	68842	.patterns = &.{
68843	68843	.{ .src = .{ .to_mem, .none, .none } },
...	...	@@ -68940,7 +68940,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
68940	68940	.{ ._, ._, .neg, .dst0d, ._, ._, ._ },
68941	68941	} },
68942	68942	}, .{
68943		.required_features = .{ .@"64bit", .false_deps_lzcnt_tzcnt, .lzcnt, null },
	68943	.required_features = .{ .@"64bit", .false_deps_lzcnt, .lzcnt, null },
68944	68944	.src_constraints = .{ .{ .remainder_int = .{ .of = .xword, .is = .xword } }, .any, .any },
68945	68945	.patterns = &.{
68946	68946	.{ .src = .{ .to_mem, .none, .none } },
...	...	@@ -70096,7 +70096,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
70096	70096	.{ ._, ._nz, .j, .@"0b", ._, ._, ._ },
70097	70097	} },
70098	70098	}, .{
70099		.required_features = .{ .@"64bit", .false_deps_lzcnt_tzcnt, .lzcnt, null },
	70099	.required_features = .{ .@"64bit", .false_deps_lzcnt, .lzcnt, null },
70100	70100	.src_constraints = .{ .{ .scalar_remainder_int = .{ .of = .xword, .is = .qword } }, .any, .any },
70101	70101	.dst_constraints = .{ .{ .scalar_int_is = .byte }, .any },
70102	70102	.patterns = &.{
...	...	@@ -70220,7 +70220,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
70220	70220	.{ ._, ._nc, .j, .@"0b", ._, ._, ._ },
70221	70221	} },
70222	70222	}, .{
70223		.required_features = .{ .@"64bit", .false_deps_lzcnt_tzcnt, .lzcnt, null },
	70223	.required_features = .{ .@"64bit", .false_deps_lzcnt, .lzcnt, null },
70224	70224	.src_constraints = .{ .{ .scalar_remainder_int = .{ .of = .xword, .is = .xword } }, .any, .any },
70225	70225	.dst_constraints = .{ .{ .scalar_int_is = .byte }, .any },
70226	70226	.patterns = &.{
...	...	@@ -70344,7 +70344,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
70344	70344	.{ ._, ._nc, .j, .@"0b", ._, ._, ._ },
70345	70345	} },
70346	70346	}, .{
70347		.required_features = .{ .@"64bit", .false_deps_lzcnt_tzcnt, .lzcnt, null },
	70347	.required_features = .{ .@"64bit", .false_deps_lzcnt, .lzcnt, null },
70348	70348	.src_constraints = .{ .{ .scalar_remainder_int = .{ .of = .xword, .is = .qword } }, .any, .any },
70349	70349	.dst_constraints = .{ .{ .scalar_int_is = .word }, .any },
70350	70350	.patterns = &.{
...	...	@@ -70468,7 +70468,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
70468	70468	.{ ._, ._nc, .j, .@"0b", ._, ._, ._ },
70469	70469	} },
70470	70470	}, .{
70471		.required_features = .{ .@"64bit", .false_deps_lzcnt_tzcnt, .lzcnt, null },
	70471	.required_features = .{ .@"64bit", .false_deps_lzcnt, .lzcnt, null },
70472	70472	.src_constraints = .{ .{ .scalar_remainder_int = .{ .of = .xword, .is = .xword } }, .any, .any },
70473	70473	.dst_constraints = .{ .{ .scalar_int_is = .word }, .any },
70474	70474	.patterns = &.{
...	...	@@ -70653,7 +70653,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
70653	70653	.{ ._, ._, .tzcnt, .dst0w, .src0w, ._, ._ },
70654	70654	} },
70655	70655	}, .{
70656		.required_features = .{ .bmi, .false_deps_lzcnt_tzcnt, null, null },
	70656	.required_features = .{ .bmi, .false_deps_tzcnt, null, null },
70657	70657	.src_constraints = .{ .{ .exact_int = 32 }, .any, .any },
70658	70658	.patterns = &.{
70659	70659	.{ .src = .{ .mem, .none, .none } },
...	...	@@ -70690,7 +70690,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
70690	70690	.{ ._, ._, .tzcnt, .dst0d, .src0d, ._, ._ },
70691	70691	} },
70692	70692	}, .{
70693		.required_features = .{ .@"64bit", .bmi, .false_deps_lzcnt_tzcnt, null },
	70693	.required_features = .{ .@"64bit", .bmi, .false_deps_tzcnt, null },
70694	70694	.src_constraints = .{ .{ .exact_int = 64 }, .any, .any },
70695	70695	.patterns = &.{
70696	70696	.{ .src = .{ .mem, .none, .none } },
...	...	@@ -70867,7 +70867,7 @@ fn genBody(cg: *CodeGen, body: []const Air.Inst.Index) InnerError!void {
70867	70867	.{ ._, ._f, .bs, .dst0q, .src0q, ._, ._ },
70868	70868	} },
70869	70869	}, .{
70870		.required_features = .{ .@"64bit", .bmi, .false_deps_lzcnt_tzcnt, null },
	70870	.required_features = .{ .@"64bit", .bmi, .false_deps_tzcnt, null },
70871	70871	.src_constraints = .{ .{ .remainder_int = .{ .of = .qword, .is = .qword } }, .any, .any },
70872	70872	.patterns = &.{
70873	70873	.{ .src = .{ .to_mem, .none, .none } },
...	...	@@ -182371,12 +182371,13 @@ fn hasFeature(cg: *CodeGen, feature: std.Target.x86.Feature) bool {
182371	182371	.x86_64 => true,
182372	182372	},
182373	182373	.false_deps_getmant,
182374		.false_deps_lzcnt_tzcnt,
	182374	.false_deps_lzcnt,
182375	182375	.false_deps_mulc,
182376	182376	.false_deps_mullq,
182377	182377	.false_deps_perm,
182378	182378	.false_deps_popcnt,
182379	182379	.false_deps_range,
	182380	.false_deps_tzcnt,
182380	182381	.slow_3ops_lea,
182381	182382	.slow_incdec,
182382	182383	.slow_lea,

src/link/Wasm.zig+3

...	...	@@ -3254,6 +3254,7 @@ pub const Feature = packed struct(u8) {
3254	3254	@"tail-call",
3255	3255	@"shared-mem",
3256	3256	@"wide-arithmetic",
	3257	@"relaxed-atomics",
3257	3258
3258	3259	pub fn fromCpuFeature(feature: std.Target.wasm.Feature) Tag {
3259	3260	return switch (feature) {
...	...	@@ -3271,6 +3272,7 @@ pub const Feature = packed struct(u8) {
3271	3272	.nontrapping_bulk_memory_len0 => .@"nontrapping-bulk-memory-len0", // Zig extension.
3272	3273	.nontrapping_fptoint => .@"nontrapping-fptoint",
3273	3274	.reference_types => .@"reference-types",
	3275	.relaxed_atomics => .@"relaxed-atomics",
3274	3276	.relaxed_simd => .@"relaxed-simd",
3275	3277	.sign_ext => .@"sign-ext",
3276	3278	.simd128 => .simd128,
...	...	@@ -3296,6 +3298,7 @@ pub const Feature = packed struct(u8) {
3296	3298	.@"nontrapping-bulk-memory-len0" => .nontrapping_bulk_memory_len0, // Zig extension.
3297	3299	.@"nontrapping-fptoint" => .nontrapping_fptoint,
3298	3300	.@"reference-types" => .reference_types,
	3301	.@"relaxed-atomics" => .relaxed_atomics,
3299	3302	.@"relaxed-simd" => .relaxed_simd,
3300	3303	.@"sign-ext" => .sign_ext,
3301	3304	.simd128 => .simd128,

src/target.zig+3

...	...	@@ -415,6 +415,9 @@ pub fn hasDebugInfo(target: *const std.Target) bool {
415	415	.ptx87,
416	416	.ptx88,
417	417	.ptx90,
	418	.ptx91,
	419	.ptx92,
	420	.ptx93,
418	421	}),
419	422	.bpfel, .bpfeb => false,
420	423	else => true,

test/tests.zig+12-1

...	...	@@ -2357,13 +2357,17 @@ pub fn isNative(actual_target: const std.Build.ResolvedTarget, host: const std
2357	2357	.@"16bit_mode",
2358	2358	.@"32bit_mode",
2359	2359	.@"64bit",
	2360	.false_deps_bls,
	2361	.false_deps_compress,
	2362	.false_deps_expand,
2360	2363	.false_deps_getmant,
2361		.false_deps_lzcnt_tzcnt,
	2364	.false_deps_lzcnt,
2362	2365	.false_deps_mulc,
2363	2366	.false_deps_mullq,
2364	2367	.false_deps_perm,
2365	2368	.false_deps_popcnt,
2366	2369	.false_deps_range,
	2370	.false_deps_tzcnt,
2367	2371	.fast_11bytenop,
2368	2372	.fast_15bytenop,
2369	2373	.fast_7bytenop,
...	...	@@ -2391,10 +2395,12 @@ pub fn isNative(actual_target: const std.Build.ResolvedTarget, host: const std
2391	2395	.prefer_legacy_setcc,
2392	2396	.prefer_mask_registers,
2393	2397	.prefer_movmsk_over_vtest,
	2398	.prefer_ndd_mem,
2394	2399	.prefer_no_gather,
2395	2400	.prefer_no_scatter,
2396	2401	.slow_3ops_lea,
2397	2402	.slow_incdec,
	2403	.slow_indirect_call,
2398	2404	.slow_lea,
2399	2405	.slow_pmaddwd,
2400	2406	.slow_pmulld,
...	...	@@ -2406,10 +2412,13 @@ pub fn isNative(actual_target: const std.Build.ResolvedTarget, host: const std
2406	2412	}),
2407	2413	.aarch64, .aarch64_be => std.Target.aarch64.featureSet(&.{
2408	2414	.addr_lsl_slow_14,
	2415	.align_cmp_csel_pairs,
2409	2416	.alu_lsl_fast,
2410	2417	.avoid_ldapur,
2411	2418	.disable_fast_inc_vl,
2412	2419	.exynos_cheap_as_move,
	2420	.fast_ld1_single,
	2421	.fixed_load_latency_4,
2413	2422	.fuse_address,
2414	2423	.fuse_addsub_2reg_const1,
2415	2424	.fuse_adrp_add,
...	...	@@ -2418,7 +2427,9 @@ pub fn isNative(actual_target: const std.Build.ResolvedTarget, host: const std
2418	2427	.fuse_crypto_eor,
2419	2428	.fuse_csel,
2420	2429	.fuse_cset,
	2430	.fuse_fcsel,
2421	2431	.fuse_literals,
	2432	.has_limited_64bit_vector_mul_bandwidth,
2422	2433	.predictable_select_expensive,
2423	2434	.slow_misaligned_128store,
2424	2435	.slow_paired_128,

tools/update_cpu_features.zig+10-99

...	...	@@ -131,6 +131,14 @@ const targets = [_]ArchTarget{
131	131	.zig_name = "contextidr_el2",
132	132	.desc = "Enable RW operand Context ID Register (EL2)",
133	133	},
	134	.{
	135	.llvm_name = "hip12",
	136	.flatten = true,
	137	},
	138	.{
	139	.llvm_name = "mca-streaming-sched",
	140	.omit = true,
	141	},
134	142	.{
135	143	.llvm_name = "neoversee1",
136	144	.flatten = true,
...	...	@@ -1553,103 +1561,6 @@ const targets = [_]ArchTarget{
1553	1561	.zig_name = "baseline_rv64",
1554	1562	.features = &.{ "64bit", "a", "c", "d", "f", "i", "m" },
1555	1563	},
1556		.{
1557		.llvm_name = null,
1558		.zig_name = "spacemit_a100",
1559		.features = &.{
1560		"64bit",
1561		"a",
1562		"b",
1563		"c",
1564		"dlen_factor_2",
1565		"i",
1566		"m",
1567		"optimized_nf2_segment_load_store",
1568		"optimized_nf3_segment_load_store",
1569		"optimized_nf4_segment_load_store",
1570		"smepmp",
1571		"smnpm",
1572		"smstateen",
1573		"ssccptr",
1574		"sscofpmf",
1575		"sscounterenw",
1576		"ssnpm",
1577		"sspm",
1578		"sstc",
1579		"sstvala",
1580		"sstvecd",
1581		"ssu64xl",
1582		"supm",
1583		"svade",
1584		"svbare",
1585		"svinval",
1586		"svnapot",
1587		"svpbmt",
1588		"unaligned_scalar_mem",
1589		"v",
1590		"vxrm_pipeline_flush",
1591		"za64rs",
1592		"zawrs",
1593		"zbc",
1594		"zbkc",
1595		"zcb",
1596		"zcmop",
1597		"zfa",
1598		"zfh",
1599		"zic64b",
1600		"zicbom",
1601		"zicbop",
1602		"zicboz",
1603		"ziccamoa",
1604		"ziccif",
1605		"zicclsm",
1606		"ziccrse",
1607		"zicntr",
1608		"zicond",
1609		"zifencei",
1610		"zihintntl",
1611		"zihintpause",
1612		"zihpm",
1613		"zimop",
1614		"zkt",
1615		"zvbb",
1616		"zvfbfwma",
1617		"zvfh",
1618		"zvkng",
1619		"zvknha",
1620		"zvksc",
1621		"zvksg",
1622		"zvl1024b",
1623		},
1624		},
1625		.{
1626		.llvm_name = null,
1627		.zig_name = "spacemit_x100",
1628		.features = &.{
1629		"dlen_factor_2",
1630		"optimized_nf2_segment_load_store",
1631		"optimized_nf3_segment_load_store",
1632		"optimized_nf4_segment_load_store",
1633		"rva23s64",
1634		"smepmp",
1635		"smnpm",
1636		"smstateen",
1637		"sspm",
1638		"unaligned_scalar_mem",
1639		"vxrm_pipeline_flush",
1640		"xsmtvdot",
1641		"zbc",
1642		"zbkc",
1643		"zfh",
1644		"zvfbfwma",
1645		"zvfh",
1646		"zvkng",
1647		"zvknha",
1648		"zvksc",
1649		"zvksg",
1650		"zvl256b",
1651		},
1652		},
1653	1564	},
1654	1565	},
1655	1566	.{
...	...	@@ -2450,7 +2361,7 @@ fn processOneTargetInner(io: Io, job: Job) !void {
2450	2361	for (all_features.items) \|feature\| {
2451	2362	if (feature.llvm_name) \|llvm_name\| {
2452	2363	try w.print(
2453		\\ result[@intFromEnum(Feature.{f})] = .{{
	2364	\\ result[@backingInt(Feature.{f})] = .{{
2454	2365	\\ .llvm_name = "{f}",
2455	2366	\\ .description = "{f}",
2456	2367	\\ .dependencies = featureSet(&[_]Feature{{
...	...	@@ -2463,7 +2374,7 @@ fn processOneTargetInner(io: Io, job: Job) !void {
2463	2374	);
2464	2375	} else {
2465	2376	try w.print(
2466		\\ result[@intFromEnum(Feature.{f})] = .{{
	2377	\\ result[@backingInt(Feature.{f})] = .{{
2467	2378	\\ .llvm_name = null,
2468	2379	\\ .description = "{f}",
2469	2380	\\ .dependencies = featureSet(&[_]Feature{{