Merge pull request #8669 from unknownbrackets/vulkan-tex

Use out/outPitch for remaining formats in Vulkan

Core/HW/MediaEngine.cpp

@@ -673,13 +673,12 @@ int MediaEngine::writeVideoImage(u32 bufferPtr, int frameWidth, int videoPixelMo
 	}
 
 	if (swizzle) {
-		const u32 pitch = videoLineSize / 4;
 		const int bxc = videoLineSize / 16;
 		int byc = (height + 7) / 8;
 		if (byc == 0)
 			byc = 1;
 
-		DoSwizzleTex16((const u32 *)imgbuf, buffer, bxc, byc, pitch, videoLineSize);
+		DoSwizzleTex16((const u32 *)imgbuf, buffer, bxc, byc, videoLineSize);
 		delete [] imgbuf;
 	}
 
@@ -789,13 +788,12 @@ int MediaEngine::writeVideoImageWithRange(u32 bufferPtr, int frameWidth, int vid
 	if (swizzle) {
 		WARN_LOG_REPORT_ONCE(vidswizzle, ME, "Swizzling Video with range");
 
-		const u32 pitch = videoLineSize / 4;
 		const int bxc = videoLineSize / 16;
 		int byc = (height + 7) / 8;
 		if (byc == 0)
 			byc = 1;
 
-		DoSwizzleTex16((const u32 *)imgbuf, buffer, bxc, byc, pitch, videoLineSize);
+		DoSwizzleTex16((const u32 *)imgbuf, buffer, bxc, byc, videoLineSize);
 		delete [] imgbuf;
 	}
 

GPU/Common/TextureCacheCommon.cpp

@@ -47,10 +47,10 @@ TextureCacheCommon::TextureCacheCommon()
 	memset(clutBufRaw_, 0, 1024 * sizeof(u32));
 	memset(clutBufConverted_, 0, 1024 * sizeof(u32));
 
-	// This is 5MB of temporary storage. Might be possible to shrink it.
-	tmpTexBuf32.resize(1024 * 512);  // 2MB
-	tmpTexBuf16.resize(1024 * 512);  // 1MB
-	tmpTexBufRearrange.resize(1024 * 512);   // 2MB
+	// These buffers will grow if necessary, but most won't need more than this.
+	tmpTexBuf32.resize(512 * 512);  // 1MB
+	tmpTexBuf16.resize(512 * 512);  // 0.5MB
+	tmpTexBufRearrange.resize(512 * 512);   // 1MB
 }
 
 TextureCacheCommon::~TextureCacheCommon() {
@@ -364,58 +364,24 @@ void TextureCacheCommon::LoadClut(u32 clutAddr, u32 loadBytes) {
 	clutMaxBytes_ = std::max(clutMaxBytes_, loadBytes);
 }
 
-void TextureCacheCommon::UnswizzleFromMem(u32 *dest, const u8 *texptr, u32 bufw, u32 height, u32 bytesPerPixel) {
+void TextureCacheCommon::UnswizzleFromMem(u32 *dest, u32 destPitch, const u8 *texptr, u32 bufw, u32 height, u32 bytesPerPixel) {
+	// Note: bufw is always aligned to 16 bytes, so rowWidth is always >= 16.
 	const u32 rowWidth = (bytesPerPixel > 0) ? (bufw * bytesPerPixel) : (bufw / 2);
-	const u32 pitch = rowWidth / 4;
+	// A visual mapping of unswizzling, where each letter is 16-byte and 8 letters is a block:
+	//
+	// ABCDEFGH IJKLMNOP
+	//      ->
+	// AI
+	// BJ
+	// CK
+	// ...
+	//
+	// bxc is the number of blocks in the x direction, and byc the number in the y direction.
 	const int bxc = rowWidth / 16;
+	// The height is not always aligned to 8, but rounds up.
 	int byc = (height + 7) / 8;
-	if (byc == 0)
-		byc = 1;
 
-	u32 ydest = 0;
-	if (rowWidth >= 16) {
-		// The most common one, so it gets an optimized implementation.
-		DoUnswizzleTex16(texptr, dest, bxc, byc, pitch, rowWidth);
-	} else if (rowWidth == 8) {
-		const u32 *src = (const u32 *) texptr;
-		for (int by = 0; by < byc; by++) {
-			for (int n = 0; n < 8; n++, ydest += 2) {
-				dest[ydest + 0] = *src++;
-				dest[ydest + 1] = *src++;
-				src += 2; // skip two u32
-			}
-		}
-	} else if (rowWidth == 4) {
-		const u32 *src = (const u32 *) texptr;
-		for (int by = 0; by < byc; by++) {
-			for (int n = 0; n < 8; n++, ydest++) {
-				dest[ydest] = *src++;
-				src += 3;
-			}
-		}
-	} else if (rowWidth == 2) {
-		const u16 *src = (const u16 *) texptr;
-		for (int by = 0; by < byc; by++) {
-			for (int n = 0; n < 4; n++, ydest++) {
-				u16 n1 = src[0];
-				u16 n2 = src[8];
-				dest[ydest] = (u32)n1 | ((u32)n2 << 16);
-				src += 16;
-			}
-		}
-	} else if (rowWidth == 1) {
-		const u8 *src = (const u8 *) texptr;
-		for (int by = 0; by < byc; by++) {
-			for (int n = 0; n < 2; n++, ydest++) {
-				u8 n1 = src[ 0];
-				u8 n2 = src[16];
-				u8 n3 = src[32];
-				u8 n4 = src[48];
-				dest[ydest] = (u32)n1 | ((u32)n2 << 8) | ((u32)n3 << 16) | ((u32)n4 << 24);
-				src += 64;
-			}
-		}
-	}
+	DoUnswizzleTex16(texptr, dest, bxc, byc, destPitch);
 }
 
 void *TextureCacheCommon::RearrangeBuf(void *inBuf, u32 inRowBytes, u32 outRowBytes, int h, bool allowInPlace) {

GPU/Common/TextureCacheCommon.h

@@ -143,7 +143,7 @@ protected:
 	// Can't be unordered_map, we use lower_bound ... although for some reason that compiles on MSVC.
 	typedef std::map<u64, TexCacheEntry> TexCache;
 
-	void UnswizzleFromMem(u32 *dest, const u8 *texptr, u32 bufw, u32 height, u32 bytesPerPixel);
+	void UnswizzleFromMem(u32 *dest, u32 destPitch, const u8 *texptr, u32 bufw, u32 height, u32 bytesPerPixel);
 	void *RearrangeBuf(void *inBuf, u32 inRowBytes, u32 outRowBytes, int h, bool allowInPlace = true);
 
 	u32 EstimateTexMemoryUsage(const TexCacheEntry *entry);

GPU/Common/TextureDecoder.cpp

@@ -197,9 +197,14 @@ static u32 QuickTexHashBasic(const void *checkp, u32 size) {
 	return check;
 }
 
-void DoSwizzleTex16(const u32 *ysrcp, u8 *texptr, int bxc, int byc, u32 pitch, u32 rowWidth) {
+void DoSwizzleTex16(const u32 *ysrcp, u8 *texptr, int bxc, int byc, u32 pitch) {
+	// ysrcp is in 32-bits, so this is convenient.
+	const u32 pitchBy32 = pitch >> 2;
 #ifdef _M_SSE
 	__m128i *dest = (__m128i *)texptr;
+	// The pitch parameter is in bytes, so shift down for 128-bit.
+	// Note: it's always aligned to 16 bytes, so this is safe.
+	const u32 pitchBy128 = pitch >> 4;
 	for (int by = 0; by < byc; by++) {
 		const __m128i *xsrc = (const __m128i *)ysrcp;
 		for (int bx = 0; bx < bxc; bx++) {
@@ -207,13 +212,13 @@ void DoSwizzleTex16(const u32 *ysrcp, u8 *texptr, int bxc, int byc, u32 pitch, u
 			for (int n = 0; n < 2; n++) {
 				// Textures are always 16-byte aligned so this is fine.
 				__m128i temp1 = _mm_load_si128(src);
-				src += pitch >> 2;
+				src += pitchBy128;
 				__m128i temp2 = _mm_load_si128(src);
-				src += pitch >> 2;
+				src += pitchBy128;
 				__m128i temp3 = _mm_load_si128(src);
-				src += pitch >> 2;
+				src += pitchBy128;
 				__m128i temp4 = _mm_load_si128(src);
-				src += pitch >> 2;
+				src += pitchBy128;
 
 				_mm_store_si128(dest, temp1);
 				_mm_store_si128(dest + 1, temp2);
@@ -223,7 +228,7 @@ void DoSwizzleTex16(const u32 *ysrcp, u8 *texptr, int bxc, int byc, u32 pitch, u
 			}
 			xsrc++;
 		}
-		ysrcp += (rowWidth * 8) / 4;
+		ysrcp += pitchBy32 * 8;
 	}
 #else
 	u32 *dest = (u32 *)texptr;
@@ -233,19 +238,24 @@ void DoSwizzleTex16(const u32 *ysrcp, u8 *texptr, int bxc, int byc, u32 pitch, u
 			const u32 *src = xsrc;
 			for (int n = 0; n < 8; n++) {
 				memcpy(dest, src, 16);
-				src += pitch;
+				src += pitchBy32;
 				dest += 4;
 			}
 			xsrc += 4;
 		}
-		ysrcp += (rowWidth * 8) / 4;
+		ysrcp += pitchBy32 * 8;
 	}
 #endif
 }
 
-void DoUnswizzleTex16Basic(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 pitch, u32 rowWidth) {
+void DoUnswizzleTex16Basic(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 pitch) {
+	// ydestp is in 32-bits, so this is convenient.
+	const u32 pitchBy32 = pitch >> 2;
 #ifdef _M_SSE
 	const __m128i *src = (const __m128i *)texptr;
+	// The pitch parameter is in bytes, so shift down for 128-bit.
+	// Note: it's always aligned to 16 bytes, so this is safe.
+	const u32 pitchBy128 = pitch >> 4;
 	for (int by = 0; by < byc; by++) {
 		__m128i *xdest = (__m128i *)ydestp;
 		for (int bx = 0; bx < bxc; bx++) {
@@ -257,18 +267,18 @@ void DoUnswizzleTex16Basic(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32
 				__m128i temp3 = _mm_load_si128(src + 2);
 				__m128i temp4 = _mm_load_si128(src + 3);
 				_mm_store_si128(dest, temp1);
-				dest += pitch >> 2;
+				dest += pitchBy128;
 				_mm_store_si128(dest, temp2);
-				dest += pitch >> 2;
+				dest += pitchBy128;
 				_mm_store_si128(dest, temp3);
-				dest += pitch >> 2;
+				dest += pitchBy128;
 				_mm_store_si128(dest, temp4);
-				dest += pitch >> 2;
+				dest += pitchBy128;
 				src += 4;
 			}
 			xdest++;
 		}
-		ydestp += (rowWidth * 8) / 4;
+		ydestp += pitchBy32 * 8;
 	}
 #else
 	const u32 *src = (const u32 *)texptr;
@@ -278,12 +288,12 @@ void DoUnswizzleTex16Basic(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32
 			u32 *dest = xdest;
 			for (int n = 0; n < 8; n++) {
 				memcpy(dest, src, 16);
-				dest += pitch;
+				dest += pitchBy32;
 				src += 4;
 			}
 			xdest += 4;
 		}
-		ydestp += (rowWidth * 8) / 4;
+		ydestp += pitchBy32 * 8;
 	}
 #endif
 }

GPU/Common/TextureDecoder.h

@@ -32,14 +32,16 @@ enum CheckAlphaResult {
 
 void SetupTextureDecoder();
 
-void DoSwizzleTex16(const u32 *ysrcp, u8 *texptr, int bxc, int byc, u32 pitch, u32 rowWidth);
+// Pitch must be aligned to 16 bits (as is the case on a PSP)
+void DoSwizzleTex16(const u32 *ysrcp, u8 *texptr, int bxc, int byc, u32 pitch);
 
 // For SSE, we statically link the SSE2 algorithms.
 #if defined(_M_SSE)
 u32 QuickTexHashSSE2(const void *checkp, u32 size);
 #define DoQuickTexHash QuickTexHashSSE2
 
-void DoUnswizzleTex16Basic(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 pitch, u32 rowWidth);
+// Pitch must be aligned to 16 bits (as is the case on a PSP)
+void DoUnswizzleTex16Basic(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 pitch);
 #define DoUnswizzleTex16 DoUnswizzleTex16Basic
 
 #include "ext/xxhash.h"
@@ -70,7 +72,7 @@ typedef u64 ReliableHashType;
 typedef u32 (*QuickTexHashFunc)(const void *checkp, u32 size);
 extern QuickTexHashFunc DoQuickTexHash;
 
-typedef void (*UnswizzleTex16Func)(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 pitch, u32 rowWidth);
+typedef void (*UnswizzleTex16Func)(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 pitch);
 extern UnswizzleTex16Func DoUnswizzleTex16;
 
 typedef u32 (*ReliableHash32Func)(const void *input, size_t len, u32 seed);

GPU/Common/TextureDecoderNEON.cpp

@@ -118,7 +118,10 @@ u32 QuickTexHashNEON(const void *checkp, u32 size) {
 	return check;
 }
 
-void DoUnswizzleTex16NEON(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 pitch, u32 rowWidth) {
+void DoUnswizzleTex16NEON(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 pitch) {
+	// ydestp is in 32-bits, so this is convenient.
+	const u32 pitchBy32 = pitch >> 2;
+
 	__builtin_prefetch(texptr, 0, 0);
 	__builtin_prefetch(ydestp, 1, 1);
 
@@ -134,18 +137,18 @@ void DoUnswizzleTex16NEON(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 p
 				uint32x4_t temp3 = vld1q_u32(src + 8);
 				uint32x4_t temp4 = vld1q_u32(src + 12);
 				vst1q_u32(dest, temp1);
-				dest += pitch;
+				dest += pitchBy32;
 				vst1q_u32(dest, temp2);
-				dest += pitch;
+				dest += pitchBy32;
 				vst1q_u32(dest, temp3);
-				dest += pitch;
+				dest += pitchBy32;
 				vst1q_u32(dest, temp4);
-				dest += pitch;
+				dest += pitchBy32;
 				src += 16;
 			}
 			xdest += 4;
 		}
-		ydestp += (rowWidth * 8) / 4;
+		ydestp += pitchBy32 * 8;
 	}
 }
 

GPU/Common/TextureDecoderNEON.h

@@ -18,7 +18,7 @@
 #include "GPU/Common/TextureDecoder.h"
 
 u32 QuickTexHashNEON(const void *checkp, u32 size);
-void DoUnswizzleTex16NEON(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 pitch, u32 rowWidth);
+void DoUnswizzleTex16NEON(const u8 *texptr, u32 *ydestp, int bxc, int byc, u32 pitch);
 u32 ReliableHash32NEON(const void *input, size_t len, u32 seed);
 
 CheckAlphaResult CheckAlphaRGBA8888NEON(const u32 *pixelData, int stride, int w, int h);

GPU/Common/TextureScalerCommon.cpp

@@ -495,18 +495,21 @@ TextureScaler::~TextureScaler() {
 bool TextureScaler::IsEmptyOrFlat(u32* data, int pixels, int fmt) {
 	int pixelsPerWord = 4 / BytesPerPixel(fmt);
 	u32 ref = data[0];
+	if (pixelsPerWord > 1 && (ref & 0x0000FFFF) != (ref >> 16)) {
+		return false;
+	}
 	for (int i = 0; i < pixels / pixelsPerWord; ++i) {
 		if (data[i] != ref) return false;
 	}
 	return true;
 }
 
-void TextureScaler::Scale(u32* &data, u32 &dstFmt, int &width, int &height, int factor) {
+bool TextureScaler::Scale(u32* &data, u32 &dstFmt, int &width, int &height, int factor) {
 	// prevent processing empty or flat textures (this happens a lot in some games)
 	// doesn't hurt the standard case, will be very quick for textures with actual texture
 	if (IsEmptyOrFlat(data, width*height, dstFmt)) {
 		INFO_LOG(G3D, "TextureScaler: early exit -- empty/flat texture");
-		return;
+		return false;
 	}
 
 #ifdef SCALING_MEASURE_TIME
@@ -559,6 +562,8 @@ void TextureScaler::Scale(u32* &data, u32 &dstFmt, int &width, int &height, int
 			width*height, t, (width*height) / (t * 1000 * 1000));
 	}
 #endif
+
+	return true;
 }
 
 void TextureScaler::ScaleXBRZ(int factor, u32* source, u32* dest, int width, int height) {

GPU/Common/TextureScalerCommon.h

@@ -27,11 +27,11 @@ public:
 	TextureScaler();
 	~TextureScaler();
 
-	void Scale(u32* &data, u32 &dstfmt, int &width, int &height, int factor);
+	bool Scale(u32* &data, u32 &dstfmt, int &width, int &height, int factor);
 
 	enum { XBRZ = 0, HYBRID = 1, BICUBIC = 2, HYBRID_BICUBIC = 3 };
 
-private:
+protected:
 	virtual void ConvertTo8888(u32 format, u32* source, u32* &dest, int width, int height) = 0;
 	virtual int BytesPerPixel(u32 format) = 0;
 	virtual u32 Get8888Format() = 0;

GPU/Directx9/TextureCacheDX9.cpp

@@ -377,7 +377,7 @@ void *TextureCacheDX9::ReadIndexedTex(int level, const u8 *texptr, int bytesPerI
 			}
 		} else {
 			tmpTexBuf32.resize(std::max(bufw, w) * h);
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, bytesPerIndex);
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw * bytesPerIndex, texptr, bufw, h, bytesPerIndex);
 			switch (bytesPerIndex) {
 			case 1:
 				DeIndexTexture(tmpTexBuf16.data(), (u8 *) tmpTexBuf32.data(), length, clut);
@@ -417,7 +417,7 @@ void *TextureCacheDX9::ReadIndexedTex(int level, const u8 *texptr, int bytesPerI
 			}
 			buf = tmpTexBuf32.data();
 		} else {
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, bytesPerIndex);
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw * bytesPerIndex, texptr, bufw, h, bytesPerIndex);
 			// Since we had to unswizzle to tmpTexBuf32, let's output to tmpTexBuf16.
 			tmpTexBuf16.resize(std::max(bufw, w) * h * 2);
 			u32 *dest32 = (u32 *) tmpTexBuf16.data();
@@ -1366,7 +1366,7 @@ void *TextureCacheDX9::DecodeTextureLevel(GETextureFormat format, GEPaletteForma
 				}
 			} else {
 				tmpTexBuf32.resize(std::max(bufw, w) * h);
-				UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 0);
+				UnswizzleFromMem(tmpTexBuf32.data(), bufw / 2, texptr, bufw, h, 0);
 				if (clutAlphaLinear_ && mipmapShareClut) {
 					DeIndexTexture4OptimalRev(tmpTexBuf16.data(), (const u8 *)tmpTexBuf32.data(), bufw * h, clutAlphaLinearColor_);
 				} else {
@@ -1386,7 +1386,7 @@ void *TextureCacheDX9::DecodeTextureLevel(GETextureFormat format, GEPaletteForma
 				DeIndexTexture4(tmpTexBuf32.data(), texptr, bufw * h, clut);
 				finalBuf = tmpTexBuf32.data();
 			} else {
-				UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 0);
+				UnswizzleFromMem(tmpTexBuf32.data(), bufw / 2, texptr, bufw, h, 0);
 				// Let's reuse tmpTexBuf16, just need double the space.
 				tmpTexBuf16.resize(std::max(bufw, w) * h * 2);
 				DeIndexTexture4((u32 *)tmpTexBuf16.data(), (u8 *)tmpTexBuf32.data(), bufw * h, clut);
@@ -1431,7 +1431,7 @@ void *TextureCacheDX9::DecodeTextureLevel(GETextureFormat format, GEPaletteForma
 		}
 		else {
 			tmpTexBuf32.resize(std::max(bufw, w) * h);
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 2);
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw * 2, texptr, bufw, h, 2);
 			finalBuf = tmpTexBuf32.data();
 		}
 		break;
@@ -1451,7 +1451,7 @@ void *TextureCacheDX9::DecodeTextureLevel(GETextureFormat format, GEPaletteForma
 			}
 		} else {
 			tmpTexBuf32.resize(std::max(bufw, w) * h);
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 4);
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw * 4, texptr, bufw, h, 4);
 			finalBuf = tmpTexBuf32.data();
 		}
 		break;

GPU/GLES/Framebuffer.cpp

@@ -2024,9 +2024,9 @@ bool FramebufferManager::GetDepthbuffer(u32 fb_address, int fb_stride, u32 z_add
 	}
 
 	if (gstate_c.Supports(GPU_SCALE_DEPTH_FROM_24BIT_TO_16BIT)) {
-		buffer.Allocate(vfb->renderWidth, vfb->renderHeight, GPU_DBG_FORMAT_FLOAT_DIV_256, false);
+		buffer.Allocate(vfb->renderWidth, vfb->renderHeight, GPU_DBG_FORMAT_FLOAT_DIV_256, !useBufferedRendering_);
 	} else {
-		buffer.Allocate(vfb->renderWidth, vfb->renderHeight, GPU_DBG_FORMAT_FLOAT, false);
+		buffer.Allocate(vfb->renderWidth, vfb->renderHeight, GPU_DBG_FORMAT_FLOAT, !useBufferedRendering_);
 	}
 	if (vfb->fbo)
 		fbo_bind_for_read(vfb->fbo);
@@ -2052,7 +2052,7 @@ bool FramebufferManager::GetStencilbuffer(u32 fb_address, int fb_stride, GPUDebu
 	}
 
 #ifndef USING_GLES2
-	buffer.Allocate(vfb->renderWidth, vfb->renderHeight, GPU_DBG_FORMAT_8BIT, false);
+	buffer.Allocate(vfb->renderWidth, vfb->renderHeight, GPU_DBG_FORMAT_8BIT, !useBufferedRendering_);
 	if (vfb->fbo)
 		fbo_bind_for_read(vfb->fbo);
 	if (gl_extensions.GLES3 || !gl_extensions.IsGLES)

GPU/GLES/TextureCache.cpp

@@ -375,7 +375,7 @@ void *TextureCache::ReadIndexedTex(int level, const u8 *texptr, int bytesPerInde
 			}
 		} else {
 			tmpTexBuf32.resize(std::max(bufw, w) * h);
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, bytesPerIndex);
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw * bytesPerIndex, texptr, bufw, h, bytesPerIndex);
 			switch (bytesPerIndex) {
 			case 1:
 				DeIndexTexture(tmpTexBuf16.data(), (u8 *) tmpTexBuf32.data(), length, clut);
@@ -415,7 +415,7 @@ void *TextureCache::ReadIndexedTex(int level, const u8 *texptr, int bytesPerInde
 			}
 			buf = tmpTexBuf32.data();
 		} else {
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, bytesPerIndex);
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw * bytesPerIndex, texptr, bufw, h, bytesPerIndex);
 			// Since we had to unswizzle to tmpTexBuf32, let's output to tmpTexBuf16.
 			tmpTexBuf16.resize(std::max(bufw, w) * h * 2);
 			u32 *dest32 = (u32 *) tmpTexBuf16.data();
@@ -1501,7 +1501,7 @@ void *TextureCache::DecodeTextureLevel(GETextureFormat format, GEPaletteFormat c
 				}
 			} else {
 				tmpTexBuf32.resize(std::max(bufw, w) * h);
-				UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 0);
+				UnswizzleFromMem(tmpTexBuf32.data(), bufw / 2, texptr, bufw, h, 0);
 				if (clutAlphaLinear_ && mipmapShareClut) {
 					DeIndexTexture4Optimal(tmpTexBuf16.data(), (const u8 *)tmpTexBuf32.data(), bufw * h, clutAlphaLinearColor_);
 				} else {
@@ -1521,7 +1521,7 @@ void *TextureCache::DecodeTextureLevel(GETextureFormat format, GEPaletteFormat c
 				DeIndexTexture4(tmpTexBuf32.data(), texptr, bufw * h, clut);
 				finalBuf = tmpTexBuf32.data();
 			} else {
-				UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 0);
+				UnswizzleFromMem(tmpTexBuf32.data(), bufw / 2, texptr, bufw, h, 0);
 				// Let's reuse tmpTexBuf16, just need double the space.
 				tmpTexBuf16.resize(std::max(bufw, w) * h * 2);
 				DeIndexTexture4((u32 *)tmpTexBuf16.data(), (u8 *)tmpTexBuf32.data(), bufw * h, clut);
@@ -1565,7 +1565,7 @@ void *TextureCache::DecodeTextureLevel(GETextureFormat format, GEPaletteFormat c
 			ConvertColors(finalBuf, texptr, dstFmt, bufw * h);
 		} else {
 			tmpTexBuf32.resize(std::max(bufw, w) * h);
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 2);
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw * 2, texptr, bufw, h, 2);
 			finalBuf = tmpTexBuf32.data();
 			ConvertColors(finalBuf, finalBuf, dstFmt, bufw * h);
 		}
@@ -1590,7 +1590,7 @@ void *TextureCache::DecodeTextureLevel(GETextureFormat format, GEPaletteFormat c
 			}
 		} else {
 			tmpTexBuf32.resize(std::max(bufw, w) * h);
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 4);
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw * 4, texptr, bufw, h, 4);
 			finalBuf = tmpTexBuf32.data();
 			ConvertColors(finalBuf, finalBuf, dstFmt, bufw * h);
 		}

GPU/Vulkan/TextureCacheVulkan.cpp

@@ -402,104 +402,75 @@ bool TextureCacheVulkan::AttachFramebuffer(TexCacheEntry *entry, u32 address, Vi
 	return false;
 }
 
-void *TextureCacheVulkan::ReadIndexedTex(int level, const u8 *texptr, int bytesPerIndex, VkFormat dstFmt, int bufw) {
+bool TextureCacheVulkan::ReadIndexedTex(u8 *out, int outPitch, int level, const u8 *texptr, int bytesPerIndex, VkFormat dstFmt, int bufw) {
 	int w = gstate.getTextureWidth(level);
 	int h = gstate.getTextureHeight(level);
-	int length = bufw * h;
-	void *buf = NULL;
+
+	if (gstate.isTextureSwizzled()) {
+		tmpTexBuf32.resize(bufw * ((h + 7) & ~7));
+		UnswizzleFromMem(tmpTexBuf32.data(), bufw * bytesPerIndex, texptr, bufw, h, bytesPerIndex);
+		texptr = (u8 *)tmpTexBuf32.data();
+	}
+
 	switch (gstate.getClutPaletteFormat()) {
 	case GE_CMODE_16BIT_BGR5650:
 	case GE_CMODE_16BIT_ABGR5551:
 	case GE_CMODE_16BIT_ABGR4444:
 	{
-		tmpTexBuf16.resize(std::max(bufw, w) * h);
 		const u16 *clut = GetCurrentClut<u16>();
-		if (!gstate.isTextureSwizzled()) {
-			switch (bytesPerIndex) {
-			case 1:
-				DeIndexTexture(tmpTexBuf16.data(), (const u8 *)texptr, length, clut);
-				break;
-
-			case 2:
-				DeIndexTexture(tmpTexBuf16.data(), (const u16_le *)texptr, length, clut);
-				break;
-
-			case 4:
-				DeIndexTexture(tmpTexBuf16.data(), (const u32_le *)texptr, length, clut);
-				break;
+		switch (bytesPerIndex) {
+		case 1:
+			for (int y = 0; y < h; ++y) {
+				DeIndexTexture((u16 *)(out + outPitch * y), (const u8 *)texptr + bufw * y, w, clut);
 			}
-		} else {
-			tmpTexBuf32.resize(std::max(bufw, w) * h);
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, bytesPerIndex);
-			switch (bytesPerIndex) {
-			case 1:
-				DeIndexTexture(tmpTexBuf16.data(), (u8 *)tmpTexBuf32.data(), length, clut);
-				break;
+			break;
 
-			case 2:
-				DeIndexTexture(tmpTexBuf16.data(), (u16 *)tmpTexBuf32.data(), length, clut);
-				break;
-
-			case 4:
-				DeIndexTexture(tmpTexBuf16.data(), (u32 *)tmpTexBuf32.data(), length, clut);
-				break;
+		case 2:
+			for (int y = 0; y < h; ++y) {
+				DeIndexTexture((u16 *)(out + outPitch * y), (const u16_le *)texptr + bufw * y, w, clut);
 			}
+			break;
+
+		case 4:
+			for (int y = 0; y < h; ++y) {
+				DeIndexTexture((u16 *)(out + outPitch * y), (const u32_le *)texptr + bufw * y, w, clut);
+			}
+			break;
 		}
-		buf = tmpTexBuf16.data();
 	}
 	break;
 
 	case GE_CMODE_32BIT_ABGR8888:
 	{
-		tmpTexBuf32.resize(std::max(bufw, w) * h);
 		const u32 *clut = GetCurrentClut<u32>();
-		if (!gstate.isTextureSwizzled()) {
-			switch (bytesPerIndex) {
-			case 1:
-				DeIndexTexture(tmpTexBuf32.data(), (const u8 *)texptr, length, clut);
-				break;
-
-			case 2:
-				DeIndexTexture(tmpTexBuf32.data(), (const u16_le *)texptr, length, clut);
-				break;
-
-			case 4:
-				DeIndexTexture(tmpTexBuf32.data(), (const u32_le *)texptr, length, clut);
-				break;
+		switch (bytesPerIndex) {
+		case 1:
+			for (int y = 0; y < h; ++y) {
+				DeIndexTexture((u32 *)(out + outPitch * y), (const u8 *)texptr + bufw * y, w, clut);
 			}
-			buf = tmpTexBuf32.data();
-		} else {
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, bytesPerIndex);
-			// Since we had to unswizzle to tmpTexBuf32, let's output to tmpTexBuf16.
-			tmpTexBuf16.resize(std::max(bufw, w) * h * 2);
-			u32 *dest32 = (u32 *)tmpTexBuf16.data();
-			switch (bytesPerIndex) {
-			case 1:
-				DeIndexTexture(dest32, (u8 *)tmpTexBuf32.data(), length, clut);
-				buf = dest32;
-				break;
+			break;
 
-			case 2:
-				DeIndexTexture(dest32, (u16 *)tmpTexBuf32.data(), length, clut);
-				buf = dest32;
-				break;
-
-			case 4:
-				// TODO: If a game actually uses this mode, check if using dest32 or tmpTexBuf32 is faster.
-				DeIndexTexture(tmpTexBuf32.data(), tmpTexBuf32.data(), length, clut);
-				buf = tmpTexBuf32.data();
-				break;
+		case 2:
+			for (int y = 0; y < h; ++y) {
+				DeIndexTexture((u32 *)(out + outPitch * y), (const u16_le *)texptr + bufw * y, w, clut);
 			}
+			break;
+
+		case 4:
+			for (int y = 0; y < h; ++y) {
+				DeIndexTexture((u32 *)(out + outPitch * y), (const u32_le *)texptr + bufw * y, w, clut);
+			}
+			break;
 		}
 	}
 	break;
 
 	default:
-		ERROR_LOG_REPORT(G3D, "Unhandled clut texture mode %d!!!", (gstate.clutformat & 3));
-		break;
+		ERROR_LOG_REPORT(G3D, "Unhandled clut texture mode %d!!!", gstate.getClutPaletteFormat());
+		return false;
 	}
 
-	return buf;
+	return true;
 }
 
 VkFormat getClutDestFormatVulkan(GEPaletteFormat format) {
@@ -638,9 +609,9 @@ void TextureCacheVulkan::UpdateCurrentClut(GEPaletteFormat clutFormat, u32 clutB
 	if (clutFormat == GE_CMODE_16BIT_ABGR4444 && clutIndexIsSimple) {
 		const u16_le *clut = GetCurrentClut<u16_le>();
 		clutAlphaLinear_ = true;
-		clutAlphaLinearColor_ = clut[15] & 0xFFF0;
+		clutAlphaLinearColor_ = clut[15] & 0x0FFF;
 		for (int i = 0; i < 16; ++i) {
-			u16 step = clutAlphaLinearColor_ | i;
+			u16 step = clutAlphaLinearColor_ | (i << 12);
 			if (clut[i] != step) {
 				clutAlphaLinear_ = false;
 				break;
@@ -1415,9 +1386,7 @@ VkFormat TextureCacheVulkan::GetDestFormat(GETextureFormat format, GEPaletteForm
 	}
 }
 
-void *TextureCacheVulkan::DecodeTextureLevel(u8 *out, int outPitch, GETextureFormat format, GEPaletteFormat clutformat, uint32_t texaddr, int level, VkFormat dstFmt, int scaleFactor, int bufw) {
-	void *finalBuf = NULL;
-
+bool TextureCacheVulkan::DecodeTextureLevel(u8 *out, int outPitch, GETextureFormat format, GEPaletteFormat clutformat, uint32_t texaddr, int level, VkFormat dstFmt, int scaleFactor, int bufw) {
 	bool swizzled = gstate.isTextureSwizzled();
 	if ((texaddr & 0x00600000) != 0 && Memory::IsVRAMAddress(texaddr)) {
 		// This means it's in a mirror, possibly a swizzled mirror.  Let's report.
@@ -1439,66 +1408,62 @@ void *TextureCacheVulkan::DecodeTextureLevel(u8 *out, int outPitch, GETextureFor
 		const bool mipmapShareClut = gstate.isClutSharedForMipmaps();
 		const int clutSharingOffset = mipmapShareClut ? 0 : level * 16;
 
+		if (swizzled) {
+			tmpTexBuf32.resize(bufw * ((h + 7) & ~7));
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw / 2, texptr, bufw, h, 0);
+			texptr = (u8 *)tmpTexBuf32.data();
+		}
+
 		switch (clutformat) {
 		case GE_CMODE_16BIT_BGR5650:
 		case GE_CMODE_16BIT_ABGR5551:
 		case GE_CMODE_16BIT_ABGR4444:
 		{
-			tmpTexBuf16.resize(std::max(bufw, w) * h);
 			const u16 *clut = GetCurrentClut<u16>() + clutSharingOffset;
-			if (!swizzled) {
-				if (clutAlphaLinear_ && mipmapShareClut) {
-					DeIndexTexture4Optimal(tmpTexBuf16.data(), texptr, bufw * h, clutAlphaLinearColor_);
-				} else {
-					DeIndexTexture4(tmpTexBuf16.data(), texptr, bufw * h, clut);
+			if (clutAlphaLinear_ && mipmapShareClut) {
+				for (int y = 0; y < h; ++y) {
+					DeIndexTexture4OptimalRev((u16 *)(out + outPitch * y), texptr + (bufw * y) / 2, w, clutAlphaLinearColor_);
 				}
 			} else {
-				tmpTexBuf32.resize(std::max(bufw, w) * h);
-				UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 0);
-				if (clutAlphaLinear_ && mipmapShareClut) {
-					DeIndexTexture4Optimal(tmpTexBuf16.data(), (const u8 *)tmpTexBuf32.data(), bufw * h, clutAlphaLinearColor_);
-				} else {
-					DeIndexTexture4(tmpTexBuf16.data(), (const u8 *)tmpTexBuf32.data(), bufw * h, clut);
+				for (int y = 0; y < h; ++y) {
+					DeIndexTexture4((u16 *)(out + outPitch * y), texptr + (bufw * y) / 2, w, clut);
 				}
 			}
-			finalBuf = tmpTexBuf16.data();
 		}
 		break;
 
 		case GE_CMODE_32BIT_ABGR8888:
 		{
-			tmpTexBuf32.resize(std::max(bufw, w) * h);
 			const u32 *clut = GetCurrentClut<u32>() + clutSharingOffset;
-			if (!swizzled) {
-				DeIndexTexture4(tmpTexBuf32.data(), texptr, bufw * h, clut);
-				finalBuf = tmpTexBuf32.data();
-			} else {
-				UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 0);
-				// Let's reuse tmpTexBuf16, just need double the space.
-				tmpTexBuf16.resize(std::max(bufw, w) * h * 2);
-				DeIndexTexture4((u32 *)tmpTexBuf16.data(), (u8 *)tmpTexBuf32.data(), bufw * h, clut);
-				finalBuf = tmpTexBuf16.data();
+			for (int y = 0; y < h; ++y) {
+				DeIndexTexture4((u32 *)(out + outPitch * y), texptr + (bufw * y) / 2, w, clut);
 			}
 		}
 		break;
 
 		default:
 			ERROR_LOG_REPORT(G3D, "Unknown CLUT4 texture mode %d", gstate.getClutPaletteFormat());
-			return NULL;
+			return false;
 		}
 	}
 	break;
 
 	case GE_TFMT_CLUT8:
-		finalBuf = ReadIndexedTex(level, texptr, 1, dstFmt, bufw);
+		if (!ReadIndexedTex(out, outPitch, level, texptr, 1, dstFmt, bufw)) {
+			return false;
+		}
 		break;
 
 	case GE_TFMT_CLUT16:
-		finalBuf = ReadIndexedTex(level, texptr, 2, dstFmt, bufw);
+		if (!ReadIndexedTex(out, outPitch, level, texptr, 2, dstFmt, bufw)) {
+			return false;
+		}
 		break;
 
 	case GE_TFMT_CLUT32:
-		finalBuf = ReadIndexedTex(level, texptr, 4, dstFmt, bufw);
+		if (!ReadIndexedTex(out, outPitch, level, texptr, 4, dstFmt, bufw)) {
+			return false;
+		}
 		break;
 
 	case GE_TFMT_4444:
@@ -1509,15 +1474,16 @@ void *TextureCacheVulkan::DecodeTextureLevel(u8 *out, int outPitch, GETextureFor
 			for (int y = 0; y < h; ++y) {
 				memcpy(out + outPitch * y, texptr + bufw * sizeof(u16) * y, w * sizeof(u16));
 			}
-			finalBuf = out;
+		} else if (h >= 8) {
+			UnswizzleFromMem((u32 *)out, outPitch, texptr, bufw, h, 2);
 		} else {
-			tmpTexBuf32.resize(std::max(bufw, w) * h);
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 2);
+			// We don't have enough space for all rows in out, so use a temp buffer.
+			tmpTexBuf32.resize(bufw * ((h + 7) & ~7));
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw * 2, texptr, bufw, h, 2);
 			const u8 *unswizzled = (u8 *)tmpTexBuf32.data();
 			for (int y = 0; y < h; ++y) {
 				memcpy(out + outPitch * y, unswizzled + bufw * sizeof(u16) * y, w * sizeof(u16));
 			}
-			finalBuf = out;
 		}
 		break;
 
@@ -1526,15 +1492,16 @@ void *TextureCacheVulkan::DecodeTextureLevel(u8 *out, int outPitch, GETextureFor
 			for (int y = 0; y < h; ++y) {
 				memcpy(out + outPitch * y, texptr + bufw * sizeof(u32) * y, w * sizeof(u32));
 			}
-			finalBuf = out;
+		} else if (h >= 8) {
+			UnswizzleFromMem((u32 *)out, outPitch, texptr, bufw, h, 4);
 		} else {
-			tmpTexBuf32.resize(std::max(bufw, w) * h);
-			UnswizzleFromMem(tmpTexBuf32.data(), texptr, bufw, h, 4);
+			// We don't have enough space for all rows in out, so use a temp buffer.
+			tmpTexBuf32.resize(bufw * ((h + 7) & ~7));
+			UnswizzleFromMem(tmpTexBuf32.data(), bufw * 4, texptr, bufw, h, 4);
 			const u8 *unswizzled = (u8 *)tmpTexBuf32.data();
 			for (int y = 0; y < h; ++y) {
 				memcpy(out + outPitch * y, unswizzled + bufw * sizeof(u32) * y, w * sizeof(u32));
 			}
-			finalBuf = out;
 		}
 		break;
 
@@ -1554,7 +1521,6 @@ void *TextureCacheVulkan::DecodeTextureLevel(u8 *out, int outPitch, GETextureFor
 		}
 		// TODO: Not height also?
 		w = (w + 3) & ~3;
-		finalBuf = out;
 	}
 	break;
 
@@ -1574,7 +1540,6 @@ void *TextureCacheVulkan::DecodeTextureLevel(u8 *out, int outPitch, GETextureFor
 		}
 		// TODO: Not height also?
 		w = (w + 3) & ~3;
-		finalBuf = out;
 	}
 	break;
 
@@ -1594,20 +1559,15 @@ void *TextureCacheVulkan::DecodeTextureLevel(u8 *out, int outPitch, GETextureFor
 		}
 		// TODO: Not height also?
 		w = (w + 3) & ~3;
-		finalBuf = out;
 	}
 	break;
 
 	default:
 		ERROR_LOG_REPORT(G3D, "Unknown Texture Format %d!!!", format);
-		return NULL;
+		return false;
 	}
 
-	if (!finalBuf) {
-		ERROR_LOG_REPORT(G3D, "NO finalbuf! Will crash!");
-	}
-
-	return finalBuf;
+	return true;
 }
 
 TextureCacheVulkan::TexCacheEntry::Status TextureCacheVulkan::CheckAlpha(const u32 *pixelData, VkFormat dstFmt, int stride, int w, int h) {
@@ -1656,37 +1616,20 @@ void TextureCacheVulkan::LoadTextureLevel(TexCacheEntry &entry, uint8_t *writePt
 			decPitch = w * bpp;
 		}
 
-		void *finalBuf = DecodeTextureLevel((u8 *)pixelData, decPitch, tfmt, clutformat, texaddr, level, dstFmt, scaleFactor, bufw);
-		if (finalBuf == NULL) {
+		bool decSuccess = DecodeTextureLevel((u8 *)pixelData, decPitch, tfmt, clutformat, texaddr, level, dstFmt, scaleFactor, bufw);
+		if (!decSuccess) {
+			memset(writePtr, 0, rowPitch * h);
 			return;
 		}
-		if (finalBuf != pixelData) {
-			// Since we didn't write to pixelData, we must've ignored decPitch too.
-			// TODO: Can remove once all formats are decoded directly and respect decPitch.
-			decPitch = bufw * bpp;
-		}
-		pixelData = (u32 *)finalBuf;
 		rowBytes = w * bpp;
 		gpuStats.numTexturesDecoded++;
 
 		if (scaleFactor > 1) {
 			u32 fmt = dstFmt;
-			// We need to repack it sometimes.
-			// TODO: Can remove once all formats are decoded directly and respect decPitch.
-			if (decPitch != w * bpp) {
-				tmpTexBufRearrange.resize(w * h);
-				const u8 *src = (u8 *)pixelData;
-				u8 *dst = (u8 *)tmpTexBufRearrange.data();
-				for (int y = 0; y < h; y++) {
-					memcpy(dst + rowBytes * y, src + decPitch * y, rowBytes);
-				}
-				pixelData = (u32 *)dst;
-			}
-
-			scaler.Scale(pixelData, fmt, w, h, scaleFactor);
+			scaler.ScaleAlways(pixelData, fmt, w, h, scaleFactor);
 			dstFmt = (VkFormat)fmt;
 
-			// We always end up at 8888.  Other parts check for this.
+			// We always end up at 8888.  Other parts assume this.
 			assert(dstFmt == VULKAN_8888_FORMAT);
 			decPitch = w * sizeof(u32);
 			rowBytes = w * sizeof(u32);
@@ -1700,13 +1643,9 @@ void TextureCacheVulkan::LoadTextureLevel(TexCacheEntry &entry, uint8_t *writePt
 		}
 	}
 
-	if (replaceImages) {
-		// TODO: No support for texture shadows
-		// DebugBreak();
-	}
-
 	PROFILE_THIS_SCOPE("loadtex");
 	if (pixelData != (u32 *)writePtr) {
+		// This is used when texture scaling was enabled.
 		for (int y = 0; y < h; y++) {
 			memcpy(writePtr + rowPitch * y, (const uint8_t *)pixelData + decPitch * y, rowBytes);
 		}

GPU/Vulkan/TextureCacheVulkan.h

@@ -124,11 +124,11 @@ protected:
 private:
 	void Decimate();  // Run this once per frame to get rid of old textures.
 	void DeleteTexture(TexCache::iterator it);
-	void *ReadIndexedTex(int level, const u8 *texptr, int bytesPerIndex, VkFormat dstFmt, int bufw);
+	bool ReadIndexedTex(u8 *out, int outPitch, int level, const u8 *texptr, int bytesPerIndex, VkFormat dstFmt, int bufw);
 	void UpdateSamplingParams(TexCacheEntry &entry, SamplerCacheKey &key);
 	void LoadTextureLevel(TexCacheEntry &entry, uint8_t *writePtr, int rowPitch,  int level, bool replaceImages, int scaleFactor, VkFormat dstFmt);
 	VkFormat GetDestFormat(GETextureFormat format, GEPaletteFormat clutFormat) const;
-	void *DecodeTextureLevel(u8 *out, int outPitch, GETextureFormat format, GEPaletteFormat clutformat, uint32_t texaddr, int level, VkFormat dstFmt, int scaleFactor, int bufw);
+	bool DecodeTextureLevel(u8 *out, int outPitch, GETextureFormat format, GEPaletteFormat clutformat, uint32_t texaddr, int level, VkFormat dstFmt, int scaleFactor, int bufw);
 	TexCacheEntry::Status CheckAlpha(const u32 *pixelData, VkFormat dstFmt, int stride, int w, int h);
 	template <typename T>
 	const T *GetCurrentClut();

GPU/Vulkan/TextureScalerVulkan.cpp

@@ -44,6 +44,39 @@ u32 TextureScalerVulkan::Get8888Format() {
 	return VULKAN_8888_FORMAT;
 }
 
+void TextureScalerVulkan::ScaleAlways(u32 *&data, u32 &dstFmt, int &width, int &height, int factor) {
+	if (!Scale(data, dstFmt, width, height, factor)) {
+		// This means it was a flat texture.  Vulkan wants the size up front, so we need to make it happen.
+		assert(IsEmptyOrFlat(data, width * height, dstFmt));
+
+		u32 pixel;
+		// Since it's flat, one pixel is enough.  It might end up pointing to data, though.
+		u32 *pixelPointer = &pixel;
+		ConvertTo8888(dstFmt, data, pixelPointer, 1, 1);
+		if (pixelPointer != &pixel) {
+			pixel = *pixelPointer;
+		}
+
+		bufOutput.resize(width * height * factor * factor);
+		dstFmt = Get8888Format();
+		data = bufOutput.data();
+		width *= factor;
+		height *= factor;
+
+		// ABCD.  If A = D, and AB = CD, then they must all be equal (B = C, etc.)
+		if ((pixel & 0x000000FF) == (pixel >> 24) && (pixel & 0x0000FFFF) == (pixel >> 16)) {
+			memset(data, pixel & 0xFF, width * height * sizeof(u32));
+		} else {
+			// TODO: Optimize.
+			for (int y = 0; y < height; ++y) {
+				for (int x = 0; x < width; ++x) {
+					data[y * width + x] = pixel;
+				}
+			}
+		}
+	}
+}
+
 void TextureScalerVulkan::ConvertTo8888(u32 format, u32* source, u32* &dest, int width, int height) {
 	switch (format) {
 	case VULKAN_8888_FORMAT:

GPU/Vulkan/TextureScalerVulkan.h

@@ -21,6 +21,10 @@
 #include "GPU/Common/TextureScalerCommon.h"
 
 class TextureScalerVulkan : public TextureScaler {
+public:
+	void ScaleAlways(u32* &data, u32 &dstFmt, int &width, int &height, int factor);
+
+protected:
 	void ConvertTo8888(u32 format, u32* source, u32* &dest, int width, int height) override;
 	int BytesPerPixel(u32 format) override;
 	u32 Get8888Format() override;