Enable partial compressed texture uploads

examples/rhi/texuploads/texuploads.cpp

@@ -211,7 +211,7 @@ void Window::customRender()
             mipDesc.image = d.customImage;
             // The image here is smaller than the original. Use a non-zero position
             // to make it more interesting.
-            mipDesc.destinationTopLeft = QPointF(100, 20);
+            mipDesc.destinationTopLeft = QPoint(100, 20);
 
             layer.mipImages.append(mipDesc);
             desc.layers.append(layer);

src/rhi/qrhi.cpp

@@ -181,7 +181,7 @@ QRhiImplementation::~QRhiImplementation()
     qDeleteAll(resUpdPool);
 }
 
-bool QRhiImplementation::isCompressedFormat(QRhiTexture::Format format)
+bool QRhiImplementation::isCompressedFormat(QRhiTexture::Format format) const
 {
     return (format >= QRhiTexture::BC1 && format <= QRhiTexture::BC7)
             || (format >= QRhiTexture::ETC2_RGB8 && format <= QRhiTexture::ETC2_RGBA8)
@@ -189,7 +189,8 @@ bool QRhiImplementation::isCompressedFormat(QRhiTexture::Format format)
 }
 
 void QRhiImplementation::compressedFormatInfo(QRhiTexture::Format format, const QSize &size,
-                                              quint32 *bpl, quint32 *byteSize)
+                                              quint32 *bpl, quint32 *byteSize,
+                                              QSize *blockDim) const
 {
     int xdim = 4;
     int ydim = 4;
@@ -303,13 +304,15 @@ void QRhiImplementation::compressedFormatInfo(QRhiTexture::Format format, const
         *bpl = wblocks * blockSize;
     if (byteSize)
         *byteSize = wblocks * hblocks * blockSize;
+    if (blockDim)
+        *blockDim = QSize(xdim, ydim);
 }
 
 void QRhiImplementation::textureFormatInfo(QRhiTexture::Format format, const QSize &size,
-                                           quint32 *bpl, quint32 *byteSize)
+                                           quint32 *bpl, quint32 *byteSize) const
 {
     if (isCompressedFormat(format)) {
-        compressedFormatInfo(format, size, bpl, byteSize);
+        compressedFormatInfo(format, size, bpl, byteSize, nullptr);
         return;
     }
 

src/rhi/qrhi.h

@@ -269,7 +269,8 @@ struct Q_RHI_EXPORT QRhiTextureUploadDescription
             MipLevel(const QByteArray &compressedData_) : compressedData(compressedData_) { }
             QImage image;
             QByteArray compressedData;
-            QPointF destinationTopLeft; // for non-compressed only. (0, 0) by default.
+            QPoint destinationTopLeft;
+            QSize compressedPixelSize; // empty = entire subresource; ignored for non-compressed as the QImage size is used instead
         };
 
         Layer() { }
@@ -288,7 +289,7 @@ Q_DECLARE_TYPEINFO(QRhiTextureUploadDescription, Q_MOVABLE_TYPE);
 
 struct Q_RHI_EXPORT QRhiTextureCopyDescription
 {
-    QSize pixelSize; // empty = entire texture
+    QSize pixelSize; // empty = entire subresource
 
     int sourceLayer = 0;
     int sourceLevel = 0;

src/rhi/qrhi_p.h

@@ -121,13 +121,14 @@ public:
     virtual QMatrix4x4 clipSpaceCorrMatrix() const = 0;
     virtual bool isTextureFormatSupported(QRhiTexture::Format format, QRhiTexture::Flags flags) const = 0;
 
-protected:
-    bool isCompressedFormat(QRhiTexture::Format format);
+    bool isCompressedFormat(QRhiTexture::Format format) const;
     void compressedFormatInfo(QRhiTexture::Format format, const QSize &size,
-                              quint32 *bpl, quint32 *byteSize);
+                              quint32 *bpl, quint32 *byteSize,
+                              QSize *blockDim) const;
     void textureFormatInfo(QRhiTexture::Format format, const QSize &size,
-                           quint32 *bpl, quint32 *byteSize);
+                           quint32 *bpl, quint32 *byteSize) const;
 
+protected:
     QVector<QRhiResourceUpdateBatch *> resUpdPool;
     QBitArray resUpdPoolMap;
 

src/rhi/qrhid3d11.cpp

@@ -743,25 +743,39 @@ void QRhiD3D11::commitResourceUpdates(QRhiResourceUpdateBatch *resourceUpdates)
             for (int level = 0, levelCount = layerDesc.mipImages.count(); level != levelCount; ++level) {
                 const QRhiTextureUploadDescription::Layer::MipLevel mipDesc(layerDesc.mipImages[level]);
                 UINT subres = D3D11CalcSubresource(level, layer, texD->mipLevelCount);
+                const int x = mipDesc.destinationTopLeft.x();
+                const int y = mipDesc.destinationTopLeft.y();
+                D3D11_BOX box;
+                box.front = 0;
+                // back, right, bottom are exclusive
+                box.back = 1;
                 if (!mipDesc.image.isNull()) {
-                    const float x = mipDesc.destinationTopLeft.x();
-                    const float y = mipDesc.destinationTopLeft.y();
-                    D3D11_BOX box;
                     box.left = x;
                     box.top = y;
-                    box.front = 0;
-                    // back, right, bottom are exclusive
                     box.right = x + mipDesc.image.width();
                     box.bottom = y + mipDesc.image.height();
-                    box.back = 1;
                     context->UpdateSubresource(texD->tex, subres, &box,
                                                mipDesc.image.constBits(), mipDesc.image.bytesPerLine(), 0);
                 } else if (!mipDesc.compressedData.isEmpty() && isCompressedFormat(texD->m_format)) {
-                    const int w = qFloor(float(qMax(1, texD->m_pixelSize.width() >> level)));
-                    const int h = qFloor(float(qMax(1, texD->m_pixelSize.height() >> level)));
+                    int w, h;
+                    if (mipDesc.compressedPixelSize.isEmpty()) {
+                        w = qFloor(float(qMax(1, texD->m_pixelSize.width() >> level)));
+                        h = qFloor(float(qMax(1, texD->m_pixelSize.height() >> level)));
+                    } else {
+                        w = mipDesc.compressedPixelSize.width();
+                        h = mipDesc.compressedPixelSize.height();
+                    }
                     quint32 bpl = 0;
-                    compressedFormatInfo(texD->m_format, QSize(w, h), &bpl, nullptr);
-                    context->UpdateSubresource(texD->tex, subres, nullptr,
+                    QSize blockDim;
+                    compressedFormatInfo(texD->m_format, QSize(w, h), &bpl, nullptr, &blockDim);
+                    // Everything must be a multiple of the block width and
+                    // height, so e.g. a mip level of size 2x2 will be 4x4 when it
+                    // comes to the actual data.
+                    box.left = aligned(x, blockDim.width());
+                    box.top = aligned(y, blockDim.height());
+                    box.right = aligned(x + w, blockDim.width());
+                    box.bottom = aligned(y + h, blockDim.height());
+                    context->UpdateSubresource(texD->tex, subres, &box,
                                                mipDesc.compressedData.constData(), bpl, 0);
                 }
             }

src/rhi/qrhigles2.cpp

@@ -583,8 +583,14 @@ void QRhiGles2::commitResourceUpdates(QRhiResourceUpdateBatch *resourceUpdates)
                 const int x = mipDesc.destinationTopLeft.x();
                 const int y = mipDesc.destinationTopLeft.y();
                 if (isCompressed && !mipDesc.compressedData.isEmpty()) {
-                    const int w = qFloor(float(qMax(1, texD->m_pixelSize.width() >> level)));
-                    const int h = qFloor(float(qMax(1, texD->m_pixelSize.height() >> level)));
+                    int w, h;
+                    if (mipDesc.compressedPixelSize.isEmpty()) {
+                        w = qFloor(float(qMax(1, texD->m_pixelSize.width() >> level)));
+                        h = qFloor(float(qMax(1, texD->m_pixelSize.height() >> level)));
+                    } else {
+                        w = mipDesc.compressedPixelSize.width();
+                        h = mipDesc.compressedPixelSize.height();
+                    }
                     if (texD->specified) {
                         f->glCompressedTexSubImage2D(targetBase + layer, level,
                                                      x, y, w, h,

src/rhi/qrhivulkan.cpp

@@ -2001,12 +2001,14 @@ void QRhiVulkan::commitResourceUpdates(QRhiCommandBuffer *cb, QRhiResourceUpdate
                 copyInfo.imageSubresource.layerCount = 1;
                 copyInfo.imageExtent.depth = 1;
 
+                const int x = mipDesc.destinationTopLeft.x();
+                const int y = mipDesc.destinationTopLeft.y();
                 if (!mipDesc.image.isNull()) {
                     imageSizeBytes = mipDesc.image.sizeInBytes();
                     if (imageSizeBytes > 0) {
                         src = mipDesc.image.constBits();
-                        copyInfo.imageOffset.x = mipDesc.destinationTopLeft.x();
-                        copyInfo.imageOffset.y = mipDesc.destinationTopLeft.y();
+                        copyInfo.imageOffset.x = x;
+                        copyInfo.imageOffset.y = y;
                         copyInfo.imageExtent.width = mipDesc.image.width();
                         copyInfo.imageExtent.height = mipDesc.image.height();
                         copyInfos.append(copyInfo);
@@ -2015,8 +2017,25 @@ void QRhiVulkan::commitResourceUpdates(QRhiCommandBuffer *cb, QRhiResourceUpdate
                     imageSizeBytes = mipDesc.compressedData.size();
                     if (imageSizeBytes > 0) {
                         src = mipDesc.compressedData.constData();
-                        copyInfo.imageExtent.width = qFloor(float(qMax(1, utexD->m_pixelSize.width() >> level)));
-                        copyInfo.imageExtent.height = qFloor(float(qMax(1, utexD->m_pixelSize.height() >> level)));
+                        const int subresw = qFloor(float(qMax(1, utexD->m_pixelSize.width() >> level)));
+                        const int subresh = qFloor(float(qMax(1, utexD->m_pixelSize.height() >> level)));
+                        int w, h;
+                        if (mipDesc.compressedPixelSize.isEmpty()) {
+                            w = subresw;
+                            h = subresh;
+                        } else {
+                            w = mipDesc.compressedPixelSize.width();
+                            h = mipDesc.compressedPixelSize.height();
+                        }
+                        QSize blockDim;
+                        compressedFormatInfo(utexD->m_format, QSize(w, h), nullptr, nullptr, &blockDim);
+                        // x and y must be multiples of the block width and height
+                        copyInfo.imageOffset.x = aligned(x, blockDim.width());
+                        copyInfo.imageOffset.y = aligned(y, blockDim.height());
+                        // width and height must be multiples of the block width and height
+                        // or x + width and y + height must equal the subresource width and height
+                        copyInfo.imageExtent.width = x + w == subresw ? w : aligned(w, blockDim.width());
+                        copyInfo.imageExtent.height = y + h == subresh ? h : aligned(h, blockDim.height());
                         copyInfos.append(copyInfo);
                     }
                 }

todo.txt

@@ -2,7 +2,7 @@ mtl: tex upload with pos
 mtl: texcopy
 mtl: rhi without a window, offscreen frame
 mtl: readback (tex, backbuffer)
-gl, mtl: compressed textures
+mtl: compressed textures
 gl, mtl: srgb (tex, swapchain buf)
 multi window? (multi swapchain) -> trouble
 vk: rendering hangs sometimes when minimize and back on some systems?
@@ -15,6 +15,7 @@ test cubemap face as target
 face cubemap readback? (test vk/d3d, impl for gl)
 cbuffer alignment rules - some things fail to translate (to hlsl e.g. with structs), which is fine but how to mitigate
 resource import/export, what's the co-op story?
+what does image copy do for compressed formats?
 does reading back an msaa swapchain buffer work?
 msaa offscreen (msaa texture. no readback.)
 resolveimage (color and ds, only to resolve samples)
@@ -56,6 +57,7 @@ dxc for d3d as an alternative to fxc?
 hlsl -> dxc -> spirv -> spirv-cross hmmm...
 
 +++ done
+gl: compressed textures
 vk, gl: texcopy
 move cb api into QRhiCommandBuffer
 d3d: texcopy