Make AMD CPU patch optional

Keep the fan center across subgroups and reuse the existing index fetch.

src/common/emulatorConfig.cpp

@@ -63,6 +63,10 @@ bool VrEnabled() {
 	return g_config->vr_enabled;
 }
 
+bool AmdCpuEnabled() {
+	return g_config->amd_cpu_enabled;
+}
+
 uint32_t GetVblankFrequency() {
 	return std::clamp(g_config->vblank_frequency, 30u, 360u);
 }

src/common/emulatorConfig.h

@@ -48,6 +48,7 @@ struct ConfigOptions {
 	int32_t                gpu_index                   = -1;
 	bool                   fullscreen_enabled          = false;
 	bool                   vr_enabled                  = false;
+	bool                   amd_cpu_enabled             = false;
 	uint32_t               vblank_frequency            = 60;
 	uint32_t               console_language            = DEFAULT_CONSOLE_LANGUAGE;
 	bool                   vulkan_validation_enabled   = false;
@@ -83,6 +84,7 @@ PresentMode GetPresentMode();
 int32_t GetGpuIndex();
 bool     FullscreenEnabled();
 bool     VrEnabled();
+bool     AmdCpuEnabled();
 uint32_t GetVblankFrequency();
 uint32_t GetConsoleLanguage();
 bool     VulkanValidationEnabled();

src/graphics/shader/recompiler/frontend/translate/Translate.cpp

@@ -1102,26 +1102,32 @@ IR::Program TranslateProgram(const Decoder::Program& decoded, const CFG::Graph&
 			    entry_ir.BitwiseOr(wave_info, entry_ir.BitwiseOr(entry_ir.ShiftLeftLogical(
 			                                                         primitive_count, u32(8)),
 			                                                     vertex_count)));
-			// GS adjacency addresses local ES records in LDS. Strip winding alternates
-			// with the global primitive number, including across subgroup boundaries.
-			const auto parity = mesh.input_primitive ==
-			                            static_cast<uint32_t>(Prospero::PrimitiveType::kTriStrip)
-			                        ? entry_ir.BitwiseAnd(entry_ir.IAdd(primitive_chunk, local), u32(1))
-			                        : u32(0);
+			// GS adjacency addresses local ES records in LDS. Fans retain the draw's
+			// center in every subgroup; strip winding follows the global primitive.
 			const auto vertex = entry_ir.IMul(local, step);
-			const auto first  = entry_ir.IAdd(vertex, parity);
-			const auto second = mesh.InputPrimitiveSize() >= 2u
-			                        ? entry_ir.ISub(entry_ir.IAdd(vertex, u32(1)), parity)
-			                        : u32(0);
-			const auto third = mesh.InputPrimitiveSize() == 3u
-			                       ? entry_ir.IAdd(vertex, u32(2))
-			                       : u32(0);
+			auto       first  = vertex;
+			auto       second = u32(0);
+			auto       third  = u32(0);
+			if (mesh.InputPrimitiveSize() >= 2u) {
+				second = entry_ir.IAdd(vertex, u32(1));
+			}
+			if (mesh.InputPrimitiveSize() == 3u) {
+				third = entry_ir.IAdd(vertex, u32(2));
+			}
+			auto input_vertex = entry_ir.IAdd(chunk, local);
+			if (mesh.input_primitive == static_cast<uint32_t>(Prospero::PrimitiveType::kTriFan)) {
+				first = u32(0);
+				input_vertex = entry_ir.Select(entry_ir.IEqual(local, u32(0)), u32(0), input_vertex);
+			} else if (mesh.input_primitive == static_cast<uint32_t>(Prospero::PrimitiveType::kTriStrip)) {
+				const auto parity = entry_ir.BitwiseAnd(entry_ir.IAdd(primitive_chunk, local), u32(1));
+				first = entry_ir.IAdd(first, parity);
+				second = entry_ir.ISub(second, parity);
+			}
 			entry_ir.SetVectorReg(static_cast<IR::VectorReg>(0),
 			                      entry_ir.BitwiseOr(entry_ir.ShiftLeftLogical(first, u32(2)),
 			                                         entry_ir.ShiftLeftLogical(second, u32(18))));
 			entry_ir.SetVectorReg(static_cast<IR::VectorReg>(1),
 			                      entry_ir.ShiftLeftLogical(third, u32(2)));
-			const auto input_vertex = entry_ir.IAdd(chunk, local);
 			const auto index_bytes  = draw(3);
 			const auto indexed      = entry_ir.INotEqual(index_bytes, u32(0));
 			const auto index_low    = draw(4);

src/graphics/shader/shader.cpp

@@ -797,6 +797,7 @@ ShaderParams PrepareProgram(const HW::VertexShaderInfo& regs, const HW::Context&
 	const auto& group = user_config.GetGeControl();
 	if ((user_config.GetPrimType() != Prospero::PrimitiveType::kPointList &&
 	     user_config.GetPrimType() != Prospero::PrimitiveType::kLineList &&
+	     user_config.GetPrimType() != Prospero::PrimitiveType::kTriFan &&
 	     user_config.GetPrimType() != Prospero::PrimitiveType::kTriStrip &&
 	     user_config.GetPrimType() != Prospero::PrimitiveType::kTriList) ||
 	    sh.m_vgtGsOutPrimType != 2u || sh.m_vgtGsMaxVertOut < 3u ||

src/graphics/shader/shader.h

@@ -97,8 +97,11 @@ struct ShaderMeshInputInfo: ShaderWorkgroupInputInfo {
 		}
 	}
 	[[nodiscard]] constexpr uint32_t InputPrimitiveStep() const {
-		return input_primitive == static_cast<uint32_t>(Prospero::PrimitiveType::kTriStrip)
-		           ? 1u : InputPrimitiveSize();
+		switch (static_cast<Prospero::PrimitiveType>(input_primitive)) {
+			case Prospero::PrimitiveType::kTriFan:
+			case Prospero::PrimitiveType::kTriStrip: return 1u;
+			default: return InputPrimitiveSize();
+		}
 	}
 	[[nodiscard]] constexpr uint32_t InputPrimitiveCount(uint32_t vertices) const {
 		const auto size = InputPrimitiveSize();

src/launcher/forms/configuration_edit_dialog.ui

@@ -271,21 +271,19 @@
        <property name="title">
         <string>Compatibility</string>
        </property>
-       <layout class="QHBoxLayout" name="compatibilityLayout">
+       <layout class="QVBoxLayout" name="compatibilityLayout">
+        <item>
+         <widget class="QCheckBox" name="checkBox_amd_cpu">
+          <property name="toolTip"><string>Patch CPU instructions that differ between AMD and the host CPU.</string></property>
+          <property name="text"><string>AMD CPU patch (EXPERIMENTAL)</string></property>
+         </widget>
+        </item>
         <item>
          <widget class="QCheckBox" name="checkBox_red_zone_protection">
           <property name="toolTip"><string>Helps prevent crashes in some games on Windows</string></property>
           <property name="text"><string>Windows SysV red zone crash protection (Experimental)</string></property>
          </widget>
         </item>
-        <item>
-         <spacer name="compatibilitySpacer">
-          <property name="orientation"><enum>Qt::Horizontal</enum></property>
-          <property name="sizeHint" stdset="0">
-           <size><width>40</width><height>20</height></size>
-          </property>
-         </spacer>
-        </item>
        </layout>
       </widget>
      </item>

src/launcher/include/configuration.h

@@ -105,6 +105,7 @@ public:
 	QString                printf_output_file          = "_kyty.txt";
 	bool                   profiler_enabled            = false;
 	bool                   renderdoc_enabled           = false;
+	bool                   amd_cpu_enabled             = false;
 #if defined(_WIN32)
 	bool red_zone_protection_enabled = false;
 #endif
@@ -134,6 +135,7 @@ public:
 		printf_output_file          = other.printf_output_file;
 		profiler_enabled            = other.profiler_enabled;
 		renderdoc_enabled           = other.renderdoc_enabled;
+		amd_cpu_enabled             = other.amd_cpu_enabled;
 #if defined(_WIN32)
 		red_zone_protection_enabled = other.red_zone_protection_enabled;
 #endif
@@ -180,6 +182,7 @@ public:
 		KYTY_CFG_SET(printf_output_file);
 		KYTY_CFG_SET(profiler_enabled);
 		KYTY_CFG_SET(renderdoc_enabled);
+		KYTY_CFG_SET(amd_cpu_enabled);
 #if defined(_WIN32)
 		KYTY_CFG_SET(red_zone_protection_enabled);
 #endif
@@ -223,6 +226,7 @@ public:
 		KYTY_CFG_GET(printf_output_file);
 		KYTY_CFG_GET(profiler_enabled);
 		KYTY_CFG_GET(renderdoc_enabled);
+		amd_cpu_enabled = s->value("amd_cpu_enabled", false).toBool();
 #if defined(_WIN32)
 		red_zone_protection_enabled =
 		    s->value("red_zone_protection_enabled", red_zone_protection_enabled).toBool();

src/launcher/src/configurationEditDialog.cpp

@@ -239,6 +239,10 @@ void ConfigurationEditDialog::Init(const Configuration& info) {
 	m_ui->checkBox_shader_validation->setChecked(info.shader_validation_enabled);
 	m_ui->checkBox_vulkan_validation->setChecked(info.vulkan_validation_enabled);
 	m_ui->checkBox_renderdoc_capture->setChecked(info.renderdoc_enabled);
+	m_ui->checkBox_amd_cpu->setChecked(info.amd_cpu_enabled);
+#if !defined(__linux__)
+	m_ui->checkBox_amd_cpu->setVisible(false);
+#endif
 #if defined(_WIN32)
 	m_ui->checkBox_red_zone_protection->setChecked(info.red_zone_protection_enabled);
 #else
@@ -379,6 +383,7 @@ static void UpdateInfo(Configuration& info, Ui::ConfigurationEditDialog& ui) {
 	info.vulkan_validation_enabled = ui.checkBox_vulkan_validation->isChecked();
 	info.shader_validation_enabled = ui.checkBox_shader_validation->isChecked();
 	info.renderdoc_enabled         = ui.checkBox_renderdoc_capture->isChecked();
+	info.amd_cpu_enabled           = ui.checkBox_amd_cpu->isChecked();
 #if defined(_WIN32)
 	info.red_zone_protection_enabled = ui.checkBox_red_zone_protection->isChecked();
 #endif

src/launcher/src/mainDialog.cpp

@@ -254,6 +254,9 @@ static QStringList CreateEmulatorArgs(const Configuration& info) {
 		args << "--profile";
 	}
 	args << "--spirv-debug-printf" << "false";
+	if (info.amd_cpu_enabled) {
+		args << "--amd-cpu";
+	}
 #if defined(_WIN32)
 	if (info.red_zone_protection_enabled) {
 		args << "--redzone";

src/libs/libJson2.cpp

@@ -751,7 +751,11 @@ static const int64_t* KYTY_SYSV_ABI JsonValueGetInteger(const JsonValue* self) {
 	PRINT_NAME();
 
 	static const int64_t zero = 0;
-	return (self != nullptr && self->type == JsonValueTypeInteger ? &self->integer : &zero);
+	if (self != nullptr &&
+	    (self->type == JsonValueTypeInteger || self->type == JsonValueTypeUInteger)) {
+		return &self->integer;
+	}
+	return &zero;
 }
 
 static const double* KYTY_SYSV_ABI JsonValueGetReal(const JsonValue* self) {

src/loader/runtimeLinker.cpp

@@ -2089,6 +2089,16 @@ void RuntimeLinker::LoadProgramToMemory(Program* program) {
 			                     mode == Common::VirtualMemory::Mode::NoAccess);
 
 			if (Common::VirtualMemory::IsExecute(mode)) {
+				if (Config::AmdCpuEnabled()) {
+					const auto reciprocal_sqrt_count =
+					    X64InstructionEmulator::PatchReciprocalSquareRoots(segment_addr,
+					                                                        segment_file_size);
+					if (reciprocal_sqrt_count != 0) {
+						LOGF("Guest VRSQRTPS emulation: %s, instructions=%" PRIu64 "\n",
+						     Common::PathToString(program->file_name.filename()).c_str(),
+						     reciprocal_sqrt_count);
+					}
+				}
 				PatchProgram(program, segment_addr, segment_memory_size);
 #if KYTY_PLATFORM == KYTY_PLATFORM_WINDOWS
 				if (use_red_zone_protection) {

src/loader/x64InstructionEmulator.cpp

@@ -2,7 +2,13 @@
 
 #include "common/common.h"
 
+#include <Zydis/Zydis.h>
+#include <bit>
 #include <cstring>
+#if KYTY_PLATFORM == KYTY_PLATFORM_LINUX && !defined(__APPLE__)
+#include <emmintrin.h>
+#include <xmmintrin.h>
+#endif
 
 #if KYTY_PLATFORM == KYTY_PLATFORM_WINDOWS
 #include <windows.h> // IWYU pragma: keep
@@ -400,57 +406,74 @@ static bool ExecuteShaNiInsn(const ShaNiInsn& insn, const XmmWords& src2, const
 	}
 }
 
+#if !defined(__APPLE__)
+
+// Keep instruction semantics shared; only access to the saved host context differs.
+struct Context {
 #if KYTY_PLATFORM == KYTY_PLATFORM_WINDOWS
+	PCONTEXT native;
 
-static void LoadXmmWordsWin(const M128A* xmm, XmmWords& out) {
-	out.w[0] = static_cast<uint32_t>(xmm->Low);
-	out.w[1] = static_cast<uint32_t>(xmm->Low >> 32u);
-	out.w[2] = static_cast<uint32_t>(xmm->High);
-	out.w[3] = static_cast<uint32_t>(xmm->High >> 32u);
-}
+	[[nodiscard]] uint64_t Rip() const { return native->Rip; }
+	void                   Advance(size_t length) { native->Rip += length; }
+	[[nodiscard]] void*    Xmm(uint8_t index) const { return &native->Xmm0 + index; }
 
-static void StoreXmmWordsWin(M128A* xmm, const XmmWords& in) {
-	xmm->Low  = static_cast<uint64_t>(in.w[0]) | (static_cast<uint64_t>(in.w[1]) << 32u);
-	xmm->High = static_cast<uint64_t>(in.w[2]) | (static_cast<uint64_t>(in.w[3]) << 32u);
-}
+	void LoadGprs(uint64_t (&gpr)[16]) const {
+		const uint64_t registers[] = {native->Rax, native->Rcx, native->Rdx, native->Rbx,
+		                              native->Rsp, native->Rbp, native->Rsi, native->Rdi,
+		                              native->R8,  native->R9,  native->R10, native->R11,
+		                              native->R12, native->R13, native->R14, native->R15};
+		std::memcpy(gpr, registers, sizeof(gpr));
+	}
+#else
+	ucontext_t* native;
 
+	[[nodiscard]] uint64_t Rip() const {
+		return static_cast<uint64_t>(native->uc_mcontext.gregs[REG_RIP]);
+	}
+	void Advance(size_t length) {
+		native->uc_mcontext.gregs[REG_RIP] += static_cast<greg_t>(length);
+	}
+	[[nodiscard]] void* Xmm(uint8_t index) const {
+		if (native->uc_mcontext.fpregs == nullptr) {
+			return nullptr;
+		}
+		return native->uc_mcontext.fpregs->_xmm[index].element;
+	}
+
+	void LoadGprs(uint64_t (&gpr)[16]) const {
+		constexpr int registers[] = {REG_RAX, REG_RCX, REG_RDX, REG_RBX, REG_RSP, REG_RBP,
+		                             REG_RSI, REG_RDI, REG_R8,  REG_R9,  REG_R10, REG_R11,
+		                             REG_R12, REG_R13, REG_R14, REG_R15};
+		for (size_t i = 0; i < 16; ++i) {
+			gpr[i] = static_cast<uint64_t>(native->uc_mcontext.gregs[registers[i]]);
+		}
+	}
+
+	void ClearUpperYmm(uint8_t index) const {
+		// Linux signal frames use the standard XSAVE layout. An absent AVX component
+		// already restores the architectural initial value (all zeroes).
+		auto*    state    = reinterpret_cast<uint8_t*>(native->uc_mcontext.fpregs);
+		uint32_t magic    = 0;
+		uint32_t size     = 0;
+		uint64_t features = 0;
+		std::memcpy(&magic, state + 464, sizeof(magic));
+		if (magic != 0x46505853) {
+			return;
+		}
+		std::memcpy(&size, state + 480, sizeof(size));
+		if (size < 832) {
+			return;
+		}
+		std::memcpy(&features, state + 512, sizeof(features));
+		if ((features & 4) != 0) {
+			std::memset(state + 576 + index * 16, 0, 16);
+		}
+	}
 #endif
+};
 
-#if KYTY_PLATFORM == KYTY_PLATFORM_WINDOWS
-
-static M128A* GetContextXmm(PCONTEXT context, uint8_t index) {
-	if (context == nullptr || index >= 16) {
-		return nullptr;
-	}
-
-	return &context->Xmm0 + index;
-}
-
-static void LoadContextGprsWin(PCONTEXT context, uint64_t (&gpr)[16]) {
-	gpr[0]  = context->Rax;
-	gpr[1]  = context->Rcx;
-	gpr[2]  = context->Rdx;
-	gpr[3]  = context->Rbx;
-	gpr[4]  = context->Rsp;
-	gpr[5]  = context->Rbp;
-	gpr[6]  = context->Rsi;
-	gpr[7]  = context->Rdi;
-	gpr[8]  = context->R8;
-	gpr[9]  = context->R9;
-	gpr[10] = context->R10;
-	gpr[11] = context->R11;
-	gpr[12] = context->R12;
-	gpr[13] = context->R13;
-	gpr[14] = context->R14;
-	gpr[15] = context->R15;
-}
-
-static bool TryEmulateShaNi(PCONTEXT context) {
-	if (context == nullptr) {
-		return false;
-	}
-
-	const auto* rip = reinterpret_cast<const uint8_t*>(context->Rip);
+static bool TryEmulateShaNi(Context& context) {
+	const auto* rip = reinterpret_cast<const uint8_t*>(context.Rip());
 	ShaNiInsn   insn {};
 	if (!DecodeShaNiInsn(rip, insn)) {
 		return false;
@@ -458,8 +481,8 @@ static bool TryEmulateShaNi(PCONTEXT context) {
 
 	const uint8_t modrm_byte = rip[insn.modrm_offset];
 	const uint8_t dest_index = ShaNiRegIndex(modrm_byte, insn.rex, true);
-	auto*         dest_xmm   = GetContextXmm(context, dest_index);
-	auto*         xmm0       = GetContextXmm(context, 0);
+	auto*         dest_xmm   = context.Xmm(dest_index);
+	auto*         xmm0       = context.Xmm(0);
 	if (dest_xmm == nullptr || xmm0 == nullptr) {
 		return false;
 	}
@@ -467,20 +490,20 @@ static bool TryEmulateShaNi(PCONTEXT context) {
 	XmmWords dest {};
 	XmmWords src2 {};
 	XmmWords xmm0_words {};
-	LoadXmmWordsWin(dest_xmm, dest);
-	LoadXmmWordsWin(xmm0, xmm0_words);
+	std::memcpy(&dest, dest_xmm, sizeof(dest));
+	std::memcpy(&xmm0_words, xmm0, sizeof(xmm0_words));
 
 	if (ShaNiModrmIsRegister(modrm_byte)) {
 		const uint8_t src_index = ShaNiRegIndex(modrm_byte, insn.rex, false);
-		auto*         src_xmm   = GetContextXmm(context, src_index);
+		auto*         src_xmm   = context.Xmm(src_index);
 		if (src_xmm == nullptr) {
 			return false;
 		}
-		LoadXmmWordsWin(src_xmm, src2);
+		std::memcpy(&src2, src_xmm, sizeof(src2));
 	} else {
 		uint64_t    gpr[16] {};
 		const void* source = nullptr;
-		LoadContextGprsWin(context, gpr);
+		context.LoadGprs(gpr);
 		if (!ResolveShaNiMemoryAddress(rip, insn, gpr, source)) {
 			return false;
 		}
@@ -491,18 +514,13 @@ static bool TryEmulateShaNi(PCONTEXT context) {
 		return false;
 	}
 
-	StoreXmmWordsWin(dest_xmm, dest);
-	context->Rip += insn.length;
+	std::memcpy(dest_xmm, &dest, sizeof(dest));
+	context.Advance(insn.length);
 	return true;
 }
 
-static bool TryEmulateSse4a(PCONTEXT context) {
-	if (context == nullptr) {
-		return false;
-	}
-
-	const auto* rip = reinterpret_cast<const uint8_t*>(context->Rip);
-
+static bool TryEmulateSse4a(Context& context) {
+	const auto*   rip    = reinterpret_cast<const uint8_t*>(context.Rip());
 	const uint8_t prefix = rip[0];
 	if (prefix != 0x66 && prefix != 0xf2) {
 		return false;
@@ -511,210 +529,13 @@ static bool TryEmulateSse4a(PCONTEXT context) {
 	size_t  offset = 1;
 	uint8_t rex    = 0;
 	if ((rip[offset] & 0xf0u) == 0x40u) {
-		rex = rip[offset];
-		offset++;
+		rex = rip[offset++];
 	}
-
 	if (rip[offset] != 0x0f || rip[offset + 1] != 0x78) {
 		return false;
 	}
 
-	auto modrm = rip[offset + 2];
-	if ((modrm & 0xc0u) != 0xc0u) {
-		return false;
-	}
-
-	const uint8_t reg    = ((modrm >> 3u) & 0x07u) | ((rex & 0x04u) << 1u);
-	const uint8_t rm     = (modrm & 0x07u) | ((rex & 0x01u) << 3u);
-	const uint8_t length = rip[offset + 3];
-	const uint8_t index  = rip[offset + 4];
-
-	// AMD SSE4a immediate-form EXTRQ/INSERTQ. PS5 code can execute these natively on AMD hardware,
-	// while Intel hosts raise an illegal-instruction exception.
-	if (prefix == 0x66) {
-		auto* dst = GetContextXmm(context, rm);
-		if (dst == nullptr) {
-			return false;
-		}
-
-		dst->Low  = ExtractBitField(dst->Low, length, index);
-		dst->High = 0;
-		context->Rip += offset + 5;
-		return true;
-	}
-
-	auto* dst = GetContextXmm(context, reg);
-	auto* src = GetContextXmm(context, rm);
-	if (dst == nullptr || src == nullptr) {
-		return false;
-	}
-
-	dst->Low = InsertBitField(dst->Low, src->Low, length, index);
-	context->Rip += offset + 5;
-	return true;
-}
-
-static bool TryEmulateMonitorxMwaitx(PCONTEXT context) {
-	if (context == nullptr) {
-		return false;
-	}
-
-	const auto* rip = reinterpret_cast<const uint8_t*>(context->Rip);
-	if (rip[0] != 0x0f || rip[1] != 0x01 || (rip[2] != 0xfa && rip[2] != 0xfb)) {
-		return false;
-	}
-
-	// AMD MONITORX/MWAITX are used by PS5 code in wait loops. Intel hosts can raise an illegal-
-	// instruction exception, so approximate them as a no-op/yield pair.
-	if (rip[2] == 0xfb) {
-		SwitchToThread();
-	}
-	context->Rip += 3;
-	return true;
-}
-
-#elif !defined(__APPLE__)
-
-// Linux signal contexts expose registers through ucontext_t.
-
-static uint32_t* GetContextXmm(ucontext_t* context, uint8_t index) {
-	if (context == nullptr || index >= 16) {
-		return nullptr;
-	}
-
-	auto* fpregs = context->uc_mcontext.fpregs;
-	if (fpregs == nullptr) {
-		return nullptr;
-	}
-
-	return static_cast<uint32_t*>(fpregs->_xmm[index].element);
-}
-
-static void LoadContextGprsLin(ucontext_t* context, uint64_t (&gpr)[16]) {
-	gpr[0]  = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_RAX]);
-	gpr[1]  = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_RCX]);
-	gpr[2]  = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_RDX]);
-	gpr[3]  = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_RBX]);
-	gpr[4]  = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_RSP]);
-	gpr[5]  = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_RBP]);
-	gpr[6]  = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_RSI]);
-	gpr[7]  = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_RDI]);
-	gpr[8]  = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_R8]);
-	gpr[9]  = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_R9]);
-	gpr[10] = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_R10]);
-	gpr[11] = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_R11]);
-	gpr[12] = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_R12]);
-	gpr[13] = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_R13]);
-	gpr[14] = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_R14]);
-	gpr[15] = static_cast<uint64_t>(context->uc_mcontext.gregs[REG_R15]);
-}
-
-static void LoadXmmWordsLin(const uint32_t* xmm, XmmWords& out) {
-	out.w[0] = xmm[0];
-	out.w[1] = xmm[1];
-	out.w[2] = xmm[2];
-	out.w[3] = xmm[3];
-}
-
-static void StoreXmmWordsLin(uint32_t* xmm, const XmmWords& in) {
-	xmm[0] = in.w[0];
-	xmm[1] = in.w[1];
-	xmm[2] = in.w[2];
-	xmm[3] = in.w[3];
-}
-
-static bool TryEmulateShaNi(ucontext_t* context) {
-	if (context == nullptr) {
-		return false;
-	}
-
-	auto&       rip_reg = context->uc_mcontext.gregs[REG_RIP];
-	const auto* rip     = reinterpret_cast<const uint8_t*>(rip_reg);
-	ShaNiInsn   insn {};
-	if (!DecodeShaNiInsn(rip, insn)) {
-		return false;
-	}
-
-	const uint8_t modrm_byte = rip[insn.modrm_offset];
-	const uint8_t dest_index = ShaNiRegIndex(modrm_byte, insn.rex, true);
-	auto*         dest_xmm   = GetContextXmm(context, dest_index);
-	auto*         xmm0       = GetContextXmm(context, 0);
-	if (dest_xmm == nullptr || xmm0 == nullptr) {
-		return false;
-	}
-
-	XmmWords dest {};
-	XmmWords src2 {};
-	XmmWords xmm0_words {};
-	LoadXmmWordsLin(dest_xmm, dest);
-	LoadXmmWordsLin(xmm0, xmm0_words);
-
-	if (ShaNiModrmIsRegister(modrm_byte)) {
-		const uint8_t src_index = ShaNiRegIndex(modrm_byte, insn.rex, false);
-		auto*         src_xmm   = GetContextXmm(context, src_index);
-		if (src_xmm == nullptr) {
-			return false;
-		}
-		LoadXmmWordsLin(src_xmm, src2);
-	} else {
-		uint64_t    gpr[16] {};
-		const void* source = nullptr;
-		LoadContextGprsLin(context, gpr);
-		if (!ResolveShaNiMemoryAddress(rip, insn, gpr, source)) {
-			return false;
-		}
-		std::memcpy(&src2, source, sizeof(src2));
-	}
-
-	if (!ExecuteShaNiInsn(insn, src2, xmm0_words, dest)) {
-		return false;
-	}
-
-	StoreXmmWordsLin(dest_xmm, dest);
-	rip_reg += static_cast<greg_t>(insn.length);
-	return true;
-}
-
-static uint64_t GetXmmLow(const uint32_t* xmm) {
-	return static_cast<uint64_t>(xmm[0]) | (static_cast<uint64_t>(xmm[1]) << 32u);
-}
-
-static void SetXmmLow(uint32_t* xmm, uint64_t value) {
-	xmm[0] = static_cast<uint32_t>(value);
-	xmm[1] = static_cast<uint32_t>(value >> 32u);
-}
-
-static void SetXmmHigh(uint32_t* xmm, uint64_t value) {
-	xmm[2] = static_cast<uint32_t>(value);
-	xmm[3] = static_cast<uint32_t>(value >> 32u);
-}
-
-static bool TryEmulateSse4a(ucontext_t* context) {
-	if (context == nullptr) {
-		return false;
-	}
-
-	auto& rip_reg = context->uc_mcontext.gregs[REG_RIP];
-
-	const auto* rip = reinterpret_cast<const uint8_t*>(rip_reg);
-
-	const uint8_t prefix = rip[0];
-	if (prefix != 0x66 && prefix != 0xf2) {
-		return false;
-	}
-
-	size_t  offset = 1;
-	uint8_t rex    = 0;
-	if ((rip[offset] & 0xf0u) == 0x40u) {
-		rex = rip[offset];
-		offset++;
-	}
-
-	if (rip[offset] != 0x0f || rip[offset + 1] != 0x78) {
-		return false;
-	}
-
-	auto modrm = rip[offset + 2];
+	const uint8_t modrm = rip[offset + 2];
 	if ((modrm & 0xc0u) != 0xc0u) {
 		return false;
 	}
@@ -725,58 +546,168 @@ static bool TryEmulateSse4a(ucontext_t* context) {
 	const uint8_t index  = rip[offset + 4];
 
 	// AMD SSE4a immediate-form EXTRQ/INSERTQ.
+	uint8_t dest_index = reg;
 	if (prefix == 0x66) {
-		auto* dst = GetContextXmm(context, rm);
-		if (dst == nullptr) {
-			return false;
-		}
-
-		SetXmmLow(dst, ExtractBitField(GetXmmLow(dst), length, index));
-		SetXmmHigh(dst, 0);
-		rip_reg += static_cast<greg_t>(offset + 5);
-		return true;
+		dest_index = rm;
 	}
-
-	auto* dst = GetContextXmm(context, reg);
-	auto* src = GetContextXmm(context, rm);
-	if (dst == nullptr || src == nullptr) {
+	auto* dest_xmm = context.Xmm(dest_index);
+	auto* src_xmm  = context.Xmm(rm);
+	if (dest_xmm == nullptr || src_xmm == nullptr) {
 		return false;
 	}
-
-	SetXmmLow(dst, InsertBitField(GetXmmLow(dst), GetXmmLow(src), length, index));
-	rip_reg += static_cast<greg_t>(offset + 5);
+	uint64_t dest[2] {};
+	uint64_t source = 0;
+	std::memcpy(dest, dest_xmm, sizeof(dest));
+	std::memcpy(&source, src_xmm, sizeof(source));
+	if (prefix == 0x66) {
+		dest[0] = ExtractBitField(dest[0], length, index);
+		dest[1] = 0;
+	} else {
+		dest[0] = InsertBitField(dest[0], source, length, index);
+	}
+	std::memcpy(dest_xmm, dest, sizeof(dest));
+	context.Advance(offset + 5);
 	return true;
 }
 
-static bool TryEmulateMonitorxMwaitx(ucontext_t* context) {
-	if (context == nullptr) {
-		return false;
-	}
-
-	auto& rip_reg = context->uc_mcontext.gregs[REG_RIP];
-
-	const auto* rip = reinterpret_cast<const uint8_t*>(rip_reg);
+static bool TryEmulateMonitorxMwaitx(Context& context) {
+	const auto* rip = reinterpret_cast<const uint8_t*>(context.Rip());
 	if (rip[0] != 0x0f || rip[1] != 0x01 || (rip[2] != 0xfa && rip[2] != 0xfb)) {
 		return false;
 	}
 
 	// Approximate AMD MONITORX/MWAITX as no-op/yield.
 	if (rip[2] == 0xfb) {
+#if KYTY_PLATFORM == KYTY_PLATFORM_WINDOWS
+		SwitchToThread();
+#else
 		::sched_yield();
+#endif
 	}
-	rip_reg += 3;
+	context.Advance(3);
+	return true;
+}
+
+#if KYTY_PLATFORM == KYTY_PLATFORM_LINUX
+
+static uint32_t ReciprocalSquareRoot(uint32_t bits) {
+	const uint32_t magnitude = bits & 0x7fffffffu;
+	const uint32_t exponent  = magnitude & 0x7f800000u;
+	if (exponent == 0) {
+		// RSQRT treats denormals as signed zero regardless of MXCSR.DAZ.
+		return (bits & 0x80000000u) | 0x7f800000u;
+	}
+	if (magnitude > 0x7f800000u) {
+		return bits | 0x00400000u; // Quiet NaNs without raising an exception.
+	}
+	if ((bits & 0x80000000u) != 0) {
+		return 0xffc00000u;
+	}
+	if (magnitude == 0x7f800000u) {
+		return 0;
+	}
+
+	// A deterministic accurate estimate meets the instruction's relative-error
+	// bound without relying on the host vendor's approximation table.
+	const __m128d input  = _mm_set_sd(static_cast<double>(std::bit_cast<float>(bits)));
+	const __m128d result = _mm_div_sd(_mm_set_sd(1.0), _mm_sqrt_sd(input, input));
+	return std::bit_cast<uint32_t>(_mm_cvtss_f32(_mm_cvtsd_ss(_mm_setzero_ps(), result)));
+}
+
+static bool TryEmulateReciprocalSquareRoot(Context& context) {
+	const auto* rip            = reinterpret_cast<const uint8_t*>(context.Rip());
+	size_t      prefix_size    = 0;
+	uint8_t     dest_extension = 0;
+	uint8_t     src_extension  = 0;
+	if (rip[0] == 0xc5 && (rip[1] & 0x7fu) == 0x70u) {
+		prefix_size    = 2;
+		dest_extension = (~rip[1] & 0x80u) >> 4u;
+	} else if (rip[0] == 0xc4 && (rip[1] & 0x1fu) == 1 && (rip[2] & 0x7fu) == 0x70u) {
+		prefix_size    = 3;
+		dest_extension = (~rip[1] & 0x80u) >> 4u;
+		src_extension  = (~rip[1] & 0x20u) >> 2u;
+	} else {
+		return false;
+	}
+	if (rip[prefix_size] != 0x52 || (rip[prefix_size + 1] & 0xc0u) != 0xc0u) {
+		return false;
+	}
+
+	const uint8_t modrm    = rip[prefix_size + 1];
+	const uint8_t dest     = ((modrm >> 3u) & 7u) | dest_extension;
+	const uint8_t source   = (modrm & 7u) | src_extension;
+	auto*         dest_xmm = context.Xmm(dest);
+	auto*         src_xmm  = context.Xmm(source);
+	if (dest_xmm == nullptr || src_xmm == nullptr) {
+		return false;
+	}
+	XmmWords result {};
+	std::memcpy(&result, src_xmm, sizeof(result));
+	// RSQRT ignores the rounding mode and never changes guest exception flags.
+	// Mask host exceptions while calculating, then restore the handler's state.
+	const uint32_t mxcsr = _mm_getcsr();
+	_mm_setcsr(0x1f80);
+	for (auto& word: result.w) {
+		word = ReciprocalSquareRoot(word);
+	}
+	_mm_setcsr(mxcsr);
+	std::memcpy(dest_xmm, &result, sizeof(result));
+	context.ClearUpperYmm(dest);
+	context.Advance(prefix_size + 2);
 	return true;
 }
 
 #endif
+#endif
+
+uint64_t PatchReciprocalSquareRoots(uint64_t address, uint64_t size) {
+	uint64_t patched = 0;
+#if KYTY_PLATFORM == KYTY_PLATFORM_LINUX && !defined(__APPLE__)
+	ZydisDecoder decoder {};
+	if (!ZYAN_SUCCESS(
+	        ZydisDecoderInit(&decoder, ZYDIS_MACHINE_MODE_LONG_64, ZYDIS_STACK_WIDTH_64))) {
+		return 0;
+	}
+	for (uint64_t offset = 0; offset < size;) {
+		auto*                   code = reinterpret_cast<uint8_t*>(address + offset);
+		ZydisDecodedInstruction instruction {};
+		ZydisDecodedOperand     operands[ZYDIS_MAX_OPERAND_COUNT] {};
+		if (!ZYAN_SUCCESS(
+		        ZydisDecoderDecodeFull(&decoder, code, size - offset, &instruction, operands))) {
+			++offset;
+			continue;
+		}
+		if (instruction.mnemonic == ZYDIS_MNEMONIC_VRSQRTPS &&
+		    instruction.encoding == ZYDIS_INSTRUCTION_ENCODING_VEX &&
+		    instruction.raw.vex.offset == 0 && operands[0].size == 128 &&
+		    operands[1].type == ZYDIS_OPERAND_TYPE_REGISTER) {
+			// vvvv is reserved (must be 1111b). Clear one bit to route this
+			// otherwise intact instruction through the existing SIGILL emulator.
+			code[instruction.raw.vex.size - 1] &= ~0x08u;
+			++patched;
+		}
+		offset += instruction.length;
+	}
+#else
+	(void)address;
+	(void)size;
+#endif
+	return patched;
+}
 
 bool TryEmulate(void* native_context) {
+#if !defined(__APPLE__)
+	if (native_context == nullptr) {
+		return false;
+	}
 #if KYTY_PLATFORM == KYTY_PLATFORM_WINDOWS
-	auto* context = static_cast<PCONTEXT>(native_context);
-	return TryEmulateMonitorxMwaitx(context) || TryEmulateSse4a(context) ||
-	       TryEmulateShaNi(context);
-#elif !defined(__APPLE__)
-	auto* context = static_cast<ucontext_t*>(native_context);
+	Context context {static_cast<PCONTEXT>(native_context)};
+#else
+	Context context {static_cast<ucontext_t*>(native_context)};
+	if (TryEmulateReciprocalSquareRoot(context)) {
+		return true;
+	}
+#endif
 	return TryEmulateMonitorxMwaitx(context) || TryEmulateSse4a(context) ||
 	       TryEmulateShaNi(context);
 #else

src/loader/x64InstructionEmulator.h

@@ -1,8 +1,11 @@
 #ifndef KYTY_LOADER_X64_INSTRUCTION_EMULATOR_H_
 #define KYTY_LOADER_X64_INSTRUCTION_EMULATOR_H_
 
+#include <cstdint>
+
 namespace Loader::X64InstructionEmulator {
 
+uint64_t           PatchReciprocalSquareRoots(uint64_t address, uint64_t size);
 [[nodiscard]] bool TryEmulate(void* native_context);
 
 } // namespace Loader::X64InstructionEmulator

src/main.cpp

@@ -56,6 +56,7 @@ static void PrintUsage() {
 	    "  --gpu <index>                        Vulkan physical device index. Default: auto.\n");
 	::printf("  --fullscreen                         Run in borderless desktop fullscreen.\n");
 	::printf("  --vr                                 Enable the virtual VR headset.\n");
+	::printf("  --amd-cpu                            Apply AMD CPU instruction patches (Linux).\n");
 	::printf("  --vblank-frequency <num>             Virtual vblank frequency. Default: 60.\n");
 	::printf("  --console-language <0-29>            Console language. Default: 1 (English US).\n");
 	::printf("  --vulkan-validation <true|false>     Enable Vulkan validation.\n");
@@ -168,6 +169,11 @@ static bool ParseArgs(int argc, char* argv[], RunOptions& options, bool& show_he
 			continue;
 		}
 
+		if (arg == "--amd-cpu") {
+			options.config.amd_cpu_enabled = true;
+			continue;
+		}
+
 		if (arg == "--playgo-hack") {
 			options.config.playgo_hack_enabled = true;
 			continue;

tests/VirtualMemoryAllocationTests.cpp

@@ -10,8 +10,12 @@
 #include "loader/redZonePatcher.h"
 #include "loader/runtimeLinker.h"
 #include "loader/systemContent.h"
+#include "loader/x64InstructionEmulator.h"
 
+#include <algorithm>
 #include <array>
+#include <bit>
+#include <cmath>
 #include <cinttypes>
 #include <cstdint>
 #include <cstdio>
@@ -21,8 +25,12 @@
 #include <vector>
 
 #if defined(__linux__)
+#include <csignal>
+#include <immintrin.h>
 #include <sys/uio.h>
+#include <ucontext.h>
 #include <unistd.h>
+#include <xbyak/xbyak.h>
 #endif
 
 #if KYTY_PLATFORM == KYTY_PLATFORM_WINDOWS
@@ -2538,15 +2546,190 @@ void TestModuleRelocationUsesWritableHostMapping() {
 	std::printf("[host]    %-48s ok\n", test);
 }
 
+#if defined(__linux__)
+volatile sig_atomic_t g_rsqrt_traps = 0;
+
+void ReciprocalSquareRootHandler(int, siginfo_t*, void* context) {
+	const auto host_mxcsr = _mm_getcsr();
+	_mm_setcsr(0x7fe1); // Exercise the handler under a distinct rounding mode and raised flags.
+	if (!Loader::X64InstructionEmulator::TryEmulate(context)) {
+		_exit(190);
+	}
+	if (_mm_getcsr() != 0x7fe1) {
+		_exit(191);
+	}
+	_mm_setcsr(host_mxcsr);
+	g_rsqrt_traps = g_rsqrt_traps + 1;
+}
+
+void TestPackedReciprocalSquareRoot() {
+	const char* test = "PackedReciprocalSquareRoot";
+	constexpr uint64_t code_size = 0x4000;
+	const auto mapping = Libs::LibKernel::Memory::AllocateProgramMemory(
+	    0x902000000, code_size, Common::VirtualMemory::Mode::ExecuteReadWrite, "rsqrt_test");
+	Check(test, mapping != 0, "failed to allocate instruction test code");
+	struct RestoreState {
+		uint64_t mapping;
+		uint64_t size;
+		uint32_t mxcsr;
+		struct sigaction previous {};
+		bool installed = false;
+		~RestoreState() {
+			_mm_setcsr(mxcsr);
+			if (installed) {
+				sigaction(SIGILL, &previous, nullptr);
+			}
+			Libs::LibKernel::Memory::FreeGuestMemory(mapping, size);
+		}
+	} restore {mapping, code_size, _mm_getcsr()};
+	struct sigaction action {};
+	action.sa_sigaction = ReciprocalSquareRootHandler;
+	action.sa_flags = SA_SIGINFO;
+	sigemptyset(&action.sa_mask);
+	restore.installed = sigaction(SIGILL, &action, &restore.previous) == 0;
+	Check(test, restore.installed, "failed to install instruction handler");
+	using GuestFunction = void(KYTY_SYSV_ABI*)(const uint32_t*, uint32_t*);
+	const auto function = reinterpret_cast<GuestFunction>(mapping);
+	const auto refine = [](float estimate) {
+		return (estimate * 0.5f) * std::fma(-estimate, estimate, 3.0f);
+	};
+	std::array<uint32_t, 16> input {};
+	std::array<uint32_t, 16> output {};
+	input.fill(0x3f800000);
+	for (const auto registers: {std::array {2, 1}, std::array {10, 9}, std::array {1, 1}}) {
+		const auto destination = registers[0];
+		const auto source = registers[1];
+		input[0] = 0x3f800000;
+		uint32_t expected = 0x3f800000;
+		if (source == 9) {
+			input[0] = 0x40800000;
+			expected = 0x3f000000;
+		}
+		Xbyak::CodeGenerator code(code_size, reinterpret_cast<void*>(mapping));
+		if (source == 9) {
+			code.vmovups(Xbyak::Ymm(1), code.ptr[code.rdi + 32]);
+		}
+		code.vmovups(Xbyak::Ymm(source), code.ptr[code.rdi]);
+		if (destination != source) {
+			code.vmovups(Xbyak::Ymm(destination), code.ptr[code.rdi + 32]);
+		}
+		code.vrsqrtps(Xbyak::Xmm(destination), Xbyak::Xmm(source));
+		code.vmovups(code.ptr[code.rsi], Xbyak::Ymm(destination));
+		code.vmovups(code.ptr[code.rsi + 32], Xbyak::Ymm(source));
+		code.vzeroupper();
+		code.ret();
+		const std::vector<uint8_t> original(code.getCode(), code.getCurr());
+		function(input.data(), output.data());
+		const float native = std::bit_cast<float>(output[0]);
+		Check(test, std::isfinite(native) && std::abs(native - std::bit_cast<float>(expected)) < 0.001f,
+		      "native reciprocal root is outside its error bound");
+		Check(test, Loader::X64InstructionEmulator::PatchReciprocalSquareRoots(mapping, code.getSize()) == 1,
+		      "instruction pass did not patch exactly one packed reciprocal root");
+		const auto* patched = reinterpret_cast<const uint8_t*>(mapping);
+		size_t changed = 0;
+		for (size_t i = 0; i < original.size(); ++i) {
+			changed += original[i] != patched[i];
+		}
+		Check(test, changed == 1, "instruction pass changed unrelated code bytes");
+		Common::VirtualMemory::FlushInstructionCache(mapping, code.getSize());
+		const auto before = g_rsqrt_traps;
+		function(input.data(), output.data());
+		Check(test, g_rsqrt_traps == before + 1, "patched instruction did not execute its handler");
+		Check(test, output[0] == expected, "patched instruction read or wrote the wrong register");
+		if (source != 9) {
+			Check(test, refine(refine(std::bit_cast<float>(output[0]))) == 1.0f,
+			      "identity quaternion normalization drifted below one");
+		}
+		for (size_t i = 0; i < 4; ++i) {
+			Check(test, output[4 + i] == 0, "128-bit VEX destination retained upper YMM lanes");
+			const auto expected_upper = destination == source ? 0u : input[4 + i];
+			Check(test, output[12 + i] == expected_upper, "source upper YMM lanes were corrupted");
+		}
+		std::printf("[host]    rsqrt xmm%d,xmm%d native=%08x patched=%08x\n",
+		            destination, source, std::bit_cast<uint32_t>(native), output[0]);
+		if (destination != 2) {
+			continue;
+		}
+		constexpr std::array<std::array<uint32_t, 8>, 4> cases {{
+		    {0x00000000, 0x80000000, 0x00000001, 0x80000001,
+		     0x7f800000, 0xff800000, 0x7f800000, 0xff800000},
+		    {0x7f800000, 0xff800000, 0xbf800000, 0x7fc12345,
+		     0x00000000, 0xffc00000, 0xffc00000, 0x7fc12345},
+		    {0x7f812345, 0xff812345, 0x40800000, 0x40000000,
+		     0x7fc12345, 0xffc12345, 0x3f000000, 0x3f3504f3},
+		    {0x00800000, 0x7f7fffff, 0x3f800000, 0x41800000,
+		     0x5f000000, 0x1f800000, 0x3f800000, 0x3e800000},
+		}};
+		for (const auto& values: cases) {
+			std::copy_n(values.begin(), 4, input.begin());
+			for (uint32_t controls = 0; controls < 8; ++controls) {
+				const uint32_t mxcsr = 0x1fa1u | ((controls & 3u) << 13u) | ((controls & 4u) << 4u);
+				_mm_setcsr(mxcsr);
+				function(input.data(), output.data());
+				const auto result_mxcsr = _mm_getcsr();
+				_mm_setcsr(restore.mxcsr);
+				Check(test, result_mxcsr == mxcsr, "instruction changed MXCSR controls or exception flags");
+				Check(test, std::equal(output.begin(), output.begin() + 4, values.begin() + 4),
+				      "special values or round-independent reciprocal roots differ from ISA semantics");
+			}
+		}
+		input.fill(0x3f800000);
+	}
+	std::array<uint8_t, 16> unknown {0x0f, 0x0b};
+	ucontext_t context {};
+	context.uc_mcontext.gregs[REG_RIP] = reinterpret_cast<greg_t>(unknown.data());
+	Check(test, !Loader::X64InstructionEmulator::TryEmulate(&context), "unrecognized UD2 was swallowed");
+	_libc_fpstate fpstate {};
+	context.uc_mcontext.fpregs = &fpstate;
+	const auto emulate = [&](std::array<uint8_t, 16> instruction, size_t length) {
+		context.uc_mcontext.gregs[REG_RIP] = reinterpret_cast<greg_t>(instruction.data());
+		Check(test, Loader::X64InstructionEmulator::TryEmulate(&context), "existing instruction emulation was rejected");
+		Check(test, context.uc_mcontext.gregs[REG_RIP] == reinterpret_cast<greg_t>(instruction.data() + length),
+		      "existing instruction emulation advanced RIP incorrectly");
+	};
+	fpstate._xmm[8].element[0] = 0x1234;
+	fpstate._xmm[8].element[2] = 0xdeadbeef;
+	emulate({0x66, 0x41, 0x0f, 0x78, 0xc0, 8, 4}, 7); // extrq xmm8, 8, 4
+	Check(test, fpstate._xmm[8].element[0] == 0x23 && fpstate._xmm[8].element[2] == 0,
+	      "SSE4a extraction lost its extended register or upper-half semantics");
+	fpstate._xmm[8].element[0] = 0x1111;
+	fpstate._xmm[8].element[2] = 0xdeadbeef;
+	fpstate._xmm[9].element[0] = 0xab;
+	emulate({0xf2, 0x45, 0x0f, 0x78, 0xc1, 8, 4}, 7); // insertq xmm8, xmm9, 8, 4
+	Check(test, fpstate._xmm[8].element[0] == 0x1ab1 && fpstate._xmm[8].element[2] == 0xdeadbeef,
+	      "SSE4a insertion corrupted its destination lanes");
+	std::array<uint32_t, 8> sha_source {0, 0, 0, 0, 5, 6, 7, 8};
+	context.uc_mcontext.gregs[REG_RDI] = reinterpret_cast<greg_t>(sha_source.data());
+	for (uint32_t lane = 0; lane < 4; ++lane) {
+		fpstate._xmm[8].element[lane] = lane + 1;
+	}
+	emulate({0x44, 0x0f, 0x38, 0xc9, 0x47, 0x10}, 6); // sha1msg1 xmm8, [rdi+16]
+	constexpr std::array<uint32_t, 4> sha_expected {6, 10, 2, 6};
+	Check(test, std::equal(sha_expected.begin(), sha_expected.end(), fpstate._xmm[8].element),
+	      "SHA emulation lost its memory operand or extended destination register");
+	emulate({0x0f, 0x01, 0xfa}, 3); // monitorx
+	std::printf("[host]    %-48s ok\n", test);
+}
+#endif
+
 } // namespace
 
 int main(int argc, char** argv) {
 	InitSubsystems();
+#if defined(__linux__)
+	if (argc == 2 && std::strcmp(argv[1], "--rsqrt-only") == 0) {
+		RunTest(TestPackedReciprocalSquareRoot);
+		return g_failed_tests == 0 ? 0 : 1;
+	}
+#endif
 	if (argc == 2 && std::strcmp(argv[1], "--red-zone-patcher-only") == 0) {
 		RunTest(TestWindowsGuestRedZoneStaticPatcher);
 		return g_failed_tests == 0 ? 0 : 1;
 	}
 
+#if defined(__linux__)
+	RunTest(TestPackedReciprocalSquareRoot);
+#endif
 	RunTest(TestWindowsGuestRedZoneStaticPatcher);
 	RunTest(TestProsperoArgumentAndInfoSizeContracts);
 	RunTest(TestGuestAddressSpaceOwnsReservationsBeforeBacking);

tests/shaderCfgTests.cpp

@@ -9606,6 +9606,15 @@ void TestMergedShaderUserDataSnapshot() {
                          params.user_data.begin() + 8),
           "monolithic NGG shader used stale GS-back state or lost its s8 user data");
   }
+  context.SetMaxOutputPerSubgroup(256);
+  context.SetGsMaxVertOut(8);
+  user_config.SetPrimitiveType(Prospero::PrimitiveType::kTriFan);
+  user_config.SetGeControl({32, 32});
+  ShaderVertexInputInfo fan_input{};
+  PrepareProgram(regs, context, user_config, fan_input);
+  Check(fan_input.mesh.primitives_per_group == 30 && fan_input.mesh.vertices_per_group == 32 &&
+            fan_input.mesh.max_vertices == 256 && fan_input.mesh.max_primitives == 192,
+        "captured triangle-fan GS configuration lost its subgroup assembly limits");
 }
 
 void TestEmbeddedFetchPreservesSharedScalarLoad() {
@@ -9791,6 +9800,18 @@ void TestMeshInputAssembly() {
        0x40000305, 1, 2, 3, 0, 15, false},
       {Prospero::PrimitiveType::kTriStrip, 5, 8, 0, 1, 0, 0, 11,
        0x40000305, 2, 1, 3, 0, 12, false},
+      {Prospero::PrimitiveType::kTriFan, 32, 6, 0, 3, 2, 0x1002, 0,
+       0x40000406, 0, 4, 5, 8, 0x0123, true},
+      {Prospero::PrimitiveType::kTriFan, 5, 8, 1, 0, 2, 0x1002, 0,
+       0x40000305, 0, 1, 2, 0, 0xabcd, true},
+      {Prospero::PrimitiveType::kTriFan, 5, 8, 1, 1, 0, 0, 11,
+       0x40000305, 0, 2, 3, 0, 15, false},
+      {Prospero::PrimitiveType::kTriFan, 5, 7, 1, 3, 2, 0x1002, 11,
+       0x40000204, 0, 4, 5, 12, 0xabd8, true},
+      {Prospero::PrimitiveType::kTriFan, 5, 7, 1, 4, 2, 0x1002, 0,
+       0x40000204, 0, 5, 6, 16, 0, false},
+      {Prospero::PrimitiveType::kTriFan, 5, 8, 1, 64, 4, 0x1000, 0,
+       0x41000000, 0, 65, 66, 268, 0, false},
       {Prospero::PrimitiveType::kLineList, 5, 17, 3, 0, 0, 0, 11,
        0x40000204, 0, 1, 0, 0, 23, false},
       {Prospero::PrimitiveType::kLineList, 6, 17, 2, 4, 2, 0x1002, 11,